[OpenACC] Implement new OpenACC device_type restriction

The OpenACC standard is going to change to clarify that init, shutdown,
    and set should only have a single architecture in each 'device_type'
    clause.  This patch implements that restriction.

    See: https://github.com/OpenACC/openacc-spec/pull/550

clang/include/clang/Basic/DiagnosticSemaKinds.td

@@ -13529,7 +13529,7 @@ def err_acc_invalid_modifier
 def err_acc_invalid_default_type
     : Error<"invalid value %0 in '%1' clause; valid values are %2">;
 def err_acc_device_type_multiple_archs
-    : Error<"OpenACC 'device_type' clause on a 'set' construct only permits "
+    : Error<"OpenACC 'device_type' clause on a '%0' construct only permits "
             "one architecture">;
 def warn_acc_var_referenced_non_const_array
     : Warning<"variable of array type %0 referenced in OpenACC '%1' clause "

clang/lib/Sema/SemaOpenACCClause.cpp

@@ -1054,13 +1054,17 @@ OpenACCClause *SemaOpenACCClauseVisitor::VisitWaitClause(
 OpenACCClause *SemaOpenACCClauseVisitor::VisitDeviceTypeClause(
     SemaOpenACC::OpenACCParsedClause &Clause) {
 
-  // Based on discussions, having more than 1 'architecture' on a 'set' is
-  // nonsensical, so we're going to fix the standard to reflect this.  Implement
-  // the limitation, since the Dialect requires this.
-  if (Clause.getDirectiveKind() == OpenACCDirectiveKind::Set &&
+  // OpenACC Pull #550 (https://github.com/OpenACC/openacc-spec/pull/550)
+  // clarified that Init, Shutdown, and Set only support a single architecture.
+  // Though the dialect only requires it for 'set' as far as we know, we'll just
+  // implement all 3 here.
+  if ((Clause.getDirectiveKind() == OpenACCDirectiveKind::Init ||
+       Clause.getDirectiveKind() == OpenACCDirectiveKind::Shutdown ||
+       Clause.getDirectiveKind() == OpenACCDirectiveKind::Set) &&
       Clause.getDeviceTypeArchitectures().size() > 1) {
     SemaRef.Diag(Clause.getDeviceTypeArchitectures()[1].getLoc(),
-                 diag::err_acc_device_type_multiple_archs);
+                 diag::err_acc_device_type_multiple_archs)
+        << Clause.getDirectiveKind();
     return nullptr;
   }
 

clang/test/CIR/CodeGenOpenACC/init.c

@@ -11,12 +11,8 @@ void acc_init(int cond) {
   // CHECK-NEXT: acc.init attributes {device_types = [#acc.device_type<star>]}
 #pragma acc init device_type(nvidia)
   // CHECK-NEXT: acc.init attributes {device_types = [#acc.device_type<nvidia>]}
-#pragma acc init device_type(host, multicore)
-  // CHECK-NEXT: acc.init attributes {device_types = [#acc.device_type<host>, #acc.device_type<multicore>]}
 #pragma acc init device_type(NVIDIA)
   // CHECK-NEXT: acc.init attributes {device_types = [#acc.device_type<nvidia>]}
-#pragma acc init device_type(HoSt, MuLtIcORe)
-  // CHECK-NEXT: acc.init attributes {device_types = [#acc.device_type<host>, #acc.device_type<multicore>]}
 #pragma acc init device_type(HoSt) device_type(MuLtIcORe)
   // CHECK-NEXT: acc.init attributes {device_types = [#acc.device_type<host>, #acc.device_type<multicore>]}
 

clang/test/CIR/CodeGenOpenACC/shutdown.c

@@ -11,12 +11,8 @@ void acc_shutdown(int cond) {
   // CHECK-NEXT: acc.shutdown attributes {device_types = [#acc.device_type<star>]}
 #pragma acc shutdown device_type(nvidia)
   // CHECK-NEXT: acc.shutdown attributes {device_types = [#acc.device_type<nvidia>]}
-#pragma acc shutdown device_type(host, multicore)
-  // CHECK-NEXT: acc.shutdown attributes {device_types = [#acc.device_type<host>, #acc.device_type<multicore>]}
 #pragma acc shutdown device_type(NVIDIA)
   // CHECK-NEXT: acc.shutdown attributes {device_types = [#acc.device_type<nvidia>]}
-#pragma acc shutdown device_type(HoSt, MuLtIcORe)
-  // CHECK-NEXT: acc.shutdown attributes {device_types = [#acc.device_type<host>, #acc.device_type<multicore>]}
 #pragma acc shutdown device_type(HoSt) device_type(MuLtIcORe)
   // CHECK-NEXT: acc.shutdown attributes {device_types = [#acc.device_type<host>, #acc.device_type<multicore>]}
 

clang/test/SemaOpenACC/init-construct.cpp

@@ -34,6 +34,12 @@ void uses() {
   // expected-error@+2{{OpenACC integer expression requires explicit conversion from 'struct ExplicitConvertOnly' to 'int'}}
   // expected-note@#EXPL_CONV{{conversion to integral type 'int'}}
 #pragma acc init device_num(Explicit)
+
+  // expected-error@+1{{OpenACC 'device_type' clause on a 'init' construct only permits one architecture}}
+#pragma acc init device_type(nvidia, radeon)
+
+  // expected-error@+1{{OpenACC 'device_type' clause on a 'init' construct only permits one architecture}}
+#pragma acc init  device_type(nonsense, nvidia, radeon)
 }
 
 template<typename T>

clang/test/SemaOpenACC/shutdown-construct.cpp

@@ -34,6 +34,12 @@ void uses() {
   // expected-error@+2{{OpenACC integer expression requires explicit conversion from 'struct ExplicitConvertOnly' to 'int'}}
   // expected-note@#EXPL_CONV{{conversion to integral type 'int'}}
 #pragma acc shutdown device_num(Explicit)
+
+  // expected-error@+1{{OpenACC 'device_type' clause on a 'shutdown' construct only permits one architecture}}
+#pragma acc shutdown device_type(nvidia, radeon)
+
+  // expected-error@+1{{OpenACC 'device_type' clause on a 'shutdown' construct only permits one architecture}}
+#pragma acc shutdown device_type(nonsense, nvidia, radeon)
 }
 
 template<typename T>