[OpenACC] 'collapse' clause 'force' tag #110906
Merged
Conversation
This file contains hidden or bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
The 'force' tag permits intervening code on the applicable 'loop' construct, so this implements the restriction when the 'force' tag isn't present.
llvmbot
added
clang
|
@llvm/pr-subscribers-clang Author: Erich Keane (erichkeane) ChangesThe 'force' tag permits intervening code on the applicable 'loop' construct, Full diff: https://github.com/llvm/llvm-project/pull/110906.diff 3 Files Affected:
diff --git a/clang/include/clang/Basic/DiagnosticSemaKinds.td b/clang/include/clang/Basic/DiagnosticSemaKinds.td
index dae357eb2d370e..066c4aa13c467d 100644
--- a/clang/include/clang/Basic/DiagnosticSemaKinds.td
+++ b/clang/include/clang/Basic/DiagnosticSemaKinds.td
@@ -12628,6 +12628,9 @@ def err_acc_collapse_multiple_loops
def err_acc_collapse_insufficient_loops
: Error<"'collapse' clause specifies a loop count greater than the number "
"of available loops">;
+def err_acc_collapse_intervening_code
+ : Error<"inner loops must be tightly nested inside a 'collapse' clause on "
+ "a 'loop' construct">;
// AMDGCN builtins diagnostics
def err_amdgcn_global_load_lds_size_invalid_value : Error<"invalid size value">;
diff --git a/clang/lib/Sema/SemaOpenACC.cpp b/clang/lib/Sema/SemaOpenACC.cpp
index 47b4bd77d86d18..dfaf726891cfc8 100644
--- a/clang/lib/Sema/SemaOpenACC.cpp
+++ b/clang/lib/Sema/SemaOpenACC.cpp
@@ -1770,11 +1770,61 @@ void SemaOpenACC::ActOnForStmtBegin(SourceLocation ForLoc) {
CollapseInfo.CurLevelHasLoopAlready = false;
}
+namespace {
+SourceLocation FindInterveningCodeInCollapseLoop(const Stmt *CurStmt) {
+ // We should diagnose on anything except `CompoundStmt`, `NullStmt`,
+ // `ForStmt`, `CXXForRangeStmt`, since those are legal, and `WhileStmt` and
+ // `DoStmt`, as those are caught as a violation elsewhere.
+ // For `CompoundStmt` we need to search inside of it.
+ if (!CurStmt ||
+ isa<ForStmt, NullStmt, ForStmt, CXXForRangeStmt, WhileStmt, DoStmt>(
+ CurStmt))
+ return SourceLocation{};
+
+ // Any other construct is an error anyway, so it has already been diagnosed.
+ if (isa<OpenACCConstructStmt>(CurStmt))
+ return SourceLocation{};
+
+ // Search inside the compound statement, this allows for arbitrary nesting
+ // of compound statements, as long as there isn't any code inside.
+ if (const auto *CS = dyn_cast<CompoundStmt>(CurStmt)) {
+ for (const auto *ChildStmt : CS->children()) {
+ SourceLocation ChildStmtLoc =
+ FindInterveningCodeInCollapseLoop(ChildStmt);
+ if (ChildStmtLoc.isValid())
+ return ChildStmtLoc;
+ }
+ // Empty/not invalid compound statements are legal.
+ return SourceLocation{};
+ }
+ return CurStmt->getBeginLoc();
+}
+} // namespace
+
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;
+
+ if (!Body.isUsable())
+ return;
+
+ if (CollapseInfo.CurCollapseCount && *CollapseInfo.CurCollapseCount > 0 &&
+ !CollapseInfo.ActiveCollapse->hasForce()) {
+ // OpenACC 3.3: 2.9.1
+ // If the 'force' modifier does not appear, then the associated loops must
+ // be tightly nested. If the force modifier appears, then any intervening
+ // code may be executed multiple times as needed to perform the collapse.
+
+ SourceLocation OtherStmtLoc = FindInterveningCodeInCollapseLoop(Body.get());
+
+ if (OtherStmtLoc.isValid()) {
+ Diag(OtherStmtLoc, diag::err_acc_collapse_intervening_code);
+ Diag(CollapseInfo.ActiveCollapse->getBeginLoc(),
+ diag::note_acc_collapse_clause_here);
+ }
+ }
}
bool SemaOpenACC::ActOnStartStmtDirective(OpenACCDirectiveKind K,
diff --git a/clang/test/SemaOpenACC/loop-construct-collapse-clause.cpp b/clang/test/SemaOpenACC/loop-construct-collapse-clause.cpp
index bbb7abf24ffdf6..953775a423cdba 100644
--- a/clang/test/SemaOpenACC/loop-construct-collapse-clause.cpp
+++ b/clang/test/SemaOpenACC/loop-construct-collapse-clause.cpp
@@ -336,3 +336,146 @@ void no_other_directives() {
}
}
+void call();
+
+template<unsigned Two>
+void intervening_without_force_templ() {
+ // expected-note@+1{{active 'collapse' clause defined here}}
+#pragma acc loop collapse(2)
+ for(;;) {
+ // expected-error@+1{{inner loops must be tightly nested inside a 'collapse' clause on a 'loop' construct}}
+ call();
+ for(;;){}
+ }
+
+ // expected-note@+1{{active 'collapse' clause defined here}}
+#pragma acc loop collapse(Two)
+ for(;;) {
+ // expected-error@+1{{inner loops must be tightly nested inside a 'collapse' clause on a 'loop' construct}}
+ call();
+ for(;;){}
+ }
+
+ // expected-note@+1{{active 'collapse' clause defined here}}
+#pragma acc loop collapse(2)
+ for(;;) {
+ for(;;){}
+ // expected-error@+1{{inner loops must be tightly nested inside a 'collapse' clause on a 'loop' construct}}
+ call();
+ }
+
+#pragma acc loop collapse(force:2)
+ for(;;) {
+ call();
+ for(;;){}
+ }
+
+#pragma acc loop collapse(force:Two)
+ for(;;) {
+ call();
+ for(;;){}
+ }
+
+
+#pragma acc loop collapse(force:2)
+ for(;;) {
+ for(;;){}
+ call();
+ }
+
+#pragma acc loop collapse(force:Two)
+ for(;;) {
+ for(;;){}
+ call();
+ }
+
+#pragma acc loop collapse(Two)
+ for(;;) {
+ for(;;){
+ call();
+ }
+ }
+
+#pragma acc loop collapse(Two)
+ for(;;) {
+ {
+ {
+ for(;;){
+ call();
+ }
+ }
+ }
+ }
+
+#pragma acc loop collapse(force:Two)
+ for(;;) {
+ for(;;){
+ call();
+ }
+ }
+
+ // expected-note@+1{{active 'collapse' clause defined here}}
+#pragma acc loop collapse(Two)
+ for(;;) {
+ for(;;){}
+ // expected-error@+1{{inner loops must be tightly nested inside a 'collapse' clause on a 'loop' construct}}
+ call();
+ }
+}
+
+void intervening_without_force() {
+ intervening_without_force_templ<2>(); // expected-note{{in instantiation of function template specialization}}
+ // expected-note@+1{{active 'collapse' clause defined here}}
+#pragma acc loop collapse(2)
+ for(;;) {
+ // expected-error@+1{{inner loops must be tightly nested inside a 'collapse' clause on a 'loop' construct}}
+ call();
+ for(;;){}
+ }
+
+ // expected-note@+1{{active 'collapse' clause defined here}}
+#pragma acc loop collapse(2)
+ for(;;) {
+ for(;;){}
+ // expected-error@+1{{inner loops must be tightly nested inside a 'collapse' clause on a 'loop' construct}}
+ call();
+ }
+
+ // The below two are fine, as they use the 'force' tag.
+#pragma acc loop collapse(force:2)
+ for(;;) {
+ call();
+ for(;;){}
+ }
+
+#pragma acc loop collapse(force:2)
+ for(;;) {
+ for(;;){}
+ call();
+ }
+
+#pragma acc loop collapse(2)
+ for(;;) {
+ for(;;){
+ call();
+ }
+ }
+#pragma acc loop collapse(2)
+ for(;;) {
+ {
+ {
+ for(;;){
+ call();
+ }
+ }
+ }
+ }
+
+#pragma acc loop collapse(force:2)
+ for(;;) {
+ for(;;){
+ call();
+ }
+ }
+}
+
|
|
LLVM Buildbot has detected a new failure on builder Full details are available at: https://lab.llvm.org/buildbot/#/builders/64/builds/1165 Here is the relevant piece of the build log for the reference |
Sign up for free
to join this conversation on GitHub.
Already have an account?
Sign in to comment
Add this suggestion to a batch that can be applied as a single commit.
This suggestion is invalid because no changes were made to the code.
Suggestions cannot be applied while the pull request is closed.
Suggestions cannot be applied while viewing a subset of changes.
Only one suggestion per line can be applied in a batch.
Add this suggestion to a batch that can be applied as a single commit.
Applying suggestions on deleted lines is not supported.
You must change the existing code in this line in order to create a valid suggestion.
Outdated suggestions cannot be applied.
This suggestion has been applied or marked resolved.
Suggestions cannot be applied from pending reviews.
Suggestions cannot be applied on multi-line comments.
Suggestions cannot be applied while the pull request is queued to merge.
Suggestion cannot be applied right now. Please check back later.
The 'force' tag permits intervening code on the applicable 'loop' construct,
so this implements the restriction when the 'force' tag isn't present.