Merge remote-tracking branch 'origin/main' into vplan-compute-iv-end-values

Author: fhahn

clang/lib/Sema/SemaOpenACC.cpp

@@ -587,14 +587,6 @@ class SemaOpenACCClauseVisitor {
 
 OpenACCClause *SemaOpenACCClauseVisitor::VisitDefaultClause(
     SemaOpenACC::OpenACCParsedClause &Clause) {
-  // Restrictions only properly implemented on 'compute'/'combined' constructs,
-  // and 'compute'/'combined' constructs are the only construct that can do
-  // anything with this yet, so skip/treat as unimplemented in this case.
-  // Only 'data' is left.
-  if (!isOpenACCComputeDirectiveKind(Clause.getDirectiveKind()) &&
-      !isOpenACCCombinedDirectiveKind(Clause.getDirectiveKind()))
-    return isNotImplemented();
-
   // Don't add an invalid clause to the AST.
   if (Clause.getDefaultClauseKind() == OpenACCDefaultClauseKind::Invalid)
     return nullptr;
@@ -1070,11 +1062,13 @@ OpenACCClause *SemaOpenACCClauseVisitor::VisitDevicePtrClause(
 
 OpenACCClause *SemaOpenACCClauseVisitor::VisitWaitClause(
     SemaOpenACC::OpenACCParsedClause &Clause) {
-  // Restrictions only properly implemented on 'compute'/'combined' constructs,
-  // and 'compute'/'combined' constructs are the only construct that can do
-  // anything with this yet, so skip/treat as unimplemented in this case.
+  // Restrictions only properly implemented on 'compute'/'combined'/'data'
+  // constructs, and 'compute'/'combined'/'data' constructs are the only
+  // construct that can do anything with this yet, so skip/treat as
+  // unimplemented in this case.
   if (!isOpenACCComputeDirectiveKind(Clause.getDirectiveKind()) &&
-      !isOpenACCCombinedDirectiveKind(Clause.getDirectiveKind()))
+      !isOpenACCCombinedDirectiveKind(Clause.getDirectiveKind()) &&
+      !isOpenACCDataDirectiveKind(Clause.getDirectiveKind()))
     return isNotImplemented();
 
   return OpenACCWaitClause::Create(

clang/test/AST/ast-print-openacc-data-construct.cpp

@@ -6,8 +6,7 @@ void foo() {
   // them, so the 'check' lines should start to include those once we implement
   // them.  For now, they don't emit those because they are 'not implemented'.
 
-// CHECK: #pragma acc data
-// CHECK-NOT: default(none)
+// CHECK: #pragma acc data default(none)
 #pragma acc data default(none)
   ;
 
@@ -29,9 +28,10 @@ void foo() {
   ;
 
   int i;
+  int *iPtr;
   int array[5];
 
-// CHECK: #pragma acc data if(i == array[1])
+// CHECK: #pragma acc data default(none) if(i == array[1])
 #pragma acc data default(none) if(i == array[1])
   ;
 // CHECK: #pragma acc enter data if(i == array[1])
@@ -44,11 +44,39 @@ void foo() {
 #pragma acc host_data use_device(Var) if(i == array[1])
   ;
 
-// CHECK: #pragma acc data async(i)
+// CHECK: #pragma acc data default(none) async(i)
 #pragma acc data default(none) async(i)
   ;
 // CHECK: #pragma acc enter data async(i)
 #pragma acc enter data copyin(i) async(i)
 // CHECK: #pragma acc exit data async
 #pragma acc exit data copyout(i) async
+
+// CHECK: #pragma acc data default(none) wait
+#pragma acc data default(none) wait()
+  ;
+
+// CHECK: #pragma acc enter data wait()
+#pragma acc enter data copyin(Var) wait()
+
+// CHECK: #pragma acc exit data wait(*iPtr, i)
+#pragma acc exit data copyout(Var) wait(*iPtr, i)
+
+// CHECK: #pragma acc data default(none) wait(queues: *iPtr, i)
+#pragma acc data default(none) wait(queues:*iPtr, i)
+  ;
+
+// CHECK: #pragma acc enter data wait(devnum: i : *iPtr, i)
+#pragma acc enter data copyin(Var) wait(devnum:i:*iPtr, i)
+
+// CHECK: #pragma acc exit data wait(devnum: i : queues: *iPtr, i)
+#pragma acc exit data copyout(Var) wait(devnum:i:queues:*iPtr, i)
+
+// CHECK: #pragma acc data default(none)
+#pragma acc data default(none)
+  ;
+
+// CHECK: #pragma acc data default(present)
+#pragma acc data default(present)
+  ;
 }

clang/test/ParserOpenACC/parse-clauses.c

@@ -19,7 +19,6 @@ void func() {
   // expected-error@+1{{invalid OpenACC clause 'invalid'}}
 #pragma acc enter data finalize invalid invalid finalize
 
-  // expected-warning@+2{{OpenACC clause 'wait' not yet implemented, clause ignored}}
   // expected-warning@+1{{OpenACC clause 'finalize' not yet implemented, clause ignored}}
 #pragma acc enter data wait finalize
 

clang/test/SemaOpenACC/combined-construct-default-ast.cpp

@@ -1,4 +1,3 @@
-
 // RUN: %clang_cc1 %s -fopenacc -ast-dump | FileCheck %s
 
 // Test this with PCH.

clang/test/SemaOpenACC/combined-construct-default-clause.c

@@ -28,7 +28,6 @@ void SingleOnly() {
   #pragma acc kernels loop default(none)
   for(int i = 0; i < 5; ++i);
 
-  // expected-warning@+1{{OpenACC clause 'default' not yet implemented}}
   #pragma acc data default(none)
   while(0);
 

clang/test/SemaOpenACC/compute-construct-default-clause.c

@@ -28,7 +28,6 @@ void SingleOnly() {
   #pragma acc kernels default(none)
   for(int i = 0; i < 5; ++i);
 
-  // expected-warning@+1{{OpenACC clause 'default' not yet implemented}}
   #pragma acc data default(none)
   while(0);
 

clang/test/SemaOpenACC/data-construct-ast.cpp

@@ -22,6 +22,7 @@ void NormalFunc() {
 #pragma acc data default(none)
   while (Var);
   // CHECK-NEXT: OpenACCDataConstruct{{.*}}data
+  // CHECK-NEXT: default(none)
   // CHECK-NEXT: WhileStmt
   // CHECK: NullStmt
 #pragma acc enter data copyin(Var)
@@ -49,6 +50,7 @@ void TemplFunc() {
 #pragma acc data default(none)
   while (Var);
   // CHECK-NEXT: OpenACCDataConstruct{{.*}}data
+  // CHECK-NEXT: default(none)
   // CHECK-NEXT: WhileStmt
   // CHECK: NullStmt
 #pragma acc enter data copyin(Var)
@@ -71,6 +73,7 @@ void TemplFunc() {
   // CHECK-NEXT: VarDecl
 
   // CHECK-NEXT: OpenACCDataConstruct{{.*}}data
+  // CHECK-NEXT: default(none)
   // CHECK-NEXT: WhileStmt
   // CHECK: NullStmt
 

clang/test/SemaOpenACC/data-construct-default-ast.cpp

@@ -0,0 +1,68 @@
+// 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
+void NormalFunc() {
+  // CHECK-LABEL: NormalFunc
+  // CHECK-NEXT: CompoundStmt
+  // CHECK-NEXT: OpenACCDataConstruct {{.*}}data
+  // CHECK-NEXT: default(none)
+#pragma acc data  default(none)
+  // CHECK: OpenACCDataConstruct {{.*}}data
+  // CHECK-NEXT: default(present)
+#pragma acc data default(present)
+    ;
+}
+template<typename T>
+void TemplFunc() {
+#pragma acc data default(none)
+  for (unsigned i = 0; i < 5; ++i) {
+    typename T::type I;
+  }
+
+#pragma acc data default(present)
+  for (unsigned i = 0; i < 5; ++i) {
+    typename T::type I;
+  }
+
+  // CHECK-LABEL: FunctionTemplateDecl {{.*}}TemplFunc
+  // CHECK-NEXT: TemplateTypeParmDecl
+
+  // Template Pattern:
+  // CHECK-NEXT: FunctionDecl
+  // CHECK-NEXT: CompoundStmt
+  // CHECK-NEXT: OpenACCDataConstruct {{.*}}data
+  // CHECK-NEXT: default(none)
+  // CHECK: VarDecl{{.*}} I 'typename T::type'
+
+  // CHECK-NEXT: OpenACCDataConstruct {{.*}}data
+  // CHECK-NEXT: default(present)
+  // CHECK: VarDecl{{.*}} I 'typename T::type'
+
+  // Check instantiation.
+  // CHECK-LABEL: FunctionDecl{{.*}} used TemplFunc 'void ()' implicit_instantiation
+  // CHECK-NEXT: TemplateArgument type 'S'
+  // CHECK-NEXT: RecordType
+  // CHECK-NEXT: CXXRecord
+  // CHECK-NEXT: CompoundStmt
+  // CHECK-NEXT: OpenACCDataConstruct {{.*}}data
+  // CHECK-NEXT: default(none)
+  // CHECK: VarDecl{{.*}} I 'typename S::type':'int'
+  // CHECK-NEXT: OpenACCDataConstruct {{.*}}data
+  // CHECK-NEXT: default(present)
+  // CHECK: VarDecl{{.*}} I 'typename S::type':'int'
+
+}
+struct S {
+  using type = int;
+};
+
+void use() {
+  TemplFunc<S>();
+}
+
+#endif

clang/test/SemaOpenACC/data-construct-default-clause.c

@@ -0,0 +1,24 @@
+// RUN: %clang_cc1 %s -fopenacc -verify
+
+void use() {
+  // expected-error@+1{{invalid value for 'default' clause; expected 'present' or 'none'}}
+#pragma acc data default(garbage)
+  ;
+#pragma acc data default(present)
+  ;
+#pragma acc data default(none)
+  ;
+  // expected-error@+2{{OpenACC 'default' clause cannot appear more than once on a 'data' directive}}
+  // expected-note@+1{{previous clause is here}}
+#pragma acc data default(none) default(present)
+  ;
+  // expected-error@+1{{OpenACC 'default' clause is not valid on 'enter data' directive}}
+#pragma acc enter data default(present)
+  ;
+  // expected-error@+1{{OpenACC 'default' clause is not valid on 'exit data' directive}}
+#pragma acc exit data default(none)
+  ;
+  // expected-error@+1{{OpenACC 'default' clause is not valid on 'host_data' directive}}
+#pragma acc host_data default(present)
+  ;
+}

clang/test/SemaOpenACC/data-construct-device_type-clause.c

@@ -4,7 +4,6 @@ void uses() {
   int Var;
 #pragma acc data device_type(foo) async
   ;
-  // expected-warning@+1{{OpenACC clause 'wait' not yet implemented}}
 #pragma acc data device_type(foo) wait
   ;
 #pragma acc data device_type(foo) dtype(false)