[OpenACC] Implement sema for 'async' on 'data' constructs

This also is a clause that doesn't have any special rules, so this patch
enables it and adds tests.

Author: erichkeane

clang/lib/Sema/SemaOpenACC.cpp

@@ -876,11 +876,13 @@ OpenACCClause *SemaOpenACCClauseVisitor::VisitVectorLengthClause(
 
 OpenACCClause *SemaOpenACCClauseVisitor::VisitAsyncClause(
     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();
 
   // There is no prose in the standard that says duplicates aren't allowed,

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

@@ -43,4 +43,12 @@ void foo() {
 // CHECK: #pragma acc host_data if(i == array[1])
 #pragma acc host_data use_device(Var) if(i == array[1])
   ;
+
+// CHECK: #pragma acc data 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
 }

clang/test/ParserOpenACC/parse-clauses.c

@@ -521,25 +521,21 @@ void VarListClauses() {
 #pragma acc serial firstprivate(s.array[s.value : 5], s.value), self
   for(int i = 0; i < 5;++i) {}
 
-  // expected-error@+3{{expected ','}}
-  // expected-warning@+2{{OpenACC clause 'delete' not yet implemented, clause ignored}}
-  // expected-warning@+1{{OpenACC clause 'async' not yet implemented, clause ignored}}
+  // expected-error@+2{{expected ','}}
+  // expected-warning@+1{{OpenACC clause 'delete' not yet implemented, clause ignored}}
 #pragma acc exit data delete(s.array[s.value] s.array[s.value :5] ) async
   for(int i = 0; i < 5;++i) {}
 
-  // expected-warning@+2{{OpenACC clause 'delete' not yet implemented, clause ignored}}
-  // expected-warning@+1{{OpenACC clause 'async' not yet implemented, clause ignored}}
+  // expected-warning@+1{{OpenACC clause 'delete' not yet implemented, clause ignored}}
 #pragma acc exit data delete(s.array[s.value : 5], s.value),async
   for(int i = 0; i < 5;++i) {}
 
-  // expected-error@+3{{expected ','}}
-  // expected-warning@+2{{OpenACC clause 'use_device' not yet implemented, clause ignored}}
-  // expected-warning@+1{{OpenACC clause 'async' not yet implemented, clause ignored}}
+  // expected-error@+2{{expected ','}}
+  // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented, clause ignored}}
 #pragma acc exit data use_device(s.array[s.value] s.array[s.value :5] ),async
   for(int i = 0; i < 5;++i) {}
 
-  // expected-warning@+2{{OpenACC clause 'use_device' not yet implemented, clause ignored}}
-  // expected-warning@+1{{OpenACC clause 'async' not yet implemented, clause ignored}}
+  // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented, clause ignored}}
 #pragma acc exit data use_device(s.array[s.value : 5], s.value), async
   for(int i = 0; i < 5;++i) {}
 

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

@@ -0,0 +1,61 @@
+// 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
+
+int some_int();
+
+template<typename T>
+void TemplUses() {
+  // CHECK: FunctionTemplateDecl{{.*}}TemplUses
+  // CHECK-NEXT: TemplateTypeParmDecl{{.*}}T
+  // CHECK-NEXT: FunctionDecl{{.*}}TemplUses
+  // CHECK-NEXT: CompoundStmt
+
+#pragma acc data async(some_int())
+  ;
+  // CHECK-NEXT: OpenACCDataConstruct{{.*}}data
+  // CHECK-NEXT: async clause
+  // CHECK-NEXT: CallExpr{{.*}}'int'
+  // CHECK-NEXT: ImplicitCastExpr
+  // CHECK-NEXT: DeclRefExpr{{.*}}'some_int' 'int ()'
+  // CHECK-NEXT: NullStmt
+#pragma acc enter data async(T{})
+  // CHECK-NEXT: OpenACCEnterDataConstruct{{.*}}enter data
+  // CHECK-NEXT: async clause
+  // CHECK-NEXT: CXXUnresolvedConstructExpr{{.*}} 'T' 'T' list
+  // CHECK-NEXT: InitListExpr{{.*}}'void'
+#pragma acc exit data async
+  // CHECK-NEXT: OpenACCExitDataConstruct{{.*}}exit data
+  // CHECK-NEXT: async clause
+
+  // Instantiations
+  // CHECK-NEXT: FunctionDecl{{.*}} TemplUses 'void ()' implicit_instantiation
+  // CHECK-NEXT: TemplateArgument type 'int'
+  // CHECK-NEXT: BuiltinType{{.*}} 'int'
+  // CHECK-NEXT: CompoundStmt
+
+  // CHECK-NEXT: OpenACCDataConstruct{{.*}}data
+  // CHECK-NEXT: async clause
+  // CHECK-NEXT: CallExpr{{.*}}'int'
+  // CHECK-NEXT: ImplicitCastExpr
+  // CHECK-NEXT: DeclRefExpr{{.*}}'some_int' 'int ()'
+  // CHECK-NEXT: NullStmt
+
+  // CHECK-NEXT: OpenACCEnterDataConstruct{{.*}}enter data
+  // CHECK-NEXT: async clause
+  // CHECK-NEXT: CXXFunctionalCastExpr
+  // CHECK-NEXT: InitListExpr{{.*}}'int'
+
+  // CHECK-NEXT: OpenACCExitDataConstruct{{.*}}exit data
+  // CHECK-NEXT: async clause
+}
+void Inst() {
+  TemplUses<int>();
+}
+
+
+#endif // PCH_HELPER

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

@@ -0,0 +1,44 @@
+// RUN: %clang_cc1 %s -fopenacc -verify
+
+void Test() {
+  int I;
+  struct NotConvertible{} NC;
+  // No special rules for this clause on the data constructs, so not much to
+  // test that isn't covered by combined/compute.
+  // expected-warning@+1{{OpenACC clause 'copyin' not yet implemented}}
+#pragma acc data copyin(I) async(I)
+  ;
+  // expected-warning@+1{{OpenACC clause 'copyin' not yet implemented}}
+#pragma acc enter data copyin(I) async(I)
+  // expected-warning@+1{{OpenACC clause 'copyout' not yet implemented}}
+#pragma acc exit data copyout(I) async(I)
+  // expected-warning@+2{{OpenACC clause 'use_device' not yet implemented}}
+  // expected-error@+1{{OpenACC 'async' clause is not valid on 'host_data' directive}}
+#pragma acc host_data use_device(I) async(I)
+  ;
+
+  // expected-warning@+2{{OpenACC clause 'copyin' not yet implemented}}
+  // expected-error@+1{{OpenACC clause 'async' requires expression of integer type ('struct NotConvertible' invalid)}}
+#pragma acc data copyin(NC) async(NC)
+  ;
+  // expected-warning@+2{{OpenACC clause 'copyin' not yet implemented}}
+  // expected-error@+1{{OpenACC clause 'async' requires expression of integer type ('struct NotConvertible' invalid)}}
+#pragma acc enter data copyin(NC) async(NC)
+  // expected-warning@+2{{OpenACC clause 'copyout' not yet implemented}}
+  // expected-error@+1{{OpenACC clause 'async' requires expression of integer type ('struct NotConvertible' invalid)}}
+#pragma acc exit data copyout(NC) async(NC)
+  // expected-warning@+2{{OpenACC clause 'use_device' not yet implemented}}
+  // expected-error@+1{{OpenACC clause 'async' requires expression of integer type ('struct NotConvertible' invalid)}}
+#pragma acc host_data use_device(NC) async(NC)
+  ;
+
+  // expected-warning@+3{{OpenACC clause 'copyin' not yet implemented}}
+  // expected-error@+2{{OpenACC 'async' clause cannot appear more than once on a 'data' directive}}
+  // expected-note@+1{{previous clause is here}}
+#pragma acc data copyin(I) async(I) async(I)
+  ;
+  // expected-warning@+3{{OpenACC clause 'copyin' not yet implemented}}
+  // expected-error@+2{{expected ')'}}
+  // expected-note@+1{{to match this '('}}
+#pragma acc enter data copyin(I) async(I, I)
+}

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

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

clang/test/SemaOpenACC/data-construct.cpp

@@ -55,7 +55,6 @@ void AtLeastOneOf() {
 #pragma acc data if(Var)
   ;
 
-  // expected-warning@+1{{OpenACC clause 'async' not yet implemented}}
 #pragma acc data async
   ;
 
@@ -80,7 +79,6 @@ void AtLeastOneOf() {
   // they don't have at least one of the above clauses.
 
 #pragma acc enter data if(Var)
-  // expected-warning@+1{{OpenACC clause 'async' not yet implemented}}
 #pragma acc enter data async
   // expected-warning@+1{{OpenACC clause 'wait' not yet implemented}}
 #pragma acc enter data wait
@@ -98,7 +96,6 @@ void AtLeastOneOf() {
   // they don't have at least one of the above clauses.
 
 #pragma acc exit data if(Var)
-  // expected-warning@+1{{OpenACC clause 'async' not yet implemented}}
 #pragma acc exit data async
   // expected-warning@+1{{OpenACC clause 'wait' not yet implemented}}
 #pragma acc exit data wait
@@ -167,7 +164,6 @@ void DataRules() {
   // expected-note@+1{{previous clause is here}}
 #pragma acc data device_type(*) if(Var)
   ;
-  // expected-warning@+1{{OpenACC clause 'async' not yet implemented}}
 #pragma acc data device_type(*) async
   ;
   // expected-warning@+1{{OpenACC clause 'wait' not yet implemented}}