[OpenACC] Implement Compute Construct 'goto' in/out logic (#83326)

Compute Constructs do not permit jumping in/out of them, so this patch
implements this for 'goto' as a followup to the other patches that have
done the same thing.

It does this by modifying the JumpDiagnostics to work with this, plus
setting the function to needing jump diagnostics if we discover a goto
or label inside of a Compute Construct.

Author: erichkeane

clang/include/clang/Basic/DiagnosticSemaKinds.td

@@ -12214,4 +12214,8 @@ def err_acc_construct_appertainment
 def err_acc_branch_in_out_compute_construct
     : Error<"invalid %select{branch|return}0 %select{out of|into}1 OpenACC "
             "Compute Construct">;
+def note_acc_branch_into_compute_construct
+    : Note<"invalid branch into OpenACC Compute Construct">;
+def note_acc_branch_out_of_compute_construct
+    : Note<"invalid branch out of OpenACC Compute Construct">;
 } // end of sema component.

clang/lib/Sema/JumpDiagnostics.cpp

@@ -604,6 +604,16 @@ void JumpScopeChecker::BuildScopeInformation(Stmt *S,
     break;
   }
 
+  case Stmt::OpenACCComputeConstructClass: {
+    unsigned NewParentScope = Scopes.size();
+    OpenACCComputeConstruct *CC = cast<OpenACCComputeConstruct>(S);
+    Scopes.push_back(GotoScope(
+        ParentScope, diag::note_acc_branch_into_compute_construct,
+        diag::note_acc_branch_out_of_compute_construct, CC->getBeginLoc()));
+    BuildScopeInformation(CC->getStructuredBlock(), NewParentScope);
+    return;
+  }
+
   default:
     if (auto *ED = dyn_cast<OMPExecutableDirective>(S)) {
       if (!ED->isStandaloneDirective()) {
@@ -936,11 +946,16 @@ void JumpScopeChecker::CheckJump(Stmt *From, Stmt *To, SourceLocation DiagLoc,
       if (Scopes[I].InDiag == diag::note_protected_by_seh_finally) {
         S.Diag(From->getBeginLoc(), diag::warn_jump_out_of_seh_finally);
         break;
-      }
-      if (Scopes[I].InDiag == diag::note_omp_protected_structured_block) {
+      } else if (Scopes[I].InDiag ==
+                 diag::note_omp_protected_structured_block) {
         S.Diag(From->getBeginLoc(), diag::err_goto_into_protected_scope);
         S.Diag(To->getBeginLoc(), diag::note_omp_exits_structured_block);
         break;
+      } else if (Scopes[I].InDiag ==
+                 diag::note_acc_branch_into_compute_construct) {
+        S.Diag(From->getBeginLoc(), diag::err_goto_into_protected_scope);
+        S.Diag(Scopes[I].Loc, diag::note_acc_branch_out_of_compute_construct);
+        return;
       }
     }
   }

clang/lib/Sema/SemaStmt.cpp

@@ -567,6 +567,11 @@ Sema::ActOnLabelStmt(SourceLocation IdentLoc, LabelDecl *TheDecl,
     Diag(IdentLoc, diag::warn_reserved_extern_symbol)
         << TheDecl << static_cast<int>(Status);
 
+  // If this label is in a compute construct scope, we need to make sure we
+  // check gotos in/out.
+  if (getCurScope()->isInOpenACCComputeConstructScope())
+    setFunctionHasBranchProtectedScope();
+
   // Otherwise, things are good.  Fill in the declaration and return it.
   LabelStmt *LS = new (Context) LabelStmt(IdentLoc, TheDecl, SubStmt);
   TheDecl->setStmt(LS);
@@ -3304,6 +3309,12 @@ StmtResult Sema::ActOnGotoStmt(SourceLocation GotoLoc,
                                SourceLocation LabelLoc,
                                LabelDecl *TheDecl) {
   setFunctionHasBranchIntoScope();
+
+  // If this goto is in a compute construct scope, we need to make sure we check
+  // gotos in/out.
+  if (getCurScope()->isInOpenACCComputeConstructScope())
+    setFunctionHasBranchProtectedScope();
+
   TheDecl->markUsed(Context);
   return new (Context) GotoStmt(TheDecl, GotoLoc, LabelLoc);
 }
@@ -3332,6 +3343,11 @@ Sema::ActOnIndirectGotoStmt(SourceLocation GotoLoc, SourceLocation StarLoc,
 
   setFunctionHasIndirectGoto();
 
+  // If this goto is in a compute construct scope, we need to make sure we
+  // check gotos in/out.
+  if (getCurScope()->isInOpenACCComputeConstructScope())
+    setFunctionHasBranchProtectedScope();
+
   return new (Context) IndirectGotoStmt(GotoLoc, StarLoc, E);
 }
 

clang/test/SemaOpenACC/no-branch-in-out.c

@@ -113,3 +113,200 @@ void Return() {
     }
   }
 }
+
+void Goto() {
+  int j;
+#pragma acc parallel // expected-note{{invalid branch out of OpenACC Compute Construct}}
+  while(j) {
+    if (j <3)
+      goto LABEL; // expected-error{{cannot jump from this goto statement to its label}}
+  }
+
+LABEL:
+  {}
+
+  goto LABEL_IN; // expected-error{{cannot jump from this goto statement to its label}}
+
+#pragma acc parallel // expected-note{{invalid branch into OpenACC Compute Construct}}
+  for(int i = 0; i < 5; ++i) {
+LABEL_IN:
+    {}
+  }
+
+#pragma acc parallel
+  for(int i = 0; i < 5; ++i) {
+LABEL_NOT_CALLED:
+    {}
+  }
+
+#pragma acc parallel
+  {
+    goto ANOTHER_LOOP; // expected-error{{cannot jump from this goto statement to its label}}
+
+  }
+#pragma acc parallel// expected-note{{invalid branch into OpenACC Compute Construct}}
+
+  {
+ANOTHER_LOOP:
+    {}
+  }
+
+#pragma acc parallel
+  {
+  while (j) {
+    --j;
+    if (j < 3)
+      goto LABEL2;
+
+    if (j > 4)
+      break;
+  }
+LABEL2:
+  {}
+  }
+
+#pragma acc parallel
+  do {
+    if (j < 3)
+      goto LABEL3;
+
+    if (j > 4)
+      break; // expected-error{{invalid branch out of OpenACC Compute Construct}}
+
+LABEL3:
+  {}
+  } while (j);
+
+LABEL4:
+  {}
+#pragma acc parallel// expected-note{{invalid branch out of OpenACC Compute Construct}}
+  {
+    goto LABEL4;// expected-error{{cannot jump from this goto statement to its label}}
+  }
+
+#pragma acc parallel// expected-note{{invalid branch into OpenACC Compute Construct}}
+
+  {
+LABEL5:
+    {}
+  }
+
+  {
+    goto LABEL5;// expected-error{{cannot jump from this goto statement to its label}}
+  }
+
+#pragma acc parallel
+  {
+LABEL6:
+    {}
+    goto LABEL6;
+
+  }
+
+#pragma acc parallel
+  goto LABEL7; // expected-error{{cannot jump from this goto statement to its label}}
+#pragma acc parallel// expected-note{{invalid branch into OpenACC Compute Construct}}
+  {
+LABEL7:{}
+  }
+
+#pragma acc parallel
+  LABEL8:{}
+#pragma acc parallel// expected-note{{invalid branch out of OpenACC Compute Construct}}
+  {
+    goto LABEL8;// expected-error{{cannot jump from this goto statement to its label}}
+  }
+
+
+#pragma acc parallel// expected-note{{invalid branch into OpenACC Compute Construct}}
+  {
+LABEL9:{}
+  }
+
+  ({goto LABEL9;});// expected-error{{cannot jump from this goto statement to its label}}
+
+#pragma acc parallel// expected-note{{invalid branch out of OpenACC Compute Construct}}
+  {
+  ({goto LABEL10;});// expected-error{{cannot jump from this goto statement to its label}}
+  }
+
+LABEL10:{}
+
+  ({goto LABEL11;});// expected-error{{cannot jump from this goto statement to its label}}
+#pragma acc parallel// expected-note{{invalid branch into OpenACC Compute Construct}}
+  {
+LABEL11:{}
+  }
+
+LABEL12:{}
+#pragma acc parallel// expected-note{{invalid branch out of OpenACC Compute Construct}}
+  {
+  ({goto LABEL12;});// expected-error{{cannot jump from this goto statement to its label}}
+  }
+
+#pragma acc parallel
+  {
+  ({goto LABEL13;});
+LABEL13:{}
+  }
+
+#pragma acc parallel
+  {
+  LABEL14:{}
+  ({goto LABEL14;});
+  }
+}
+
+void IndirectGoto1() {
+  void* ptr;
+#pragma acc parallel
+  {
+LABEL1:{}
+    ptr = &&LABEL1;
+
+    goto *ptr;
+
+  }
+}
+
+void IndirectGoto2() {
+  void* ptr;
+LABEL2:{} // #GOTOLBL2
+    ptr = &&LABEL2;
+#pragma acc parallel // #GOTOPAR2
+  {
+// expected-error@+3{{cannot jump from this indirect goto statement to one of its possible targets}}
+// expected-note@#GOTOLBL2{{possible target of indirect goto statement}}
+// expected-note@#GOTOPAR2{{invalid branch out of OpenACC Compute Construct}}
+    goto *ptr;
+  }
+}
+
+void IndirectGoto3() {
+  void* ptr;
+#pragma acc parallel // #GOTOPAR3
+  {
+LABEL3:{} // #GOTOLBL3
+    ptr = &&LABEL3;
+  }
+// expected-error@+3{{cannot jump from this indirect goto statement to one of its possible targets}}
+// expected-note@#GOTOLBL3{{possible target of indirect goto statement}}
+// expected-note@#GOTOPAR3{{invalid branch into OpenACC Compute Construct}}
+  goto *ptr;
+}
+
+void IndirectGoto4() {
+  void* ptr;
+#pragma acc parallel // #GOTOPAR4
+  {
+LABEL4:{}
+    ptr = &&LABEL4;
+// expected-error@+3{{cannot jump from this indirect goto statement to one of its possible targets}}
+// expected-note@#GOTOLBL5{{possible target of indirect goto statement}}
+// expected-note@#GOTOPAR4{{invalid branch out of OpenACC Compute Construct}}
+    goto *ptr;
+  }
+LABEL5:// #GOTOLBL5
+
+  ptr=&&LABEL5;
+}