[clang] 5e92bfe - [OpenACC] Check Loop counts for 'collapse' clause. (#110851)
via cfe-commits
cfe-commits at lists.llvm.org
Wed Oct 2 10:33:52 PDT 2024
Author: Erich Keane
Date: 2024-10-02T10:33:49-07:00
New Revision: 5e92bfe97fe0f72f3052df53f813d8dcbb7038d3
URL: https://github.com/llvm/llvm-project/commit/5e92bfe97fe0f72f3052df53f813d8dcbb7038d3
DIFF: https://github.com/llvm/llvm-project/commit/5e92bfe97fe0f72f3052df53f813d8dcbb7038d3.diff
LOG: [OpenACC] Check Loop counts for 'collapse' clause. (#110851)
OpenACC Spec requires that each loop associated with a 'collapse' have
exactly 1 loop/nest. This is implemented in 2 parts: 1- diagnosing when
we see a 2nd loop at any level with an applicable 'collapse'
2- Diagnosing if we didn't see enough 'depth' of loops.
Other loops (non-for) are diagnosed by a previous patch, and other
intervening code will be diagnosed in a followup.
Added:
Modified:
clang/include/clang/Basic/DiagnosticSemaKinds.td
clang/include/clang/Sema/SemaOpenACC.h
clang/lib/Sema/SemaOpenACC.cpp
clang/test/SemaOpenACC/loop-construct-collapse-clause.cpp
Removed:
################################################################################
diff --git a/clang/include/clang/Basic/DiagnosticSemaKinds.td b/clang/include/clang/Basic/DiagnosticSemaKinds.td
index 880dda75ffb110..dae357eb2d370e 100644
--- a/clang/include/clang/Basic/DiagnosticSemaKinds.td
+++ b/clang/include/clang/Basic/DiagnosticSemaKinds.td
@@ -12622,6 +12622,12 @@ def err_acc_invalid_in_collapse_loop
"in intervening code of a 'loop' with a 'collapse' clause">;
def note_acc_collapse_clause_here
: Note<"active 'collapse' clause defined here">;
+def err_acc_collapse_multiple_loops
+ : Error<"more than one for-loop in a loop associated with OpenACC 'loop' "
+ "construct with a 'collapse' clause">;
+def err_acc_collapse_insufficient_loops
+ : Error<"'collapse' clause specifies a loop count greater than the number "
+ "of available loops">;
// AMDGCN builtins diagnostics
def err_amdgcn_global_load_lds_size_invalid_value : Error<"invalid size value">;
diff --git a/clang/include/clang/Sema/SemaOpenACC.h b/clang/include/clang/Sema/SemaOpenACC.h
index 43b1b1b8e870aa..26564835fa1af6 100644
--- a/clang/include/clang/Sema/SemaOpenACC.h
+++ b/clang/include/clang/Sema/SemaOpenACC.h
@@ -64,7 +64,20 @@ class SemaOpenACC : public SemaBase {
/// 'collapse inner loop not a 'for' loop' diagnostic.
LLVM_PREFERRED_TYPE(bool)
unsigned TopLevelLoopSeen : 1;
- } CollapseInfo{nullptr, std::nullopt, false};
+
+ /// Records whether this 'tier' of the loop has already seen a 'for' loop,
+ /// used to diagnose if there are multiple 'for' loops at any one level.
+ LLVM_PREFERRED_TYPE(bool)
+ unsigned CurLevelHasLoopAlready : 1;
+
+ /// Records whether we've hit a CurCollapseCount of '0' on the way down,
+ /// which allows us to diagnose if the value of 'N' is too large for the
+ /// current number of 'for' loops.
+ LLVM_PREFERRED_TYPE(bool)
+ unsigned CollapseDepthSatisfied : 1;
+ } CollapseInfo{nullptr, std::nullopt, /*TopLevelLoopSeen=*/false,
+ /*CurLevelHasLoopAlready=*/false,
+ /*CollapseDepthSatisfied=*/true};
public:
// Redeclaration of the version in OpenACCClause.h.
@@ -515,11 +528,26 @@ class SemaOpenACC : public SemaBase {
class LoopInConstructRAII {
SemaOpenACC &SemaRef;
CollapseCheckingInfo OldCollapseInfo;
+ bool PreserveDepth;
public:
- LoopInConstructRAII(SemaOpenACC &SemaRef)
- : SemaRef(SemaRef), OldCollapseInfo(SemaRef.CollapseInfo) {}
- ~LoopInConstructRAII() { SemaRef.CollapseInfo = OldCollapseInfo; }
+ LoopInConstructRAII(SemaOpenACC &SemaRef, bool PreserveDepth = true)
+ : SemaRef(SemaRef), OldCollapseInfo(SemaRef.CollapseInfo),
+ PreserveDepth(PreserveDepth) {}
+ ~LoopInConstructRAII() {
+ // The associated-statement level of this should NOT preserve this, as it
+ // is a new construct, but other loop uses need to preserve the depth so
+ // it makes it to the 'top level' for diagnostics.
+ bool CollapseDepthSatisified =
+ PreserveDepth ? SemaRef.CollapseInfo.CollapseDepthSatisfied
+ : OldCollapseInfo.CollapseDepthSatisfied;
+ bool CurLevelHasLoopAlready =
+ PreserveDepth ? SemaRef.CollapseInfo.CurLevelHasLoopAlready
+ : OldCollapseInfo.CurLevelHasLoopAlready;
+ SemaRef.CollapseInfo = OldCollapseInfo;
+ SemaRef.CollapseInfo.CollapseDepthSatisfied = CollapseDepthSatisified;
+ SemaRef.CollapseInfo.CurLevelHasLoopAlready = CurLevelHasLoopAlready;
+ }
};
/// Helper type for the registration/assignment of constructs that need to
diff --git a/clang/lib/Sema/SemaOpenACC.cpp b/clang/lib/Sema/SemaOpenACC.cpp
index 58ed3f3cb0777f..47b4bd77d86d18 100644
--- a/clang/lib/Sema/SemaOpenACC.cpp
+++ b/clang/lib/Sema/SemaOpenACC.cpp
@@ -1078,7 +1078,7 @@ SemaOpenACC::AssociatedStmtRAII::AssociatedStmtRAII(
ArrayRef<const OpenACCClause *> UnInstClauses,
ArrayRef<OpenACCClause *> Clauses)
: SemaRef(S), WasInsideComputeConstruct(S.InsideComputeConstruct),
- DirKind(DK), LoopRAII(SemaRef) {
+ DirKind(DK), LoopRAII(SemaRef, /*PreserveDepth=*/false) {
// Compute constructs end up taking their 'loop'.
if (DirKind == OpenACCDirectiveKind::Parallel ||
DirKind == OpenACCDirectiveKind::Serial ||
@@ -1093,6 +1093,11 @@ SemaOpenACC::AssociatedStmtRAII::AssociatedStmtRAII(
void SemaOpenACC::AssociatedStmtRAII::SetCollapseInfoBeforeAssociatedStmt(
ArrayRef<const OpenACCClause *> UnInstClauses,
ArrayRef<OpenACCClause *> Clauses) {
+
+ // Reset this checking for loops that aren't covered in a RAII object.
+ SemaRef.CollapseInfo.CurLevelHasLoopAlready = false;
+ SemaRef.CollapseInfo.CollapseDepthSatisfied = true;
+
// We make sure to take an optional list of uninstantiated clauses, so that
// we can check to make sure we don't 'double diagnose' in the event that
// the value of 'N' was not dependent in a template. We also ensure during
@@ -1125,6 +1130,7 @@ void SemaOpenACC::AssociatedStmtRAII::SetCollapseInfoBeforeAssociatedStmt(
->isInstantiationDependent())
return;
+ SemaRef.CollapseInfo.CollapseDepthSatisfied = false;
SemaRef.CollapseInfo.CurCollapseCount =
cast<ConstantExpr>(LoopCount)->getResultAsAPSInt();
}
@@ -1737,13 +1743,38 @@ void SemaOpenACC::ActOnForStmtBegin(SourceLocation ForLoc) {
// Enable the while/do-while checking.
CollapseInfo.TopLevelLoopSeen = true;
- if (CollapseInfo.CurCollapseCount && *CollapseInfo.CurCollapseCount > 0)
- --(*CollapseInfo.CurCollapseCount);
+ if (CollapseInfo.CurCollapseCount && *CollapseInfo.CurCollapseCount > 0) {
+
+ // OpenACC 3.3 2.9.1:
+ // Each associated loop, except the innermost, must contain exactly one loop
+ // or loop nest.
+ // This checks for more than 1 loop at the current level, the
+ // 'depth'-satisifed checking manages the 'not zero' case.
+ if (CollapseInfo.CurLevelHasLoopAlready) {
+ Diag(ForLoc, diag::err_acc_collapse_multiple_loops);
+ assert(CollapseInfo.ActiveCollapse && "No collapse object?");
+ Diag(CollapseInfo.ActiveCollapse->getBeginLoc(),
+ diag::note_acc_collapse_clause_here);
+ } else {
+ --(*CollapseInfo.CurCollapseCount);
+
+ // Once we've hit zero here, we know we have deep enough 'for' loops to
+ // get to the bottom.
+ if (*CollapseInfo.CurCollapseCount == 0)
+ CollapseInfo.CollapseDepthSatisfied = true;
+ }
+ }
+
+ // Set this to 'false' for the body of this loop, so that the next level
+ // checks independently.
+ CollapseInfo.CurLevelHasLoopAlready = false;
}
void SemaOpenACC::ActOnForStmtEnd(SourceLocation ForLoc, StmtResult Body) {
if (!getLangOpts().OpenACC)
return;
+ // Set this to 'true' so if we find another one at this level we can diagnose.
+ CollapseInfo.CurLevelHasLoopAlready = true;
}
bool SemaOpenACC::ActOnStartStmtDirective(OpenACCDirectiveKind K,
@@ -1827,12 +1858,23 @@ StmtResult SemaOpenACC::ActOnAssociatedStmt(SourceLocation DirectiveLoc,
// the 'structured block'.
return AssocStmt;
case OpenACCDirectiveKind::Loop:
- if (AssocStmt.isUsable() &&
- !isa<CXXForRangeStmt, ForStmt>(AssocStmt.get())) {
+ if (!AssocStmt.isUsable())
+ return StmtError();
+
+ if (!isa<CXXForRangeStmt, ForStmt>(AssocStmt.get())) {
Diag(AssocStmt.get()->getBeginLoc(), diag::err_acc_loop_not_for_loop);
Diag(DirectiveLoc, diag::note_acc_construct_here) << K;
return StmtError();
}
+
+ if (!CollapseInfo.CollapseDepthSatisfied) {
+ Diag(DirectiveLoc, diag::err_acc_collapse_insufficient_loops);
+ assert(CollapseInfo.ActiveCollapse && "Collapse count without object?");
+ Diag(CollapseInfo.ActiveCollapse->getBeginLoc(),
+ diag::note_acc_collapse_clause_here);
+ return StmtError();
+ }
+
// TODO OpenACC: 2.9 ~ line 2010 specifies that the associated loop has some
// restrictions when there is a 'seq' clause in place. We probably need to
// implement that, including piping in the clauses here.
diff --git a/clang/test/SemaOpenACC/loop-construct-collapse-clause.cpp b/clang/test/SemaOpenACC/loop-construct-collapse-clause.cpp
index 6e4a7e9468aef8..bbb7abf24ffdf6 100644
--- a/clang/test/SemaOpenACC/loop-construct-collapse-clause.cpp
+++ b/clang/test/SemaOpenACC/loop-construct-collapse-clause.cpp
@@ -115,10 +115,42 @@ void negative_constexpr(int i) {
negative_constexpr_templ<int, 1>(); // #NCET1
}
+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 loop collapse(Val)
+ for(;;)
+ for(;;);
+}
+
+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 loop collapse(3)
+ for(;;)
+ for(;;);
+
+ // 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 loop collapse(three())
+ for(;;)
+ for(;;);
+
+ // 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 loop collapse(ConvertsThree{})
+ for(;;)
+ for(;;);
+}
+
template<typename T, unsigned Three>
void not_single_loop_templ() {
T Arr[5];
- // expected-note at +1{{active 'collapse' clause defined here}}
+ // 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 loop collapse(3)
for(auto x : Arr) {
for(auto y : Arr){
@@ -126,7 +158,8 @@ void not_single_loop_templ() {
}
}
- // expected-note at +1{{active 'collapse' clause defined here}}
+ // 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 loop collapse(Three)
for(;;) {
for(;;){
@@ -142,7 +175,8 @@ void not_single_loop_templ() {
}
}
}
- // expected-note at +1{{active 'collapse' clause defined here}}
+ // 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 loop collapse(Three)
for(auto x : Arr) {
for(auto y: Arr) {
@@ -181,15 +215,16 @@ void not_single_loop() {
do{}while(true); // expected-error{{do loop cannot appear in intervening code of a 'loop' with a 'collapse' clause}}
}
- // expected-note at +1{{active 'collapse' clause defined here}}
+ // 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 loop collapse(3)
for(;;) {
for(;;){
while(true); // expected-error{{while loop cannot appear in intervening code of a 'loop' with a 'collapse' clause}}
}
}
-
- // expected-note at +1{{active 'collapse' clause defined here}}
+ // 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 loop collapse(3)
for(;;) {
for(;;){
@@ -211,8 +246,8 @@ void not_single_loop() {
}
int Arr[5];
-
- // expected-note at +1{{active 'collapse' clause defined here}}
+ // 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 loop collapse(3)
for(auto x : Arr) {
for(auto y : Arr){
@@ -220,6 +255,33 @@ void not_single_loop() {
}
}
+ // expected-note at +1 {{active 'collapse' clause defined here}}
+#pragma acc loop collapse(3)
+ for (;;) {
+ for (;;) {
+ for(;;);
+ }
+ // expected-error at +1{{more than one for-loop in a loop associated with OpenACC 'loop' construct with a 'collapse' clause}}
+ for(;;);
+ }
+
+ // expected-note at +1 {{active 'collapse' clause defined here}}
+#pragma acc loop collapse(3)
+ for (;;) {
+ for (;;) {
+ for(;;);
+ // expected-error at +1{{more than one for-loop in a loop associated with OpenACC 'loop' construct with a 'collapse' clause}}
+ for(;;);
+ }
+ }
+
+ for(;;);
+#pragma acc loop collapse(3)
+ for (;;) {
+ for (;;) {
+ for (;;);
+ }
+ }
}
template<unsigned Two, unsigned Three>
@@ -227,7 +289,8 @@ void no_other_directives() {
#pragma acc loop collapse(Two)
for(;;) {
for (;;) { // last loop associated with the top level.
-#pragma acc loop collapse(Three) // expected-note{{active 'collapse' clause defined here}}
+ // expected-error at +1{{'collapse' clause specifies a loop count greater than the number of available loops}}
+#pragma acc loop collapse(Three) // expected-note 2{{active 'collapse' clause defined here}}
for(;;) {
for(;;) {
// expected-error at +1{{OpenACC 'serial' construct cannot appear in intervening code of a 'loop' with a 'collapse' clause}}
@@ -243,6 +306,7 @@ void no_other_directives() {
#pragma acc loop collapse(Three)
for(;;) {
for(;;) {
+ for(;;);
}
}
}
More information about the cfe-commits
mailing list