[OpenACC] Implement copyin, copyout, create clauses for compute construct

Like 'copy', these also have alternate names, so this implements that as
well.  Additionally, these have an optional tag of either 'readonly' or
'zero' depending on the clause.

Otherwise, this is a pretty rote implementation of the clause, as there
aren't any special rules for it.

22 files changed, 1268 insertions, 74 deletions

diff --git a/clang/include/clang/AST/OpenACCClause.h b/clang/include/clang/AST/OpenACCClause.h
index ba20b2c811a8..abbd7c3989bc 100644
--- a/clang/include/clang/AST/OpenACCClause.h
+++ b/clang/include/clang/AST/OpenACCClause.h
@@ -375,6 +375,87 @@ public:
          ArrayRef<Expr *> VarList, SourceLocation EndLoc);
 };
 
+class OpenACCCopyInClause final
+    : public OpenACCClauseWithVarList,
+      public llvm::TrailingObjects<OpenACCCopyInClause, Expr *> {
+  bool IsReadOnly;
+
+  OpenACCCopyInClause(OpenACCClauseKind Spelling, SourceLocation BeginLoc,
+                      SourceLocation LParenLoc, bool IsReadOnly,
+                      ArrayRef<Expr *> VarList, SourceLocation EndLoc)
+      : OpenACCClauseWithVarList(Spelling, BeginLoc, LParenLoc, EndLoc),
+        IsReadOnly(IsReadOnly) {
+    assert((Spelling == OpenACCClauseKind::CopyIn ||
+            Spelling == OpenACCClauseKind::PCopyIn ||
+            Spelling == OpenACCClauseKind::PresentOrCopyIn) &&
+           "Invalid clause kind for copyin-clause");
+    std::uninitialized_copy(VarList.begin(), VarList.end(),
+                            getTrailingObjects<Expr *>());
+    setExprs(MutableArrayRef(getTrailingObjects<Expr *>(), VarList.size()));
+  }
+
+public:
+  bool isReadOnly() const { return IsReadOnly; }
+  static OpenACCCopyInClause *
+  Create(const ASTContext &C, OpenACCClauseKind Spelling,
+         SourceLocation BeginLoc, SourceLocation LParenLoc, bool IsReadOnly,
+         ArrayRef<Expr *> VarList, SourceLocation EndLoc);
+};
+
+class OpenACCCopyOutClause final
+    : public OpenACCClauseWithVarList,
+      public llvm::TrailingObjects<OpenACCCopyOutClause, Expr *> {
+  bool IsZero;
+
+  OpenACCCopyOutClause(OpenACCClauseKind Spelling, SourceLocation BeginLoc,
+                       SourceLocation LParenLoc, bool IsZero,
+                       ArrayRef<Expr *> VarList, SourceLocation EndLoc)
+      : OpenACCClauseWithVarList(Spelling, BeginLoc, LParenLoc, EndLoc),
+        IsZero(IsZero) {
+    assert((Spelling == OpenACCClauseKind::CopyOut ||
+            Spelling == OpenACCClauseKind::PCopyOut ||
+            Spelling == OpenACCClauseKind::PresentOrCopyOut) &&
+           "Invalid clause kind for copyout-clause");
+    std::uninitialized_copy(VarList.begin(), VarList.end(),
+                            getTrailingObjects<Expr *>());
+    setExprs(MutableArrayRef(getTrailingObjects<Expr *>(), VarList.size()));
+  }
+
+public:
+  bool isZero() const { return IsZero; }
+  static OpenACCCopyOutClause *
+  Create(const ASTContext &C, OpenACCClauseKind Spelling,
+         SourceLocation BeginLoc, SourceLocation LParenLoc, bool IsZero,
+         ArrayRef<Expr *> VarList, SourceLocation EndLoc);
+};
+
+class OpenACCCreateClause final
+    : public OpenACCClauseWithVarList,
+      public llvm::TrailingObjects<OpenACCCreateClause, Expr *> {
+  bool IsZero;
+
+  OpenACCCreateClause(OpenACCClauseKind Spelling, SourceLocation BeginLoc,
+                      SourceLocation LParenLoc, bool IsZero,
+                      ArrayRef<Expr *> VarList, SourceLocation EndLoc)
+      : OpenACCClauseWithVarList(Spelling, BeginLoc, LParenLoc, EndLoc),
+        IsZero(IsZero) {
+    assert((Spelling == OpenACCClauseKind::Create ||
+            Spelling == OpenACCClauseKind::PCreate ||
+            Spelling == OpenACCClauseKind::PresentOrCreate) &&
+           "Invalid clause kind for create-clause");
+    std::uninitialized_copy(VarList.begin(), VarList.end(),
+                            getTrailingObjects<Expr *>());
+    setExprs(MutableArrayRef(getTrailingObjects<Expr *>(), VarList.size()));
+  }
+
+public:
+  bool isZero() const { return IsZero; }
+  static OpenACCCreateClause *
+  Create(const ASTContext &C, OpenACCClauseKind Spelling,
+         SourceLocation BeginLoc, SourceLocation LParenLoc, bool IsZero,
+         ArrayRef<Expr *> VarList, SourceLocation EndLoc);
+};
+
 template <class Impl> class OpenACCClauseVisitor {
   Impl &getDerived() { return static_cast<Impl &>(*this); }
 
diff --git a/clang/include/clang/Basic/OpenACCClauses.def b/clang/include/clang/Basic/OpenACCClauses.def
index 5cb421975c95..06c7a379b222 100644
--- a/clang/include/clang/Basic/OpenACCClauses.def
+++ b/clang/include/clang/Basic/OpenACCClauses.def
@@ -24,6 +24,15 @@
 VISIT_CLAUSE(Copy)
 CLAUSE_ALIAS(PCopy, Copy)
 CLAUSE_ALIAS(PresentOrCopy, Copy)
+VISIT_CLAUSE(CopyIn)
+CLAUSE_ALIAS(PCopyIn, CopyIn)
+CLAUSE_ALIAS(PresentOrCopyIn, CopyIn)
+VISIT_CLAUSE(CopyOut)
+CLAUSE_ALIAS(PCopyOut, CopyOut)
+CLAUSE_ALIAS(PresentOrCopyOut, CopyOut)
+VISIT_CLAUSE(Create)
+CLAUSE_ALIAS(PCreate, Create)
+CLAUSE_ALIAS(PresentOrCreate, Create)
 VISIT_CLAUSE(Default)
 VISIT_CLAUSE(FirstPrivate)
 VISIT_CLAUSE(If)
diff --git a/clang/include/clang/Basic/OpenACCKinds.h b/clang/include/clang/Basic/OpenACCKinds.h
index e39f9e876d54..0e38a04e7164 100644
--- a/clang/include/clang/Basic/OpenACCKinds.h
+++ b/clang/include/clang/Basic/OpenACCKinds.h
@@ -228,12 +228,27 @@ enum class OpenACCClauseKind {
   /// 'copyout' clause, allowed on Compute and Combined constructs, plus 'data',
   /// 'exit data', and 'declare'.
   CopyOut,
+  /// 'copyout' clause alias 'pcopyout'.  Preserved for diagnostic purposes.
+  PCopyOut,
+  /// 'copyout' clause alias 'present_or_copyout'.  Preserved for diagnostic
+  /// purposes.
+  PresentOrCopyOut,
   /// 'copyin' clause, allowed on Compute and Combined constructs, plus 'data',
   /// 'enter data', and 'declare'.
   CopyIn,
-  /// 'copyin' clause, allowed on Compute and Combined constructs, plus 'data',
+  /// 'copyin' clause alias 'pcopyin'.  Preserved for diagnostic purposes.
+  PCopyIn,
+  /// 'copyin' clause alias 'present_or_copyin'.  Preserved for diagnostic
+  /// purposes.
+  PresentOrCopyIn,
+  /// 'create' clause, allowed on Compute and Combined constructs, plus 'data',
   /// 'enter data', and 'declare'.
   Create,
+  /// 'create' clause alias 'pcreate'.  Preserved for diagnostic purposes.
+  PCreate,
+  /// 'create' clause alias 'present_or_create'.  Preserved for diagnostic
+  /// purposes.
+  PresentOrCreate,
   /// 'reduction' clause, allowed on Parallel, Serial, Loop, and the combined
   /// constructs.
   Reduction,
@@ -362,12 +377,30 @@ inline StreamTy &printOpenACCClauseKind(StreamTy &Out, OpenACCClauseKind K) {
   case OpenACCClauseKind::CopyOut:
     return Out << "copyout";
 
+  case OpenACCClauseKind::PCopyOut:
+    return Out << "pcopyout";
+
+  case OpenACCClauseKind::PresentOrCopyOut:
+    return Out << "present_or_copyout";
+
   case OpenACCClauseKind::CopyIn:
     return Out << "copyin";
 
+  case OpenACCClauseKind::PCopyIn:
+    return Out << "pcopyin";
+
+  case OpenACCClauseKind::PresentOrCopyIn:
+    return Out << "present_or_copyin";
+
   case OpenACCClauseKind::Create:
     return Out << "create";
 
+  case OpenACCClauseKind::PCreate:
+    return Out << "pcreate";
+
+  case OpenACCClauseKind::PresentOrCreate:
+    return Out << "present_or_create";
+
   case OpenACCClauseKind::Reduction:
     return Out << "reduction";
 
diff --git a/clang/include/clang/Sema/SemaOpenACC.h b/clang/include/clang/Sema/SemaOpenACC.h
index 1c5127e79a66..5f77ec90d0b6 100644
--- a/clang/include/clang/Sema/SemaOpenACC.h
+++ b/clang/include/clang/Sema/SemaOpenACC.h
@@ -50,6 +50,8 @@ public:
 
     struct VarListDetails {
       SmallVector<Expr *> VarList;
+      bool IsReadOnly;
+      bool IsZero;
     };
 
     std::variant<std::monostate, DefaultDetails, ConditionDetails,
@@ -123,6 +125,15 @@ public:
               ClauseKind == OpenACCClauseKind::Copy ||
               ClauseKind == OpenACCClauseKind::PCopy ||
               ClauseKind == OpenACCClauseKind::PresentOrCopy ||
+              ClauseKind == OpenACCClauseKind::CopyIn ||
+              ClauseKind == OpenACCClauseKind::PCopyIn ||
+              ClauseKind == OpenACCClauseKind::PresentOrCopyIn ||
+              ClauseKind == OpenACCClauseKind::CopyOut ||
+              ClauseKind == OpenACCClauseKind::PCopyOut ||
+              ClauseKind == OpenACCClauseKind::PresentOrCopyOut ||
+              ClauseKind == OpenACCClauseKind::Create ||
+              ClauseKind == OpenACCClauseKind::PCreate ||
+              ClauseKind == OpenACCClauseKind::PresentOrCreate ||
               ClauseKind == OpenACCClauseKind::FirstPrivate) &&
              "Parsed clause kind does not have a var-list");
       return std::get<VarListDetails>(Details).VarList;
@@ -132,6 +143,25 @@ public:
       return const_cast<OpenACCParsedClause *>(this)->getVarList();
     }
 
+    bool isReadOnly() const {
+      assert((ClauseKind == OpenACCClauseKind::CopyIn ||
+              ClauseKind == OpenACCClauseKind::PCopyIn ||
+              ClauseKind == OpenACCClauseKind::PresentOrCopyIn) &&
+             "Only copyin accepts 'readonly:' tag");
+      return std::get<VarListDetails>(Details).IsReadOnly;
+    }
+
+    bool isZero() const {
+      assert((ClauseKind == OpenACCClauseKind::CopyOut ||
+              ClauseKind == OpenACCClauseKind::PCopyOut ||
+              ClauseKind == OpenACCClauseKind::PresentOrCopyOut ||
+              ClauseKind == OpenACCClauseKind::Create ||
+              ClauseKind == OpenACCClauseKind::PCreate ||
+              ClauseKind == OpenACCClauseKind::PresentOrCreate) &&
+             "Only copyout/create accepts 'zero' tag");
+      return std::get<VarListDetails>(Details).IsZero;
+    }
+
     void setLParenLoc(SourceLocation EndLoc) { LParenLoc = EndLoc; }
     void setEndLoc(SourceLocation EndLoc) { ClauseRange.setEnd(EndLoc); }
 
@@ -170,28 +200,71 @@ public:
       Details = IntExprDetails{std::move(IntExprs)};
     }
 
-    void setVarListDetails(ArrayRef<Expr *> VarList) {
+    void setVarListDetails(ArrayRef<Expr *> VarList, bool IsReadOnly,
+                           bool IsZero) {
       assert((ClauseKind == OpenACCClauseKind::Private ||
               ClauseKind == OpenACCClauseKind::NoCreate ||
               ClauseKind == OpenACCClauseKind::Present ||
               ClauseKind == OpenACCClauseKind::Copy ||
               ClauseKind == OpenACCClauseKind::PCopy ||
               ClauseKind == OpenACCClauseKind::PresentOrCopy ||
+              ClauseKind == OpenACCClauseKind::CopyIn ||
+              ClauseKind == OpenACCClauseKind::PCopyIn ||
+              ClauseKind == OpenACCClauseKind::PresentOrCopyIn ||
+              ClauseKind == OpenACCClauseKind::CopyOut ||
+              ClauseKind == OpenACCClauseKind::PCopyOut ||
+              ClauseKind == OpenACCClauseKind::PresentOrCopyOut ||
+              ClauseKind == OpenACCClauseKind::Create ||
+              ClauseKind == OpenACCClauseKind::PCreate ||
+              ClauseKind == OpenACCClauseKind::PresentOrCreate ||
               ClauseKind == OpenACCClauseKind::FirstPrivate) &&
              "Parsed clause kind does not have a var-list");
-      Details = VarListDetails{{VarList.begin(), VarList.end()}};
+      assert((!IsReadOnly || ClauseKind == OpenACCClauseKind::CopyIn ||
+              ClauseKind == OpenACCClauseKind::PCopyIn ||
+              ClauseKind == OpenACCClauseKind::PresentOrCopyIn) &&
+             "readonly: tag only valid on copyin");
+      assert((!IsZero || ClauseKind == OpenACCClauseKind::CopyOut ||
+              ClauseKind == OpenACCClauseKind::PCopyOut ||
+              ClauseKind == OpenACCClauseKind::PresentOrCopyOut ||
+              ClauseKind == OpenACCClauseKind::Create ||
+              ClauseKind == OpenACCClauseKind::PCreate ||
+              ClauseKind == OpenACCClauseKind::PresentOrCreate) &&
+             "zero: tag only valid on copyout/create");
+      Details =
+          VarListDetails{{VarList.begin(), VarList.end()}, IsReadOnly, IsZero};
     }
 
-    void setVarListDetails(llvm::SmallVector<Expr *> &&VarList) {
+    void setVarListDetails(llvm::SmallVector<Expr *> &&VarList, bool IsReadOnly,
+                           bool IsZero) {
       assert((ClauseKind == OpenACCClauseKind::Private ||
               ClauseKind == OpenACCClauseKind::NoCreate ||
               ClauseKind == OpenACCClauseKind::Present ||
               ClauseKind == OpenACCClauseKind::Copy ||
               ClauseKind == OpenACCClauseKind::PCopy ||
               ClauseKind == OpenACCClauseKind::PresentOrCopy ||
+              ClauseKind == OpenACCClauseKind::CopyIn ||
+              ClauseKind == OpenACCClauseKind::PCopyIn ||
+              ClauseKind == OpenACCClauseKind::PresentOrCopyIn ||
+              ClauseKind == OpenACCClauseKind::CopyOut ||
+              ClauseKind == OpenACCClauseKind::PCopyOut ||
+              ClauseKind == OpenACCClauseKind::PresentOrCopyOut ||
+              ClauseKind == OpenACCClauseKind::Create ||
+              ClauseKind == OpenACCClauseKind::PCreate ||
+              ClauseKind == OpenACCClauseKind::PresentOrCreate ||
               ClauseKind == OpenACCClauseKind::FirstPrivate) &&
              "Parsed clause kind does not have a var-list");
-      Details = VarListDetails{std::move(VarList)};
+      assert((!IsReadOnly || ClauseKind == OpenACCClauseKind::CopyIn ||
+              ClauseKind == OpenACCClauseKind::PCopyIn ||
+              ClauseKind == OpenACCClauseKind::PresentOrCopyIn) &&
+             "readonly: tag only valid on copyin");
+      assert((!IsZero || ClauseKind == OpenACCClauseKind::CopyOut ||
+              ClauseKind == OpenACCClauseKind::PCopyOut ||
+              ClauseKind == OpenACCClauseKind::PresentOrCopyOut ||
+              ClauseKind == OpenACCClauseKind::Create ||
+              ClauseKind == OpenACCClauseKind::PCreate ||
+              ClauseKind == OpenACCClauseKind::PresentOrCreate) &&
+             "zero: tag only valid on copyout/create");
+      Details = VarListDetails{std::move(VarList), IsReadOnly, IsZero};
     }
   };
 
diff --git a/clang/lib/AST/OpenACCClause.cpp b/clang/lib/AST/OpenACCClause.cpp
index 2a12a74899cd..f682100e35d3 100644
--- a/clang/lib/AST/OpenACCClause.cpp
+++ b/clang/lib/AST/OpenACCClause.cpp
@@ -186,6 +186,39 @@ OpenACCCopyClause::Create(const ASTContext &C, OpenACCClauseKind Spelling,
       OpenACCCopyClause(Spelling, BeginLoc, LParenLoc, VarList, EndLoc);
 }
 
+OpenACCCopyInClause *
+OpenACCCopyInClause::Create(const ASTContext &C, OpenACCClauseKind Spelling,
+                            SourceLocation BeginLoc, SourceLocation LParenLoc,
+                            bool IsReadOnly, ArrayRef<Expr *> VarList,
+                            SourceLocation EndLoc) {
+  void *Mem =
+      C.Allocate(OpenACCCopyInClause::totalSizeToAlloc<Expr *>(VarList.size()));
+  return new (Mem) OpenACCCopyInClause(Spelling, BeginLoc, LParenLoc,
+                                       IsReadOnly, VarList, EndLoc);
+}
+
+OpenACCCopyOutClause *
+OpenACCCopyOutClause::Create(const ASTContext &C, OpenACCClauseKind Spelling,
+                             SourceLocation BeginLoc, SourceLocation LParenLoc,
+                             bool IsZero, ArrayRef<Expr *> VarList,
+                             SourceLocation EndLoc) {
+  void *Mem = C.Allocate(
+      OpenACCCopyOutClause::totalSizeToAlloc<Expr *>(VarList.size()));
+  return new (Mem) OpenACCCopyOutClause(Spelling, BeginLoc, LParenLoc, IsZero,
+                                        VarList, EndLoc);
+}
+
+OpenACCCreateClause *
+OpenACCCreateClause::Create(const ASTContext &C, OpenACCClauseKind Spelling,
+                            SourceLocation BeginLoc, SourceLocation LParenLoc,
+                            bool IsZero, ArrayRef<Expr *> VarList,
+                            SourceLocation EndLoc) {
+  void *Mem =
+      C.Allocate(OpenACCCreateClause::totalSizeToAlloc<Expr *>(VarList.size()));
+  return new (Mem) OpenACCCreateClause(Spelling, BeginLoc, LParenLoc, IsZero,
+                                       VarList, EndLoc);
+}
+
 //===----------------------------------------------------------------------===//
 //  OpenACC clauses printing methods
 //===----------------------------------------------------------------------===//
@@ -269,3 +302,30 @@ void OpenACCClausePrinter::VisitCopyClause(const OpenACCCopyClause &C) {
                         [&](const Expr *E) { printExpr(E); });
   OS << ")";
 }
+
+void OpenACCClausePrinter::VisitCopyInClause(const OpenACCCopyInClause &C) {
+  OS << C.getClauseKind() << '(';
+  if (C.isReadOnly())
+    OS << "readonly: ";
+  llvm::interleaveComma(C.getVarList(), OS,
+                        [&](const Expr *E) { printExpr(E); });
+  OS << ")";
+}
+
+void OpenACCClausePrinter::VisitCopyOutClause(const OpenACCCopyOutClause &C) {
+  OS << C.getClauseKind() << '(';
+  if (C.isZero())
+    OS << "zero: ";
+  llvm::interleaveComma(C.getVarList(), OS,
+                        [&](const Expr *E) { printExpr(E); });
+  OS << ")";
+}
+
+void OpenACCClausePrinter::VisitCreateClause(const OpenACCCreateClause &C) {
+  OS << C.getClauseKind() << '(';
+  if (C.isZero())
+    OS << "zero: ";
+  llvm::interleaveComma(C.getVarList(), OS,
+                        [&](const Expr *E) { printExpr(E); });
+  OS << ")";
+}
diff --git a/clang/lib/AST/StmtProfile.cpp b/clang/lib/AST/StmtProfile.cpp
index 6fe4f673a59d..b97c351f83db 100644
--- a/clang/lib/AST/StmtProfile.cpp
+++ b/clang/lib/AST/StmtProfile.cpp
@@ -2496,6 +2496,23 @@ void OpenACCClauseProfiler::VisitCopyClause(const OpenACCCopyClause &Clause) {
   for (auto *E : Clause.getVarList())
     Profiler.VisitStmt(E);
 }
+void OpenACCClauseProfiler::VisitCopyInClause(
+    const OpenACCCopyInClause &Clause) {
+  for (auto *E : Clause.getVarList())
+    Profiler.VisitStmt(E);
+}
+
+void OpenACCClauseProfiler::VisitCopyOutClause(
+    const OpenACCCopyOutClause &Clause) {
+  for (auto *E : Clause.getVarList())
+    Profiler.VisitStmt(E);
+}
+
+void OpenACCClauseProfiler::VisitCreateClause(
+    const OpenACCCreateClause &Clause) {
+  for (auto *E : Clause.getVarList())
+    Profiler.VisitStmt(E);
+}
 
 void OpenACCClauseProfiler::VisitSelfClause(const OpenACCSelfClause &Clause) {
   if (Clause.hasConditionExpr())
diff --git a/clang/lib/AST/TextNodeDumper.cpp b/clang/lib/AST/TextNodeDumper.cpp
index d383d9fe1704..5aeeb80e8fe7 100644
--- a/clang/lib/AST/TextNodeDumper.cpp
+++ b/clang/lib/AST/TextNodeDumper.cpp
@@ -413,6 +413,27 @@ void TextNodeDumper::Visit(const OpenACCClause *C) {
       // but print 'clause' here so it is clear what is happening from the dump.
       OS << " clause";
       break;
+    case OpenACCClauseKind::CopyIn:
+    case OpenACCClauseKind::PCopyIn:
+    case OpenACCClauseKind::PresentOrCopyIn:
+      OS << " clause";
+      if (cast<OpenACCCopyInClause>(C)->isReadOnly())
+        OS << " : readonly";
+      break;
+    case OpenACCClauseKind::CopyOut:
+    case OpenACCClauseKind::PCopyOut:
+    case OpenACCClauseKind::PresentOrCopyOut:
+      OS << " clause";
+      if (cast<OpenACCCopyOutClause>(C)->isZero())
+        OS << " : zero";
+      break;
+    case OpenACCClauseKind::Create:
+    case OpenACCClauseKind::PCreate:
+    case OpenACCClauseKind::PresentOrCreate:
+      OS << " clause";
+      if (cast<OpenACCCreateClause>(C)->isZero())
+        OS << " : zero";
+      break;
     default:
       // Nothing to do here.
       break;
diff --git a/clang/lib/Parse/ParseOpenACC.cpp b/clang/lib/Parse/ParseOpenACC.cpp
index bc785fa7136a..c90ae8806f08 100644
--- a/clang/lib/Parse/ParseOpenACC.cpp
+++ b/clang/lib/Parse/ParseOpenACC.cpp
@@ -100,12 +100,18 @@ OpenACCClauseKind getOpenACCClauseKind(Token Tok) {
       .Case("auto", OpenACCClauseKind::Auto)
       .Case("bind", OpenACCClauseKind::Bind)
       .Case("create", OpenACCClauseKind::Create)
+      .Case("pcreate", OpenACCClauseKind::PCreate)
+      .Case("present_or_create", OpenACCClauseKind::PresentOrCreate)
       .Case("collapse", OpenACCClauseKind::Collapse)
       .Case("copy", OpenACCClauseKind::Copy)
       .Case("pcopy", OpenACCClauseKind::PCopy)
       .Case("present_or_copy", OpenACCClauseKind::PresentOrCopy)
       .Case("copyin", OpenACCClauseKind::CopyIn)
+      .Case("pcopyin", OpenACCClauseKind::PCopyIn)
+      .Case("present_or_copyin", OpenACCClauseKind::PresentOrCopyIn)
       .Case("copyout", OpenACCClauseKind::CopyOut)
+      .Case("pcopyout", OpenACCClauseKind::PCopyOut)
+      .Case("present_or_copyout", OpenACCClauseKind::PresentOrCopyOut)
       .Case("default", OpenACCClauseKind::Default)
       .Case("default_async", OpenACCClauseKind::DefaultAsync)
       .Case("delete", OpenACCClauseKind::Delete)
@@ -490,11 +496,17 @@ ClauseParensKind getClauseParensKind(OpenACCDirectiveKind DirKind,
   case OpenACCClauseKind::Default:
   case OpenACCClauseKind::If:
   case OpenACCClauseKind::Create:
+  case OpenACCClauseKind::PCreate:
+  case OpenACCClauseKind::PresentOrCreate:
   case OpenACCClauseKind::Copy:
   case OpenACCClauseKind::PCopy:
   case OpenACCClauseKind::PresentOrCopy:
   case OpenACCClauseKind::CopyIn:
+  case OpenACCClauseKind::PCopyIn:
+  case OpenACCClauseKind::PresentOrCopyIn:
   case OpenACCClauseKind::CopyOut:
+  case OpenACCClauseKind::PCopyOut:
+  case OpenACCClauseKind::PresentOrCopyOut:
   case OpenACCClauseKind::UseDevice:
   case OpenACCClauseKind::NoCreate:
   case OpenACCClauseKind::Present:
@@ -901,16 +913,26 @@ Parser::OpenACCClauseParseResult Parser::ParseOpenACCClauseParams(
       break;
     }
     case OpenACCClauseKind::CopyIn:
-      tryParseAndConsumeSpecialTokenKind(
+    case OpenACCClauseKind::PCopyIn:
+    case OpenACCClauseKind::PresentOrCopyIn: {
+      bool IsReadOnly = tryParseAndConsumeSpecialTokenKind(
           *this, OpenACCSpecialTokenKind::ReadOnly, ClauseKind);
-      ParseOpenACCVarList();
+      ParsedClause.setVarListDetails(ParseOpenACCVarList(), IsReadOnly,
+                                     /*IsZero=*/false);
       break;
+    }
     case OpenACCClauseKind::Create:
+    case OpenACCClauseKind::PCreate:
+    case OpenACCClauseKind::PresentOrCreate:
     case OpenACCClauseKind::CopyOut:
-      tryParseAndConsumeSpecialTokenKind(*this, OpenACCSpecialTokenKind::Zero,
-                                         ClauseKind);
-      ParseOpenACCVarList();
+    case OpenACCClauseKind::PCopyOut:
+    case OpenACCClauseKind::PresentOrCopyOut: {
+      bool IsZero = tryParseAndConsumeSpecialTokenKind(
+          *this, OpenACCSpecialTokenKind::Zero, ClauseKind);
+      ParsedClause.setVarListDetails(ParseOpenACCVarList(),
+                                     /*IsReadOnly=*/false, IsZero);
       break;
+    }
     case OpenACCClauseKind::Reduction:
       // If we're missing a clause-kind (or it is invalid), see if we can parse
       // the var-list anyway.
@@ -941,7 +963,8 @@ Parser::OpenACCClauseParseResult Parser::ParseOpenACCClauseParams(
     case OpenACCClauseKind::NoCreate:
     case OpenACCClauseKind::Present:
     case OpenACCClauseKind::Private:
-      ParsedClause.setVarListDetails(ParseOpenACCVarList());
+      ParsedClause.setVarListDetails(ParseOpenACCVarList(),
+                                     /*IsReadOnly=*/false, /*IsZero=*/false);
       break;
     case OpenACCClauseKind::Collapse: {
       tryParseAndConsumeSpecialTokenKind(*this, OpenACCSpecialTokenKind::Force,
diff --git a/clang/lib/Sema/SemaOpenACC.cpp b/clang/lib/Sema/SemaOpenACC.cpp
index ee73447067b1..4e1a8c1277c5 100644
--- a/clang/lib/Sema/SemaOpenACC.cpp
+++ b/clang/lib/Sema/SemaOpenACC.cpp
@@ -450,6 +450,69 @@ SemaOpenACC::ActOnClause(ArrayRef<const OpenACCClause *> ExistingClauses,
         getASTContext(), Clause.getClauseKind(), Clause.getBeginLoc(),
         Clause.getLParenLoc(), Clause.getVarList(), Clause.getEndLoc());
   }
+  case OpenACCClauseKind::PresentOrCopyIn:
+  case OpenACCClauseKind::PCopyIn:
+    Diag(Clause.getBeginLoc(), diag::warn_acc_deprecated_alias_name)
+        << Clause.getClauseKind() << OpenACCClauseKind::CopyIn;
+    LLVM_FALLTHROUGH;
+  case OpenACCClauseKind::CopyIn: {
+    // Restrictions only properly implemented on 'compute' constructs, and
+    // 'compute' constructs are the only construct that can do anything with
+    // this yet, so skip/treat as unimplemented in this case.
+    if (!isOpenACCComputeDirectiveKind(Clause.getDirectiveKind()))
+      break;
+
+    // ActOnVar ensured that everything is a valid variable reference, so there
+    // really isn't anything to do here. GCC does some duplicate-finding, though
+    // it isn't apparent in the standard where this is justified.
+
+    return OpenACCCopyInClause::Create(
+        getASTContext(), Clause.getClauseKind(), Clause.getBeginLoc(),
+        Clause.getLParenLoc(), Clause.isReadOnly(), Clause.getVarList(),
+        Clause.getEndLoc());
+  }
+  case OpenACCClauseKind::PresentOrCopyOut:
+  case OpenACCClauseKind::PCopyOut:
+    Diag(Clause.getBeginLoc(), diag::warn_acc_deprecated_alias_name)
+        << Clause.getClauseKind() << OpenACCClauseKind::CopyOut;
+    LLVM_FALLTHROUGH;
+  case OpenACCClauseKind::CopyOut: {
+    // Restrictions only properly implemented on 'compute' constructs, and
+    // 'compute' constructs are the only construct that can do anything with
+    // this yet, so skip/treat as unimplemented in this case.
+    if (!isOpenACCComputeDirectiveKind(Clause.getDirectiveKind()))
+      break;
+
+    // ActOnVar ensured that everything is a valid variable reference, so there
+    // really isn't anything to do here. GCC does some duplicate-finding, though
+    // it isn't apparent in the standard where this is justified.
+
+    return OpenACCCopyOutClause::Create(
+        getASTContext(), Clause.getClauseKind(), Clause.getBeginLoc(),
+        Clause.getLParenLoc(), Clause.isZero(), Clause.getVarList(),
+        Clause.getEndLoc());
+  }
+  case OpenACCClauseKind::PresentOrCreate:
+  case OpenACCClauseKind::PCreate:
+    Diag(Clause.getBeginLoc(), diag::warn_acc_deprecated_alias_name)
+        << Clause.getClauseKind() << OpenACCClauseKind::Create;
+    LLVM_FALLTHROUGH;
+  case OpenACCClauseKind::Create: {
+    // Restrictions only properly implemented on 'compute' constructs, and
+    // 'compute' constructs are the only construct that can do anything with
+    // this yet, so skip/treat as unimplemented in this case.
+    if (!isOpenACCComputeDirectiveKind(Clause.getDirectiveKind()))
+      break;
+
+    // ActOnVar ensured that everything is a valid variable reference, so there
+    // really isn't anything to do here. GCC does some duplicate-finding, though
+    // it isn't apparent in the standard where this is justified.
+
+    return OpenACCCreateClause::Create(getASTContext(), Clause.getClauseKind(),
+                                       Clause.getBeginLoc(),
+                                       Clause.getLParenLoc(), Clause.isZero(),
+                                       Clause.getVarList(), Clause.getEndLoc());
+  }
   default:
     break;
   }
diff --git a/clang/lib/Sema/TreeTransform.h b/clang/lib/Sema/TreeTransform.h
index 810247f8700f..edf557038746 100644
--- a/clang/lib/Sema/TreeTransform.h
+++ b/clang/lib/Sema/TreeTransform.h
@@ -11223,7 +11223,8 @@ void OpenACCClauseTransform<Derived>::VisitNumGangsClause(
 template <typename Derived>
 void OpenACCClauseTransform<Derived>::VisitPrivateClause(
     const OpenACCPrivateClause &C) {
-  ParsedClause.setVarListDetails(VisitVarList(C.getVarList()));
+  ParsedClause.setVarListDetails(VisitVarList(C.getVarList()),
+                                 /*IsReadOnly=*/false, /*IsZero=*/false);
 
   NewClause = OpenACCPrivateClause::Create(
       Self.getSema().getASTContext(), ParsedClause.getBeginLoc(),
@@ -11234,7 +11235,8 @@ void OpenACCClauseTransform<Derived>::VisitPrivateClause(
 template <typename Derived>
 void OpenACCClauseTransform<Derived>::VisitFirstPrivateClause(
     const OpenACCFirstPrivateClause &C) {
-  ParsedClause.setVarListDetails(VisitVarList(C.getVarList()));
+  ParsedClause.setVarListDetails(VisitVarList(C.getVarList()),
+                                 /*IsReadOnly=*/false, /*IsZero=*/false);
 
   NewClause = OpenACCFirstPrivateClause::Create(
       Self.getSema().getASTContext(), ParsedClause.getBeginLoc(),
@@ -11245,7 +11247,8 @@ void OpenACCClauseTransform<Derived>::VisitFirstPrivateClause(
 template <typename Derived>
 void OpenACCClauseTransform<Derived>::VisitNoCreateClause(
     const OpenACCNoCreateClause &C) {
-  ParsedClause.setVarListDetails(VisitVarList(C.getVarList()));
+  ParsedClause.setVarListDetails(VisitVarList(C.getVarList()),
+                                 /*IsReadOnly=*/false, /*IsZero=*/false);
 
   NewClause = OpenACCNoCreateClause::Create(
       Self.getSema().getASTContext(), ParsedClause.getBeginLoc(),
@@ -11256,7 +11259,8 @@ void OpenACCClauseTransform<Derived>::VisitNoCreateClause(
 template <typename Derived>
 void OpenACCClauseTransform<Derived>::VisitPresentClause(
     const OpenACCPresentClause &C) {
-  ParsedClause.setVarListDetails(VisitVarList(C.getVarList()));
+  ParsedClause.setVarListDetails(VisitVarList(C.getVarList()),
+                                 /*IsReadOnly=*/false, /*IsZero=*/false);
 
   NewClause = OpenACCPresentClause::Create(
       Self.getSema().getASTContext(), ParsedClause.getBeginLoc(),
@@ -11267,7 +11271,8 @@ void OpenACCClauseTransform<Derived>::VisitPresentClause(
 template <typename Derived>
 void OpenACCClauseTransform<Derived>::VisitCopyClause(
     const OpenACCCopyClause &C) {
-  ParsedClause.setVarListDetails(VisitVarList(C.getVarList()));
+  ParsedClause.setVarListDetails(VisitVarList(C.getVarList()),
+                                 /*IsReadOnly=*/false, /*IsZero=*/false);
 
   NewClause = OpenACCCopyClause::Create(
       Self.getSema().getASTContext(), ParsedClause.getClauseKind(),
@@ -11276,6 +11281,45 @@ void OpenACCClauseTransform<Derived>::VisitCopyClause(
 }
 
 template <typename Derived>
+void OpenACCClauseTransform<Derived>::VisitCopyInClause(
+    const OpenACCCopyInClause &C) {
+  ParsedClause.setVarListDetails(VisitVarList(C.getVarList()), C.isReadOnly(),
+                                 /*IsZero=*/false);
+
+  NewClause = OpenACCCopyInClause::Create(
+      Self.getSema().getASTContext(), ParsedClause.getClauseKind(),
+      ParsedClause.getBeginLoc(), ParsedClause.getLParenLoc(),
+      ParsedClause.isReadOnly(), ParsedClause.getVarList(),
+      ParsedClause.getEndLoc());
+}
+
+template <typename Derived>
+void OpenACCClauseTransform<Derived>::VisitCopyOutClause(
+    const OpenACCCopyOutClause &C) {
+  ParsedClause.setVarListDetails(VisitVarList(C.getVarList()),
+                                 /*IsReadOnly=*/false, C.isZero());
+
+  NewClause = OpenACCCopyOutClause::Create(
+      Self.getSema().getASTContext(), ParsedClause.getClauseKind(),
+      ParsedClause.getBeginLoc(), ParsedClause.getLParenLoc(),
+      ParsedClause.isZero(), ParsedClause.getVarList(),
+      ParsedClause.getEndLoc());
+}
+
+template <typename Derived>
+void OpenACCClauseTransform<Derived>::VisitCreateClause(
+    const OpenACCCreateClause &C) {
+  ParsedClause.setVarListDetails(VisitVarList(C.getVarList()),
+                                 /*IsReadOnly=*/false, C.isZero());
+
+  NewClause = OpenACCCreateClause::Create(
+      Self.getSema().getASTContext(), ParsedClause.getClauseKind(),
+      ParsedClause.getBeginLoc(), ParsedClause.getLParenLoc(),
+      ParsedClause.isZero(), ParsedClause.getVarList(),
+      ParsedClause.getEndLoc());
+}
+
+template <typename Derived>
 void OpenACCClauseTransform<Derived>::VisitNumWorkersClause(
     const OpenACCNumWorkersClause &C) {
   Expr *IntExpr = const_cast<Expr *>(C.getIntExpr());
diff --git a/clang/lib/Serialization/ASTReader.cpp b/clang/lib/Serialization/ASTReader.cpp
index 9f4397906b8c..6d121f4ab80e 100644
--- a/clang/lib/Serialization/ASTReader.cpp
+++ b/clang/lib/Serialization/ASTReader.cpp
@@ -11861,6 +11861,34 @@ OpenACCClause *ASTRecordReader::readOpenACCClause() {
     return OpenACCCopyClause::Create(getContext(), ClauseKind, BeginLoc,
                                      LParenLoc, VarList, EndLoc);
   }
+  case OpenACCClauseKind::CopyIn:
+  case OpenACCClauseKind::PCopyIn:
+  case OpenACCClauseKind::PresentOrCopyIn: {
+    SourceLocation LParenLoc = readSourceLocation();
+    bool IsReadOnly = readBool();
+    llvm::SmallVector<Expr *> VarList = readOpenACCVarList();
+    return OpenACCCopyInClause::Create(getContext(), ClauseKind, BeginLoc,
+                                       LParenLoc, IsReadOnly, VarList, EndLoc);
+  }
+  case OpenACCClauseKind::CopyOut:
+  case OpenACCClauseKind::PCopyOut:
+  case OpenACCClauseKind::PresentOrCopyOut: {
+    SourceLocation LParenLoc = readSourceLocation();
+    bool IsZero = readBool();
+    llvm::SmallVector<Expr *> VarList = readOpenACCVarList();
+    return OpenACCCopyOutClause::Create(getContext(), ClauseKind, BeginLoc,
+                                        LParenLoc, IsZero, VarList, EndLoc);
+  }
+  case OpenACCClauseKind::Create:
+  case OpenACCClauseKind::PCreate:
+  case OpenACCClauseKind::PresentOrCreate: {
+    SourceLocation LParenLoc = readSourceLocation();
+    bool IsZero = readBool();
+    llvm::SmallVector<Expr *> VarList = readOpenACCVarList();
+    return OpenACCCreateClause::Create(getContext(), ClauseKind, BeginLoc,
+                                       LParenLoc, IsZero, VarList, EndLoc);
+  }
+
   case OpenACCClauseKind::Finalize:
   case OpenACCClauseKind::IfPresent:
   case OpenACCClauseKind::Seq:
@@ -11878,9 +11906,6 @@ OpenACCClause *ASTRecordReader::readOpenACCClause() {
   case OpenACCClauseKind::DeviceResident:
   case OpenACCClauseKind::Host:
   case OpenACCClauseKind::Link:
-  case OpenACCClauseKind::CopyOut:
-  case OpenACCClauseKind::CopyIn:
-  case OpenACCClauseKind::Create:
   case OpenACCClauseKind::Reduction:
   case OpenACCClauseKind::Collapse:
   case OpenACCClauseKind::Bind:
diff --git a/clang/lib/Serialization/ASTWriter.cpp b/clang/lib/Serialization/ASTWriter.cpp
index 118e4395d971..9712ed6e8391 100644
--- a/clang/lib/Serialization/ASTWriter.cpp
+++ b/clang/lib/Serialization/ASTWriter.cpp
@@ -7813,6 +7813,33 @@ void ASTRecordWriter::writeOpenACCClause(const OpenACCClause *C) {
     writeOpenACCVarList(CC);
     return;
   }
+  case OpenACCClauseKind::CopyIn:
+  case OpenACCClauseKind::PCopyIn:
+  case OpenACCClauseKind::PresentOrCopyIn: {
+    const auto *CIC = cast<OpenACCCopyInClause>(C);
+    writeSourceLocation(CIC->getLParenLoc());
+    writeBool(CIC->isReadOnly());
+    writeOpenACCVarList(CIC);
+    return;
+  }
+  case OpenACCClauseKind::CopyOut:
+  case OpenACCClauseKind::PCopyOut:
+  case OpenACCClauseKind::PresentOrCopyOut: {
+    const auto *COC = cast<OpenACCCopyOutClause>(C);
+    writeSourceLocation(COC->getLParenLoc());
+    writeBool(COC->isZero());
+    writeOpenACCVarList(COC);
+    return;
+  }
+  case OpenACCClauseKind::Create:
+  case OpenACCClauseKind::PCreate:
+  case OpenACCClauseKind::PresentOrCreate: {
+    const auto *CC = cast<OpenACCCreateClause>(C);
+    writeSourceLocation(CC->getLParenLoc());
+    writeBool(CC->isZero());
+    writeOpenACCVarList(CC);
+    return;
+  }
 
   case OpenACCClauseKind::Finalize:
   case OpenACCClauseKind::IfPresent:
@@ -7831,9 +7858,6 @@ void ASTRecordWriter::writeOpenACCClause(const OpenACCClause *C) {
   case OpenACCClauseKind::DeviceResident:
   case OpenACCClauseKind::Host:
   case OpenACCClauseKind::Link:
-  case OpenACCClauseKind::CopyOut:
-  case OpenACCClauseKind::CopyIn:
-  case OpenACCClauseKind::Create:
   case OpenACCClauseKind::Reduction:
   case OpenACCClauseKind::Collapse:
   case OpenACCClauseKind::Bind:
diff --git a/clang/test/AST/ast-print-openacc-compute-construct.cpp b/clang/test/AST/ast-print-openacc-compute-construct.cpp
index 895660ae1641..112f328f5cb9 100644
--- a/clang/test/AST/ast-print-openacc-compute-construct.cpp
+++ b/clang/test/AST/ast-print-openacc-compute-construct.cpp
@@ -1,4 +1,4 @@
-// RUN: %clang_cc1 -fopenacc -ast-print %s -o - | FileCheck %s
+// RUN: %clang_cc1 -fopenacc -Wno-openacc-deprecated-clause-alias -ast-print %s -o - | FileCheck %s
 
 void foo() {
   int i;
@@ -53,5 +53,17 @@ void foo() {
 // CHECK: #pragma acc parallel no_create(i, array[1], array, array[1:2]) present(i, array[1], array, array[1:2])
 #pragma acc parallel no_create(i, array[1], array, array[1:2]) present(i, array[1], array, array[1:2])
   while(true);
+
+// CHECK: #pragma acc parallel copyin(i, array[1], array, array[1:2]) pcopyin(readonly: i, array[1], array, array[1:2]) present_or_copyin(i, array[1], array, array[1:2])
+#pragma acc parallel copyin(i, array[1], array, array[1:2]) pcopyin(readonly:i, array[1], array, array[1:2]) present_or_copyin(i, array[1], array, array[1:2])
+  while(true);
+
+// CHECK: #pragma acc parallel copyout(i, array[1], array, array[1:2]) pcopyout(zero: i, array[1], array, array[1:2]) present_or_copyout(i, array[1], array, array[1:2])
+#pragma acc parallel copyout(i, array[1], array, array[1:2]) pcopyout(zero: i, array[1], array, array[1:2]) present_or_copyout(i, array[1], array, array[1:2])
+  while(true);
+
+// CHECK: #pragma acc parallel create(i, array[1], array, array[1:2]) pcreate(zero: i, array[1], array, array[1:2]) present_or_create(i, array[1], array, array[1:2])
+#pragma acc parallel create(i, array[1], array, array[1:2]) pcreate(zero: i, array[1], array, array[1:2]) present_or_create(i, array[1], array, array[1:2])
+  while(true);
 }
 
diff --git a/clang/test/ParserOpenACC/parse-clauses.c b/clang/test/ParserOpenACC/parse-clauses.c
index c1287c329bc8..65247e4db63e 100644
--- a/clang/test/ParserOpenACC/parse-clauses.c
+++ b/clang/test/ParserOpenACC/parse-clauses.c
@@ -665,161 +665,158 @@ void VarListClauses() {
 #pragma acc serial device(s.array[s.value : 5], s.value), seq
   for(;;){}
 
-  // expected-error@+3{{expected ','}}
-  // expected-warning@+2{{OpenACC clause 'copyout' not yet implemented, clause ignored}}
+  // expected-error@+2{{expected ','}}
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial copyout(s.array[s.value] s.array[s.value :5] ), seq
   for(;;){}
 
-  // expected-warning@+2{{OpenACC clause 'copyout' not yet implemented, clause ignored}}
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial copyout(s.array[s.value : 5], s.value), seq
   for(;;){}
 
-  // expected-warning@+2{{OpenACC clause 'copyout' not yet implemented, clause ignored}}
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial copyout(zero:s.array[s.value : 5], s.value), seq
   for(;;){}
 
-  // expected-warning@+2{{OpenACC clause 'copyout' not yet implemented, clause ignored}}
+  // expected-warning@+1{{OpenACC clause name 'pcopyout' is a deprecated clause name and is now an alias for 'copyout'}}
+#pragma acc serial pcopyout(s.array[s.value : 5], s.value)
+  for(;;){}
+
+  // expected-warning@+1{{OpenACC clause name 'present_or_copyout' is a deprecated clause name and is now an alias for 'copyout'}}
+#pragma acc serial present_or_copyout(zero:s.array[s.value : 5], s.value)
+  for(;;){}
+
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial copyout(zero : s.array[s.value : 5], s.value), seq
   for(;;){}
 
-  // expected-error@+4{{use of undeclared identifier 'zero'}}
-  // expected-error@+3{{expected ','}}
-  // expected-warning@+2{{OpenACC clause 'copyout' not yet implemented, clause ignored}}
+  // expected-error@+3{{use of undeclared identifier 'zero'}}
+  // expected-error@+2{{expected ','}}
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial copyout(zero s.array[s.value : 5], s.value), seq
   for(;;){}
 
-  // expected-error@+3{{invalid tag 'readonly' on 'copyout' clause}}
-  // expected-warning@+2{{OpenACC clause 'copyout' not yet implemented, clause ignored}}
+  // expected-error@+2{{invalid tag 'readonly' on 'copyout' clause}}
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial copyout(readonly:s.array[s.value : 5], s.value), seq
   for(;;){}
 
-  // expected-error@+3{{invalid tag 'invalid' on 'copyout' clause}}
-  // expected-warning@+2{{OpenACC clause 'copyout' not yet implemented, clause ignored}}
+  // expected-error@+2{{invalid tag 'invalid' on 'copyout' clause}}
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial copyout(invalid:s.array[s.value : 5], s.value), seq
   for(;;){}
 
-  // expected-error@+3{{invalid tag 'invalid' on 'copyout' clause}}
-  // expected-warning@+2{{OpenACC clause 'copyout' not yet implemented, clause ignored}}
+  // expected-error@+2{{invalid tag 'invalid' on 'copyout' clause}}
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial copyout(invalid:s.array[s.value : 5], s.value), seq
   for(;;){}
 
-  // expected-error@+4{{use of undeclared identifier 'invalid'}}
-  // expected-error@+3{{expected ','}}
-  // expected-warning@+2{{OpenACC clause 'copyout' not yet implemented, clause ignored}}
+  // expected-error@+3{{use of undeclared identifier 'invalid'}}
+  // expected-error@+2{{expected ','}}
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial copyout(invalid s.array[s.value : 5], s.value), seq
   for(;;){}
 
-  // expected-error@+3{{expected ','}}
-  // expected-warning@+2{{OpenACC clause 'create' not yet implemented, clause ignored}}
+  // expected-error@+2{{expected ','}}
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial create(s.array[s.value] s.array[s.value :5] ), seq
   for(;;){}
 
-  // expected-warning@+2{{OpenACC clause 'create' not yet implemented, clause ignored}}
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial create(s.array[s.value : 5], s.value), seq
   for(;;){}
 
-  // expected-warning@+2{{OpenACC clause 'create' not yet implemented, clause ignored}}
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial create(zero:s.array[s.value : 5], s.value), seq
   for(;;){}
 
-  // expected-warning@+2{{OpenACC clause 'create' not yet implemented, clause ignored}}
+  // expected-warning@+1{{OpenACC clause name 'pcreate' is a deprecated clause name and is now an alias for 'create'}}
+#pragma acc serial pcreate(s.array[s.value : 5], s.value)
+  for(;;){}
+
+  // expected-warning@+1{{OpenACC clause name 'present_or_create' is a deprecated clause name and is now an alias for 'create'}}
+#pragma acc serial present_or_create(zero:s.array[s.value : 5], s.value)
+  for(;;){}
+
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial create(zero : s.array[s.value : 5], s.value), seq
   for(;;){}
 
-  // expected-error@+4{{use of undeclared identifier 'zero'}}
-  // expected-error@+3{{expected ','}}
-  // expected-warning@+2{{OpenACC clause 'create' not yet implemented, clause ignored}}
+  // expected-error@+3{{use of undeclared identifier 'zero'}}
+  // expected-error@+2{{expected ','}}
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial create(zero s.array[s.value : 5], s.value), seq
   for(;;){}
 
-  // expected-error@+3{{invalid tag 'readonly' on 'create' clause}}
-  // expected-warning@+2{{OpenACC clause 'create' not yet implemented, clause ignored}}
+  // expected-error@+2{{invalid tag 'readonly' on 'create' clause}}
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial create(readonly:s.array[s.value : 5], s.value), seq
   for(;;){}
 
-  // expected-error@+3{{invalid tag 'invalid' on 'create' clause}}
-  // expected-warning@+2{{OpenACC clause 'create' not yet implemented, clause ignored}}
+  // expected-error@+2{{invalid tag 'invalid' on 'create' clause}}
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial create(invalid:s.array[s.value : 5], s.value), seq
   for(;;){}
 
-  // expected-error@+3{{invalid tag 'invalid' on 'create' clause}}
-  // expected-warning@+2{{OpenACC clause 'create' not yet implemented, clause ignored}}
+  // expected-error@+2{{invalid tag 'invalid' on 'create' clause}}
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial create(invalid:s.array[s.value : 5], s.value), seq
   for(;;){}
 
-  // expected-error@+4{{use of undeclared identifier 'invalid'}}
-  // expected-error@+3{{expected ','}}
-  // expected-warning@+2{{OpenACC clause 'create' not yet implemented, clause ignored}}
+  // expected-error@+3{{use of undeclared identifier 'invalid'}}
+  // expected-error@+2{{expected ','}}
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial create(invalid s.array[s.value : 5], s.value), seq
   for(;;){}
 
-  // expected-error@+3{{expected ','}}
-  // expected-warning@+2{{OpenACC clause 'copyin' not yet implemented, clause ignored}}
+  // expected-error@+2{{expected ','}}
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial copyin(s.array[s.value] s.array[s.value :5] ), seq
   for(;;){}
 
-  // expected-warning@+2{{OpenACC clause 'copyin' not yet implemented, clause ignored}}
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial copyin(s.array[s.value : 5], s.value), seq
   for(;;){}
 
-  // expected-warning@+2{{OpenACC clause 'copyin' not yet implemented, clause ignored}}
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial copyin(readonly:s.array[s.value : 5], s.value), seq
   for(;;){}
 
-  // expected-warning@+2{{OpenACC clause 'copyin' not yet implemented, clause ignored}}
+  // expected-warning@+1{{OpenACC clause name 'pcopyin' is a deprecated clause name and is now an alias for 'copyin'}}
+#pragma acc serial pcopyin(s.array[s.value : 5], s.value)
+  for(;;){}
+
+  // expected-warning@+1{{OpenACC clause name 'present_or_copyin' is a deprecated clause name and is now an alias for 'copyin'}}
+#pragma acc serial present_or_copyin(readonly:s.array[s.value : 5], s.value)
+  for(;;){}
+
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial copyin(readonly : s.array[s.value : 5], s.value), seq
   for(;;){}
 
-  // expected-error@+4{{use of undeclared identifier 'readonly'}}
-  // expected-error@+3{{expected ','}}
-  // expected-warning@+2{{OpenACC clause 'copyin' not yet implemented, clause ignored}}
+  // expected-error@+3{{use of undeclared identifier 'readonly'}}
+  // expected-error@+2{{expected ','}}
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial copyin(readonly s.array[s.value : 5], s.value), seq
   for(;;){}
 
-  // expected-error@+3{{invalid tag 'zero' on 'copyin' clause}}
-  // expected-warning@+2{{OpenACC clause 'copyin' not yet implemented, clause ignored}}
+  // expected-error@+2{{invalid tag 'zero' on 'copyin' clause}}
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial copyin(zero :s.array[s.value : 5], s.value), seq
   for(;;){}
 
-  // expected-error@+3{{invalid tag 'invalid' on 'copyin' clause}}
-  // expected-warning@+2{{OpenACC clause 'copyin' not yet implemented, clause ignored}}
+  // expected-error@+2{{invalid tag 'invalid' on 'copyin' clause}}
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial copyin(invalid:s.array[s.value : 5], s.value), seq
   for(;;){}
 
-  // expected-error@+3{{invalid tag 'invalid' on 'copyin' clause}}
-  // expected-warning@+2{{OpenACC clause 'copyin' not yet implemented, clause ignored}}
+  // expected-error@+2{{invalid tag 'invalid' on 'copyin' clause}}
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial copyin(invalid:s.array[s.value : 5], s.value), seq
   for(;;){}
 
-  // expected-error@+4{{use of undeclared identifier 'invalid'}}
-  // expected-error@+3{{expected ','}}
-  // expected-warning@+2{{OpenACC clause 'copyin' not yet implemented, clause ignored}}
+  // expected-error@+3{{use of undeclared identifier 'invalid'}}
+  // expected-error@+2{{expected ','}}
   // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
 #pragma acc serial copyin(invalid s.array[s.value : 5], s.value), seq
   for(;;){}
diff --git a/clang/test/SemaOpenACC/compute-construct-copyin-clause.c b/clang/test/SemaOpenACC/compute-construct-copyin-clause.c
new file mode 100644
index 000000000000..6f200b357f52
--- /dev/null
+++ b/clang/test/SemaOpenACC/compute-construct-copyin-clause.c
@@ -0,0 +1,68 @@
+// RUN: %clang_cc1 %s -fopenacc -verify
+
+typedef struct IsComplete {
+  struct S { int A; } CompositeMember;
+  int ScalarMember;
+  float ArrayMember[5];
+  void *PointerMember;
+} Complete;
+void uses(int IntParam, short *PointerParam, float ArrayParam[5], Complete CompositeParam) {
+  int LocalInt;
+  short *LocalPointer;
+  float LocalArray[5];
+  Complete LocalComposite;
+  // Check Appertainment:
+#pragma acc parallel copyin(LocalInt)
+  while(1);
+#pragma acc serial copyin(LocalInt)
+  while(1);
+#pragma acc kernels copyin(LocalInt)
+  while(1);
+
+  // expected-warning@+1{{OpenACC clause name 'pcopyin' is a deprecated clause name and is now an alias for 'copyin'}}
+#pragma acc parallel pcopyin(LocalInt)
+  while(1);
+
+  // expected-warning@+1{{OpenACC clause name 'present_or_copyin' is a deprecated clause name and is now an alias for 'copyin'}}
+#pragma acc parallel present_or_copyin(LocalInt)
+  while(1);
+
+  // Valid cases:
+#pragma acc parallel copyin(LocalInt, LocalPointer, LocalArray)
+  while(1);
+#pragma acc parallel copyin(LocalArray[2:1])
+  while(1);
+#pragma acc parallel copyin(readonly:LocalArray[2:1])
+  while(1);
+
+#pragma acc parallel copyin(LocalComposite.ScalarMember, LocalComposite.ScalarMember)
+  while(1);
+
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel copyin(1 + IntParam)
+  while(1);
+
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel copyin(+IntParam)
+  while(1);
+
+  // expected-error@+1{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is not an array}}
+#pragma acc parallel copyin(PointerParam[2:])
+  while(1);
+
+  // expected-error@+1{{OpenACC sub-array specified range [2:5] would be out of the range of the subscripted array size of 5}}
+#pragma acc parallel copyin(ArrayParam[2:5])
+  while(1);
+
+  // expected-error@+2{{OpenACC sub-array specified range [2:5] would be out of the range of the subscripted array size of 5}}
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel copyin((float*)ArrayParam[2:5])
+  while(1);
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel copyin((float)ArrayParam[2])
+  while(1);
+  // expected-error@+2{{invalid tag 'invalid' on 'copyin' clause}}
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel copyin(invalid:(float)ArrayParam[2])
+  while(1);
+}
diff --git a/clang/test/SemaOpenACC/compute-construct-copyin-clause.cpp b/clang/test/SemaOpenACC/compute-construct-copyin-clause.cpp
new file mode 100644
index 000000000000..79275e701161
--- /dev/null
+++ b/clang/test/SemaOpenACC/compute-construct-copyin-clause.cpp
@@ -0,0 +1,112 @@
+// RUN: %clang_cc1 %s -fopenacc -verify
+
+enum SomeE{};
+typedef struct IsComplete {
+  struct S { int A; } CompositeMember;
+  int ScalarMember;
+  float ArrayMember[5];
+  SomeE EnumMember;
+  char *PointerMember;
+} Complete;
+
+void uses(int IntParam, char *PointerParam, float ArrayParam[5], Complete CompositeParam, int &IntParamRef) {
+  int LocalInt;
+  char *LocalPointer;
+  float LocalArray[5];
+  // Check Appertainment:
+#pragma acc parallel copyin(LocalInt)
+  while(1);
+#pragma acc serial copyin(LocalInt)
+  while(1);
+#pragma acc kernels copyin(LocalInt)
+  while(1);
+
+  // Valid cases:
+#pragma acc parallel copyin(LocalInt, LocalPointer, LocalArray)
+  while(1);
+#pragma acc parallel copyin(LocalArray[2:1])
+  while(1);
+
+  Complete LocalComposite2;
+#pragma acc parallel copyin(LocalComposite2.ScalarMember, LocalComposite2.ScalarMember)
+  while(1);
+
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel copyin(1 + IntParam)
+  while(1);
+
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel copyin(+IntParam)
+  while(1);
+
+  // expected-error@+1{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is not an array}}
+#pragma acc parallel copyin(PointerParam[2:])
+  while(1);
+
+  // expected-error@+1{{OpenACC sub-array specified range [2:5] would be out of the range of the subscripted array size of 5}}
+#pragma acc parallel copyin(ArrayParam[2:5])
+  while(1);
+
+  // expected-error@+2{{OpenACC sub-array specified range [2:5] would be out of the range of the subscripted array size of 5}}
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel copyin((float*)ArrayParam[2:5])
+  while(1);
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel copyin((float)ArrayParam[2])
+  while(1);
+}
+
+template<typename T, unsigned I, typename V>
+void TemplUses(T t, T (&arrayT)[I], V TemplComp) {
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel copyin(+t)
+  while(true);
+
+  // NTTP's are only valid if it is a reference to something.
+  // expected-error@+2{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+  // expected-note@#TEMPL_USES_INST{{in instantiation of}}
+#pragma acc parallel copyin(I)
+  while(true);
+
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel copyin(t, I)
+  while(true);
+
+#pragma acc parallel copyin(arrayT)
+  while(true);
+
+#pragma acc parallel copyin(TemplComp)
+  while(true);
+
+#pragma acc parallel copyin(TemplComp.PointerMember[5])
+  while(true);
+ int *Pointer;
+#pragma acc parallel copyin(Pointer[:I])
+  while(true);
+#pragma acc parallel copyin(Pointer[:t])
+  while(true);
+  // expected-error@+1{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is not an array}}
+#pragma acc parallel copyin(Pointer[1:])
+  while(true);
+}
+
+template<unsigned I, auto &NTTP_REF>
+void NTTP() {
+  // NTTP's are only valid if it is a reference to something.
+  // expected-error@+2{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+  // expected-note@#NTTP_INST{{in instantiation of}}
+#pragma acc parallel copyin(I)
+  while(true);
+
+#pragma acc parallel copyin(NTTP_REF)
+  while(true);
+}
+
+void Inst() {
+  static constexpr int NTTP_REFed = 1;
+  int i;
+  int Arr[5];
+  Complete C;
+  TemplUses(i, Arr, C); // #TEMPL_USES_INST
+  NTTP<5, NTTP_REFed>(); // #NTTP_INST
+}
diff --git a/clang/test/SemaOpenACC/compute-construct-copyout-clause.c b/clang/test/SemaOpenACC/compute-construct-copyout-clause.c
new file mode 100644
index 000000000000..38a50f8373e8
--- /dev/null
+++ b/clang/test/SemaOpenACC/compute-construct-copyout-clause.c
@@ -0,0 +1,68 @@
+// RUN: %clang_cc1 %s -fopenacc -verify
+
+typedef struct IsComplete {
+  struct S { int A; } CompositeMember;
+  int ScalarMember;
+  float ArrayMember[5];
+  void *PointerMember;
+} Complete;
+void uses(int IntParam, short *PointerParam, float ArrayParam[5], Complete CompositeParam) {
+  int LocalInt;
+  short *LocalPointer;
+  float LocalArray[5];
+  Complete LocalComposite;
+  // Check Appertainment:
+#pragma acc parallel copyout(LocalInt)
+  while(1);
+#pragma acc serial copyout(LocalInt)
+  while(1);
+#pragma acc kernels copyout(LocalInt)
+  while(1);
+
+  // expected-warning@+1{{OpenACC clause name 'pcopyout' is a deprecated clause name and is now an alias for 'copyout'}}
+#pragma acc parallel pcopyout(LocalInt)
+  while(1);
+
+  // expected-warning@+1{{OpenACC clause name 'present_or_copyout' is a deprecated clause name and is now an alias for 'copyout'}}
+#pragma acc parallel present_or_copyout(LocalInt)
+  while(1);
+
+  // Valid cases:
+#pragma acc parallel copyout(LocalInt, LocalPointer, LocalArray)
+  while(1);
+#pragma acc parallel copyout(LocalArray[2:1])
+  while(1);
+#pragma acc parallel copyout(zero:LocalArray[2:1])
+  while(1);
+
+#pragma acc parallel copyout(LocalComposite.ScalarMember, LocalComposite.ScalarMember)
+  while(1);
+
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel copyout(1 + IntParam)
+  while(1);
+
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel copyout(+IntParam)
+  while(1);
+
+  // expected-error@+1{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is not an array}}
+#pragma acc parallel copyout(PointerParam[2:])
+  while(1);
+
+  // expected-error@+1{{OpenACC sub-array specified range [2:5] would be out of the range of the subscripted array size of 5}}
+#pragma acc parallel copyout(ArrayParam[2:5])
+  while(1);
+
+  // expected-error@+2{{OpenACC sub-array specified range [2:5] would be out of the range of the subscripted array size of 5}}
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel copyout((float*)ArrayParam[2:5])
+  while(1);
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel copyout((float)ArrayParam[2])
+  while(1);
+  // expected-error@+2{{invalid tag 'invalid' on 'copyout' clause}}
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel copyout(invalid:(float)ArrayParam[2])
+  while(1);
+}
diff --git a/clang/test/SemaOpenACC/compute-construct-copyout-clause.cpp b/clang/test/SemaOpenACC/compute-construct-copyout-clause.cpp
new file mode 100644
index 000000000000..3d05a5670092
--- /dev/null
+++ b/clang/test/SemaOpenACC/compute-construct-copyout-clause.cpp
@@ -0,0 +1,112 @@
+// RUN: %clang_cc1 %s -fopenacc -verify
+
+enum SomeE{};
+typedef struct IsComplete {
+  struct S { int A; } CompositeMember;
+  int ScalarMember;
+  float ArrayMember[5];
+  SomeE EnumMember;
+  char *PointerMember;
+} Complete;
+
+void uses(int IntParam, char *PointerParam, float ArrayParam[5], Complete CompositeParam, int &IntParamRef) {
+  int LocalInt;
+  char *LocalPointer;
+  float LocalArray[5];
+  // Check Appertainment:
+#pragma acc parallel copyout(LocalInt)
+  while(1);
+#pragma acc serial copyout(LocalInt)
+  while(1);
+#pragma acc kernels copyout(LocalInt)
+  while(1);
+
+  // Valid cases:
+#pragma acc parallel copyout(LocalInt, LocalPointer, LocalArray)
+  while(1);
+#pragma acc parallel copyout(LocalArray[2:1])
+  while(1);
+
+  Complete LocalComposite2;
+#pragma acc parallel copyout(LocalComposite2.ScalarMember, LocalComposite2.ScalarMember)
+  while(1);
+
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel copyout(1 + IntParam)
+  while(1);
+
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel copyout(+IntParam)
+  while(1);
+
+  // expected-error@+1{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is not an array}}
+#pragma acc parallel copyout(PointerParam[2:])
+  while(1);
+
+  // expected-error@+1{{OpenACC sub-array specified range [2:5] would be out of the range of the subscripted array size of 5}}
+#pragma acc parallel copyout(ArrayParam[2:5])
+  while(1);
+
+  // expected-error@+2{{OpenACC sub-array specified range [2:5] would be out of the range of the subscripted array size of 5}}
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel copyout((float*)ArrayParam[2:5])
+  while(1);
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel copyout((float)ArrayParam[2])
+  while(1);
+}
+
+template<typename T, unsigned I, typename V>
+void TemplUses(T t, T (&arrayT)[I], V TemplComp) {
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel copyout(+t)
+  while(true);
+
+  // NTTP's are only valid if it is a reference to something.
+  // expected-error@+2{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+  // expected-note@#TEMPL_USES_INST{{in instantiation of}}
+#pragma acc parallel copyout(I)
+  while(true);
+
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel copyout(t, I)
+  while(true);
+
+#pragma acc parallel copyout(arrayT)
+  while(true);
+
+#pragma acc parallel copyout(TemplComp)
+  while(true);
+
+#pragma acc parallel copyout(TemplComp.PointerMember[5])
+  while(true);
+ int *Pointer;
+#pragma acc parallel copyout(Pointer[:I])
+  while(true);
+#pragma acc parallel copyout(Pointer[:t])
+  while(true);
+  // expected-error@+1{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is not an array}}
+#pragma acc parallel copyout(Pointer[1:])
+  while(true);
+}
+
+template<unsigned I, auto &NTTP_REF>
+void NTTP() {
+  // NTTP's are only valid if it is a reference to something.
+  // expected-error@+2{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+  // expected-note@#NTTP_INST{{in instantiation of}}
+#pragma acc parallel copyout(I)
+  while(true);
+
+#pragma acc parallel copyout(NTTP_REF)
+  while(true);
+}
+
+void Inst() {
+  static constexpr int NTTP_REFed = 1;
+  int i;
+  int Arr[5];
+  Complete C;
+  TemplUses(i, Arr, C); // #TEMPL_USES_INST
+  NTTP<5, NTTP_REFed>(); // #NTTP_INST
+}
diff --git a/clang/test/SemaOpenACC/compute-construct-create-clause.c b/clang/test/SemaOpenACC/compute-construct-create-clause.c
new file mode 100644
index 000000000000..9c94e3a1a407
--- /dev/null
+++ b/clang/test/SemaOpenACC/compute-construct-create-clause.c
@@ -0,0 +1,69 @@
+
+// RUN: %clang_cc1 %s -fopenacc -verify
+
+typedef struct IsComplete {
+  struct S { int A; } CompositeMember;
+  int ScalarMember;
+  float ArrayMember[5];
+  void *PointerMember;
+} Complete;
+void uses(int IntParam, short *PointerParam, float ArrayParam[5], Complete CompositeParam) {
+  int LocalInt;
+  short *LocalPointer;
+  float LocalArray[5];
+  Complete LocalComposite;
+  // Check Appertainment:
+#pragma acc parallel create(LocalInt)
+  while(1);
+#pragma acc serial create(LocalInt)
+  while(1);
+#pragma acc kernels create(LocalInt)
+  while(1);
+
+  // expected-warning@+1{{OpenACC clause name 'pcreate' is a deprecated clause name and is now an alias for 'create'}}
+#pragma acc parallel pcreate(LocalInt)
+  while(1);
+
+  // expected-warning@+1{{OpenACC clause name 'present_or_create' is a deprecated clause name and is now an alias for 'create'}}
+#pragma acc parallel present_or_create(LocalInt)
+  while(1);
+
+  // Valid cases:
+#pragma acc parallel create(LocalInt, LocalPointer, LocalArray)
+  while(1);
+#pragma acc parallel create(LocalArray[2:1])
+  while(1);
+#pragma acc parallel create(zero:LocalArray[2:1])
+  while(1);
+
+#pragma acc parallel create(LocalComposite.ScalarMember, LocalComposite.ScalarMember)
+  while(1);
+
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel create(1 + IntParam)
+  while(1);
+
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel create(+IntParam)
+  while(1);
+
+  // expected-error@+1{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is not an array}}
+#pragma acc parallel create(PointerParam[2:])
+  while(1);
+
+  // expected-error@+1{{OpenACC sub-array specified range [2:5] would be out of the range of the subscripted array size of 5}}
+#pragma acc parallel create(ArrayParam[2:5])
+  while(1);
+
+  // expected-error@+2{{OpenACC sub-array specified range [2:5] would be out of the range of the subscripted array size of 5}}
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel create((float*)ArrayParam[2:5])
+  while(1);
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel create((float)ArrayParam[2])
+  while(1);
+  // expected-error@+2{{invalid tag 'invalid' on 'create' clause}}
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel create(invalid:(float)ArrayParam[2])
+  while(1);
+}
diff --git a/clang/test/SemaOpenACC/compute-construct-create-clause.cpp b/clang/test/SemaOpenACC/compute-construct-create-clause.cpp
new file mode 100644
index 000000000000..d0323620b8f7
--- /dev/null
+++ b/clang/test/SemaOpenACC/compute-construct-create-clause.cpp
@@ -0,0 +1,112 @@
+// RUN: %clang_cc1 %s -fopenacc -verify
+
+enum SomeE{};
+typedef struct IsComplete {
+  struct S { int A; } CompositeMember;
+  int ScalarMember;
+  float ArrayMember[5];
+  SomeE EnumMember;
+  char *PointerMember;
+} Complete;
+
+void uses(int IntParam, char *PointerParam, float ArrayParam[5], Complete CompositeParam, int &IntParamRef) {
+  int LocalInt;
+  char *LocalPointer;
+  float LocalArray[5];
+  // Check Appertainment:
+#pragma acc parallel create(LocalInt)
+  while(1);
+#pragma acc serial create(LocalInt)
+  while(1);
+#pragma acc kernels create(LocalInt)
+  while(1);
+
+  // Valid cases:
+#pragma acc parallel create(LocalInt, LocalPointer, LocalArray)
+  while(1);
+#pragma acc parallel create(LocalArray[2:1])
+  while(1);
+
+  Complete LocalComposite2;
+#pragma acc parallel create(LocalComposite2.ScalarMember, LocalComposite2.ScalarMember)
+  while(1);
+
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel create(1 + IntParam)
+  while(1);
+
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel create(+IntParam)
+  while(1);
+
+  // expected-error@+1{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is not an array}}
+#pragma acc parallel create(PointerParam[2:])
+  while(1);
+
+  // expected-error@+1{{OpenACC sub-array specified range [2:5] would be out of the range of the subscripted array size of 5}}
+#pragma acc parallel create(ArrayParam[2:5])
+  while(1);
+
+  // expected-error@+2{{OpenACC sub-array specified range [2:5] would be out of the range of the subscripted array size of 5}}
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel create((float*)ArrayParam[2:5])
+  while(1);
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel create((float)ArrayParam[2])
+  while(1);
+}
+
+template<typename T, unsigned I, typename V>
+void TemplUses(T t, T (&arrayT)[I], V TemplComp) {
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel create(+t)
+  while(true);
+
+  // NTTP's are only valid if it is a reference to something.
+  // expected-error@+2{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+  // expected-note@#TEMPL_USES_INST{{in instantiation of}}
+#pragma acc parallel create(I)
+  while(true);
+
+  // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel create(t, I)
+  while(true);
+
+#pragma acc parallel create(arrayT)
+  while(true);
+
+#pragma acc parallel create(TemplComp)
+  while(true);
+
+#pragma acc parallel create(TemplComp.PointerMember[5])
+  while(true);
+ int *Pointer;
+#pragma acc parallel create(Pointer[:I])
+  while(true);
+#pragma acc parallel create(Pointer[:t])
+  while(true);
+  // expected-error@+1{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is not an array}}
+#pragma acc parallel create(Pointer[1:])
+  while(true);
+}
+
+template<unsigned I, auto &NTTP_REF>
+void NTTP() {
+  // NTTP's are only valid if it is a reference to something.
+  // expected-error@+2{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+  // expected-note@#NTTP_INST{{in instantiation of}}
+#pragma acc parallel create(I)
+  while(true);
+
+#pragma acc parallel create(NTTP_REF)
+  while(true);
+}
+
+void Inst() {
+  static constexpr int NTTP_REFed = 1;
+  int i;
+  int Arr[5];
+  Complete C;
+  TemplUses(i, Arr, C); // #TEMPL_USES_INST
+  NTTP<5, NTTP_REFed>(); // #NTTP_INST
+}
diff --git a/clang/test/SemaOpenACC/compute-construct-varlist-ast.cpp b/clang/test/SemaOpenACC/compute-construct-varlist-ast.cpp
index d3de9f146d8d..d35f62adfe07 100644
--- a/clang/test/SemaOpenACC/compute-construct-varlist-ast.cpp
+++ b/clang/test/SemaOpenACC/compute-construct-varlist-ast.cpp
@@ -109,6 +109,57 @@ void NormalUses(float *PointerParam) {
   // CHECK-NEXT: CXXBoolLiteralExpr
   // CHECK-NEXT: NullStmt
 
+#pragma acc parallel copyin(GlobalArray) pcopyin(readonly: PointerParam[Global]) present_or_copyin(Global)
+  while(true);
+  // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+  // CHECK-NEXT: copyin clause
+  // CHECK-NEXT: DeclRefExpr{{.*}}'short[5]' lvalue Var{{.*}}'GlobalArray' 'short[5]'
+  // CHECK-NEXT: pcopyin clause : readonly
+  // CHECK-NEXT: ArraySubscriptExpr{{.*}}'float' lvalue
+  // CHECK-NEXT: ImplicitCastExpr{{.*}} 'float *' <LValueToRValue>
+  // CHECK-NEXT: DeclRefExpr{{.*}}'float *' lvalue ParmVar{{.*}}'PointerParam' 'float *'
+  // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+  // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int'
+  // CHECK-NEXT: present_or_copyin clause
+  // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int'
+  // CHECK-NEXT: WhileStmt
+  // CHECK-NEXT: CXXBoolLiteralExpr
+  // CHECK-NEXT: NullStmt
+
+#pragma acc parallel copyout(GlobalArray) pcopyout(zero:PointerParam[Global]) present_or_copyout(Global)
+  while(true);
+  // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+  // CHECK-NEXT: copyout clause
+  // CHECK-NEXT: DeclRefExpr{{.*}}'short[5]' lvalue Var{{.*}}'GlobalArray' 'short[5]'
+  // CHECK-NEXT: pcopyout clause : zero
+  // CHECK-NEXT: ArraySubscriptExpr{{.*}}'float' lvalue
+  // CHECK-NEXT: ImplicitCastExpr{{.*}} 'float *' <LValueToRValue>
+  // CHECK-NEXT: DeclRefExpr{{.*}}'float *' lvalue ParmVar{{.*}}'PointerParam' 'float *'
+  // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+  // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int'
+  // CHECK-NEXT: present_or_copyout clause
+  // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int'
+  // CHECK-NEXT: WhileStmt
+  // CHECK-NEXT: CXXBoolLiteralExpr
+  // CHECK-NEXT: NullStmt
+
+#pragma acc parallel create(GlobalArray) pcreate(zero:PointerParam[Global]) present_or_create(Global)
+  while(true);
+  // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+  // CHECK-NEXT: create clause
+  // CHECK-NEXT: DeclRefExpr{{.*}}'short[5]' lvalue Var{{.*}}'GlobalArray' 'short[5]'
+  // CHECK-NEXT: pcreate clause : zero
+  // CHECK-NEXT: ArraySubscriptExpr{{.*}}'float' lvalue
+  // CHECK-NEXT: ImplicitCastExpr{{.*}} 'float *' <LValueToRValue>
+  // CHECK-NEXT: DeclRefExpr{{.*}}'float *' lvalue ParmVar{{.*}}'PointerParam' 'float *'
+  // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+  // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int'
+  // CHECK-NEXT: present_or_create clause
+  // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int'
+  // CHECK-NEXT: WhileStmt
+  // CHECK-NEXT: CXXBoolLiteralExpr
+  // CHECK-NEXT: NullStmt
+
 #pragma acc parallel private(GlobalArray, PointerParam[Global : Global])
   while(true);
   // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
@@ -277,6 +328,57 @@ void TemplUses(T t, U u, T*PointerParam) {
   // CHECK-NEXT: CXXBoolLiteralExpr
   // CHECK-NEXT: NullStmt
 
+#pragma acc parallel copyin(t) pcopyin(readonly:NTTP, u) present_or_copyin(u[0:t])
+  while(true);
+  // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+  // CHECK-NEXT: copyin clause
+  // CHECK-NEXT: DeclRefExpr{{.*}}'T' lvalue ParmVar{{.*}} 't' 'T'
+  // CHECK-NEXT: pcopyin clause : readonly
+  // CHECK-NEXT: DeclRefExpr{{.*}}'auto' lvalue NonTypeTemplateParm{{.*}} 'NTTP' 'auto &'
+  // CHECK-NEXT: DeclRefExpr{{.*}}'U' lvalue ParmVar{{.*}} 'u' 'U'
+  // CHECK-NEXT: present_or_copyin clause
+  // CHECK-NEXT: ArraySectionExpr
+  // CHECK-NEXT: DeclRefExpr{{.*}}'U' lvalue ParmVar{{.*}} 'u' 'U'
+  // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+  // CHECK-NEXT: DeclRefExpr{{.*}}'T' lvalue ParmVar{{.*}} 't' 'T'
+  // CHECK-NEXT: WhileStmt
+  // CHECK-NEXT: CXXBoolLiteralExpr
+  // CHECK-NEXT: NullStmt
+
+#pragma acc parallel copyout(t) pcopyout(zero:NTTP, u) present_or_copyout(u[0:t])
+  while(true);
+  // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+  // CHECK-NEXT: copyout clause
+  // CHECK-NEXT: DeclRefExpr{{.*}}'T' lvalue ParmVar{{.*}} 't' 'T'
+  // CHECK-NEXT: pcopyout clause : zero
+  // CHECK-NEXT: DeclRefExpr{{.*}}'auto' lvalue NonTypeTemplateParm{{.*}} 'NTTP' 'auto &'
+  // CHECK-NEXT: DeclRefExpr{{.*}}'U' lvalue ParmVar{{.*}} 'u' 'U'
+  // CHECK-NEXT: present_or_copyout clause
+  // CHECK-NEXT: ArraySectionExpr
+  // CHECK-NEXT: DeclRefExpr{{.*}}'U' lvalue ParmVar{{.*}} 'u' 'U'
+  // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+  // CHECK-NEXT: DeclRefExpr{{.*}}'T' lvalue ParmVar{{.*}} 't' 'T'
+  // CHECK-NEXT: WhileStmt
+  // CHECK-NEXT: CXXBoolLiteralExpr
+  // CHECK-NEXT: NullStmt
+
+#pragma acc parallel create(t) pcreate(zero: NTTP, u) present_or_create(u[0:t])
+  while(true);
+  // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+  // CHECK-NEXT: create clause
+  // CHECK-NEXT: DeclRefExpr{{.*}}'T' lvalue ParmVar{{.*}} 't' 'T'
+  // CHECK-NEXT: pcreate clause : zero
+  // CHECK-NEXT: DeclRefExpr{{.*}}'auto' lvalue NonTypeTemplateParm{{.*}} 'NTTP' 'auto &'
+  // CHECK-NEXT: DeclRefExpr{{.*}}'U' lvalue ParmVar{{.*}} 'u' 'U'
+  // CHECK-NEXT: present_or_create clause
+  // CHECK-NEXT: ArraySectionExpr
+  // CHECK-NEXT: DeclRefExpr{{.*}}'U' lvalue ParmVar{{.*}} 'u' 'U'
+  // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+  // CHECK-NEXT: DeclRefExpr{{.*}}'T' lvalue ParmVar{{.*}} 't' 'T'
+  // CHECK-NEXT: WhileStmt
+  // CHECK-NEXT: CXXBoolLiteralExpr
+  // CHECK-NEXT: NullStmt
+
 #pragma acc parallel private(u[0])
   while(true);
   // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
@@ -418,6 +520,66 @@ void TemplUses(T t, U u, T*PointerParam) {
   // CHECK-NEXT: CXXBoolLiteralExpr
   // CHECK-NEXT: NullStmt
 
+//#pragma acc parallel copyin(t) pcopyin(readonly:NTTP, u) present_or_copyin(u[0:t])
+  // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+  // CHECK-NEXT: copyin clause
+  // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}} 't' 'int'
+  // CHECK-NEXT: pcopyin clause : readonly
+  // CHECK-NEXT: SubstNonTypeTemplateParmExpr{{.*}}'const unsigned int' lvalue
+  // CHECK-NEXT: NonTypeTemplateParmDecl{{.*}} referenced 'auto &' depth 0 index 0 NTTP
+  // CHECK-NEXT: DeclRefExpr{{.*}}'const unsigned int' lvalue Var{{.*}} 'CEVar' 'const unsigned int'
+  // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *'
+  // CHECK-NEXT: present_or_copyin clause
+  // CHECK-NEXT: ArraySectionExpr
+  // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <LValueToRValue>
+  // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *'
+  // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+  // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+  // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}} 't' 'int'
+  // CHECK-NEXT: WhileStmt
+  // CHECK-NEXT: CXXBoolLiteralExpr
+  // CHECK-NEXT: NullStmt
+
+//#pragma acc parallel copyout(t) pcopyout(zero:NTTP, u) present_or_copyout(u[0:t])
+  // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+  // CHECK-NEXT: copyout clause
+  // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}} 't' 'int'
+  // CHECK-NEXT: pcopyout clause : zero
+  // CHECK-NEXT: SubstNonTypeTemplateParmExpr{{.*}}'const unsigned int' lvalue
+  // CHECK-NEXT: NonTypeTemplateParmDecl{{.*}} referenced 'auto &' depth 0 index 0 NTTP
+  // CHECK-NEXT: DeclRefExpr{{.*}}'const unsigned int' lvalue Var{{.*}} 'CEVar' 'const unsigned int'
+  // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *'
+  // CHECK-NEXT: present_or_copyout clause
+  // CHECK-NEXT: ArraySectionExpr
+  // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <LValueToRValue>
+  // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *'
+  // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+  // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+  // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}} 't' 'int'
+  // CHECK-NEXT: WhileStmt
+  // CHECK-NEXT: CXXBoolLiteralExpr
+  // CHECK-NEXT: NullStmt
+
+//#pragma acc parallel create(t) pcreate(zero: NTTP, u) present_or_create(u[0:t])
+  // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+  // CHECK-NEXT: create clause
+  // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}} 't' 'int'
+  // CHECK-NEXT: pcreate clause : zero
+  // CHECK-NEXT: SubstNonTypeTemplateParmExpr{{.*}}'const unsigned int' lvalue
+  // CHECK-NEXT: NonTypeTemplateParmDecl{{.*}} referenced 'auto &' depth 0 index 0 NTTP
+  // CHECK-NEXT: DeclRefExpr{{.*}}'const unsigned int' lvalue Var{{.*}} 'CEVar' 'const unsigned int'
+  // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *'
+  // CHECK-NEXT: present_or_create clause
+  // CHECK-NEXT: ArraySectionExpr
+  // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <LValueToRValue>
+  // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *'
+  // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+  // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+  // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}} 't' 'int'
+  // CHECK-NEXT: WhileStmt
+  // CHECK-NEXT: CXXBoolLiteralExpr
+  // CHECK-NEXT: NullStmt
+
 // #pragma acc parallel private(u[0])
   // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
   // CHECK-NEXT: private clause
diff --git a/clang/tools/libclang/CIndex.cpp b/clang/tools/libclang/CIndex.cpp
index c02545f76ed2..487f57859576 100644
--- a/clang/tools/libclang/CIndex.cpp
+++ b/clang/tools/libclang/CIndex.cpp
@@ -2831,6 +2831,15 @@ void OpenACCClauseEnqueue::VisitNoCreateClause(const OpenACCNoCreateClause &C) {
 void OpenACCClauseEnqueue::VisitCopyClause(const OpenACCCopyClause &C) {
   VisitVarList(C);
 }
+void OpenACCClauseEnqueue::VisitCopyInClause(const OpenACCCopyInClause &C) {
+  VisitVarList(C);
+}
+void OpenACCClauseEnqueue::VisitCopyOutClause(const OpenACCCopyOutClause &C) {
+  VisitVarList(C);
+}
+void OpenACCClauseEnqueue::VisitCreateClause(const OpenACCCreateClause &C) {
+  VisitVarList(C);
+}
 } // namespace
 
 void EnqueueVisitor::EnqueueChildren(const OpenACCClause *C) {