[OpenACC] Start enforcing 'device_type' clause values

Researching in prep of doing the implementation for lowering, I found
that the source of the valid identifiers list from flang is in the
frontend.  This patch adds the same list to the frontend, but does it as
a sema diagnostic, so we still parse it as an identifier/identifier-like
thing, but then diagnose it as invalid later.

Author: erichkeane

clang/include/clang/Basic/DiagnosticSemaKinds.td

@@ -13061,6 +13061,8 @@ def err_acc_decl_for_routine
     : Error<"expected function or lambda declaration for 'routine' construct">;
 def err_acc_invalid_modifier
     : Error<"OpenACC '%0' modifier not valid on '%1' clause">;
+def err_acc_invalid_default_type
+    : Error<"invalid value %0 in '%1' clause; valid values are %2">;
 
 // AMDGCN builtins diagnostics
 def err_amdgcn_load_lds_size_invalid_value : Error<"invalid size value">;

clang/lib/Sema/SemaOpenACCClause.cpp

@@ -1338,13 +1338,50 @@ OpenACCClause *SemaOpenACCClauseVisitor::VisitDeviceTypeClause(
       checkAlreadyHasClauseOfKind(SemaRef, ExistingClauses, Clause))
     return nullptr;
 
-  // TODO OpenACC: Once we get enough of the CodeGen implemented that we have
-  // a source for the list of valid architectures, we need to warn on unknown
-  // identifiers here.
+  // The list of valid device_type values. Flang also has these hardcoded in
+  // openacc_parsers.cpp, as there does not seem to be a reliable backend
+  // source. The list below is sourced from Flang, though NVC++ supports only
+  // 'nvidia', 'host', 'multicore', and 'default'.
+  const std::array<llvm::StringLiteral, 6> ValidValues{
+      "default", "nvidia", "acc_device_nvidia", "radeon", "host", "multicore"};
+  // As an optimization, we have a manually maintained list of valid values
+  // below, rather than trying to calculate from above. These should be kept in
+  // sync if/when the above list ever changes.
+  std::string ValidValuesString =
+      "'default', 'nvidia', 'acc_device_nvidia', 'radeon', 'host', 'multicore'";
+
+  llvm::SmallVector<DeviceTypeArgument> Architectures{
+      Clause.getDeviceTypeArchitectures()};
+
+  // The parser has ensured that we either have a single entry of just '*'
+  // (represented by a nullptr IdentifierInfo), or a list.
+
+  bool Diagnosed = false;
+  auto FilterPred = [&](const DeviceTypeArgument &Arch) {
+    // The '*' case.
+    if (!Arch.first)
+      return false;
+    return llvm::find_if(ValidValues, [&](StringRef RHS) {
+             return Arch.first->getName().equals_insensitive(RHS);
+           }) == ValidValues.end();
+  };
+
+  auto Diagnose = [&](const DeviceTypeArgument &Arch) {
+    Diagnosed = SemaRef.Diag(Arch.second, diag::err_acc_invalid_default_type)
+                << Arch.first << Clause.getClauseKind() << ValidValuesString;
+  };
+
+  // There aren't stable enumertor versions of 'for-each-then-erase', so do it
+  // here.  We DO keep track of whether we diagnosed something to make sure we
+  // don't do the 'erase_if' in the event that the first list didn't find
+  // anything.
+  llvm::for_each(llvm::make_filter_range(Architectures, FilterPred), Diagnose);
+  if (Diagnosed)
+    llvm::erase_if(Architectures, FilterPred);
 
   return OpenACCDeviceTypeClause::Create(
       Ctx, Clause.getClauseKind(), Clause.getBeginLoc(), Clause.getLParenLoc(),
-      Clause.getDeviceTypeArchitectures(), Clause.getEndLoc());
+      Architectures, Clause.getEndLoc());
 }
 
 OpenACCClause *SemaOpenACCClauseVisitor::VisitAutoClause(

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

@@ -37,40 +37,40 @@ void foo() {
   bool SomeB;
   struct SomeStruct{} SomeStructImpl;
 
-//CHECK: #pragma acc parallel loop dtype(SomeB)
+//CHECK: #pragma acc parallel loop dtype(default)
 // CHECK-NEXT: for (int i = 0; i < 5; ++i)
 // CHECK-NEXT: ;
-#pragma acc parallel loop dtype(SomeB)
+#pragma acc parallel loop dtype(default)
   for(int i = 0;i<5;++i);
 
-//CHECK: #pragma acc serial loop device_type(SomeStruct)
+//CHECK: #pragma acc serial loop device_type(radeon, host)
 // CHECK-NEXT: for (int i = 0; i < 5; ++i)
 // CHECK-NEXT: ;
-#pragma acc serial loop device_type(SomeStruct)
+#pragma acc serial loop device_type(radeon, host)
   for(int i = 0;i<5;++i);
 
-//CHECK: #pragma acc kernels loop device_type(int)
+//CHECK: #pragma acc kernels loop device_type(nvidia)
 // CHECK-NEXT: for (int i = 0; i < 5; ++i)
 // CHECK-NEXT: ;
-#pragma acc kernels loop device_type(int)
+#pragma acc kernels loop device_type(nvidia)
   for(int i = 0;i<5;++i);
 
-//CHECK: #pragma acc parallel loop dtype(bool)
+//CHECK: #pragma acc parallel loop dtype(multicore)
 // CHECK-NEXT: for (int i = 0; i < 5; ++i)
 // CHECK-NEXT: ;
-#pragma acc parallel loop dtype(bool)
+#pragma acc parallel loop dtype(multicore)
   for(int i = 0;i<5;++i);
 
-//CHECK: #pragma acc serial loop device_type(SomeStructImpl)
+//CHECK: #pragma acc serial loop device_type(default)
 // CHECK-NEXT: for (int i = 0; i < 5; ++i)
 // CHECK-NEXT: ;
-#pragma acc serial loop device_type (SomeStructImpl)
+#pragma acc serial loop device_type (default)
   for(int i = 0;i<5;++i);
 
-// CHECK: #pragma acc kernels loop dtype(AnotherIdent)
+// CHECK: #pragma acc kernels loop dtype(host)
 // CHECK-NEXT: for (int i = 0; i < 5; ++i)
 // CHECK-NEXT: ;
-#pragma acc kernels loop dtype(AnotherIdent)
+#pragma acc kernels loop dtype(host)
   for(int i = 0;i<5;++i);
 
   int i;

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

@@ -115,24 +115,24 @@ void foo() {
   bool SomeB;
   struct SomeStruct{} SomeStructImpl;
 
-//CHECK: #pragma acc parallel dtype(SomeB)
-#pragma acc parallel dtype(SomeB)
+//CHECK: #pragma acc parallel dtype(default)
+#pragma acc parallel dtype(default)
   while(true);
 
-//CHECK: #pragma acc parallel device_type(SomeStruct)
-#pragma acc parallel device_type(SomeStruct)
+//CHECK: #pragma acc parallel device_type(radeon)
+#pragma acc parallel device_type(radeon)
   while(true);
 
-//CHECK: #pragma acc parallel device_type(int)
-#pragma acc parallel device_type(int)
+//CHECK: #pragma acc parallel device_type(nvidia)
+#pragma acc parallel device_type(nvidia)
   while(true);
 
-//CHECK: #pragma acc parallel dtype(bool)
-#pragma acc parallel dtype(bool)
+//CHECK: #pragma acc parallel dtype(multicore)
+#pragma acc parallel dtype(multicore)
   while(true);
 
-//CHECK: #pragma acc parallel device_type(SomeStructImpl)
-#pragma acc parallel device_type (SomeStructImpl)
+//CHECK: #pragma acc parallel device_type(host)
+#pragma acc parallel device_type (host)
   while(true);
 
 //CHECK: #pragma acc parallel reduction(+: iPtr)

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

@@ -2,16 +2,12 @@
 
 void foo() {
   int Var;
-  // TODO OpenACC: These are only legal if they have one of a list of clauses on
-  // 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 default(none)
 #pragma acc data default(none)
   ;
 
-// CHECK: #pragma acc data default(none) device_type(int)
-#pragma acc data default(none) device_type(int)
+// CHECK: #pragma acc data default(none) device_type(radeon)
+#pragma acc data default(none) device_type(radeon)
   ;
 
 // CHECK: #pragma acc enter data copyin(Var)

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

@@ -9,6 +9,6 @@ void uses() {
 #pragma acc init device_type(*) device_num(Int)
 // CHECK: #pragma acc init device_type(*) if(Int == 5)
 #pragma acc init device_type(*) if (Int == 5)
-// CHECK: #pragma acc init device_type(SomeName)
-#pragma acc init device_type(SomeName)
+// CHECK: #pragma acc init device_type(radeon)
+#pragma acc init device_type(radeon)
 }

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

@@ -10,28 +10,28 @@ void foo() {
 #pragma acc loop
   for(int i = 0;i<5;++i);
 
-// CHECK: #pragma acc loop device_type(SomeStruct)
+// CHECK: #pragma acc loop device_type(default)
 // CHECK-NEXT: for (int i = 0; i < 5; ++i)
 // CHECK-NEXT: ;
-#pragma acc loop device_type(SomeStruct)
+#pragma acc loop device_type(default)
   for(int i = 0;i<5;++i);
 
-// CHECK: #pragma acc loop device_type(int)
+// CHECK: #pragma acc loop device_type(nvidia)
 // CHECK-NEXT: for (int i = 0; i < 5; ++i)
 // CHECK-NEXT: ;
-#pragma acc loop device_type(int)
+#pragma acc loop device_type(nvidia)
   for(int i = 0;i<5;++i);
 
-// CHECK: #pragma acc loop dtype(bool)
+// CHECK: #pragma acc loop dtype(radeon)
 // CHECK-NEXT: for (int i = 0; i < 5; ++i)
 // CHECK-NEXT: ;
-#pragma acc loop dtype(bool)
+#pragma acc loop dtype(radeon)
   for(int i = 0;i<5;++i);
 
-// CHECK: #pragma acc loop dtype(AnotherIdent)
+// CHECK: #pragma acc loop dtype(host)
 // CHECK-NEXT: for (int i = 0; i < 5; ++i)
 // CHECK-NEXT: ;
-#pragma acc loop dtype(AnotherIdent)
+#pragma acc loop dtype(host)
   for(int i = 0;i<5;++i);
 
 // CHECK: #pragma acc loop independent

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

@@ -8,10 +8,10 @@ auto Lambda = [](){};
 int function();
 #pragma acc routine (function) vector nohost bind("string")
 // CHECK: #pragma acc routine(function) vector nohost bind("string")
-#pragma acc routine(function) device_type(Something) seq
-// CHECK-NEXT: #pragma acc routine(function) device_type(Something) seq
-#pragma acc routine(function) dtype(Something) seq
-// CHECK-NEXT: #pragma acc routine(function) dtype(Something) seq
+#pragma acc routine(function) device_type(multicore) seq
+// CHECK-NEXT: #pragma acc routine(function) device_type(multicore) seq
+#pragma acc routine(function) dtype(radeon) seq
+// CHECK-NEXT: #pragma acc routine(function) dtype(radeon) seq
 
 #pragma acc routine nohost vector
 int function2();
@@ -136,8 +136,8 @@ struct DepS {
 
 #pragma acc routine (MemFunc) worker dtype(*)
 // CHECK-NEXT: #pragma acc routine(MemFunc) worker dtype(*)
-#pragma acc routine (MemFunc) device_type(Lambda) vector
-// CHECK-NEXT: #pragma acc routine(MemFunc) device_type(Lambda) vector
+#pragma acc routine (MemFunc) device_type(nvidia) vector
+// CHECK-NEXT: #pragma acc routine(MemFunc) device_type(nvidia) vector
 };
 
 // CHECK: #pragma acc routine(DepS<int>::Lambda) gang bind("string")

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

@@ -3,22 +3,22 @@
 unsigned Int;
 
 void uses() {
-// CHECK: #pragma acc set default_async(Int) if(Int == 5) device_type(I) device_num(Int)
-#pragma acc set default_async(Int) if (Int == 5) device_type(I) device_num(Int)
-// CHECK: #pragma acc set default_async(Int) device_type(I) device_num(Int)
-#pragma acc set default_async(Int) device_type(I) device_num(Int)
+// CHECK: #pragma acc set default_async(Int) if(Int == 5) device_type(default) device_num(Int)
+#pragma acc set default_async(Int) if (Int == 5) device_type(default) device_num(Int)
+// CHECK: #pragma acc set default_async(Int) device_type(nvidia) device_num(Int)
+#pragma acc set default_async(Int) device_type(nvidia) device_num(Int)
 // CHECK: #pragma acc set default_async(Int) if(Int == 5) device_num(Int)
 #pragma acc set default_async(Int) if (Int == 5) device_num(Int)
-// CHECK: #pragma acc set default_async(Int) if(Int == 5) device_type(I)
-#pragma acc set default_async(Int) if (Int == 5) device_type(I)
-// CHECK: #pragma acc set if(Int == 5) device_type(I) device_num(Int)
-#pragma acc set if (Int == 5) device_type(I) device_num(Int)
+// CHECK: #pragma acc set default_async(Int) if(Int == 5) device_type(host)
+#pragma acc set default_async(Int) if (Int == 5) device_type(host)
+// CHECK: #pragma acc set if(Int == 5) device_type(multicore) device_num(Int)
+#pragma acc set if (Int == 5) device_type(multicore) device_num(Int)
 // CHECK: #pragma acc set default_async(Int)
 #pragma acc set default_async(Int)
 // CHECK: #pragma acc set if(Int == 5)
 #pragma acc set if (Int == 5)
-// CHECK: #pragma acc set device_type(I)
-#pragma acc set device_type(I)
+// CHECK: #pragma acc set device_type(radeon)
+#pragma acc set device_type(radeon)
 // CHECK: #pragma acc set device_num(Int)
 #pragma acc set device_num(Int)
 }

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

@@ -9,6 +9,6 @@ void uses() {
 #pragma acc shutdown device_type(*) device_num(Int)
 // CHECK: #pragma acc shutdown device_type(*) if(Int == 5)
 #pragma acc shutdown device_type(*) if (Int == 5)
-// CHECK: #pragma acc shutdown device_type(SomeName)
-#pragma acc shutdown device_type(SomeName)
+// CHECK: #pragma acc shutdown device_type(radeon)
+#pragma acc shutdown device_type(radeon)
 }