From d47401e376d458fb9bb4f4f9be0e07c0fdd7a52c Mon Sep 17 00:00:00 2001
From: erichkeane <ekeane@nvidia.com>
Date: Wed, 9 Apr 2025 14:20:15 -0700
Subject: [PATCH] [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.
---
 .../clang/Basic/DiagnosticSemaKinds.td        |  2 +
 clang/lib/Sema/SemaOpenACCClause.cpp          | 45 ++++++++++--
 .../ast-print-openacc-combined-construct.cpp  | 24 +++----
 .../ast-print-openacc-compute-construct.cpp   | 20 +++---
 .../AST/ast-print-openacc-data-construct.cpp  |  8 +--
 .../AST/ast-print-openacc-init-construct.cpp  |  4 +-
 .../AST/ast-print-openacc-loop-construct.cpp  | 16 ++---
 .../ast-print-openacc-routine-construct.cpp   | 12 ++--
 .../AST/ast-print-openacc-set-construct.cpp   | 20 +++---
 .../ast-print-openacc-shutdown-construct.cpp  |  4 +-
 .../ast-print-openacc-update-construct.cpp    |  8 +--
 clang/test/ParserOpenACC/parse-clauses.c      | 16 +++++
 .../combined-construct-device_type-ast.cpp    | 70 +++++++++----------
 .../combined-construct-device_type-clause.c   | 14 ++--
 .../combined-construct-device_type-clause.cpp | 10 +--
 .../compute-construct-device_type-ast.cpp     | 42 +++++------
 .../compute-construct-device_type-clause.c    | 14 ++--
 .../compute-construct-device_type-clause.cpp  | 10 +--
 .../data-construct-device_type-ast.cpp        | 10 +--
 .../data-construct-device_type-clause.c       | 28 ++++----
 .../SemaOpenACC/device_type_valid_values.c    | 56 +++++++++++++++
 clang/test/SemaOpenACC/init-construct-ast.cpp | 20 +++---
 clang/test/SemaOpenACC/init-construct.cpp     | 10 +--
 .../loop-construct-device_type-ast.cpp        | 70 +++++++++----------
 .../loop-construct-device_type-clause.c       | 12 ++--
 .../loop-construct-device_type-clause.cpp     | 10 +--
 .../SemaOpenACC/routine-construct-ast.cpp     | 22 +++---
 .../SemaOpenACC/routine-construct-clauses.cpp |  4 +-
 clang/test/SemaOpenACC/set-construct-ast.cpp  | 10 +--
 clang/test/SemaOpenACC/set-construct.cpp      | 10 +--
 .../SemaOpenACC/shutdown-construct-ast.cpp    | 20 +++---
 clang/test/SemaOpenACC/shutdown-construct.cpp | 10 +--
 .../test/SemaOpenACC/update-construct-ast.cpp | 16 ++---
 clang/test/SemaOpenACC/update-construct.cpp   | 22 +++---
 34 files changed, 388 insertions(+), 281 deletions(-)
 create mode 100644 clang/test/SemaOpenACC/device_type_valid_values.c

diff --git a/clang/include/clang/Basic/DiagnosticSemaKinds.td b/clang/include/clang/Basic/DiagnosticSemaKinds.td
index d17519d4c415..fdf3f8816e74 100644
--- a/clang/include/clang/Basic/DiagnosticSemaKinds.td
+++ b/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">;
diff --git a/clang/lib/Sema/SemaOpenACCClause.cpp b/clang/lib/Sema/SemaOpenACCClause.cpp
index 7d10c50d404d..8aa0c52f0ce1 100644
--- a/clang/lib/Sema/SemaOpenACCClause.cpp
+++ b/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(
diff --git a/clang/test/AST/ast-print-openacc-combined-construct.cpp b/clang/test/AST/ast-print-openacc-combined-construct.cpp
index be959fb2b611..7331ec3786e9 100644
--- a/clang/test/AST/ast-print-openacc-combined-construct.cpp
+++ b/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;
diff --git a/clang/test/AST/ast-print-openacc-compute-construct.cpp b/clang/test/AST/ast-print-openacc-compute-construct.cpp
index 1fbb81a220aa..7c3ac17ec1a2 100644
--- a/clang/test/AST/ast-print-openacc-compute-construct.cpp
+++ b/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)
diff --git a/clang/test/AST/ast-print-openacc-data-construct.cpp b/clang/test/AST/ast-print-openacc-data-construct.cpp
index b7d2428fb605..98c785032b77 100644
--- a/clang/test/AST/ast-print-openacc-data-construct.cpp
+++ b/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)
diff --git a/clang/test/AST/ast-print-openacc-init-construct.cpp b/clang/test/AST/ast-print-openacc-init-construct.cpp
index 0cfcbc99ab14..e77d54494419 100644
--- a/clang/test/AST/ast-print-openacc-init-construct.cpp
+++ b/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)
 }
diff --git a/clang/test/AST/ast-print-openacc-loop-construct.cpp b/clang/test/AST/ast-print-openacc-loop-construct.cpp
index 9ddd5c81ca53..6971089e5919 100644
--- a/clang/test/AST/ast-print-openacc-loop-construct.cpp
+++ b/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
diff --git a/clang/test/AST/ast-print-openacc-routine-construct.cpp b/clang/test/AST/ast-print-openacc-routine-construct.cpp
index b3de3402a952..be8d95387d2c 100644
--- a/clang/test/AST/ast-print-openacc-routine-construct.cpp
+++ b/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")
diff --git a/clang/test/AST/ast-print-openacc-set-construct.cpp b/clang/test/AST/ast-print-openacc-set-construct.cpp
index a801ae16739e..869a3feb2dd9 100644
--- a/clang/test/AST/ast-print-openacc-set-construct.cpp
+++ b/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)
 }
diff --git a/clang/test/AST/ast-print-openacc-shutdown-construct.cpp b/clang/test/AST/ast-print-openacc-shutdown-construct.cpp
index 4e2529658d41..6f6142254a44 100644
--- a/clang/test/AST/ast-print-openacc-shutdown-construct.cpp
+++ b/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)
 }
diff --git a/clang/test/AST/ast-print-openacc-update-construct.cpp b/clang/test/AST/ast-print-openacc-update-construct.cpp
index f999e1037e90..289754f181d9 100644
--- a/clang/test/AST/ast-print-openacc-update-construct.cpp
+++ b/clang/test/AST/ast-print-openacc-update-construct.cpp
@@ -30,11 +30,11 @@ void uses(bool cond) {
 // CHECK: #pragma acc update self(I) wait(devnum: I : queues: *iPtr, I) if(I == array[I]) async(I)
 #pragma acc update self(I) wait(devnum:I:queues:*iPtr, I) if(I == array[I]) async(I)
 
-// CHECK: #pragma acc update self(I) device_type(I) dtype(H)
-#pragma acc update self(I) device_type(I) dtype(H)
+// CHECK: #pragma acc update self(I) device_type(nvidia) dtype(host)
+#pragma acc update self(I) device_type(nvidia) dtype(host)
 
-// CHECK: #pragma acc update self(I) device_type(J) dtype(K)
-#pragma acc update self(I) device_type(J) dtype(K)
+// CHECK: #pragma acc update self(I) device_type(radeon) dtype(multicore)
+#pragma acc update self(I) device_type(radeon) dtype(multicore)
 
 // CHECK: #pragma acc update self(I, iPtr, array, array[1], array[1:2])
 #pragma acc update self(I, iPtr, array, array[1], array[1:2])
diff --git a/clang/test/ParserOpenACC/parse-clauses.c b/clang/test/ParserOpenACC/parse-clauses.c
index 2319b1abb9e8..ef0bccf39029 100644
--- a/clang/test/ParserOpenACC/parse-clauses.c
+++ b/clang/test/ParserOpenACC/parse-clauses.c
@@ -1080,11 +1080,27 @@ void device_type() {
 #pragma acc parallel dtype(31, "bar")
   {}
 
+  // expected-error@+4{{invalid value 'ident' in 'device_type' clause; valid values are 'default', 'nvidia', 'acc_device_nvidia', 'radeon', 'host', 'multicore'}}
+  // expected-error@+3{{invalid value 'auto' in 'device_type' clause; valid values are 'default', 'nvidia', 'acc_device_nvidia', 'radeon', 'host', 'multicore'}}
+  // expected-error@+2{{invalid value 'int' in 'device_type' clause; valid values are 'default', 'nvidia', 'acc_device_nvidia', 'radeon', 'host', 'multicore'}}
+  // expected-error@+1{{invalid value 'float' in 'device_type' clause; valid values are 'default', 'nvidia', 'acc_device_nvidia', 'radeon', 'host', 'multicore'}}
 #pragma acc parallel device_type(ident, auto, int, float)
   {}
+  // expected-error@+4{{invalid value 'ident' in 'dtype' clause; valid values are 'default', 'nvidia', 'acc_device_nvidia', 'radeon', 'host', 'multicore'}}
+  // expected-error@+3{{invalid value 'auto' in 'dtype' clause; valid values are 'default', 'nvidia', 'acc_device_nvidia', 'radeon', 'host', 'multicore'}}
+  // expected-error@+2{{invalid value 'int' in 'dtype' clause; valid values are 'default', 'nvidia', 'acc_device_nvidia', 'radeon', 'host', 'multicore'}}
+  // expected-error@+1{{invalid value 'float' in 'dtype' clause; valid values are 'default', 'nvidia', 'acc_device_nvidia', 'radeon', 'host', 'multicore'}}
 #pragma acc parallel dtype(ident, auto, int, float)
   {}
 
+  // expected-error@+8{{invalid value 'ident' in 'device_type' clause; valid values are 'default', 'nvidia', 'acc_device_nvidia', 'radeon', 'host', 'multicore'}}
+  // expected-error@+7{{invalid value 'auto' in 'device_type' clause; valid values are 'default', 'nvidia', 'acc_device_nvidia', 'radeon', 'host', 'multicore'}}
+  // expected-error@+6{{invalid value 'int' in 'device_type' clause; valid values are 'default', 'nvidia', 'acc_device_nvidia', 'radeon', 'host', 'multicore'}}
+  // expected-error@+5{{invalid value 'float' in 'device_type' clause; valid values are 'default', 'nvidia', 'acc_device_nvidia', 'radeon', 'host', 'multicore'}}
+  // expected-error@+4{{invalid value 'ident' in 'dtype' clause; valid values are 'default', 'nvidia', 'acc_device_nvidia', 'radeon', 'host', 'multicore'}}
+  // expected-error@+3{{invalid value 'auto' in 'dtype' clause; valid values are 'default', 'nvidia', 'acc_device_nvidia', 'radeon', 'host', 'multicore'}}
+  // expected-error@+2{{invalid value 'int' in 'dtype' clause; valid values are 'default', 'nvidia', 'acc_device_nvidia', 'radeon', 'host', 'multicore'}}
+  // expected-error@+1{{invalid value 'float' in 'dtype' clause; valid values are 'default', 'nvidia', 'acc_device_nvidia', 'radeon', 'host', 'multicore'}}
 #pragma acc parallel device_type(ident, auto, int, float) dtype(ident, auto, int, float)
   {}
 }
diff --git a/clang/test/SemaOpenACC/combined-construct-device_type-ast.cpp b/clang/test/SemaOpenACC/combined-construct-device_type-ast.cpp
index abc65dfc9e52..5994f09f2e69 100644
--- a/clang/test/SemaOpenACC/combined-construct-device_type-ast.cpp
+++ b/clang/test/SemaOpenACC/combined-construct-device_type-ast.cpp
@@ -19,11 +19,11 @@ void NormalUses() {
   // CHECK-NEXT: DeclStmt
   // CHECK-NEXT: VarDecl{{.*}} SomeVar 'bool'
 
-#pragma acc parallel loop device_type(SomeS) dtype(SomeImpl)
+#pragma acc parallel loop device_type(*) dtype(default)
   for(int i = 0; i < 5; ++i){}
   // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} parallel loop
-  // CHECK-NEXT: device_type(SomeS)
-  // CHECK-NEXT: dtype(SomeImpl)
+  // CHECK-NEXT: device_type(*)
+  // CHECK-NEXT: dtype(default)
   // CHECK-NEXT: ForStmt
   // CHECK-NEXT: DeclStmt
   // CHECK-NEXT: VarDecl{{.*}} i 'int'
@@ -36,11 +36,11 @@ void NormalUses() {
   // CHECK-NEXT: UnaryOperator{{.*}}++
   // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
   // CHECK-NEXT: CompoundStmt
-#pragma acc serial loop device_type(SomeVar) dtype(int)
+#pragma acc serial loop device_type(nvidia) dtype(acc_device_nvidia)
   for(int i = 0; i < 5; ++i){}
   // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} serial loop
-  // CHECK-NEXT: device_type(SomeVar)
-  // CHECK-NEXT: dtype(int)
+  // CHECK-NEXT: device_type(nvidia)
+  // CHECK-NEXT: dtype(acc_device_nvidia)
   // CHECK-NEXT: ForStmt
   // CHECK-NEXT: DeclStmt
   // CHECK-NEXT: VarDecl{{.*}} i 'int'
@@ -53,11 +53,11 @@ void NormalUses() {
   // CHECK-NEXT: UnaryOperator{{.*}}++
   // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
   // CHECK-NEXT: CompoundStmt
-#pragma acc kernels loop device_type(private) dtype(struct)
+#pragma acc kernels loop device_type(radeon) dtype(host)
   for(int i = 0; i < 5; ++i){}
   // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} kernels loop
-  // CHECK-NEXT: device_type(private)
-  // CHECK-NEXT: dtype(struct)
+  // CHECK-NEXT: device_type(radeon)
+  // CHECK-NEXT: dtype(host)
   // CHECK-NEXT: ForStmt
   // CHECK-NEXT: DeclStmt
   // CHECK-NEXT: VarDecl{{.*}} i 'int'
@@ -70,27 +70,10 @@ void NormalUses() {
   // CHECK-NEXT: UnaryOperator{{.*}}++
   // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
   // CHECK-NEXT: CompoundStmt
-#pragma acc parallel loop device_type(private) dtype(class)
+#pragma acc parallel loop device_type(multicore) dtype(*)
   for(int i = 0; i < 5; ++i){}
   // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} parallel loop
-  // CHECK-NEXT: device_type(private)
-  // CHECK-NEXT: dtype(class)
-  // CHECK-NEXT: ForStmt
-  // CHECK-NEXT: DeclStmt
-  // CHECK-NEXT: VarDecl{{.*}} i 'int'
-  // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
-  // CHECK-NEXT: <<<NULL>>>
-  // CHECK-NEXT: BinaryOperator{{.*}}'<'
-  // CHECK-NEXT: ImplicitCastExpr
-  // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
-  // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 5
-  // CHECK-NEXT: UnaryOperator{{.*}}++
-  // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
-  // CHECK-NEXT: CompoundStmt
-#pragma acc serial loop device_type(float) dtype(*)
-  for(int i = 0; i < 5; ++i){}
-  // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} serial loop
-  // CHECK-NEXT: device_type(float)
+  // CHECK-NEXT: device_type(multicore)
   // CHECK-NEXT: dtype(*)
   // CHECK-NEXT: ForStmt
   // CHECK-NEXT: DeclStmt
@@ -104,10 +87,27 @@ void NormalUses() {
   // CHECK-NEXT: UnaryOperator{{.*}}++
   // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
   // CHECK-NEXT: CompoundStmt
-#pragma acc kernels loop device_type(float, int) dtype(*)
+#pragma acc serial loop device_type(default, nvidia) dtype(*)
+  for(int i = 0; i < 5; ++i){}
+  // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} serial loop
+  // CHECK-NEXT: device_type(default, nvidia)
+  // CHECK-NEXT: dtype(*)
+  // CHECK-NEXT: ForStmt
+  // CHECK-NEXT: DeclStmt
+  // CHECK-NEXT: VarDecl{{.*}} i 'int'
+  // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+  // CHECK-NEXT: <<<NULL>>>
+  // CHECK-NEXT: BinaryOperator{{.*}}'<'
+  // CHECK-NEXT: ImplicitCastExpr
+  // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+  // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 5
+  // CHECK-NEXT: UnaryOperator{{.*}}++
+  // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+  // CHECK-NEXT: CompoundStmt
+#pragma acc kernels loop device_type(acc_device_nvidia, radeon) dtype(*)
   for(int i = 0; i < 5; ++i){}
   // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} kernels loop
-  // CHECK-NEXT: device_type(float, int)
+  // CHECK-NEXT: device_type(acc_device_nvidia, radeon)
   // CHECK-NEXT: dtype(*)
   // CHECK-NEXT: ForStmt
   // CHECK-NEXT: DeclStmt
@@ -129,11 +129,11 @@ void TemplUses() {
   // CHECK-NEXT: TemplateTypeParmDecl{{.*}}T
   // CHECK-NEXT: FunctionDecl{{.*}}TemplUses
   // CHECK-NEXT: CompoundStmt
-#pragma acc parallel loop device_type(T) dtype(T)
+#pragma acc parallel loop device_type(host, multicore) dtype(*)
   for(int i = 0; i < 5; ++i){}
   // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} parallel loop
-  // CHECK-NEXT: device_type(T)
-  // CHECK-NEXT: dtype(T)
+  // CHECK-NEXT: device_type(host, multicore)
+  // CHECK-NEXT: dtype(*)
   // CHECK-NEXT: ForStmt
   // CHECK-NEXT: DeclStmt
   // CHECK-NEXT: VarDecl{{.*}} i 'int'
@@ -155,8 +155,8 @@ void TemplUses() {
   // CHECK-NEXT: CompoundStmt
 
   // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} parallel loop
-  // CHECK-NEXT: device_type(T)
-  // CHECK-NEXT: dtype(T)
+  // CHECK-NEXT: device_type(host, multicore)
+  // CHECK-NEXT: dtype(*)
   // CHECK-NEXT: ForStmt
   // CHECK-NEXT: DeclStmt
   // CHECK-NEXT: VarDecl{{.*}} i 'int'
diff --git a/clang/test/SemaOpenACC/combined-construct-device_type-clause.c b/clang/test/SemaOpenACC/combined-construct-device_type-clause.c
index 38c72551997b..92174b2d55cf 100644
--- a/clang/test/SemaOpenACC/combined-construct-device_type-clause.c
+++ b/clang/test/SemaOpenACC/combined-construct-device_type-clause.c
@@ -6,22 +6,22 @@ void uses() {
   typedef struct S{} STy;
   STy SImpl;
 
-#pragma acc parallel loop device_type(I)
+#pragma acc parallel loop device_type(radeon)
   for(int i = 0; i < 5; ++i);
-#pragma acc serial loop device_type(S) dtype(STy)
+#pragma acc serial loop device_type(multicore) dtype(host)
   for(int i = 0; i < 5; ++i);
-#pragma acc kernels loop dtype(SImpl)
+#pragma acc kernels loop dtype(acc_device_nvidia)
   for(int i = 0; i < 5; ++i);
-#pragma acc kernels loop dtype(int) device_type(*)
+#pragma acc kernels loop dtype(default) device_type(*)
   for(int i = 0; i < 5; ++i);
-#pragma acc kernels loop dtype(true) device_type(false)
+#pragma acc kernels loop dtype(nvidia) device_type(radeon)
   for(int i = 0; i < 5; ++i);
 
   // expected-error@+1{{expected identifier}}
-#pragma acc kernels loop dtype(int, *)
+#pragma acc kernels loop dtype(radeon, *)
   for(int i = 0; i < 5; ++i);
 
-#pragma acc parallel loop device_type(I, int)
+#pragma acc parallel loop device_type(host, multicore)
   for(int i = 0; i < 5; ++i);
   // expected-error@+2{{expected ','}}
   // expected-error@+1{{expected identifier}}
diff --git a/clang/test/SemaOpenACC/combined-construct-device_type-clause.cpp b/clang/test/SemaOpenACC/combined-construct-device_type-clause.cpp
index b46709aa8d4c..c0ed0f90fbe9 100644
--- a/clang/test/SemaOpenACC/combined-construct-device_type-clause.cpp
+++ b/clang/test/SemaOpenACC/combined-construct-device_type-clause.cpp
@@ -2,17 +2,17 @@
 
 template<typename T>
 void TemplUses() {
-#pragma acc parallel loop device_type(I)
+#pragma acc parallel loop device_type(default)
   for(int i = 0; i < 5; ++i);
 #pragma acc serial loop dtype(*)
   for(int i = 0; i < 5; ++i);
-#pragma acc kernels loop device_type(class)
+#pragma acc kernels loop device_type(nvidia)
   for(int i = 0; i < 5; ++i);
-#pragma acc parallel loop device_type(private)
+#pragma acc parallel loop device_type(radeon)
   for(int i = 0; i < 5; ++i);
-#pragma acc serial loop device_type(bool)
+#pragma acc serial loop device_type(host)
   for(int i = 0; i < 5; ++i);
-#pragma acc kernels loop dtype(true) device_type(false)
+#pragma acc kernels loop dtype(multicore) device_type(host)
   for(int i = 0; i < 5; ++i);
   // expected-error@+2{{expected ','}}
   // expected-error@+1{{expected identifier}}
diff --git a/clang/test/SemaOpenACC/compute-construct-device_type-ast.cpp b/clang/test/SemaOpenACC/compute-construct-device_type-ast.cpp
index 8a2423f4f542..97bb0a1ee2a1 100644
--- a/clang/test/SemaOpenACC/compute-construct-device_type-ast.cpp
+++ b/clang/test/SemaOpenACC/compute-construct-device_type-ast.cpp
@@ -19,50 +19,50 @@ void NormalUses() {
   // CHECK-NEXT: DeclStmt
   // CHECK-NEXT: VarDecl{{.*}} SomeVar 'bool'
 
-#pragma acc parallel device_type(SomeS) dtype(SomeImpl)
+#pragma acc parallel device_type(default) dtype(nvidia)
   while(true){}
   // CHECK-NEXT: OpenACCComputeConstruct{{.*}}parallel
-  // CHECK-NEXT: device_type(SomeS)
-  // CHECK-NEXT: dtype(SomeImpl)
+  // CHECK-NEXT: device_type(default)
+  // CHECK-NEXT: dtype(nvidia)
   // CHECK-NEXT: WhileStmt
   // CHECK-NEXT: CXXBoolLiteralExpr
   // CHECK-NEXT: CompoundStmt
-#pragma acc parallel device_type(SomeVar) dtype(int)
+#pragma acc parallel device_type(radeon) dtype(host)
   while(true){}
   // CHECK-NEXT: OpenACCComputeConstruct{{.*}}parallel
-  // CHECK-NEXT: device_type(SomeVar)
-  // CHECK-NEXT: dtype(int)
+  // CHECK-NEXT: device_type(radeon)
+  // CHECK-NEXT: dtype(host)
   // CHECK-NEXT: WhileStmt
   // CHECK-NEXT: CXXBoolLiteralExpr
   // CHECK-NEXT: CompoundStmt
-#pragma acc parallel device_type(private) dtype(struct)
+#pragma acc parallel device_type(multicore) dtype(default)
   while(true){}
   // CHECK-NEXT: OpenACCComputeConstruct{{.*}}parallel
-  // CHECK-NEXT: device_type(private)
-  // CHECK-NEXT: dtype(struct)
+  // CHECK-NEXT: device_type(multicore)
+  // CHECK-NEXT: dtype(default)
   // CHECK-NEXT: WhileStmt
   // CHECK-NEXT: CXXBoolLiteralExpr
   // CHECK-NEXT: CompoundStmt
-#pragma acc parallel device_type(private) dtype(class)
+#pragma acc parallel device_type(nvidia) dtype(acc_device_nvidia)
   while(true){}
   // CHECK-NEXT: OpenACCComputeConstruct{{.*}}parallel
-  // CHECK-NEXT: device_type(private)
-  // CHECK-NEXT: dtype(class)
+  // CHECK-NEXT: device_type(nvidia)
+  // CHECK-NEXT: dtype(acc_device_nvidia)
   // CHECK-NEXT: WhileStmt
   // CHECK-NEXT: CXXBoolLiteralExpr
   // CHECK-NEXT: CompoundStmt
-#pragma acc parallel device_type(float) dtype(*)
+#pragma acc parallel device_type(radeon) dtype(*)
   while(true){}
   // CHECK-NEXT: OpenACCComputeConstruct{{.*}}parallel
-  // CHECK-NEXT: device_type(float)
+  // CHECK-NEXT: device_type(radeon)
   // CHECK-NEXT: dtype(*)
   // CHECK-NEXT: WhileStmt
   // CHECK-NEXT: CXXBoolLiteralExpr
   // CHECK-NEXT: CompoundStmt
-#pragma acc parallel device_type(float, int) dtype(*)
+#pragma acc parallel device_type(host, multicore) dtype(*)
   while(true){}
   // CHECK-NEXT: OpenACCComputeConstruct{{.*}}parallel
-  // CHECK-NEXT: device_type(float, int)
+  // CHECK-NEXT: device_type(host, multicore)
   // CHECK-NEXT: dtype(*)
   // CHECK-NEXT: WhileStmt
   // CHECK-NEXT: CXXBoolLiteralExpr
@@ -75,11 +75,11 @@ void TemplUses() {
   // CHECK-NEXT: TemplateTypeParmDecl{{.*}}T
   // CHECK-NEXT: FunctionDecl{{.*}}TemplUses
   // CHECK-NEXT: CompoundStmt
-#pragma acc parallel device_type(T) dtype(T)
+#pragma acc parallel device_type(host) dtype(multicore)
   while(true){}
   // CHECK-NEXT: OpenACCComputeConstruct{{.*}}parallel
-  // CHECK-NEXT: device_type(T)
-  // CHECK-NEXT: dtype(T)
+  // CHECK-NEXT: device_type(host)
+  // CHECK-NEXT: dtype(multicore)
   // CHECK-NEXT: WhileStmt
   // CHECK-NEXT: CXXBoolLiteralExpr
   // CHECK-NEXT: CompoundStmt
@@ -92,8 +92,8 @@ void TemplUses() {
   // CHECK-NEXT: CompoundStmt
 
   // CHECK-NEXT: OpenACCComputeConstruct{{.*}}parallel
-  // CHECK-NEXT: device_type(T)
-  // CHECK-NEXT: dtype(T)
+  // CHECK-NEXT: device_type(host)
+  // CHECK-NEXT: dtype(multicore)
   // CHECK-NEXT: WhileStmt
   // CHECK-NEXT: CXXBoolLiteralExpr
   // CHECK-NEXT: CompoundStmt
diff --git a/clang/test/SemaOpenACC/compute-construct-device_type-clause.c b/clang/test/SemaOpenACC/compute-construct-device_type-clause.c
index 1e74e5ac2337..16caa7205721 100644
--- a/clang/test/SemaOpenACC/compute-construct-device_type-clause.c
+++ b/clang/test/SemaOpenACC/compute-construct-device_type-clause.c
@@ -6,22 +6,22 @@ void uses() {
   typedef struct S{} STy;
   STy SImpl;
 
-#pragma acc parallel device_type(I)
+#pragma acc parallel device_type(radeon)
   while(1);
-#pragma acc serial device_type(S) dtype(STy)
+#pragma acc serial device_type(nvidia) dtype(multicore)
   while(1);
-#pragma acc kernels dtype(SImpl)
+#pragma acc kernels dtype(host)
   while(1);
-#pragma acc kernels dtype(int) device_type(*)
+#pragma acc kernels dtype(default) device_type(*)
   while(1);
-#pragma acc kernels dtype(true) device_type(false)
+#pragma acc kernels dtype(nvidia) device_type(multicore)
   while(1);
 
   // expected-error@+1{{expected identifier}}
-#pragma acc kernels dtype(int, *)
+#pragma acc kernels dtype(default, *)
   while(1);
 
-#pragma acc parallel device_type(I, int)
+#pragma acc parallel device_type(host, radeon)
   while(1);
   // expected-error@+2{{expected ','}}
   // expected-error@+1{{expected identifier}}
diff --git a/clang/test/SemaOpenACC/compute-construct-device_type-clause.cpp b/clang/test/SemaOpenACC/compute-construct-device_type-clause.cpp
index ed40e8bbceae..1b235fd80667 100644
--- a/clang/test/SemaOpenACC/compute-construct-device_type-clause.cpp
+++ b/clang/test/SemaOpenACC/compute-construct-device_type-clause.cpp
@@ -2,17 +2,17 @@
 
 template<typename T>
 void TemplUses() {
-#pragma acc parallel device_type(I)
+#pragma acc parallel device_type(default)
   while(true);
 #pragma acc parallel dtype(*)
   while(true);
-#pragma acc parallel device_type(class)
+#pragma acc parallel device_type(nvidia)
   while(true);
-#pragma acc parallel device_type(private)
+#pragma acc parallel device_type(radeon)
   while(true);
-#pragma acc parallel device_type(bool)
+#pragma acc parallel device_type(host)
   while(true);
-#pragma acc kernels dtype(true) device_type(false)
+#pragma acc kernels dtype(multicore) device_type(host)
   while(true);
   // expected-error@+2{{expected ','}}
   // expected-error@+1{{expected identifier}}
diff --git a/clang/test/SemaOpenACC/data-construct-device_type-ast.cpp b/clang/test/SemaOpenACC/data-construct-device_type-ast.cpp
index 1f497bb7afcc..e558785fe38a 100644
--- a/clang/test/SemaOpenACC/data-construct-device_type-ast.cpp
+++ b/clang/test/SemaOpenACC/data-construct-device_type-ast.cpp
@@ -13,12 +13,12 @@ void TemplUses() {
   // CHECK-NEXT: FunctionDecl{{.*}}TemplUses
   // CHECK-NEXT: CompoundStmt
 
-#pragma acc data default(none) device_type(T) dtype(T)
+#pragma acc data default(none) device_type(host) dtype(multicore)
   ;
   // CHECK-NEXT: OpenACCDataConstruct{{.*}} data
   // CHECK-NEXT: default(none)
-  // CHECK-NEXT: device_type(T)
-  // CHECK-NEXT: dtype(T)
+  // CHECK-NEXT: device_type(host)
+  // CHECK-NEXT: dtype(multicore)
   // CHECK-NEXT: NullStmt
 
   // Instantiations
@@ -30,8 +30,8 @@ void TemplUses() {
   // Argument to 'device-type' is just an identifier, so we don't transform it.
   // CHECK-NEXT: OpenACCDataConstruct{{.*}} data
   // CHECK-NEXT: default(none)
-  // CHECK-NEXT: device_type(T)
-  // CHECK-NEXT: dtype(T)
+  // CHECK-NEXT: device_type(host)
+  // CHECK-NEXT: dtype(multicore)
   // CHECK-NEXT: NullStmt
 }
 void Inst() {
diff --git a/clang/test/SemaOpenACC/data-construct-device_type-clause.c b/clang/test/SemaOpenACC/data-construct-device_type-clause.c
index a467287b93e0..e450e320fa53 100644
--- a/clang/test/SemaOpenACC/data-construct-device_type-clause.c
+++ b/clang/test/SemaOpenACC/data-construct-device_type-clause.c
@@ -2,54 +2,54 @@
 
 void uses() {
   int Var;
-#pragma acc data default(none) device_type(foo) async
+#pragma acc data default(none) device_type(radeon) async
   ;
-#pragma acc data default(none) device_type(foo) wait
+#pragma acc data default(none) device_type(radeon) wait
   ;
-#pragma acc data default(none) device_type(foo) dtype(false)
+#pragma acc data default(none) device_type(radeon) dtype(nvidia)
   ;
-#pragma acc data default(none) dtype(foo) device_type(false)
+#pragma acc data default(none) dtype(radeon) device_type(nvidia)
   ;
 
   // expected-error@+2{{OpenACC clause 'if' may not follow a 'device_type' clause in a 'data' construct}}
   // expected-note@+1{{previous clause is here}}
-#pragma acc data default(none) device_type(foo) if(1)
+#pragma acc data default(none) device_type(radeon) if(1)
   ;
   // expected-error@+2{{OpenACC clause 'copy' may not follow a 'device_type' clause in a 'data' construct}}
   // expected-note@+1{{previous clause is here}}
-#pragma acc data default(none) device_type(foo) copy(Var)
+#pragma acc data default(none) device_type(radeon) copy(Var)
   ;
   // expected-error@+2{{OpenACC clause 'copyin' may not follow a 'device_type' clause in a 'data' construct}}
   // expected-note@+1{{previous clause is here}}
-#pragma acc data default(none) device_type(foo) copyin(Var)
+#pragma acc data default(none) device_type(radeon) copyin(Var)
   ;
   // expected-error@+2{{OpenACC clause 'copyout' may not follow a 'device_type' clause in a 'data' construct}}
   // expected-note@+1{{previous clause is here}}
-#pragma acc data default(none) device_type(foo) copyout(Var)
+#pragma acc data default(none) device_type(radeon) copyout(Var)
   ;
   // expected-error@+2{{OpenACC clause 'create' may not follow a 'device_type' clause in a 'data' construct}}
   // expected-note@+1{{previous clause is here}}
-#pragma acc data default(none) device_type(foo) create(Var)
+#pragma acc data default(none) device_type(radeon) create(Var)
   ;
   // expected-error@+2{{OpenACC clause 'no_create' may not follow a 'device_type' clause in a 'data' construct}}
   // expected-note@+1{{previous clause is here}}
-#pragma acc data default(none) device_type(foo) no_create(Var)
+#pragma acc data default(none) device_type(radeon) no_create(Var)
   ;
   // expected-error@+2{{OpenACC clause 'present' may not follow a 'device_type' clause in a 'data' construct}}
   // expected-note@+1{{previous clause is here}}
-#pragma acc data default(none) device_type(foo) present(Var)
+#pragma acc data default(none) device_type(radeon) present(Var)
   ;
   // expected-error@+2{{OpenACC clause 'deviceptr' may not follow a 'device_type' clause in a 'data' construct}}
   // expected-note@+1{{previous clause is here}}
-#pragma acc data default(none) device_type(foo) deviceptr(Var)
+#pragma acc data default(none) device_type(radeon) deviceptr(Var)
   ;
   // expected-error@+2{{OpenACC clause 'attach' may not follow a 'device_type' clause in a 'data' construct}}
   // expected-note@+1{{previous clause is here}}
-#pragma acc data default(none) device_type(foo) attach(Var)
+#pragma acc data default(none) device_type(radeon) attach(Var)
   ;
   // expected-error@+3{{OpenACC 'data' construct must have at least one 'copy', 'copyin', 'copyout', 'create', 'no_create', 'present', 'deviceptr', 'attach' or 'default' clause}}
   // expected-error@+2{{OpenACC clause 'default' may not follow a 'device_type' clause in a 'data' construct}}
   // expected-note@+1{{previous clause is here}}
-#pragma acc data device_type(foo) default(none)
+#pragma acc data device_type(radeon) default(none)
   ;
 }
diff --git a/clang/test/SemaOpenACC/device_type_valid_values.c b/clang/test/SemaOpenACC/device_type_valid_values.c
new file mode 100644
index 000000000000..7cbe4a67fa46
--- /dev/null
+++ b/clang/test/SemaOpenACC/device_type_valid_values.c
@@ -0,0 +1,56 @@
+// RUN: %clang_cc1 %s -fopenacc -verify 
+
+void foo(int i) {
+#pragma acc parallel device_type(*)
+  ;
+#pragma acc serial device_type(default)
+  ;
+#pragma acc kernels device_type(Default)
+  ;
+#pragma acc data copy(i) device_type(nvidia)
+  ;
+#pragma acc init device_type(acc_device_nvidia)
+  ;
+#pragma acc shutdown device_type(radeon)
+  ;
+#pragma acc set device_type(host)
+  ;
+#pragma acc update self(i) device_type(multicore)
+  ;
+#pragma acc loop device_type(DEFAULT)
+  for(unsigned I = 0; I < 5; ++I);
+#pragma acc parallel loop device_type(HOST)
+  for(unsigned I = 0; I < 5; ++I);
+#pragma acc serial loop device_type(MuLtIcOrE)
+  for(unsigned I = 0; I < 5; ++I);
+#pragma acc kernels loop device_type(radEon)
+  for(unsigned I = 0; I < 5; ++I);
+}
+
+#pragma acc routine(foo) seq device_type(host)
+
+#pragma acc routine seq device_type(acc_device_nvidia)
+void bar(){}
+
+// Anything outside of the above list is invalid.
+void invalid() {
+  // expected-error@+1{{invalid value 'other' in 'device_type' clause; valid values are 'default', 'nvidia', 'acc_device_nvidia', 'radeon', 'host', 'multicore'}}
+#pragma acc parallel device_type(other)
+  ;
+  // expected-error@+1{{invalid value 'invalid' in 'device_type' clause; valid values are 'default', 'nvidia', 'acc_device_nvidia', 'radeon', 'host', 'multicore'}}
+#pragma acc kernels device_type(invalid)
+  ;
+  // expected-error@+1{{invalid value 'invalid' in 'device_type' clause; valid values are 'default', 'nvidia', 'acc_device_nvidia', 'radeon', 'host', 'multicore'}}
+#pragma acc kernels device_type(invalid, nvidia)
+  ;
+  // expected-error@+1{{invalid value 'invalid' in 'device_type' clause; valid values are 'default', 'nvidia', 'acc_device_nvidia', 'radeon', 'host', 'multicore'}}
+#pragma acc kernels device_type(nvidia, invalid)
+  ;
+  // expected-error@+1{{invalid value 'invalid' in 'dtype' clause; valid values are 'default', 'nvidia', 'acc_device_nvidia', 'radeon', 'host', 'multicore'}}
+#pragma acc kernels dtype(nvidia, invalid, radeon)
+  ;
+  // expected-error@+2{{invalid value 'invalid' in 'device_type' clause; valid values are 'default', 'nvidia', 'acc_device_nvidia', 'radeon', 'host', 'multicore'}}
+  // expected-error@+1{{invalid value 'invalid2' in 'device_type' clause; valid values are 'default', 'nvidia', 'acc_device_nvidia', 'radeon', 'host', 'multicore'}}
+#pragma acc kernels device_type(nvidia, invalid, radeon, invalid2)
+  ;
+}
diff --git a/clang/test/SemaOpenACC/init-construct-ast.cpp b/clang/test/SemaOpenACC/init-construct-ast.cpp
index ba0ca1ec3632..64ef8b142894 100644
--- a/clang/test/SemaOpenACC/init-construct-ast.cpp
+++ b/clang/test/SemaOpenACC/init-construct-ast.cpp
@@ -34,10 +34,10 @@ _Pragma("acc init")
   // CHECK-NEXT: CallExpr
   // CHECK-NEXT: ImplicitCastExpr
   // CHECK-NEXT: DeclRefExpr{{.*}}'some_int' 'int ()'
-#pragma acc init device_type(T)
+#pragma acc init device_type(default)
   // CHECK-NEXT: OpenACCInitConstruct{{.*}}init
-  // CHECK-NEXT: device_type(T)
-#pragma acc init if (some_int() < some_long()) device_type(T) device_num(some_int())
+  // CHECK-NEXT: device_type(default)
+#pragma acc init if (some_int() < some_long()) device_type(nvidia) device_num(some_int())
   // CHECK-NEXT: OpenACCInitConstruct{{.*}}init
   // CHECK-NEXT: if clause
   // CHECK-NEXT: BinaryOperator{{.*}} 'bool' '<'
@@ -48,7 +48,7 @@ _Pragma("acc init")
   // CHECK-NEXT: CallExpr
   // CHECK-NEXT: ImplicitCastExpr
   // CHECK-NEXT: DeclRefExpr{{.*}}'some_long' 'long ()'
-  // CHECK-NEXT: device_type(T)
+  // CHECK-NEXT: device_type(nvidia)
   // CHECK-NEXT: device_num clause
   // CHECK-NEXT: CallExpr
   // CHECK-NEXT: ImplicitCastExpr
@@ -76,17 +76,17 @@ void TemplFunc(T t) {
   // CHECK-NEXT: OpenACCInitConstruct{{.*}}init
   // CHECK-NEXT: device_num clause
   // CHECK-NEXT: DeclRefExpr{{.*}} 't' 'T'
-#pragma acc init device_type(T)
+#pragma acc init device_type(radeon)
   // CHECK-NEXT: OpenACCInitConstruct{{.*}}init
-  // CHECK-NEXT: device_type(T)
-#pragma acc init if (T::value > t) device_type(T) device_num(t)
+  // CHECK-NEXT: device_type(radeon)
+#pragma acc init if (T::value > t) device_type(host) device_num(t)
   // CHECK-NEXT: OpenACCInitConstruct{{.*}}init
   // CHECK-NEXT: if clause
   // CHECK-NEXT: BinaryOperator{{.*}} '<dependent type>' '>'
   // CHECK-NEXT: DependentScopeDeclRefExpr
   // CHECK-NEXT: NestedNameSpecifier{{.*}} 'T'
   // CHECK-NEXT: DeclRefExpr{{.*}} 't' 'T'
-  // CHECK-NEXT: device_type(T)
+  // CHECK-NEXT: device_type(host)
   // CHECK-NEXT: device_num clause
   // CHECK-NEXT: DeclRefExpr{{.*}} 't' 'T'
 
@@ -119,7 +119,7 @@ void TemplFunc(T t) {
   // CHECK-NEXT: DeclRefExpr{{.*}} 't' 'SomeStruct'
 
   // CHECK-NEXT: OpenACCInitConstruct{{.*}}init
-  // CHECK-NEXT: device_type(T)
+  // CHECK-NEXT: device_type(radeon)
 
   // CHECK-NEXT: OpenACCInitConstruct{{.*}}init
   // CHECK-NEXT: if clause
@@ -131,7 +131,7 @@ void TemplFunc(T t) {
   // CHECK-NEXT: CXXMemberCallExpr{{.*}} 'unsigned int'
   // CHECK-NEXT: MemberExpr{{.*}} .operator unsigned int
   // CHECK-NEXT: DeclRefExpr{{.*}} 't' 'SomeStruct'
-  // CHECK-NEXT: device_type(T)
+  // CHECK-NEXT: device_type(host)
   // CHECK-NEXT: device_num clause
   // CHECK-NEXT: ImplicitCastExpr{{.*}}'unsigned int'
   // CHECK-NEXT: CXXMemberCallExpr{{.*}}'unsigned int'
diff --git a/clang/test/SemaOpenACC/init-construct.cpp b/clang/test/SemaOpenACC/init-construct.cpp
index e4491ee905f3..2cbe5f1ded04 100644
--- a/clang/test/SemaOpenACC/init-construct.cpp
+++ b/clang/test/SemaOpenACC/init-construct.cpp
@@ -18,9 +18,9 @@ void uses() {
 #pragma acc init
 #pragma acc init if (getI() < getS())
 #pragma acc init device_num(getI())
-#pragma acc init device_type(SOMETHING) device_num(getI())
-#pragma acc init device_type(SOMETHING) if (getI() < getS())
-#pragma acc init device_type(SOMETHING) device_num(getI()) if (getI() < getS())
+#pragma acc init device_type(host) device_num(getI())
+#pragma acc init device_type(nvidia) if (getI() < getS())
+#pragma acc init device_type(multicore) device_num(getI()) if (getI() < getS())
 
   // expected-error@+1{{value of type 'struct NotConvertible' is not contextually convertible to 'bool'}}
 #pragma acc init if (NC)
@@ -41,8 +41,8 @@ void TestInst() {
   T t;
 #pragma acc init
 #pragma acc init if (T::value < T{})
-#pragma acc init device_type(SOMETHING) device_num(getI()) if (getI() < getS())
-#pragma acc init device_type(SOMETHING) device_type(T) device_num(t) if (t < T::value) device_num(getI()) if (getI() < getS())
+#pragma acc init device_type(radeon) device_num(getI()) if (getI() < getS())
+#pragma acc init device_type(multicore) device_type(host) device_num(t) if (t < T::value) device_num(getI()) if (getI() < getS())
 
   // expected-error@+1{{value of type 'const NotConvertible' is not contextually convertible to 'bool'}}
 #pragma acc init if (T::NCValue)
diff --git a/clang/test/SemaOpenACC/loop-construct-device_type-ast.cpp b/clang/test/SemaOpenACC/loop-construct-device_type-ast.cpp
index 0030750a20db..c69fd49ee00e 100644
--- a/clang/test/SemaOpenACC/loop-construct-device_type-ast.cpp
+++ b/clang/test/SemaOpenACC/loop-construct-device_type-ast.cpp
@@ -19,11 +19,11 @@ void NormalUses() {
   // CHECK-NEXT: DeclStmt
   // CHECK-NEXT: VarDecl{{.*}} SomeVar 'bool'
 
-#pragma acc loop device_type(SomeS) dtype(SomeImpl)
+#pragma acc loop device_type(default) dtype(nvidia)
   for(int i = 0; i < 5; ++i){}
   // CHECK-NEXT: OpenACCLoopConstruct{{.*}}
-  // CHECK-NEXT: device_type(SomeS)
-  // CHECK-NEXT: dtype(SomeImpl)
+  // CHECK-NEXT: device_type(default)
+  // CHECK-NEXT: dtype(nvidia)
   // CHECK-NEXT: ForStmt
   // CHECK-NEXT: DeclStmt
   // CHECK-NEXT: VarDecl{{.*}} i 'int'
@@ -36,11 +36,11 @@ void NormalUses() {
   // CHECK-NEXT: UnaryOperator{{.*}}++
   // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
   // CHECK-NEXT: CompoundStmt
-#pragma acc loop device_type(SomeVar) dtype(int)
+#pragma acc loop device_type(acc_device_nvidia) dtype(radeon)
   for(int i = 0; i < 5; ++i){}
   // CHECK-NEXT: OpenACCLoopConstruct{{.*}}
-  // CHECK-NEXT: device_type(SomeVar)
-  // CHECK-NEXT: dtype(int)
+  // CHECK-NEXT: device_type(acc_device_nvidia)
+  // CHECK-NEXT: dtype(radeon)
   // CHECK-NEXT: ForStmt
   // CHECK-NEXT: DeclStmt
   // CHECK-NEXT: VarDecl{{.*}} i 'int'
@@ -53,11 +53,11 @@ void NormalUses() {
   // CHECK-NEXT: UnaryOperator{{.*}}++
   // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
   // CHECK-NEXT: CompoundStmt
-#pragma acc loop device_type(private) dtype(struct)
+#pragma acc loop device_type(host) dtype(multicore)
   for(int i = 0; i < 5; ++i){}
   // CHECK-NEXT: OpenACCLoopConstruct{{.*}}
-  // CHECK-NEXT: device_type(private)
-  // CHECK-NEXT: dtype(struct)
+  // CHECK-NEXT: device_type(host)
+  // CHECK-NEXT: dtype(multicore)
   // CHECK-NEXT: ForStmt
   // CHECK-NEXT: DeclStmt
   // CHECK-NEXT: VarDecl{{.*}} i 'int'
@@ -70,27 +70,10 @@ void NormalUses() {
   // CHECK-NEXT: UnaryOperator{{.*}}++
   // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
   // CHECK-NEXT: CompoundStmt
-#pragma acc loop device_type(private) dtype(class)
+#pragma acc loop device_type(default) dtype(*)
   for(int i = 0; i < 5; ++i){}
   // CHECK-NEXT: OpenACCLoopConstruct{{.*}}
-  // CHECK-NEXT: device_type(private)
-  // CHECK-NEXT: dtype(class)
-  // CHECK-NEXT: ForStmt
-  // CHECK-NEXT: DeclStmt
-  // CHECK-NEXT: VarDecl{{.*}} i 'int'
-  // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
-  // CHECK-NEXT: <<<NULL>>>
-  // CHECK-NEXT: BinaryOperator{{.*}}'<'
-  // CHECK-NEXT: ImplicitCastExpr
-  // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
-  // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 5
-  // CHECK-NEXT: UnaryOperator{{.*}}++
-  // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
-  // CHECK-NEXT: CompoundStmt
-#pragma acc loop device_type(float) dtype(*)
-  for(int i = 0; i < 5; ++i){}
-  // CHECK-NEXT: OpenACCLoopConstruct{{.*}}
-  // CHECK-NEXT: device_type(float)
+  // CHECK-NEXT: device_type(default)
   // CHECK-NEXT: dtype(*)
   // CHECK-NEXT: ForStmt
   // CHECK-NEXT: DeclStmt
@@ -104,10 +87,27 @@ void NormalUses() {
   // CHECK-NEXT: UnaryOperator{{.*}}++
   // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
   // CHECK-NEXT: CompoundStmt
-#pragma acc loop device_type(float, int) dtype(*)
+#pragma acc loop device_type(nvidia) dtype(*)
   for(int i = 0; i < 5; ++i){}
   // CHECK-NEXT: OpenACCLoopConstruct{{.*}}
-  // CHECK-NEXT: device_type(float, int)
+  // CHECK-NEXT: device_type(nvidia)
+  // CHECK-NEXT: dtype(*)
+  // CHECK-NEXT: ForStmt
+  // CHECK-NEXT: DeclStmt
+  // CHECK-NEXT: VarDecl{{.*}} i 'int'
+  // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+  // CHECK-NEXT: <<<NULL>>>
+  // CHECK-NEXT: BinaryOperator{{.*}}'<'
+  // CHECK-NEXT: ImplicitCastExpr
+  // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+  // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 5
+  // CHECK-NEXT: UnaryOperator{{.*}}++
+  // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+  // CHECK-NEXT: CompoundStmt
+#pragma acc loop device_type(radeon, host) dtype(*)
+  for(int i = 0; i < 5; ++i){}
+  // CHECK-NEXT: OpenACCLoopConstruct{{.*}}
+  // CHECK-NEXT: device_type(radeon, host)
   // CHECK-NEXT: dtype(*)
   // CHECK-NEXT: ForStmt
   // CHECK-NEXT: DeclStmt
@@ -129,11 +129,11 @@ void TemplUses() {
   // CHECK-NEXT: TemplateTypeParmDecl{{.*}}T
   // CHECK-NEXT: FunctionDecl{{.*}}TemplUses
   // CHECK-NEXT: CompoundStmt
-#pragma acc loop device_type(T) dtype(T)
+#pragma acc loop device_type(default) dtype(multicore)
   for(int i = 0; i < 5; ++i){}
   // CHECK-NEXT: OpenACCLoopConstruct{{.*}}
-  // CHECK-NEXT: device_type(T)
-  // CHECK-NEXT: dtype(T)
+  // CHECK-NEXT: device_type(default)
+  // CHECK-NEXT: dtype(multicore)
   // CHECK-NEXT: ForStmt
   // CHECK-NEXT: DeclStmt
   // CHECK-NEXT: VarDecl{{.*}} i 'int'
@@ -155,8 +155,8 @@ void TemplUses() {
   // CHECK-NEXT: CompoundStmt
 
   // CHECK-NEXT: OpenACCLoopConstruct{{.*}}
-  // CHECK-NEXT: device_type(T)
-  // CHECK-NEXT: dtype(T)
+  // CHECK-NEXT: device_type(default)
+  // CHECK-NEXT: dtype(multicore)
   // CHECK-NEXT: ForStmt
   // CHECK-NEXT: DeclStmt
   // CHECK-NEXT: VarDecl{{.*}} i 'int'
diff --git a/clang/test/SemaOpenACC/loop-construct-device_type-clause.c b/clang/test/SemaOpenACC/loop-construct-device_type-clause.c
index 5c5e76881a88..24da7ed88ff7 100644
--- a/clang/test/SemaOpenACC/loop-construct-device_type-clause.c
+++ b/clang/test/SemaOpenACC/loop-construct-device_type-clause.c
@@ -6,22 +6,22 @@ void uses() {
   typedef struct S{} STy;
   STy SImpl;
 
-#pragma acc loop device_type(I)
+#pragma acc loop device_type(radeon)
   for(int i = 0; i < 5; ++i);
-#pragma acc loop device_type(S) dtype(STy)
+#pragma acc loop device_type(host) dtype(multicore)
   for(int i = 0; i < 5; ++i);
-#pragma acc loop dtype(SImpl)
+#pragma acc loop dtype(nvidia)
   for(int i = 0; i < 5; ++i);
-#pragma acc loop dtype(int) device_type(*)
+#pragma acc loop dtype(default) device_type(*)
   for(int i = 0; i < 5; ++i);
-#pragma acc loop dtype(true) device_type(false)
+#pragma acc loop dtype(radeon) device_type(nvidia)
   for(int i = 0; i < 5; ++i);
 
   // expected-error@+1{{expected identifier}}
 #pragma acc loop dtype(int, *)
   for(int i = 0; i < 5; ++i);
 
-#pragma acc loop device_type(I, int)
+#pragma acc loop device_type(default, nvidia)
   for(int i = 0; i < 5; ++i);
   // expected-error@+2{{expected ','}}
   // expected-error@+1{{expected identifier}}
diff --git a/clang/test/SemaOpenACC/loop-construct-device_type-clause.cpp b/clang/test/SemaOpenACC/loop-construct-device_type-clause.cpp
index 57c70dd79d5e..257e0cc75fad 100644
--- a/clang/test/SemaOpenACC/loop-construct-device_type-clause.cpp
+++ b/clang/test/SemaOpenACC/loop-construct-device_type-clause.cpp
@@ -2,17 +2,17 @@
 
 template<typename T>
 void TemplUses() {
-#pragma acc loop device_type(I)
+#pragma acc loop device_type(host)
   for(int i = 0; i < 5; ++i);
 #pragma acc loop dtype(*)
   for(int i = 0; i < 5; ++i);
-#pragma acc loop device_type(class)
+#pragma acc loop device_type(default)
   for(int i = 0; i < 5; ++i);
-#pragma acc loop device_type(private)
+#pragma acc loop device_type(multicore)
   for(int i = 0; i < 5; ++i);
-#pragma acc loop device_type(bool)
+#pragma acc loop device_type(host)
   for(int i = 0; i < 5; ++i);
-#pragma acc kernels dtype(true) device_type(false)
+#pragma acc kernels dtype(radeon) device_type(nvidia)
   for(int i = 0; i < 5; ++i);
   // expected-error@+2{{expected ','}}
   // expected-error@+1{{expected identifier}}
diff --git a/clang/test/SemaOpenACC/routine-construct-ast.cpp b/clang/test/SemaOpenACC/routine-construct-ast.cpp
index 068d99123522..dfdcc321b94d 100644
--- a/clang/test/SemaOpenACC/routine-construct-ast.cpp
+++ b/clang/test/SemaOpenACC/routine-construct-ast.cpp
@@ -24,16 +24,16 @@ int function();
 // CHECK-NEXT: vector clause
 // CHECK-NEXT: bind clause identifier 'identifier'
 
-#pragma acc routine(function) device_type(Something) seq
+#pragma acc routine(function) device_type(acc_device_nvidia) seq
 // CHECK-NEXT: OpenACCRoutineDecl{{.*}} routine
 // CHECK-NEXT: DeclRefExpr{{.*}} 'function' 'int ()'
-// CHECK-NEXT: device_type(Something)
+// CHECK-NEXT: device_type(acc_device_nvidia)
 // CHECK-NEXT: seq clause
-#pragma acc routine(function) nohost dtype(Something) vector
+#pragma acc routine(function) nohost dtype(nvidia) vector
 // CHECK-NEXT: OpenACCRoutineDecl{{.*}} routine
 // CHECK-NEXT: DeclRefExpr{{.*}} 'function' 'int ()'
 // CHECK-NEXT: nohost clause
-// CHECK-NEXT: dtype(Something)
+// CHECK-NEXT: dtype(nvidia)
 // CHECK-NEXT: vector clause
 
 #pragma acc routine nohost vector
@@ -172,15 +172,15 @@ struct S {
 // CHECK-NEXT: OpenACCRoutineDecl{{.*}} routine
 // CHECK-NEXT: DeclRefExpr{{.*}} 'Lambda' 'const S::(lambda at
 // CHECK-NEXT: worker clause
-#pragma acc routine(Lambda) worker device_type(Lambda)
+#pragma acc routine(Lambda) worker device_type(host)
 // CHECK-NEXT: OpenACCRoutineDecl{{.*}} routine
 // CHECK-NEXT: DeclRefExpr{{.*}} 'Lambda' 'const S::(lambda at
 // CHECK-NEXT: worker clause
-// CHECK-NEXT: device_type(Lambda)
-#pragma acc routine(Lambda) dtype(Lambda) vector
+// CHECK-NEXT: device_type(host)
+#pragma acc routine(Lambda) dtype(radeon) vector
 // CHECK-NEXT: OpenACCRoutineDecl{{.*}} routine
 // CHECK-NEXT: DeclRefExpr{{.*}} 'Lambda' 'const S::(lambda at
-// CHECK-NEXT: dtype(Lambda)
+// CHECK-NEXT: dtype(radeon)
 // CHECK-NEXT: vector clause
 };
 
@@ -294,12 +294,12 @@ struct DepS {
 // CHECK-NEXT: DeclRefExpr{{.*}} 'StaticMemFunc' 'T ()'
 // CHECK-NEXT: NestedNameSpecifier{{.*}} 'DepS<T>'
 // CHECK-NEXT: worker clause
-#pragma acc routine(DepS<T>::StaticMemFunc) worker device_type(T)
+#pragma acc routine(DepS<T>::StaticMemFunc) worker device_type(multicore)
 // CHECK-NEXT: OpenACCRoutineDecl{{.*}} routine 
 // CHECK-NEXT: DeclRefExpr{{.*}} 'StaticMemFunc' 'T ()'
 // CHECK-NEXT: NestedNameSpecifier{{.*}} 'DepS<T>'
 // CHECK-NEXT: worker clause
-// CHECK-NEXT: device_type(T)
+// CHECK-NEXT: device_type(multicore)
 
 // Instantiation:
 // CHECK: ClassTemplateSpecializationDecl{{.*}}struct DepS
@@ -398,7 +398,7 @@ struct DepS {
 // CHECK-NEXT: DeclRefExpr{{.*}} 'StaticMemFunc' 'int ()'
 // CHECK-NEXT: NestedNameSpecifier{{.*}} 'DepS<int>'
 // CHECK-NEXT: worker clause
-// CHECK-NEXT: device_type(T)
+// CHECK-NEXT: device_type(multicore)
 };
 
 #pragma acc routine(DepS<int>::Lambda) gang(dim:1)
diff --git a/clang/test/SemaOpenACC/routine-construct-clauses.cpp b/clang/test/SemaOpenACC/routine-construct-clauses.cpp
index 78554fa70292..b9fc46aeff09 100644
--- a/clang/test/SemaOpenACC/routine-construct-clauses.cpp
+++ b/clang/test/SemaOpenACC/routine-construct-clauses.cpp
@@ -139,8 +139,8 @@ void Inst() {
   DependentT<HasFuncs> T;// expected-note{{in instantiation of}}
 }
 
-#pragma acc routine(Func) gang device_type(Inst)
-#pragma acc routine(Func) gang dtype(Inst)
+#pragma acc routine(Func) gang device_type(host)
+#pragma acc routine(Func) gang dtype(multicore)
 #pragma acc routine(Func) device_type(*) worker
 #pragma acc routine(Func) dtype(*) worker
 
diff --git a/clang/test/SemaOpenACC/set-construct-ast.cpp b/clang/test/SemaOpenACC/set-construct-ast.cpp
index c5eed944ada7..69a750fb0284 100644
--- a/clang/test/SemaOpenACC/set-construct-ast.cpp
+++ b/clang/test/SemaOpenACC/set-construct-ast.cpp
@@ -13,7 +13,7 @@ void NormalFunc() {
   // CHECK-LABEL: NormalFunc
   // CHECK-NEXT: CompoundStmt
 
-#pragma acc set default_async(some_int()) device_num(some_long()) device_type(DT) if (some_int() < some_long())
+#pragma acc set default_async(some_int()) device_num(some_long()) device_type(acc_device_nvidia) if (some_int() < some_long())
   // CHECK-NEXT: OpenACCSetConstruct{{.*}}set
   // CHECK-NEXT: default_async clause
   // CHECK-NEXT: CallExpr{{.*}}'int'
@@ -23,7 +23,7 @@ void NormalFunc() {
   // CHECK-NEXT: CallExpr{{.*}} 'long'
   // CHECK-NEXT: ImplicitCastExpr
   // CHECK-NEXT: DeclRefExpr{{.*}}'some_long' 'long ()'
-  // CHECK-NEXT: device_type(DT)
+  // CHECK-NEXT: device_type(acc_device_nvidia)
   // CHECK-NEXT: if clause
   // CHECK-NEXT: BinaryOperator{{.*}}'bool' '<'
   // CHECK-NEXT: ImplicitCastExpr{{.*}}'long'
@@ -43,14 +43,14 @@ void TemplFunc(T t) {
   // CHECK-NEXT: ParmVarDecl{{.*}} t 'T'
   // CHECK-NEXT: CompoundStmt
 
-#pragma acc set default_async(T::value) device_num(t) device_type(DT) if (T::value < t)
+#pragma acc set default_async(T::value) device_num(t) device_type(multicore) if (T::value < t)
   // CHECK-NEXT: OpenACCSetConstruct{{.*}}set
   // CHECK-NEXT: default_async clause
   // CHECK-NEXT: DependentScopeDeclRefExpr{{.*}} '<dependent type>'
   // CHECK-NEXT: NestedNameSpecifier TypeSpec 'T'
   // CHECK-NEXT: device_num clause
   // CHECK-NEXT: DeclRefExpr{{.*}}'t' 'T'
-  // CHECK-NEXT: device_type(DT)
+  // CHECK-NEXT: device_type(multicore)
   // CHECK-NEXT: if clause
   // CHECK-NEXT: BinaryOperator{{.*}}'<dependent type>' '<'
   // CHECK-NEXT: DependentScopeDeclRefExpr{{.*}} '<dependent type>'
@@ -75,7 +75,7 @@ void TemplFunc(T t) {
   // CHECK-NEXT: CXXMemberCallExpr{{.*}}'unsigned int'
   // CHECK-NEXT: MemberExpr{{.*}}.operator unsigned int
   // CHECK-NEXT: DeclRefExpr{{.*}}'t' 'SomeStruct'
-  // CHECK-NEXT: device_type(DT)
+  // CHECK-NEXT: device_type(multicore)
   // CHECK-NEXT: if clause
   // CHECK-NEXT: BinaryOperator{{.*}}'bool' '<'
   // CHECK-NEXT: ImplicitCastExpr {{.*}}'unsigned int'
diff --git a/clang/test/SemaOpenACC/set-construct.cpp b/clang/test/SemaOpenACC/set-construct.cpp
index 23816dbab291..9a90ea67b1bd 100644
--- a/clang/test/SemaOpenACC/set-construct.cpp
+++ b/clang/test/SemaOpenACC/set-construct.cpp
@@ -17,11 +17,11 @@ struct ExplicitConvertOnly {
 void uses() {
 #pragma acc set default_async(getI())
 #pragma acc set device_num(getI())
-#pragma acc set device_type(getI)
-#pragma acc set device_type(getI) if (getI() < getS())
+#pragma acc set device_type(multicore)
+#pragma acc set device_type(host) if (getI() < getS())
 
   // expected-error@+1{{value of type 'struct NotConvertible' is not contextually convertible to 'bool'}}
-#pragma acc set if (NC) device_type(I)
+#pragma acc set if (NC) device_type(radeon)
 
   // expected-error@+2{{OpenACC 'set' construct must have at least one 'default_async', 'device_num', 'device_type' or 'if' clause}}
   // expected-error@+1{{OpenACC clause 'device_num' requires expression of integer type ('struct NotConvertible' invalid)}}
@@ -62,8 +62,8 @@ void uses() {
 
   // expected-error@+2{{'device_type' clause cannot appear more than once on a 'set' directive}}
   // expected-note@+1{{previous clause is here}}
-#pragma acc set device_type(I) device_type(I)
+#pragma acc set device_type(nvidia) device_type(default)
   // expected-error@+2{{'if' clause cannot appear more than once on a 'set' directive}}
   // expected-note@+1{{previous clause is here}}
-#pragma acc set device_type(I) if(true) if (true)
+#pragma acc set device_type(acc_device_nvidia) if(true) if (true)
 }
diff --git a/clang/test/SemaOpenACC/shutdown-construct-ast.cpp b/clang/test/SemaOpenACC/shutdown-construct-ast.cpp
index 13f345799e0c..9cc72b3af891 100644
--- a/clang/test/SemaOpenACC/shutdown-construct-ast.cpp
+++ b/clang/test/SemaOpenACC/shutdown-construct-ast.cpp
@@ -34,10 +34,10 @@ _Pragma("acc shutdown")
   // CHECK-NEXT: CallExpr
   // CHECK-NEXT: ImplicitCastExpr
   // CHECK-NEXT: DeclRefExpr{{.*}}'some_int' 'int ()'
-#pragma acc shutdown device_type(T)
+#pragma acc shutdown device_type(multicore)
   // CHECK-NEXT: OpenACCShutdownConstruct{{.*}}shutdown
-  // CHECK-NEXT: device_type(T)
-#pragma acc shutdown if (some_int() < some_long()) device_type(T) device_num(some_int())
+  // CHECK-NEXT: device_type(multicore)
+#pragma acc shutdown if (some_int() < some_long()) device_type(host) device_num(some_int())
   // CHECK-NEXT: OpenACCShutdownConstruct{{.*}}shutdown
   // CHECK-NEXT: if clause
   // CHECK-NEXT: BinaryOperator{{.*}} 'bool' '<'
@@ -48,7 +48,7 @@ _Pragma("acc shutdown")
   // CHECK-NEXT: CallExpr
   // CHECK-NEXT: ImplicitCastExpr
   // CHECK-NEXT: DeclRefExpr{{.*}}'some_long' 'long ()'
-  // CHECK-NEXT: device_type(T)
+  // CHECK-NEXT: device_type(host)
   // CHECK-NEXT: device_num clause
   // CHECK-NEXT: CallExpr
   // CHECK-NEXT: ImplicitCastExpr
@@ -76,17 +76,17 @@ void TemplFunc(T t) {
   // CHECK-NEXT: OpenACCShutdownConstruct{{.*}}shutdown
   // CHECK-NEXT: device_num clause
   // CHECK-NEXT: DeclRefExpr{{.*}} 't' 'T'
-#pragma acc shutdown device_type(T)
+#pragma acc shutdown device_type(radeon)
   // CHECK-NEXT: OpenACCShutdownConstruct{{.*}}shutdown
-  // CHECK-NEXT: device_type(T)
-#pragma acc shutdown if (T::value > t) device_type(T) device_num(t)
+  // CHECK-NEXT: device_type(radeon)
+#pragma acc shutdown if (T::value > t) device_type(acc_device_nvidia) device_num(t)
   // CHECK-NEXT: OpenACCShutdownConstruct{{.*}}shutdown
   // CHECK-NEXT: if clause
   // CHECK-NEXT: BinaryOperator{{.*}} '<dependent type>' '>'
   // CHECK-NEXT: DependentScopeDeclRefExpr
   // CHECK-NEXT: NestedNameSpecifier{{.*}} 'T'
   // CHECK-NEXT: DeclRefExpr{{.*}} 't' 'T'
-  // CHECK-NEXT: device_type(T)
+  // CHECK-NEXT: device_type(acc_device_nvidia)
   // CHECK-NEXT: device_num clause
   // CHECK-NEXT: DeclRefExpr{{.*}} 't' 'T'
 
@@ -119,7 +119,7 @@ void TemplFunc(T t) {
   // CHECK-NEXT: DeclRefExpr{{.*}} 't' 'SomeStruct'
 
   // CHECK-NEXT: OpenACCShutdownConstruct{{.*}}shutdown
-  // CHECK-NEXT: device_type(T)
+  // CHECK-NEXT: device_type(radeon)
 
   // CHECK-NEXT: OpenACCShutdownConstruct{{.*}}shutdown
   // CHECK-NEXT: if clause
@@ -131,7 +131,7 @@ void TemplFunc(T t) {
   // CHECK-NEXT: CXXMemberCallExpr{{.*}} 'unsigned int'
   // CHECK-NEXT: MemberExpr{{.*}} .operator unsigned int
   // CHECK-NEXT: DeclRefExpr{{.*}} 't' 'SomeStruct'
-  // CHECK-NEXT: device_type(T)
+  // CHECK-NEXT: device_type(acc_device_nvidia)
   // CHECK-NEXT: device_num clause
   // CHECK-NEXT: ImplicitCastExpr{{.*}}'unsigned int'
   // CHECK-NEXT: CXXMemberCallExpr{{.*}}'unsigned int'
diff --git a/clang/test/SemaOpenACC/shutdown-construct.cpp b/clang/test/SemaOpenACC/shutdown-construct.cpp
index c3390b041287..b167ac66c85d 100644
--- a/clang/test/SemaOpenACC/shutdown-construct.cpp
+++ b/clang/test/SemaOpenACC/shutdown-construct.cpp
@@ -18,9 +18,9 @@ void uses() {
 #pragma acc shutdown
 #pragma acc shutdown if (getI() < getS())
 #pragma acc shutdown device_num(getI())
-#pragma acc shutdown device_type(SOMETHING) device_num(getI())
-#pragma acc shutdown device_type(SOMETHING) if (getI() < getS())
-#pragma acc shutdown device_type(SOMETHING) device_num(getI()) if (getI() < getS())
+#pragma acc shutdown device_type(host) device_num(getI())
+#pragma acc shutdown device_type(nvidia) if (getI() < getS())
+#pragma acc shutdown device_type(radeon) device_num(getI()) if (getI() < getS())
 
   // expected-error@+1{{value of type 'struct NotConvertible' is not contextually convertible to 'bool'}}
 #pragma acc shutdown if (NC)
@@ -41,8 +41,8 @@ void TestInst() {
   T t;
 #pragma acc shutdown
 #pragma acc shutdown if (T::value < T{})
-#pragma acc shutdown device_type(SOMETHING) device_num(getI()) if (getI() < getS())
-#pragma acc shutdown device_type(SOMETHING) device_type(T) device_num(t) if (t < T::value) device_num(getI()) if (getI() < getS())
+#pragma acc shutdown device_type(multicore) device_num(getI()) if (getI() < getS())
+#pragma acc shutdown device_type(default) device_type(radeon) device_num(t) if (t < T::value) device_num(getI()) if (getI() < getS())
 
   // expected-error@+1{{value of type 'const NotConvertible' is not contextually convertible to 'bool'}}
 #pragma acc shutdown if (T::NCValue)
diff --git a/clang/test/SemaOpenACC/update-construct-ast.cpp b/clang/test/SemaOpenACC/update-construct-ast.cpp
index 4238f93d90a3..de7f4da589ed 100644
--- a/clang/test/SemaOpenACC/update-construct-ast.cpp
+++ b/clang/test/SemaOpenACC/update-construct-ast.cpp
@@ -33,15 +33,15 @@ void NormalFunc() {
   // CHECK-NEXT: ImplicitCastExpr
   // CHECK-NEXT: DeclRefExpr{{.*}}'some_long' 'long ()'
 
-#pragma acc update self(Global) wait async device_type(A) dtype(B)
+#pragma acc update self(Global) wait async device_type(nvidia) dtype(radeon)
   // CHECK-NEXT: OpenACCUpdateConstruct{{.*}}update
   // CHECK-NEXT: self clause
   // CHECK-NEXT: DeclRefExpr{{.*}}'Global' 'int'
   // CHECK-NEXT: wait clause
   // CHECK-NEXT: <<<NULL>>>
   // CHECK-NEXT: async clause
-  // CHECK-NEXT: device_type(A)
-  // CHECK-NEXT: dtype(B)
+  // CHECK-NEXT: device_type(nvidia)
+  // CHECK-NEXT: dtype(radeon)
 #pragma acc update self(Global) wait(some_int(), some_long()) async(some_int())
   // CHECK-NEXT: OpenACCUpdateConstruct{{.*}}update
   // CHECK-NEXT: self clause
@@ -147,15 +147,15 @@ void TemplFunc(T t) {
   // CHECK-NEXT: NestedNameSpecifier TypeSpec 'T'
   // CHECK-NEXT: DeclRefExpr{{.*}}'t' 'T'
 
-#pragma acc update self(t) wait async device_type(T) dtype(U)
+#pragma acc update self(t) wait async device_type(default) dtype(host)
   // CHECK-NEXT: OpenACCUpdateConstruct{{.*}}update
   // CHECK-NEXT: self clause
   // CHECK-NEXT: DeclRefExpr{{.*}} 't' 'T'
   // CHECK-NEXT: wait clause
   // CHECK-NEXT: <<<NULL>>>
   // CHECK-NEXT: async clause
-  // CHECK-NEXT: device_type(T)
-  // CHECK-NEXT: dtype(U)
+  // CHECK-NEXT: device_type(default)
+  // CHECK-NEXT: dtype(host)
 #pragma acc update self(t) wait(T::value, t) async(T::value)
   // CHECK-NEXT: OpenACCUpdateConstruct{{.*}}update
   // CHECK-NEXT: self clause
@@ -261,8 +261,8 @@ void TemplFunc(T t) {
   // CHECK-NEXT: wait clause
   // CHECK-NEXT: <<<NULL>>>
   // CHECK-NEXT: async clause
-  // CHECK-NEXT: device_type(T)
-  // CHECK-NEXT: dtype(U)
+  // CHECK-NEXT: device_type(default)
+  // CHECK-NEXT: dtype(host)
 
   // CHECK-NEXT: OpenACCUpdateConstruct{{.*}}update
   // CHECK-NEXT: self clause
diff --git a/clang/test/SemaOpenACC/update-construct.cpp b/clang/test/SemaOpenACC/update-construct.cpp
index 92a8f3a0e2d4..f0372736452d 100644
--- a/clang/test/SemaOpenACC/update-construct.cpp
+++ b/clang/test/SemaOpenACC/update-construct.cpp
@@ -6,7 +6,7 @@ void uses() {
   int Var;
 #pragma acc update async self(Var)
 #pragma acc update wait self(Var)
-#pragma acc update self(Var) device_type(I)
+#pragma acc update self(Var) device_type(host)
 #pragma acc update if(true) self(Var)
 #pragma acc update if_present self(Var)
 #pragma acc update self(Var)
@@ -15,24 +15,24 @@ void uses() {
 
   // expected-error@+2{{OpenACC clause 'if' may not follow a 'device_type' clause in a 'update' construct}}
   // expected-note@+1{{previous clause is here}}
-#pragma acc update self(Var) device_type(I) if(true)
+#pragma acc update self(Var) device_type(default) if(true)
   // expected-error@+2{{OpenACC clause 'if_present' may not follow a 'device_type' clause in a 'update' construct}}
   // expected-note@+1{{previous clause is here}}
-#pragma acc update self(Var) device_type(I) if_present
+#pragma acc update self(Var) device_type(multicore) if_present
   // expected-error@+2{{OpenACC clause 'self' may not follow a 'device_type' clause in a 'update' construct}}
   // expected-note@+1{{previous clause is here}}
-#pragma acc update self(Var) device_type(I) self(Var)
+#pragma acc update self(Var) device_type(nvidia) self(Var)
   // expected-error@+2{{OpenACC clause 'host' may not follow a 'device_type' clause in a 'update' construct}}
   // expected-note@+1{{previous clause is here}}
-#pragma acc update self(Var) device_type(I) host(Var)
+#pragma acc update self(Var) device_type(radeon) host(Var)
   // expected-error@+2{{OpenACC clause 'device' may not follow a 'device_type' clause in a 'update' construct}}
   // expected-note@+1{{previous clause is here}}
-#pragma acc update self(Var) device_type(I) device(Var)
+#pragma acc update self(Var) device_type(multicore) device(Var)
   // These 2 are OK.
-#pragma acc update self(Var) device_type(I) async
-#pragma acc update self(Var) device_type(I) wait
+#pragma acc update self(Var) device_type(default) async
+#pragma acc update self(Var) device_type(nvidia) wait
   // Unless otherwise specified, we assume 'device_type' can happen after itself.
-#pragma acc update self(Var) device_type(I) device_type(I)
+#pragma acc update self(Var) device_type(acc_device_nvidia) device_type(multicore)
 
   // These diagnose because there isn't at least 1 of 'self', 'host', or
   // 'device'.
@@ -41,14 +41,14 @@ void uses() {
   // expected-error@+1{{OpenACC 'update' construct must have at least one 'self', 'host' or 'device' clause}}
 #pragma acc update wait
   // expected-error@+1{{OpenACC 'update' construct must have at least one 'self', 'host' or 'device' clause}}
-#pragma acc update device_type(I)
+#pragma acc update device_type(host)
   // expected-error@+1{{OpenACC 'update' construct must have at least one 'self', 'host' or 'device' clause}}
 #pragma acc update if(true)
   // expected-error@+1{{OpenACC 'update' construct must have at least one 'self', 'host' or 'device' clause}}
 #pragma acc update if_present
 
   // expected-error@+1{{value of type 'struct NotConvertible' is not contextually convertible to 'bool'}}
-#pragma acc update self(Var) if (NC) device_type(I)
+#pragma acc update self(Var) if (NC) device_type(radeon)
 
   // expected-error@+2{{OpenACC 'if' clause cannot appear more than once on a 'update' directive}}
   // expected-note@+1{{previous clause is here}}