[clang] 0aef005 - [OpenACC] Implement 'collapse' for combined constructs.
via cfe-commits
cfe-commits at lists.llvm.org
Mon Dec 2 12:49:55 PST 2024
Author: erichkeane
Date: 2024-12-02T12:49:49-08:00
New Revision: 0aef005c94f91a91edc77a4f43337950e66bc462
URL: https://github.com/llvm/llvm-project/commit/0aef005c94f91a91edc77a4f43337950e66bc462
DIFF: https://github.com/llvm/llvm-project/commit/0aef005c94f91a91edc77a4f43337950e66bc462.diff
LOG: [OpenACC] Implement 'collapse' for combined constructs.
Most of the restrictions on 'collapse' involve the contents of the loop,
so this extends enforcement of all of that to all combined construct
loops, and alters the diagnostics to better reflect the construct it is
associated with.
Added:
clang/test/SemaOpenACC/combined-construct-collapse-ast.cpp
clang/test/SemaOpenACC/combined-construct-collapse-clause.cpp
Modified:
clang/include/clang/Basic/DiagnosticSemaKinds.td
clang/include/clang/Sema/SemaOpenACC.h
clang/lib/Sema/SemaOpenACC.cpp
clang/test/AST/ast-print-openacc-combined-construct.cpp
clang/test/SemaOpenACC/combined-construct-auto_seq_independent-clauses.c
clang/test/SemaOpenACC/combined-construct-device_type-clause.c
Removed:
################################################################################
diff --git a/clang/include/clang/Basic/DiagnosticSemaKinds.td b/clang/include/clang/Basic/DiagnosticSemaKinds.td
index 8495884dcd058f..8020be6c57bcfb 100644
--- a/clang/include/clang/Basic/DiagnosticSemaKinds.td
+++ b/clang/include/clang/Basic/DiagnosticSemaKinds.td
@@ -12733,19 +12733,19 @@ def err_acc_size_expr_value
"OpenACC 'tile' clause size expression must be %select{an asterisk "
"or a constant expression|positive integer value, evaluated to %1}0">;
def err_acc_invalid_in_loop
- : Error<"%select{OpenACC '%2' construct|while loop|do loop}0 cannot appear "
- "in intervening code of a 'loop' with a '%1' clause">;
+ : Error<"%select{OpenACC '%3' construct|while loop|do loop}0 cannot appear "
+ "in intervening code of a '%1' with a '%2' clause">;
def note_acc_active_clause_here
: Note<"active '%0' clause defined here">;
def err_acc_clause_multiple_loops
- : Error<"more than one for-loop in a loop associated with OpenACC 'loop' "
- "construct with a '%select{collapse|tile}0' clause">;
+ : Error<"more than one for-loop in a loop associated with OpenACC '%0' "
+ "construct with a '%1' clause">;
def err_acc_insufficient_loops
: Error<"'%0' clause specifies a loop count greater than the number "
"of available loops">;
def err_acc_intervening_code
: Error<"inner loops must be tightly nested inside a '%0' clause on "
- "a 'loop' construct">;
+ "a '%1' construct">;
def err_acc_gang_multiple_elt
: Error<"OpenACC 'gang' clause may have at most one %select{unnamed or "
"'num'|'dim'|'static'}0 argument">;
diff --git a/clang/include/clang/Sema/SemaOpenACC.h b/clang/include/clang/Sema/SemaOpenACC.h
index e8f022362ea168..b968ff779c89b0 100644
--- a/clang/include/clang/Sema/SemaOpenACC.h
+++ b/clang/include/clang/Sema/SemaOpenACC.h
@@ -87,6 +87,10 @@ class SemaOpenACC : public SemaBase {
/// which allows us to diagnose if the value of 'N' is too large for the
/// current number of 'for' loops.
bool CollapseDepthSatisfied = true;
+
+ /// Records the kind of the directive that this clause is attached to, which
+ /// allows us to use it in diagnostics.
+ OpenACCDirectiveKind DirectiveKind = OpenACCDirectiveKind::Invalid;
} CollapseInfo;
/// The 'tile' clause requires a bit of additional checking as well, so like
diff --git a/clang/lib/Sema/SemaOpenACC.cpp b/clang/lib/Sema/SemaOpenACC.cpp
index 8775327a25f66a..7585bfca059567 100644
--- a/clang/lib/Sema/SemaOpenACC.cpp
+++ b/clang/lib/Sema/SemaOpenACC.cpp
@@ -1504,10 +1504,6 @@ OpenACCClause *SemaOpenACCClauseVisitor::VisitReductionClause(
OpenACCClause *SemaOpenACCClauseVisitor::VisitCollapseClause(
SemaOpenACC::OpenACCParsedClause &Clause) {
- // TODO: Remove this check once we implement this for combined constructs.
- if (!isOpenACCComputeDirectiveKind(Clause.getDirectiveKind()) &&
- Clause.getDirectiveKind() != OpenACCDirectiveKind::Loop)
- return isNotImplemented();
// Duplicates here are not really sensible. We could possible permit
// multiples if they all had the same value, but there isn't really a good
// reason to do so. Also, this simplifies the suppression of duplicates, in
@@ -1701,6 +1697,7 @@ void SemaOpenACC::AssociatedStmtRAII::SetCollapseInfoBeforeAssociatedStmt(
SemaRef.CollapseInfo.CollapseDepthSatisfied = false;
SemaRef.CollapseInfo.CurCollapseCount =
cast<ConstantExpr>(LoopCount)->getResultAsAPSInt();
+ SemaRef.CollapseInfo.DirectiveKind = DirKind;
}
void SemaOpenACC::AssociatedStmtRAII::SetTileInfoBeforeAssociatedStmt(
@@ -2597,7 +2594,8 @@ void SemaOpenACC::ActOnWhileStmt(SourceLocation WhileLoc) {
if (CollapseInfo.CurCollapseCount && *CollapseInfo.CurCollapseCount > 0) {
Diag(WhileLoc, diag::err_acc_invalid_in_loop)
- << /*while loop*/ 1 << OpenACCClauseKind::Collapse;
+ << /*while loop*/ 1 << CollapseInfo.DirectiveKind
+ << OpenACCClauseKind::Collapse;
assert(CollapseInfo.ActiveCollapse && "Collapse count without object?");
Diag(CollapseInfo.ActiveCollapse->getBeginLoc(),
diag::note_acc_active_clause_here)
@@ -2610,7 +2608,8 @@ void SemaOpenACC::ActOnWhileStmt(SourceLocation WhileLoc) {
if (TileInfo.CurTileCount && *TileInfo.CurTileCount > 0) {
Diag(WhileLoc, diag::err_acc_invalid_in_loop)
- << /*while loop*/ 1 << OpenACCClauseKind::Tile;
+ << /*while loop*/ 1 << OpenACCDirectiveKind::Loop
+ << OpenACCClauseKind::Tile;
assert(TileInfo.ActiveTile && "tile count without object?");
Diag(TileInfo.ActiveTile->getBeginLoc(), diag::note_acc_active_clause_here)
<< OpenACCClauseKind::Tile;
@@ -2630,7 +2629,8 @@ void SemaOpenACC::ActOnDoStmt(SourceLocation DoLoc) {
if (CollapseInfo.CurCollapseCount && *CollapseInfo.CurCollapseCount > 0) {
Diag(DoLoc, diag::err_acc_invalid_in_loop)
- << /*do loop*/ 2 << OpenACCClauseKind::Collapse;
+ << /*do loop*/ 2 << CollapseInfo.DirectiveKind
+ << OpenACCClauseKind::Collapse;
assert(CollapseInfo.ActiveCollapse && "Collapse count without object?");
Diag(CollapseInfo.ActiveCollapse->getBeginLoc(),
diag::note_acc_active_clause_here)
@@ -2643,7 +2643,8 @@ void SemaOpenACC::ActOnDoStmt(SourceLocation DoLoc) {
if (TileInfo.CurTileCount && *TileInfo.CurTileCount > 0) {
Diag(DoLoc, diag::err_acc_invalid_in_loop)
- << /*do loop*/ 2 << OpenACCClauseKind::Tile;
+ << /*do loop*/ 2 << OpenACCDirectiveKind::Loop
+ << OpenACCClauseKind::Tile;
assert(TileInfo.ActiveTile && "tile count without object?");
Diag(TileInfo.ActiveTile->getBeginLoc(), diag::note_acc_active_clause_here)
<< OpenACCClauseKind::Tile;
@@ -2670,7 +2671,8 @@ void SemaOpenACC::ForStmtBeginHelper(SourceLocation ForLoc,
// This checks for more than 1 loop at the current level, the
// 'depth'-satisifed checking manages the 'not zero' case.
if (LoopInfo.CurLevelHasLoopAlready) {
- Diag(ForLoc, diag::err_acc_clause_multiple_loops) << /*Collapse*/ 0;
+ Diag(ForLoc, diag::err_acc_clause_multiple_loops)
+ << CollapseInfo.DirectiveKind << OpenACCClauseKind::Collapse;
assert(CollapseInfo.ActiveCollapse && "No collapse object?");
Diag(CollapseInfo.ActiveCollapse->getBeginLoc(),
diag::note_acc_active_clause_here)
@@ -2689,7 +2691,8 @@ void SemaOpenACC::ForStmtBeginHelper(SourceLocation ForLoc,
C.check();
if (LoopInfo.CurLevelHasLoopAlready) {
- Diag(ForLoc, diag::err_acc_clause_multiple_loops) << /*Tile*/ 1;
+ Diag(ForLoc, diag::err_acc_clause_multiple_loops)
+ << OpenACCDirectiveKind::Loop << OpenACCClauseKind::Tile;
assert(TileInfo.ActiveTile && "No tile object?");
Diag(TileInfo.ActiveTile->getBeginLoc(),
diag::note_acc_active_clause_here)
@@ -3192,7 +3195,7 @@ void SemaOpenACC::ActOnForStmtEnd(SourceLocation ForLoc, StmtResult Body) {
if (OtherStmtLoc.isValid() && IsActiveCollapse) {
Diag(OtherStmtLoc, diag::err_acc_intervening_code)
- << OpenACCClauseKind::Collapse;
+ << OpenACCClauseKind::Collapse << CollapseInfo.DirectiveKind;
Diag(CollapseInfo.ActiveCollapse->getBeginLoc(),
diag::note_acc_active_clause_here)
<< OpenACCClauseKind::Collapse;
@@ -3200,7 +3203,7 @@ void SemaOpenACC::ActOnForStmtEnd(SourceLocation ForLoc, StmtResult Body) {
if (OtherStmtLoc.isValid() && IsActiveTile) {
Diag(OtherStmtLoc, diag::err_acc_intervening_code)
- << OpenACCClauseKind::Tile;
+ << OpenACCClauseKind::Tile << OpenACCDirectiveKind::Loop;
Diag(TileInfo.ActiveTile->getBeginLoc(),
diag::note_acc_active_clause_here)
<< OpenACCClauseKind::Tile;
@@ -3220,7 +3223,8 @@ bool SemaOpenACC::ActOnStartStmtDirective(OpenACCDirectiveKind K,
// ALL constructs are ill-formed if there is an active 'collapse'
if (CollapseInfo.CurCollapseCount && *CollapseInfo.CurCollapseCount > 0) {
Diag(StartLoc, diag::err_acc_invalid_in_loop)
- << /*OpenACC Construct*/ 0 << OpenACCClauseKind::Collapse << K;
+ << /*OpenACC Construct*/ 0 << CollapseInfo.DirectiveKind
+ << OpenACCClauseKind::Collapse << K;
assert(CollapseInfo.ActiveCollapse && "Collapse count without object?");
Diag(CollapseInfo.ActiveCollapse->getBeginLoc(),
diag::note_acc_active_clause_here)
@@ -3228,7 +3232,8 @@ bool SemaOpenACC::ActOnStartStmtDirective(OpenACCDirectiveKind K,
}
if (TileInfo.CurTileCount && *TileInfo.CurTileCount > 0) {
Diag(StartLoc, diag::err_acc_invalid_in_loop)
- << /*OpenACC Construct*/ 0 << OpenACCClauseKind::Tile << K;
+ << /*OpenACC Construct*/ 0 << OpenACCDirectiveKind::Loop
+ << OpenACCClauseKind::Tile << K;
assert(TileInfo.ActiveTile && "Tile count without object?");
Diag(TileInfo.ActiveTile->getBeginLoc(), diag::note_acc_active_clause_here)
<< OpenACCClauseKind::Tile;
diff --git a/clang/test/AST/ast-print-openacc-combined-construct.cpp b/clang/test/AST/ast-print-openacc-combined-construct.cpp
index 114c0bb1c739af..44ec0bb282f00e 100644
--- a/clang/test/AST/ast-print-openacc-combined-construct.cpp
+++ b/clang/test/AST/ast-print-openacc-combined-construct.cpp
@@ -185,4 +185,29 @@ void foo() {
// CHECK: #pragma acc parallel loop create(i, array[1], array, array[1:2]) pcreate(zero: i, array[1], array, array[1:2]) present_or_create(i, array[1], array, array[1:2])
#pragma acc parallel loop create(i, array[1], array, array[1:2]) pcreate(zero: i, array[1], array, array[1:2]) present_or_create(i, array[1], array, array[1:2])
for(int i = 0;i<5;++i);
+
+// CHECK: #pragma acc parallel loop collapse(1)
+// CHECK-NEXT: for (int i = 0; i < 5; ++i)
+// CHECK-NEXT: ;
+#pragma acc parallel loop collapse(1)
+ for(int i = 0;i<5;++i);
+// CHECK: #pragma acc serial loop collapse(force:1)
+// CHECK-NEXT: for (int i = 0; i < 5; ++i)
+// CHECK-NEXT: ;
+#pragma acc serial loop collapse(force:1)
+ for(int i = 0;i<5;++i);
+// CHECK: #pragma acc kernels loop collapse(2)
+// CHECK-NEXT: for (int i = 0; i < 5; ++i)
+// CHECK-NEXT: for (int i = 0; i < 5; ++i)
+// CHECK-NEXT: ;
+#pragma acc kernels loop collapse(2)
+ for(int i = 0;i<5;++i)
+ for(int i = 0;i<5;++i);
+// CHECK: #pragma acc parallel loop collapse(force:2)
+// CHECK-NEXT: for (int i = 0; i < 5; ++i)
+// CHECK-NEXT: for (int i = 0; i < 5; ++i)
+// CHECK-NEXT: ;
+#pragma acc parallel loop collapse(force:2)
+ for(int i = 0;i<5;++i)
+ for(int i = 0;i<5;++i);
}
diff --git a/clang/test/SemaOpenACC/combined-construct-auto_seq_independent-clauses.c b/clang/test/SemaOpenACC/combined-construct-auto_seq_independent-clauses.c
index c614bbbd1d1769..9a12c1870e7379 100644
--- a/clang/test/SemaOpenACC/combined-construct-auto_seq_independent-clauses.c
+++ b/clang/test/SemaOpenACC/combined-construct-auto_seq_independent-clauses.c
@@ -129,7 +129,6 @@ void uses() {
// expected-warning at +1{{OpenACC clause 'reduction' not yet implemented}}
#pragma acc parallel loop auto reduction(+:Var)
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'collapse' not yet implemented}}
#pragma acc parallel loop auto collapse(1)
for(unsigned i = 0; i < 5; ++i);
// expected-warning at +1{{OpenACC clause 'bind' not yet implemented}}
@@ -258,7 +257,6 @@ void uses() {
// expected-warning at +1{{OpenACC clause 'reduction' not yet implemented}}
#pragma acc parallel loop reduction(+:Var) auto
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'collapse' not yet implemented}}
#pragma acc parallel loop collapse(1) auto
for(unsigned i = 0; i < 5; ++i);
// expected-warning at +1{{OpenACC clause 'bind' not yet implemented}}
@@ -388,7 +386,6 @@ void uses() {
// expected-warning at +1{{OpenACC clause 'reduction' not yet implemented}}
#pragma acc parallel loop independent reduction(+:Var)
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'collapse' not yet implemented}}
#pragma acc parallel loop independent collapse(1)
for(unsigned i = 0; i < 5; ++i);
// expected-warning at +1{{OpenACC clause 'bind' not yet implemented}}
@@ -517,7 +514,6 @@ void uses() {
// expected-warning at +1{{OpenACC clause 'reduction' not yet implemented}}
#pragma acc parallel loop reduction(+:Var) independent
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'collapse' not yet implemented}}
#pragma acc parallel loop collapse(1) independent
for(unsigned i = 0; i < 5; ++i);
// expected-warning at +1{{OpenACC clause 'bind' not yet implemented}}
@@ -653,7 +649,6 @@ void uses() {
// expected-warning at +1{{OpenACC clause 'reduction' not yet implemented}}
#pragma acc parallel loop seq reduction(+:Var)
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'collapse' not yet implemented}}
#pragma acc parallel loop seq collapse(1)
for(unsigned i = 0; i < 5; ++i);
// expected-warning at +1{{OpenACC clause 'bind' not yet implemented}}
@@ -788,7 +783,6 @@ void uses() {
// expected-warning at +1{{OpenACC clause 'reduction' not yet implemented}}
#pragma acc parallel loop reduction(+:Var) seq
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'collapse' not yet implemented}}
#pragma acc parallel loop collapse(1) seq
for(unsigned i = 0; i < 5; ++i);
// expected-warning at +1{{OpenACC clause 'bind' not yet implemented}}
diff --git a/clang/test/SemaOpenACC/combined-construct-collapse-ast.cpp b/clang/test/SemaOpenACC/combined-construct-collapse-ast.cpp
new file mode 100644
index 00000000000000..89ba7cb24814f6
--- /dev/null
+++ b/clang/test/SemaOpenACC/combined-construct-collapse-ast.cpp
@@ -0,0 +1,219 @@
+
+// RUN: %clang_cc1 %s -fopenacc -ast-dump | FileCheck %s
+
+// Test this with PCH.
+// RUN: %clang_cc1 %s -fopenacc -emit-pch -o %t %s
+// RUN: %clang_cc1 %s -fopenacc -include-pch %t -ast-dump-all | FileCheck %s
+#ifndef PCH_HELPER
+#define PCH_HELPER
+
+struct S {
+ constexpr S(){};
+ constexpr operator auto() {return 1;}
+};
+
+void NormalUses() {
+ // CHECK: FunctionDecl{{.*}}NormalUses
+ // CHECK-NEXT: CompoundStmt
+
+#pragma acc parallel loop collapse(1)
+ for(int i = 0; i < 5; ++i);
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} parallel loop
+ // CHECK-NEXT: collapse clause
+ // CHECK-NEXT: ConstantExpr{{.*}}'int'
+ // CHECK-NEXT: value: Int 1
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 1
+ // CHECK-NEXT: ForStmt
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} i 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: BinaryOperator{{.*}}'<'
+ // CHECK-NEXT: ImplicitCastExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 5
+ // CHECK-NEXT: UnaryOperator{{.*}}++
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: NullStmt
+
+#pragma acc serial loop collapse(force:S{})
+ for(int i = 0; i < 5; ++i);
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} serial loop
+ // CHECK-NEXT: collapse clause
+ // CHECK-NEXT: ConstantExpr{{.*}}'int'
+ // CHECK-NEXT: value: Int 1
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <UserDefinedConversion>
+ // CHECK-NEXT: CXXMemberCallExpr{{.*}} 'int'
+ // CHECK-NEXT: MemberExpr{{.*}} .operator auto
+ // CHECK-NEXT: MaterializeTemporaryExpr{{.*}}'S' lvalue
+ // CHECK-NEXT: CXXTemporaryObjectExpr{{.*}}'S' 'void ()' list
+ // CHECK-NEXT: ForStmt
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} i 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: BinaryOperator{{.*}}'<'
+ // CHECK-NEXT: ImplicitCastExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 5
+ // CHECK-NEXT: UnaryOperator{{.*}}++
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: NullStmt
+}
+
+template<typename T, unsigned Value>
+void TemplUses() {
+ // CHECK: FunctionTemplateDecl{{.*}}TemplUses
+ // CHECK-NEXT: TemplateTypeParmDecl {{.*}} referenced typename depth 0 index 0 T
+ // CHECK-NEXT: NonTypeTemplateParmDecl{{.*}} referenced 'unsigned int' depth 0 index 1 Value
+ // CHECK-NEXT: FunctionDecl{{.*}} TemplUses 'void ()'
+ // CHECK-NEXT: CompoundStmt
+
+#pragma acc parallel loop collapse(Value)
+ for(int i = 0; i < 5; ++i)
+ for(int j = 0; j < 5; ++j);
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} parallel loop
+ // CHECK-NEXT: collapse clause
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'unsigned int' NonTypeTemplateParm{{.*}} 'Value'
+ // CHECK-NEXT: ForStmt
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} i 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: BinaryOperator{{.*}}'<'
+ // CHECK-NEXT: ImplicitCastExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 5
+ // CHECK-NEXT: UnaryOperator{{.*}}++
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: ForStmt
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} j 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: BinaryOperator{{.*}}'<'
+ // CHECK-NEXT: ImplicitCastExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}}'j' 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 5
+ // CHECK-NEXT: UnaryOperator{{.*}}++
+ // CHECK-NEXT: DeclRefExpr{{.*}}'j' 'int'
+ // CHECK-NEXT: NullStmt
+
+#pragma acc kernels loop collapse(force:T{} + S{})
+ for(int i = 0; i < 5; ++i)
+ for(int j = 0; j < 5; ++j);
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} kernels loop
+ // CHECK-NEXT: collapse clause
+ // CHECK-NEXT: BinaryOperator {{.*}}'+'
+ // CHECK-NEXT: CXXUnresolvedConstructExpr{{.*}}'T' 'T' list
+ // CHECK-NEXT: InitListExpr
+ // CHECK-NEXT: CXXTemporaryObjectExpr{{.*}}'S' 'void ()' list
+ // CHECK-NEXT: ForStmt
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} i 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: BinaryOperator{{.*}}'<'
+ // CHECK-NEXT: ImplicitCastExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 5
+ // CHECK-NEXT: UnaryOperator{{.*}}++
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: ForStmt
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} j 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: BinaryOperator{{.*}}'<'
+ // CHECK-NEXT: ImplicitCastExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}}'j' 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 5
+ // CHECK-NEXT: UnaryOperator{{.*}}++
+ // CHECK-NEXT: DeclRefExpr{{.*}}'j' 'int'
+ // CHECK-NEXT: NullStmt
+
+ // Instantiation:
+ // CHECK-NEXT: FunctionDecl{{.*}} used TemplUses 'void ()' implicit_instantiation
+ // CHECK-NEXT: TemplateArgument type 'S'
+ // CHECK-NEXT: RecordType{{.*}} 'S'
+ // CHECK-NEXT: CXXRecord{{.*}} 'S'
+ // CHECK-NEXT: TemplateArgument integral '2U'
+ // CHECK-NEXT: CompoundStmt
+
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} parallel loop
+ // CHECK-NEXT: collapse clause
+ // CHECK-NEXT: ConstantExpr{{.*}}'unsigned int'
+ // CHECK-NEXT: value: Int 2
+ // CHECK-NEXT: SubstNonTypeTemplateParmExpr{{.*}} 'unsigned int'
+ // CHECK-NEXT: NonTypeTemplateParmDecl
+ // CHECK-NEXT: IntegerLiteral {{.*}} 'unsigned int' 2
+ // CHECK-NEXT: ForStmt
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} i 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: BinaryOperator{{.*}}'<'
+ // CHECK-NEXT: ImplicitCastExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 5
+ // CHECK-NEXT: UnaryOperator{{.*}}++
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: ForStmt
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} j 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: BinaryOperator{{.*}}'<'
+ // CHECK-NEXT: ImplicitCastExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}}'j' 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 5
+ // CHECK-NEXT: UnaryOperator{{.*}}++
+ // CHECK-NEXT: DeclRefExpr{{.*}}'j' 'int'
+ // CHECK-NEXT: NullStmt
+
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} kernels loop
+ // CHECK-NEXT: collapse clause
+ // CHECK-NEXT: ConstantExpr{{.*}}'int'
+ // CHECK-NEXT: value: Int 2
+ // CHECK-NEXT: BinaryOperator {{.*}}'+'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <UserDefinedConversion>
+ // CHECK-NEXT: CXXMemberCallExpr{{.*}} 'int'
+ // CHECK-NEXT: MemberExpr{{.*}} .operator auto
+ // CHECK-NEXT: MaterializeTemporaryExpr{{.*}}'S' lvalue
+ // CHECK-NEXT: CXXTemporaryObjectExpr{{.*}}'S' 'void ()' list
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <UserDefinedConversion>
+ // CHECK-NEXT: CXXMemberCallExpr{{.*}} 'int'
+ // CHECK-NEXT: MemberExpr{{.*}} .operator auto
+ // CHECK-NEXT: MaterializeTemporaryExpr{{.*}}'S' lvalue
+ // CHECK-NEXT: CXXTemporaryObjectExpr{{.*}}'S' 'void ()' list
+ // CHECK-NEXT: ForStmt
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} i 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: BinaryOperator{{.*}}'<'
+ // CHECK-NEXT: ImplicitCastExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 5
+ // CHECK-NEXT: UnaryOperator{{.*}}++
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: ForStmt
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} j 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: BinaryOperator{{.*}}'<'
+ // CHECK-NEXT: ImplicitCastExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}}'j' 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 5
+ // CHECK-NEXT: UnaryOperator{{.*}}++
+ // CHECK-NEXT: DeclRefExpr{{.*}}'j' 'int'
+ // CHECK-NEXT: NullStmt
+
+}
+
+void Inst() {
+ TemplUses<S, 2>();
+}
+
+#endif // PCH_HELPER
diff --git a/clang/test/SemaOpenACC/combined-construct-collapse-clause.cpp b/clang/test/SemaOpenACC/combined-construct-collapse-clause.cpp
new file mode 100644
index 00000000000000..c7db9669a9879b
--- /dev/null
+++ b/clang/test/SemaOpenACC/combined-construct-collapse-clause.cpp
@@ -0,0 +1,390 @@
+// RUN: %clang_cc1 %s -fopenacc -verify
+
+template<unsigned Val>
+void depth_too_high_templ() {
+ // expected-error at +2{{'collapse' clause specifies a loop count greater than the number of available loops}}
+ // expected-note at +1{{active 'collapse' clause defined here}}
+#pragma acc parallel loop collapse(Val)
+ for(unsigned i = 0; i < 5; ++i)
+ for(unsigned j = 0; j < 5; ++j);
+}
+constexpr int three() { return 3; }
+struct ConvertsThree{
+ constexpr ConvertsThree(){};
+
+ constexpr operator int(){ return 3; }
+};
+
+void depth_too_high() {
+ depth_too_high_templ<3>(); // expected-note{{in instantiation of function template specialization}}
+
+ // expected-error at +2{{'collapse' clause specifies a loop count greater than the number of available loops}}
+ // expected-note at +1{{active 'collapse' clause defined here}}
+#pragma acc serial loop collapse(3)
+ for(unsigned i = 0; i < 5; ++i)
+ for(unsigned j = 0; j < 5; ++j);
+
+ // expected-error at +2{{'collapse' clause specifies a loop count greater than the number of available loops}}
+ // expected-note at +1{{active 'collapse' clause defined here}}
+#pragma acc kernels loop collapse(three())
+ for(unsigned i = 0; i < 5; ++i)
+ for(unsigned j = 0; j < 5; ++j);
+
+ // expected-error at +2{{'collapse' clause specifies a loop count greater than the number of available loops}}
+ // expected-note at +1{{active 'collapse' clause defined here}}
+#pragma acc parallel loop collapse(ConvertsThree{})
+ for(unsigned i = 0; i < 5; ++i)
+ for(unsigned j = 0; j < 5; ++j);
+}
+
+template<typename T, unsigned Three>
+void not_single_loop_templ() {
+ T Arr[5];
+ // expected-error at +2{{'collapse' clause specifies a loop count greater than the number of available loops}}
+ // expected-note at +1 2{{active 'collapse' clause defined here}}
+#pragma acc parallel loop collapse(3)
+ for(auto x : Arr) {
+ for(auto y : Arr){
+ do{}while(true); // expected-error{{do loop cannot appear in intervening code of a 'parallel loop' with a 'collapse' clause}}
+ }
+ }
+
+ // expected-error at +2{{'collapse' clause specifies a loop count greater than the number of available loops}}
+ // expected-note at +1 2{{active 'collapse' clause defined here}}
+#pragma acc serial loop collapse(Three)
+ for(unsigned i = 0; i < 5; ++i) {
+ for(unsigned j = 0; j < 5; ++j) {
+ do{}while(true); // expected-error{{do loop cannot appear in intervening code of a 'serial loop' with a 'collapse' clause}}
+ }
+ }
+
+#pragma acc kernels loop collapse(Three)
+ for(unsigned i = 0; i < 5; ++i) {
+ for(unsigned j = 0; j < 5; ++j) {
+ for(unsigned k = 0; k < 5;++k) {
+ do{}while(true);
+ }
+ }
+ }
+ // expected-error at +2{{'collapse' clause specifies a loop count greater than the number of available loops}}
+ // expected-note at +1 2{{active 'collapse' clause defined here}}
+#pragma acc parallel loop collapse(Three)
+ for(auto x : Arr) {
+ for(auto y: Arr) {
+ do{}while(true); // expected-error{{do loop cannot appear in intervening code of a 'parallel loop' with a 'collapse' clause}}
+ }
+ }
+
+#pragma acc serial loop collapse(Three)
+ for(auto x : Arr) {
+ for(auto y: Arr) {
+ for(auto z: Arr) {
+ do{}while(true);
+ }
+ }
+ }
+}
+
+void not_single_loop() {
+ not_single_loop_templ<int, 3>(); // expected-note{{in instantiation of function template}}
+
+ // expected-note at +1{{active 'collapse' clause defined here}}
+#pragma acc kernels loop collapse(3)
+ for(unsigned i = 0; i < 5; ++i) {
+ for(unsigned j = 0; j < 5; ++j) {
+ for(unsigned k = 0; k < 5;++k);
+ }
+ while(true); // expected-error{{while loop cannot appear in intervening code of a 'kernels loop' with a 'collapse' clause}}
+ }
+
+ // expected-note at +1{{active 'collapse' clause defined here}}
+#pragma acc parallel loop collapse(3)
+ for(unsigned i = 0; i < 5; ++i) {
+ for(unsigned j = 0; j < 5; ++j) {
+ for(unsigned k = 0; k < 5;++k);
+ }
+ do{}while(true); // expected-error{{do loop cannot appear in intervening code of a 'parallel loop' with a 'collapse' clause}}
+ }
+
+ // expected-error at +2{{'collapse' clause specifies a loop count greater than the number of available loops}}
+ // expected-note at +1 2{{active 'collapse' clause defined here}}
+#pragma acc serial loop collapse(3)
+ for(unsigned i = 0; i < 5; ++i) {
+ for(unsigned j = 0; j < 5; ++j) {
+ while(true); // expected-error{{while loop cannot appear in intervening code of a 'serial loop' with a 'collapse' clause}}
+ }
+ }
+ // expected-error at +2{{'collapse' clause specifies a loop count greater than the number of available loops}}
+ // expected-note at +1 2{{active 'collapse' clause defined here}}
+#pragma acc kernels loop collapse(3)
+ for(unsigned i = 0; i < 5; ++i) {
+ for(unsigned j = 0; j < 5; ++j) {
+ do{}while(true); // expected-error{{do loop cannot appear in intervening code of a 'kernels loop' with a 'collapse' clause}}
+ }
+ }
+
+#pragma acc parallel loop collapse(2)
+ for(unsigned i = 0; i < 5; ++i) {
+ for(unsigned j = 0; j < 5; ++j) {
+ do{}while(true);
+ }
+ }
+#pragma acc serial loop collapse(2)
+ for(unsigned i = 0; i < 5; ++i) {
+ for(unsigned j = 0; j < 5; ++j) {
+ while(true);
+ }
+ }
+
+ int Arr[5];
+ // expected-error at +2{{'collapse' clause specifies a loop count greater than the number of available loops}}
+ // expected-note at +1 2{{active 'collapse' clause defined here}}
+#pragma acc kernels loop collapse(3)
+ for(auto x : Arr) {
+ for(auto y : Arr){
+ do{}while(true); // expected-error{{do loop cannot appear in intervening code of a 'kernels loop' with a 'collapse' clause}}
+ }
+ }
+
+ // expected-note at +1 {{active 'collapse' clause defined here}}
+#pragma acc parallel loop collapse(3)
+ for(unsigned i = 0; i < 5; ++i) {
+ for(unsigned j = 0; j < 5; ++j) {
+ for(unsigned k = 0; k < 5;++k);
+ }
+ // expected-error at +1{{more than one for-loop in a loop associated with OpenACC 'parallel loop' construct with a 'collapse' clause}}
+ for(unsigned k = 0; k < 5;++k);
+ }
+
+ // expected-note at +1 {{active 'collapse' clause defined here}}
+#pragma acc serial loop collapse(3)
+ for(unsigned i = 0; i < 5; ++i) {
+ for(unsigned j = 0; j < 5; ++j) {
+ for(unsigned k = 0; k < 5;++k);
+ // expected-error at +1{{more than one for-loop in a loop associated with OpenACC 'serial loop' construct with a 'collapse' clause}}
+ for(unsigned k = 0; k < 5;++k);
+ }
+ }
+
+ for(unsigned k = 0; k < 5;++k);
+#pragma acc kernels loop collapse(3)
+ for(unsigned i = 0; i < 5; ++i) {
+ for(unsigned j = 0; j < 5; ++j) {
+ for(unsigned k = 0; k < 5;++k);
+ }
+ }
+}
+
+template<unsigned Two, unsigned Three>
+void no_other_directives() {
+#pragma acc parallel loop collapse(Two)
+ for(unsigned i = 0; i < 5; ++i) {
+ for(unsigned j = 0; j < 5; ++j) {// last loop associated with the top level.
+ // expected-error at +1{{'collapse' clause specifies a loop count greater than the number of available loops}}
+#pragma acc serial loop collapse(Three) // expected-note 2{{active 'collapse' clause defined here}}
+ for(unsigned k = 0; k < 6;++k) {
+ for(unsigned l = 0; l < 5; ++l) {
+ // expected-error at +1{{OpenACC 'serial' construct cannot appear in intervening code of a 'serial loop' with a 'collapse' clause}}
+#pragma acc serial
+ ;
+ }
+ }
+ }
+ }
+#pragma acc kernels loop collapse(Two)// expected-note{{active 'collapse' clause defined here}}
+ for(unsigned i = 0; i < 5; ++i) {
+ for(unsigned j = 0; j < 5; ++j) {// last loop associated with the top level.
+#pragma acc parallel loop collapse(Three)
+ for(unsigned k = 0; k < 6;++k) {
+ for(unsigned l = 0; l < 5; ++l) {
+ for(unsigned m = 0; m < 5; ++m);
+ }
+ }
+ }
+ // expected-error at +1{{OpenACC 'serial' construct cannot appear in intervening code of a 'kernels loop' with a 'collapse' clause}}
+#pragma acc serial
+ ;
+ }
+}
+
+void no_other_directives() {
+ no_other_directives<2,3>(); // expected-note{{in instantiation of function template specialization}}
+
+ // Ok, not inside the intervening list
+#pragma acc serial loop collapse(2)
+ for(unsigned i = 0; i < 5; ++i) {
+ for(unsigned j = 0; j < 5; ++j) {
+#pragma acc data // expected-warning{{OpenACC construct 'data' not yet implemented}}
+ }
+ }
+ // expected-note at +1{{active 'collapse' clause defined here}}
+#pragma acc kernels loop collapse(2)
+ for(unsigned i = 0; i < 5; ++i) {
+ // expected-error at +1{{OpenACC 'data' construct cannot appear in intervening code of a 'kernels loop' with a 'collapse' clause}}
+#pragma acc data // expected-warning{{OpenACC construct 'data' not yet implemented}}
+ for(unsigned j = 0; j < 5; ++j) {
+ }
+ }
+}
+
+void call();
+
+template<unsigned Two>
+void intervening_without_force_templ() {
+ // expected-note at +1{{active 'collapse' clause defined here}}
+#pragma acc parallel loop collapse(2)
+ for(unsigned i = 0; i < 5; ++i) {
+ // expected-error at +1{{inner loops must be tightly nested inside a 'collapse' clause on a 'parallel loop' construct}}
+ call();
+ for(unsigned j = 0; j < 5; ++j);
+ }
+
+ // expected-note at +1{{active 'collapse' clause defined here}}
+#pragma acc serial loop collapse(Two)
+ for(unsigned i = 0; i < 5; ++i) {
+ // expected-error at +1{{inner loops must be tightly nested inside a 'collapse' clause on a 'serial loop' construct}}
+ call();
+ for(unsigned j = 0; j < 5; ++j);
+ }
+
+ // expected-note at +1{{active 'collapse' clause defined here}}
+#pragma acc kernels loop collapse(2)
+ for(unsigned i = 0; i < 5; ++i) {
+ for(unsigned j = 0; j < 5; ++j);
+ // expected-error at +1{{inner loops must be tightly nested inside a 'collapse' clause on a 'kernels loop' construct}}
+ call();
+ }
+
+#pragma acc parallel loop collapse(force:2)
+ for(unsigned i = 0; i < 5; ++i) {
+ call();
+ for(unsigned j = 0; j < 5; ++j);
+ }
+
+#pragma acc parallel loop collapse(force:Two)
+ for(unsigned i = 0; i < 5; ++i) {
+ call();
+ for(unsigned j = 0; j < 5; ++j);
+ }
+
+
+#pragma acc parallel loop collapse(force:2)
+ for(unsigned i = 0; i < 5; ++i) {
+ for(unsigned j = 0; j < 5; ++j);
+ call();
+ }
+
+#pragma acc parallel loop collapse(force:Two)
+ for(unsigned i = 0; i < 5; ++i) {
+ for(unsigned j = 0; j < 5; ++j);
+ call();
+ }
+
+#pragma acc parallel loop collapse(Two)
+ for(unsigned i = 0; i < 5; ++i) {
+ for(unsigned j = 0; j < 5; ++j) {
+ call();
+ }
+ }
+
+#pragma acc parallel loop collapse(Two)
+ for(unsigned i = 0; i < 5; ++i) {
+ {
+ {
+ for(unsigned j = 0; j < 5; ++j) {
+ call();
+ }
+ }
+ }
+ }
+
+#pragma acc parallel loop collapse(force:Two)
+ for(unsigned i = 0; i < 5; ++i) {
+ for(unsigned j = 0; j < 5; ++j) {
+ call();
+ }
+ }
+
+ // expected-note at +1{{active 'collapse' clause defined here}}
+#pragma acc parallel loop collapse(Two)
+ for(unsigned i = 0; i < 5; ++i) {
+ for(unsigned j = 0; j < 5; ++j);
+ // expected-error at +1{{inner loops must be tightly nested inside a 'collapse' clause on a 'parallel loop' construct}}
+ call();
+ }
+
+#pragma acc parallel loop collapse(2)
+ // expected-error at +2{{OpenACC 'parallel loop' construct must have a terminating condition}}
+ // expected-note at -2{{'parallel loop' construct is here}}
+ for(int i = 0;;++i)
+ // expected-error at +2{{OpenACC 'parallel loop' construct must have a terminating condition}}
+ // expected-note at -5{{'parallel loop' construct is here}}
+ for(int j = 0;;++j)
+ for(;;);
+}
+
+void intervening_without_force() {
+ intervening_without_force_templ<2>(); // expected-note{{in instantiation of function template specialization}}
+ // expected-note at +1{{active 'collapse' clause defined here}}
+#pragma acc parallel loop collapse(2)
+ for(unsigned i = 0; i < 5; ++i) {
+ // expected-error at +1{{inner loops must be tightly nested inside a 'collapse' clause on a 'parallel loop' construct}}
+ call();
+ for(unsigned j = 0; j < 5; ++j);
+ }
+
+ // expected-note at +1{{active 'collapse' clause defined here}}
+#pragma acc parallel loop collapse(2)
+ for(unsigned i = 0; i < 5; ++i) {
+ for(unsigned j = 0; j < 5; ++j);
+ // expected-error at +1{{inner loops must be tightly nested inside a 'collapse' clause on a 'parallel loop' construct}}
+ call();
+ }
+
+ // The below two are fine, as they use the 'force' tag.
+#pragma acc parallel loop collapse(force:2)
+ for(unsigned i = 0; i < 5; ++i) {
+ call();
+ for(unsigned j = 0; j < 5; ++j);
+ }
+
+#pragma acc parallel loop collapse(force:2)
+ for(unsigned i = 0; i < 5; ++i) {
+ for(unsigned j = 0; j < 5; ++j);
+ call();
+ }
+
+#pragma acc parallel loop collapse(2)
+ for(unsigned i = 0; i < 5; ++i) {
+ for(unsigned j = 0; j < 5; ++j) {
+ call();
+ }
+ }
+#pragma acc parallel loop collapse(2)
+ for(unsigned i = 0; i < 5; ++i) {
+ {
+ {
+ for(unsigned j = 0; j < 5; ++j) {
+ call();
+ }
+ }
+ }
+ }
+
+#pragma acc parallel loop collapse(force:2)
+ for(unsigned i = 0; i < 5; ++i) {
+ for(unsigned j = 0; j < 5; ++j) {
+ call();
+ }
+ }
+
+#pragma acc parallel loop collapse(2)
+ // expected-error at +2{{OpenACC 'parallel loop' construct must have a terminating condition}}
+ // expected-note at -2{{'parallel loop' construct is here}}
+ for(int i = 0;;++i)
+ // expected-error at +2{{OpenACC 'parallel loop' construct must have a terminating condition}}
+ // expected-note at -5{{'parallel loop' construct is here}}
+ for(int j = 0;;++j)
+ for(;;);
+}
+
diff --git a/clang/test/SemaOpenACC/combined-construct-device_type-clause.c b/clang/test/SemaOpenACC/combined-construct-device_type-clause.c
index 58b3cdc2820bcf..8933112a0063de 100644
--- a/clang/test/SemaOpenACC/combined-construct-device_type-clause.c
+++ b/clang/test/SemaOpenACC/combined-construct-device_type-clause.c
@@ -183,7 +183,6 @@ void uses() {
// expected-note at +1{{previous clause is here}}
#pragma acc serial loop device_type(*) reduction(+:Var)
for(int i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'collapse' not yet implemented, clause ignored}}
#pragma acc serial loop device_type(*) collapse(1)
for(int i = 0; i < 5; ++i);
// expected-error at +2{{OpenACC clause 'bind' may not follow a 'device_type' clause in a 'parallel loop' construct}}
More information about the cfe-commits
mailing list