[clang][OpenMP] Add AST nodes for opaque block/loop directivesusers/kparzysz/spr/p03-a01-clang-opaque-directive

Add two new AST nodes:
1. `OMPOpaqueBlockDirective` representing an executable directive
not associated with a loop, and
2. `OMPOpaqueLoopDirective` representing an executable directive
associated with a loop.

This is an indermediate step towards treating compound directives as
a sequece of either leaf or composite directives. The break-up of
compound directives into these constituents may be impossible in a
template, since the identity of objects referenced in clauses may
be impossible to establish.
When an executable directive is encountered in a function template, it
will be represented by one of the new AST nodes, depending on its
association. When the template is instantiated, these AST nodes will
replaced with the AST nodes used in non-templates.

34 files changed, 1135 insertions, 410 deletions

diff --git a/clang/bindings/python/clang/cindex.py b/clang/bindings/python/clang/cindex.py
index f8a20a1..3ae7c47 100644
--- a/clang/bindings/python/clang/cindex.py
+++ b/clang/bindings/python/clang/cindex.py
@@ -1410,6 +1410,12 @@ class CursorKind(BaseEnumeration):
     # OpenMP scope directive.
     OMP_SCOPE_DIRECTIVE = 306
 
+    # OpenMP opaque block-associated directive.
+    OMP_OPAQUE_BLOCK_DIRECTIVE = 310
+
+    # OpenMP opaque loop-associated directive.
+    OMP_OPAQUE_LOOP_DIRECTIVE = 311
+
     # OpenACC Compute Construct.
     OPEN_ACC_COMPUTE_DIRECTIVE = 320
 
diff --git a/clang/include/clang-c/Index.h b/clang/include/clang-c/Index.h
index 8fc0632..5d1db15 100644
--- a/clang/include/clang-c/Index.h
+++ b/clang/include/clang-c/Index.h
@@ -2158,6 +2158,14 @@ enum CXCursorKind {
    */
   CXCursor_OMPAssumeDirective = 309,
 
+  /** OpenMP opaque block-associated directive.
+   */
+  CXCursor_OMPOpaqueBlockDirective = 310,
+
+  /** OpenMP opaque loop-associated directive.
+   */
+  CXCursor_OMPOpaqueLoopDirective = 311,
+
   /** OpenACC Compute Construct.
    */
   CXCursor_OpenACCComputeConstruct = 320,
diff --git a/clang/include/clang/AST/RecursiveASTVisitor.h b/clang/include/clang/AST/RecursiveASTVisitor.h
index 76b598a..e6fe46a 100644
--- a/clang/include/clang/AST/RecursiveASTVisitor.h
+++ b/clang/include/clang/AST/RecursiveASTVisitor.h
@@ -3026,6 +3026,12 @@ RecursiveASTVisitor<Derived>::TraverseOMPLoopDirective(OMPLoopDirective *S) {
   return TraverseOMPExecutableDirective(S);
 }
 
+DEF_TRAVERSE_STMT(OMPOpaqueBlockDirective,
+                  { TRY_TO(TraverseOMPExecutableDirective(S)); })
+
+DEF_TRAVERSE_STMT(OMPOpaqueLoopDirective,
+                  { TRY_TO(TraverseOMPExecutableDirective(S)); })
+
 DEF_TRAVERSE_STMT(OMPMetaDirective,
                   { TRY_TO(TraverseOMPExecutableDirective(S)); })
 
diff --git a/clang/include/clang/AST/StmtOpenMP.h b/clang/include/clang/AST/StmtOpenMP.h
index 9ec49b8..6543496 100644
--- a/clang/include/clang/AST/StmtOpenMP.h
+++ b/clang/include/clang/AST/StmtOpenMP.h
@@ -1520,6 +1520,7 @@ public:
   static bool classof(const Stmt *T) {
     return T->getStmtClass() == OMPSimdDirectiveClass ||
            T->getStmtClass() == OMPForDirectiveClass ||
+           T->getStmtClass() == OMPOpaqueLoopDirectiveClass ||
            T->getStmtClass() == OMPForSimdDirectiveClass ||
            T->getStmtClass() == OMPParallelForDirectiveClass ||
            T->getStmtClass() == OMPParallelForSimdDirectiveClass ||
@@ -1559,6 +1560,187 @@ public:
   }
 };
 
+/// This represents any executable OpenMP directive that is not loop-
+/// associated (usually block-associated).
+class OMPOpaqueBlockDirective final : public OMPExecutableDirective {
+  friend class ASTStmtReader;
+  friend class OMPExecutableDirective;
+
+  /// true if current directive has inner cancel directive.
+  bool HasCancel = false;
+
+  /// Region for cancel/cancellation-point directives.
+  OpenMPDirectiveKind CancelRegion = llvm::omp::OMPD_unknown;
+
+  /// Name of the directive for the "critical" directive.
+  DeclarationNameInfo DirName;
+
+  /// Build directive with the given start and end location.
+  ///
+  /// \param Kind The OpenMP directive kind.
+  /// \param StartLoc Starting location of the directive kind.
+  /// \param EndLoc Ending location of the directive.
+  ///
+  OMPOpaqueBlockDirective(OpenMPDirectiveKind Kind, SourceLocation StartLoc,
+                          SourceLocation EndLoc)
+      : OMPExecutableDirective(OMPOpaqueBlockDirectiveClass, Kind, StartLoc,
+                               EndLoc) {}
+
+  /// Build an empty directive.
+  ///
+  /// \param Kind The OpenMP directive kind.
+  ///
+  explicit OMPOpaqueBlockDirective(OpenMPDirectiveKind Kind)
+      : OMPExecutableDirective(OMPOpaqueBlockDirectiveClass, Kind,
+                               SourceLocation(), SourceLocation()) {}
+
+  /// Sets special task reduction descriptor.
+  void setTaskReductionRefExpr(Expr *E) { Data->getChildren()[0] = E; }
+
+  /// Set cancel state.
+  void setHasCancel(bool Has) { HasCancel = Has; }
+
+  /// Set cancellation region.
+  void setCancelRegion(OpenMPDirectiveKind CR) { CancelRegion = CR; }
+
+  /// Set directive name.
+  void setDirectiveName(const DeclarationNameInfo &Name) { DirName = Name; }
+
+public:
+  /// Creates directive with a list of \a Clauses.
+  ///
+  /// \param C AST context.
+  /// \param StartLoc Starting location of the directive kind.
+  /// \param EndLoc Ending Location of the directive.
+  /// \param Kind The OpenMP directive kind.
+  /// \param Clauses List of clauses.
+  /// \param AssociatedStmt Statement, associated with the directive.
+  /// \param TaskRedRef Task reduction special reference expression to handle
+  /// taskgroup descriptor.
+  /// \param HasCancel true if current directive has inner cancel directive.
+  /// \param CancelRegion Cancellation region, applicable for cancel or
+  /// cancellation-point directives.
+  /// \param DirName Name of the directive (used if DKind is OMPD_critical).
+  static OMPOpaqueBlockDirective *
+  Create(const ASTContext &C, SourceLocation StartLoc, SourceLocation EndLoc,
+         OpenMPDirectiveKind DKind, ArrayRef<OMPClause *> Clauses,
+         Stmt *AssociatedStmt, Expr *TaskRedRef, bool HasCancel,
+         OpenMPDirectiveKind CancelRegion, const DeclarationNameInfo &DirName);
+
+  /// Creates an empty directive with the place for \a NumClauses
+  /// clauses.
+  ///
+  /// \param C AST context.
+  /// \param Kind The OpenMP directive kind.
+  /// \param NumClauses Number of clauses.
+  ///
+  static OMPOpaqueBlockDirective *
+  CreateEmpty(const ASTContext &C, OpenMPDirectiveKind Kind,
+              unsigned NumClauses, bool HasAssociatedStmt, EmptyShell);
+
+  /// Returns special task reduction reference expression.
+  Expr *getTaskReductionRefExpr() const {
+    return cast_or_null<Expr>(Data->getChildren()[0]);
+  }
+
+  /// Return true if current directive has inner cancel directive.
+  bool hasCancel() const { return HasCancel; }
+
+  /// Return cancellation region.
+  OpenMPDirectiveKind getCancelRegion() const { return CancelRegion; }
+
+  /// Return the directive name (valid for "critical" only).
+  DeclarationNameInfo getDirectiveName() const { return DirName; }
+
+  static bool classof(const Stmt *T) {
+    return T->getStmtClass() == OMPOpaqueBlockDirectiveClass;
+  }
+};
+
+/// This represents any executable, loop-associated OpenMP directive.
+class OMPOpaqueLoopDirective final : public OMPLoopDirective {
+  friend class ASTStmtReader;
+  friend class OMPExecutableDirective;
+
+  /// true if current directive has inner cancel directive.
+  bool HasCancel = false;
+
+  /// Build directive with the given start and end location.
+  ///
+  /// \param Kind The OpenMP directive kind.
+  /// \param StartLoc Starting location of the directive kind.
+  /// \param EndLoc Ending location of the directive.
+  /// \param CollapsedNum Number of collapsed nested loops.
+  ///
+  OMPOpaqueLoopDirective(OpenMPDirectiveKind Kind, SourceLocation StartLoc,
+                         SourceLocation EndLoc, unsigned CollapsedNum)
+      : OMPLoopDirective(OMPOpaqueLoopDirectiveClass, Kind, StartLoc, EndLoc,
+                         CollapsedNum) {}
+
+  /// Build an empty directive.
+  ///
+  /// \param Kind The OpenMP directive kind.
+  /// \param CollapsedNum Number of collapsed nested loops.
+  ///
+  explicit OMPOpaqueLoopDirective(OpenMPDirectiveKind Kind,
+                                  unsigned CollapsedNum)
+      : OMPLoopDirective(OMPOpaqueLoopDirectiveClass, Kind, SourceLocation(),
+                         SourceLocation(), CollapsedNum) {}
+
+  /// Sets special task reduction descriptor.
+  void setTaskReductionRefExpr(Expr *E) {
+    size_t Index = numLoopChildren(getLoopsNumber(), getDirectiveKind());
+    Data->getChildren()[Index] = E;
+  }
+
+  /// Set cancel state.
+  void setHasCancel(bool Has) { HasCancel = Has; }
+
+public:
+  /// Creates directive with a list of \a Clauses.
+  ///
+  /// \param C AST context.
+  /// \param StartLoc Starting location of the directive kind.
+  /// \param EndLoc Ending Location of the directive.
+  /// \param Kind The OpenMP directive kind.
+  /// \param CollapsedNum Number of collapsed loops.
+  /// \param Clauses List of clauses.
+  /// \param AssociatedStmt Statement, associated with the directive.
+  /// \param Exprs Helper expressions for CodeGen.
+  ///
+  static OMPOpaqueLoopDirective *
+  Create(const ASTContext &C, SourceLocation StartLoc, SourceLocation EndLoc,
+         OpenMPDirectiveKind Kind, unsigned CollapsedNum,
+         ArrayRef<OMPClause *> Clauses, Stmt *AssociatedStmt,
+         const HelperExprs &Exprs, Expr *TaskRedRef, bool HasCancel);
+
+  /// Creates an empty directive with the place
+  /// for \a NumClauses clauses.
+  ///
+  /// \param C AST context.
+  /// \param Kind The OpenMP directive kind.
+  /// \param CollapsedNum Number of collapsed nested loops.
+  /// \param NumClauses Number of clauses.
+  ///
+  static OMPOpaqueLoopDirective *CreateEmpty(const ASTContext &C,
+                                             OpenMPDirectiveKind Kind,
+                                             unsigned NumClauses,
+                                             unsigned CollapsedNum, EmptyShell);
+
+  /// Returns special task reduction reference expression.
+  Expr *getTaskReductionRefExpr() const {
+    size_t Index = numLoopChildren(getLoopsNumber(), getDirectiveKind());
+    return cast_or_null<Expr>(Data->getChildren()[Index]);
+  }
+
+  /// Return true if current directive has inner cancel directive.
+  bool hasCancel() const { return HasCancel; }
+
+  static bool classof(const Stmt *T) {
+    return T->getStmtClass() == OMPOpaqueLoopDirectiveClass;
+  }
+};
+
 /// This represents '#pragma omp simd' directive.
 ///
 /// \code
diff --git a/clang/include/clang/AST/TextNodeDumper.h b/clang/include/clang/AST/TextNodeDumper.h
index 988b142..6606cce 100644
--- a/clang/include/clang/AST/TextNodeDumper.h
+++ b/clang/include/clang/AST/TextNodeDumper.h
@@ -366,6 +366,8 @@ public:
   void VisitPragmaCommentDecl(const PragmaCommentDecl *D);
   void VisitPragmaDetectMismatchDecl(const PragmaDetectMismatchDecl *D);
   void VisitOMPExecutableDirective(const OMPExecutableDirective *D);
+  void VisitOMPOpaqueBlockDirective(const OMPOpaqueBlockDirective *D);
+  void VisitOMPOpaqueLoopDirective(const OMPOpaqueLoopDirective *D);
   void VisitOMPDeclareReductionDecl(const OMPDeclareReductionDecl *D);
   void VisitOMPRequiresDecl(const OMPRequiresDecl *D);
   void VisitOMPCapturedExprDecl(const OMPCapturedExprDecl *D);
diff --git a/clang/include/clang/Basic/StmtNodes.td b/clang/include/clang/Basic/StmtNodes.td
index 89f5a76..da8c910 100644
--- a/clang/include/clang/Basic/StmtNodes.td
+++ b/clang/include/clang/Basic/StmtNodes.td
@@ -225,6 +225,8 @@ def OMPExecutableDirective : StmtNode<Stmt, 1>;
 def OMPMetaDirective : StmtNode<OMPExecutableDirective>;
 def OMPLoopBasedDirective : StmtNode<OMPExecutableDirective, 1>;
 def OMPLoopDirective : StmtNode<OMPLoopBasedDirective, 1>;
+def OMPOpaqueBlockDirective : StmtNode<OMPExecutableDirective>;
+def OMPOpaqueLoopDirective : StmtNode<OMPLoopDirective>;
 def OMPParallelDirective : StmtNode<OMPExecutableDirective>;
 def OMPSimdDirective : StmtNode<OMPLoopDirective>;
 def OMPLoopTransformationDirective : StmtNode<OMPLoopBasedDirective, 1>;
diff --git a/clang/include/clang/Sema/SemaOpenMP.h b/clang/include/clang/Sema/SemaOpenMP.h
index 3d1cc4f..18103a7 100644
--- a/clang/include/clang/Sema/SemaOpenMP.h
+++ b/clang/include/clang/Sema/SemaOpenMP.h
@@ -849,6 +849,17 @@ public:
       ArrayRef<OMPInteropInfo> AppendArgs, SourceLocation AdjustArgsLoc,
       SourceLocation AppendArgsLoc, SourceRange SR);
 
+  StmtResult ActOnOpenMPOpaqueBlockDirective(
+      OpenMPDirectiveKind Kind, ArrayRef<OMPClause *> Clauses, Stmt *AStmt,
+      OpenMPDirectiveKind CancelRegion, const DeclarationNameInfo &DirName,
+      SourceLocation StartLoc, SourceLocation EndLoc);
+
+  StmtResult
+  ActOnOpenMPOpaqueLoopDirective(OpenMPDirectiveKind Kind,
+                                 ArrayRef<OMPClause *> Clauses, Stmt *AStmt,
+                                 SourceLocation StartLoc, SourceLocation EndLoc,
+                                 VarsWithInheritedDSAType &VarsWithImplicitDSA);
+
   OMPClause *ActOnOpenMPSingleExprClause(OpenMPClauseKind Kind, Expr *Expr,
                                          SourceLocation StartLoc,
                                          SourceLocation LParenLoc,
diff --git a/clang/include/clang/Serialization/ASTBitCodes.h b/clang/include/clang/Serialization/ASTBitCodes.h
index fd834c1..ac332f6 100644
--- a/clang/include/clang/Serialization/ASTBitCodes.h
+++ b/clang/include/clang/Serialization/ASTBitCodes.h
@@ -1904,6 +1904,8 @@ enum StmtCode {
   STMT_SEH_TRY,                     // SEHTryStmt
 
   // OpenMP directives
+  STMT_OMP_OPAQUE_BLOCK_DIRECTIVE,
+  STMT_OMP_OPAQUE_LOOP_DIRECTIVE,
   STMT_OMP_META_DIRECTIVE,
   STMT_OMP_CANONICAL_LOOP,
   STMT_OMP_PARALLEL_DIRECTIVE,
diff --git a/clang/lib/AST/StmtOpenMP.cpp b/clang/lib/AST/StmtOpenMP.cpp
index 4f441c2..3bd88ac 100644
--- a/clang/lib/AST/StmtOpenMP.cpp
+++ b/clang/lib/AST/StmtOpenMP.cpp
@@ -259,6 +259,99 @@ void OMPLoopDirective::setFinalsConditions(ArrayRef<Expr *> A) {
   llvm::copy(A, getFinalsConditions().begin());
 }
 
+OMPOpaqueBlockDirective *OMPOpaqueBlockDirective::Create(
+    const ASTContext &C, SourceLocation StartLoc, SourceLocation EndLoc,
+    OpenMPDirectiveKind DKind, ArrayRef<OMPClause *> Clauses,
+    Stmt *AssociatedStmt, Expr *TaskRedRef, bool HasCancel,
+    OpenMPDirectiveKind CancelRegion, const DeclarationNameInfo &DirName) {
+  // The child is TaskReductionRefExpr.
+  auto *Dir = createDirective<OMPOpaqueBlockDirective>(
+      C, Clauses, AssociatedStmt, /*NumChildren=*/1, DKind, StartLoc, EndLoc);
+  Dir->setTaskReductionRefExpr(TaskRedRef);
+  Dir->setHasCancel(HasCancel);
+  Dir->setCancelRegion(CancelRegion);
+  Dir->setDirectiveName(DirName);
+  return Dir;
+}
+
+OMPOpaqueBlockDirective *OMPOpaqueBlockDirective::CreateEmpty(
+    const ASTContext &C, OpenMPDirectiveKind DKind, unsigned NumClauses,
+    bool HasAssociatedStmt, EmptyShell) {
+  // The child is TaskReductionRefExpr.
+  return createEmptyDirective<OMPOpaqueBlockDirective>(
+      C, NumClauses, HasAssociatedStmt, /*NumChildren=*/1, DKind);
+}
+
+OMPOpaqueLoopDirective *OMPOpaqueLoopDirective::Create(
+    const ASTContext &C, SourceLocation StartLoc, SourceLocation EndLoc,
+    OpenMPDirectiveKind Kind, unsigned CollapsedNum,
+    ArrayRef<OMPClause *> Clauses, Stmt *AssociatedStmt,
+    const HelperExprs &Exprs, Expr *TaskRedRef, bool HasCancel) {
+  // The +1 in the number of children is for TaskReductionRefExpr.
+  auto *Dir = createDirective<OMPOpaqueLoopDirective>(
+      C, Clauses, AssociatedStmt, numLoopChildren(CollapsedNum, Kind) + 1, Kind,
+      StartLoc, EndLoc, CollapsedNum);
+  Dir->setIterationVariable(Exprs.IterationVarRef);
+  Dir->setLastIteration(Exprs.LastIteration);
+  Dir->setCalcLastIteration(Exprs.CalcLastIteration);
+  Dir->setPreCond(Exprs.PreCond);
+  Dir->setCond(Exprs.Cond);
+  Dir->setInit(Exprs.Init);
+  Dir->setInc(Exprs.Inc);
+  Dir->setPreInits(Exprs.PreInits);
+
+  if (isOpenMPWorksharingDirective(Kind) ||
+      isOpenMPGenericLoopDirective(Kind) || isOpenMPTaskLoopDirective(Kind) ||
+      isOpenMPDistributeDirective(Kind)) {
+    Dir->setIsLastIterVariable(Exprs.IL);
+    Dir->setLowerBoundVariable(Exprs.LB);
+    Dir->setUpperBoundVariable(Exprs.UB);
+    Dir->setStrideVariable(Exprs.ST);
+    Dir->setEnsureUpperBound(Exprs.EUB);
+    Dir->setNextLowerBound(Exprs.NLB);
+    Dir->setNextUpperBound(Exprs.NUB);
+    Dir->setNumIterations(Exprs.NumIterations);
+  }
+
+  if (isOpenMPLoopBoundSharingDirective(Kind)) {
+    Dir->setPrevLowerBoundVariable(Exprs.PrevLB);
+    Dir->setPrevUpperBoundVariable(Exprs.PrevUB);
+    Dir->setDistInc(Exprs.DistInc);
+    Dir->setPrevEnsureUpperBound(Exprs.PrevEUB);
+    Dir->setCombinedLowerBoundVariable(Exprs.DistCombinedFields.LB);
+    Dir->setCombinedUpperBoundVariable(Exprs.DistCombinedFields.UB);
+    Dir->setCombinedEnsureUpperBound(Exprs.DistCombinedFields.EUB);
+    Dir->setCombinedInit(Exprs.DistCombinedFields.Init);
+    Dir->setCombinedCond(Exprs.DistCombinedFields.Cond);
+    Dir->setCombinedNextLowerBound(Exprs.DistCombinedFields.NLB);
+    Dir->setCombinedNextUpperBound(Exprs.DistCombinedFields.NUB);
+    Dir->setCombinedDistCond(Exprs.DistCombinedFields.DistCond);
+    Dir->setCombinedParForInDistCond(Exprs.DistCombinedFields.ParForInDistCond);
+  }
+
+  Dir->setCounters(Exprs.Counters);
+  Dir->setPrivateCounters(Exprs.PrivateCounters);
+  Dir->setInits(Exprs.Inits);
+  Dir->setUpdates(Exprs.Updates);
+  Dir->setFinals(Exprs.Finals);
+  Dir->setDependentCounters(Exprs.DependentCounters);
+  Dir->setDependentInits(Exprs.DependentInits);
+  Dir->setFinalsConditions(Exprs.FinalsConditions);
+
+  Dir->setTaskReductionRefExpr(TaskRedRef);
+  Dir->setHasCancel(HasCancel);
+  return Dir;
+}
+
+OMPOpaqueLoopDirective *OMPOpaqueLoopDirective::CreateEmpty(
+    const ASTContext &C, OpenMPDirectiveKind Kind, unsigned NumClauses,
+    unsigned CollapsedNum, EmptyShell) {
+  // The +1 in the number of children is for TaskReductionRefExpr.
+  return createEmptyDirective<OMPOpaqueLoopDirective>(
+      C, NumClauses, /*HasAssociatedStmt=*/true,
+      numLoopChildren(CollapsedNum, Kind) + 1, Kind, CollapsedNum);
+}
+
 OMPMetaDirective *OMPMetaDirective::Create(const ASTContext &C,
                                            SourceLocation StartLoc,
                                            SourceLocation EndLoc,
diff --git a/clang/lib/AST/StmtPrinter.cpp b/clang/lib/AST/StmtPrinter.cpp
index 7507c9d..cd4f913 100644
--- a/clang/lib/AST/StmtPrinter.cpp
+++ b/clang/lib/AST/StmtPrinter.cpp
@@ -735,6 +735,41 @@ void StmtPrinter::PrintOMPExecutableDirective(OMPExecutableDirective *S,
     PrintStmt(S->getRawStmt());
 }
 
+void StmtPrinter::VisitOMPOpaqueBlockDirective(OMPOpaqueBlockDirective *Node) {
+  OpenMPDirectiveKind DKind = Node->getDirectiveKind();
+  bool ForceNoStmt = false;
+
+  Indent() << "#pragma omp " << llvm::omp::getOpenMPDirectiveName(DKind);
+  switch (DKind) {
+  case llvm::omp::OMPD_cancel:
+  case llvm::omp::OMPD_cancellation_point:
+    if (Node->getCancelRegion() != llvm::omp::OMPD_unknown)
+      OS << ' ' << llvm::omp::getOpenMPDirectiveName(Node->getCancelRegion());
+    break;
+  case llvm::omp::OMPD_critical:
+    if (Node->getDirectiveName().getName()) {
+      OS << " (";
+      Node->getDirectiveName().printName(OS, Policy);
+      OS << ")";
+    }
+    break;
+  case llvm::omp::OMPD_target_enter_data:
+  case llvm::omp::OMPD_target_exit_data:
+  case llvm::omp::OMPD_target_update:
+    ForceNoStmt = true;
+    break;
+  default:
+    break;
+  }
+  PrintOMPExecutableDirective(Node, ForceNoStmt);
+}
+
+void StmtPrinter::VisitOMPOpaqueLoopDirective(OMPOpaqueLoopDirective *Node) {
+  Indent() << "#pragma omp "
+           << llvm::omp::getOpenMPDirectiveName(Node->getDirectiveKind());
+  PrintOMPExecutableDirective(Node);
+}
+
 void StmtPrinter::VisitOMPMetaDirective(OMPMetaDirective *Node) {
   Indent() << "#pragma omp metadirective";
   PrintOMPExecutableDirective(Node);
diff --git a/clang/lib/AST/StmtProfile.cpp b/clang/lib/AST/StmtProfile.cpp
index 3dfbef1..a4f032c 100644
--- a/clang/lib/AST/StmtProfile.cpp
+++ b/clang/lib/AST/StmtProfile.cpp
@@ -967,6 +967,16 @@ StmtProfiler::VisitOMPExecutableDirective(const OMPExecutableDirective *S) {
       P.Visit(*I);
 }
 
+void StmtProfiler::VisitOMPOpaqueBlockDirective(
+    const OMPOpaqueBlockDirective *S) {
+  VisitOMPExecutableDirective(S);
+}
+
+void StmtProfiler::VisitOMPOpaqueLoopDirective(
+    const OMPOpaqueLoopDirective *S) {
+  VisitOMPExecutableDirective(S);
+}
+
 void StmtProfiler::VisitOMPCanonicalLoop(const OMPCanonicalLoop *L) {
   VisitStmt(L);
 }
diff --git a/clang/lib/AST/TextNodeDumper.cpp b/clang/lib/AST/TextNodeDumper.cpp
index 2552c11..09d3802 100644
--- a/clang/lib/AST/TextNodeDumper.cpp
+++ b/clang/lib/AST/TextNodeDumper.cpp
@@ -2372,6 +2372,20 @@ void TextNodeDumper::VisitOMPExecutableDirective(
     OS << " openmp_standalone_directive";
 }
 
+void TextNodeDumper::VisitOMPOpaqueBlockDirective(
+    const OMPOpaqueBlockDirective *D) {
+  VisitOMPExecutableDirective(D);
+  OpenMPDirectiveKind DKind = D->getDirectiveKind();
+  OS << " '" << llvm::omp::getOpenMPDirectiveName(DKind) << '\'';
+}
+
+void TextNodeDumper::VisitOMPOpaqueLoopDirective(
+    const OMPOpaqueLoopDirective *D) {
+  VisitOMPExecutableDirective(D);
+  OpenMPDirectiveKind DKind = D->getDirectiveKind();
+  OS << " '" << llvm::omp::getOpenMPDirectiveName(DKind) << '\'';
+}
+
 void TextNodeDumper::VisitOMPDeclareReductionDecl(
     const OMPDeclareReductionDecl *D) {
   dumpName(D);
diff --git a/clang/lib/CodeGen/CGStmt.cpp b/clang/lib/CodeGen/CGStmt.cpp
index 698baf8..42e9490 100644
--- a/clang/lib/CodeGen/CGStmt.cpp
+++ b/clang/lib/CodeGen/CGStmt.cpp
@@ -204,6 +204,11 @@ void CodeGenFunction::EmitStmt(const Stmt *S, ArrayRef<const Attr *> Attrs) {
   case Stmt::SEHTryStmtClass:
     EmitSEHTryStmt(cast<SEHTryStmt>(*S));
     break;
+  case Stmt::OMPOpaqueBlockDirectiveClass:
+  case Stmt::OMPOpaqueLoopDirectiveClass:
+    // These are catch-all nodes for executable OpenMP directives in templates.
+    // Template instantiation should replace them with specific nodes.
+    llvm_unreachable("These nodes should have been eliminated");
   case Stmt::OMPMetaDirectiveClass:
     EmitOMPMetaDirective(cast<OMPMetaDirective>(*S));
     break;
diff --git a/clang/lib/Sema/SemaExceptionSpec.cpp b/clang/lib/Sema/SemaExceptionSpec.cpp
index 6a9f43d..467eb89 100644
--- a/clang/lib/Sema/SemaExceptionSpec.cpp
+++ b/clang/lib/Sema/SemaExceptionSpec.cpp
@@ -1452,6 +1452,8 @@ CanThrowResult Sema::canThrow(const Stmt *S) {
   case Stmt::OMPMaskedTaskLoopDirectiveClass:
   case Stmt::OMPMasterTaskLoopSimdDirectiveClass:
   case Stmt::OMPMaskedTaskLoopSimdDirectiveClass:
+  case Stmt::OMPOpaqueBlockDirectiveClass:
+  case Stmt::OMPOpaqueLoopDirectiveClass:
   case Stmt::OMPOrderedDirectiveClass:
   case Stmt::OMPCanonicalLoopClass:
   case Stmt::OMPParallelDirectiveClass:
diff --git a/clang/lib/Sema/SemaOpenMP.cpp b/clang/lib/Sema/SemaOpenMP.cpp
index 66ff92f..fbed2952 100644
--- a/clang/lib/Sema/SemaOpenMP.cpp
+++ b/clang/lib/Sema/SemaOpenMP.cpp
@@ -5965,6 +5965,328 @@ static bool teamsLoopCanBeParallelFor(Stmt *AStmt, Sema &SemaRef) {
   return Checker.teamsLoopCanBeParallelFor();
 }
 
+static StmtResult createASTForDirective(
+    Sema &SemaRef, OpenMPDirectiveKind Kind, ArrayRef<OMPClause *> Clauses,
+    Stmt *AStmt, SourceLocation StartLoc, SourceLocation EndLoc,
+    const DeclarationNameInfo &DirName, OpenMPDirectiveKind CancelRegion,
+    SemaOpenMP::VarsWithInheritedDSAType &VarsWithInheritedDSA) {
+
+  SemaOpenMP &S = SemaRef.OpenMP();
+  StmtResult Res = StmtError();
+
+  switch (Kind) {
+  case OMPD_parallel:
+    Res = S.ActOnOpenMPParallelDirective(Clauses, AStmt, StartLoc, EndLoc);
+    break;
+  case OMPD_simd:
+    Res = S.ActOnOpenMPSimdDirective(Clauses, AStmt, StartLoc, EndLoc,
+                                     VarsWithInheritedDSA);
+    break;
+  case OMPD_tile:
+    Res = S.ActOnOpenMPTileDirective(Clauses, AStmt, StartLoc, EndLoc);
+    break;
+  case OMPD_unroll:
+    Res = S.ActOnOpenMPUnrollDirective(Clauses, AStmt, StartLoc, EndLoc);
+    break;
+  case OMPD_reverse:
+    assert(Clauses.empty() && "reverse directive does not support any clauses");
+    Res = S.ActOnOpenMPReverseDirective(AStmt, StartLoc, EndLoc);
+    break;
+  case OMPD_interchange:
+    Res = S.ActOnOpenMPInterchangeDirective(Clauses, AStmt, StartLoc, EndLoc);
+    break;
+  case OMPD_for:
+    Res = S.ActOnOpenMPForDirective(Clauses, AStmt, StartLoc, EndLoc,
+                                    VarsWithInheritedDSA);
+    break;
+  case OMPD_for_simd:
+    Res = S.ActOnOpenMPForSimdDirective(Clauses, AStmt, StartLoc, EndLoc,
+                                        VarsWithInheritedDSA);
+    break;
+  case OMPD_sections:
+    Res = S.ActOnOpenMPSectionsDirective(Clauses, AStmt, StartLoc, EndLoc);
+    break;
+  case OMPD_section:
+    assert(Clauses.empty() &&
+           "No clauses are allowed for 'omp section' directive");
+    Res = S.ActOnOpenMPSectionDirective(AStmt, StartLoc, EndLoc);
+    break;
+  case OMPD_single:
+    Res = S.ActOnOpenMPSingleDirective(Clauses, AStmt, StartLoc, EndLoc);
+    break;
+  case OMPD_master:
+    assert(Clauses.empty() &&
+           "No clauses are allowed for 'omp master' directive");
+    Res = S.ActOnOpenMPMasterDirective(AStmt, StartLoc, EndLoc);
+    break;
+  case OMPD_masked:
+    Res = S.ActOnOpenMPMaskedDirective(Clauses, AStmt, StartLoc, EndLoc);
+    break;
+  case OMPD_critical:
+    Res = S.ActOnOpenMPCriticalDirective(DirName, Clauses, AStmt, StartLoc,
+                                         EndLoc);
+    break;
+  case OMPD_parallel_for:
+    Res = S.ActOnOpenMPParallelForDirective(Clauses, AStmt, StartLoc, EndLoc,
+                                            VarsWithInheritedDSA);
+    break;
+  case OMPD_parallel_for_simd:
+    Res = S.ActOnOpenMPParallelForSimdDirective(Clauses, AStmt, StartLoc,
+                                                EndLoc, VarsWithInheritedDSA);
+    break;
+  case OMPD_scope:
+    Res = S.ActOnOpenMPScopeDirective(Clauses, AStmt, StartLoc, EndLoc);
+    break;
+  case OMPD_parallel_master:
+    Res =
+        S.ActOnOpenMPParallelMasterDirective(Clauses, AStmt, StartLoc, EndLoc);
+    break;
+  case OMPD_parallel_masked:
+    Res =
+        S.ActOnOpenMPParallelMaskedDirective(Clauses, AStmt, StartLoc, EndLoc);
+    break;
+  case OMPD_parallel_sections:
+    Res = S.ActOnOpenMPParallelSectionsDirective(Clauses, AStmt, StartLoc,
+                                                 EndLoc);
+    break;
+  case OMPD_task:
+    Res = S.ActOnOpenMPTaskDirective(Clauses, AStmt, StartLoc, EndLoc);
+    break;
+  case OMPD_taskyield:
+    assert(Clauses.empty() &&
+           "No clauses are allowed for 'omp taskyield' directive");
+    assert(AStmt == nullptr &&
+           "No associated statement allowed for 'omp taskyield' directive");
+    Res = S.ActOnOpenMPTaskyieldDirective(StartLoc, EndLoc);
+    break;
+  case OMPD_error:
+    assert(AStmt == nullptr &&
+           "No associated statement allowed for 'omp error' directive");
+    Res = S.ActOnOpenMPErrorDirective(Clauses, StartLoc, EndLoc);
+    break;
+  case OMPD_barrier:
+    assert(Clauses.empty() &&
+           "No clauses are allowed for 'omp barrier' directive");
+    assert(AStmt == nullptr &&
+           "No associated statement allowed for 'omp barrier' directive");
+    Res = S.ActOnOpenMPBarrierDirective(StartLoc, EndLoc);
+    break;
+  case OMPD_taskwait:
+    assert(AStmt == nullptr &&
+           "No associated statement allowed for 'omp taskwait' directive");
+    Res = S.ActOnOpenMPTaskwaitDirective(Clauses, StartLoc, EndLoc);
+    break;
+  case OMPD_taskgroup:
+    Res = S.ActOnOpenMPTaskgroupDirective(Clauses, AStmt, StartLoc, EndLoc);
+    break;
+  case OMPD_flush:
+    assert(AStmt == nullptr &&
+           "No associated statement allowed for 'omp flush' directive");
+    Res = S.ActOnOpenMPFlushDirective(Clauses, StartLoc, EndLoc);
+    break;
+  case OMPD_depobj:
+    assert(AStmt == nullptr &&
+           "No associated statement allowed for 'omp depobj' directive");
+    Res = S.ActOnOpenMPDepobjDirective(Clauses, StartLoc, EndLoc);
+    break;
+  case OMPD_scan:
+    assert(AStmt == nullptr &&
+           "No associated statement allowed for 'omp scan' directive");
+    Res = S.ActOnOpenMPScanDirective(Clauses, StartLoc, EndLoc);
+    break;
+  case OMPD_ordered:
+    Res = S.ActOnOpenMPOrderedDirective(Clauses, AStmt, StartLoc, EndLoc);
+    break;
+  case OMPD_atomic:
+    Res = S.ActOnOpenMPAtomicDirective(Clauses, AStmt, StartLoc, EndLoc);
+    break;
+  case OMPD_teams:
+    Res = S.ActOnOpenMPTeamsDirective(Clauses, AStmt, StartLoc, EndLoc);
+    break;
+  case OMPD_target:
+    Res = S.ActOnOpenMPTargetDirective(Clauses, AStmt, StartLoc, EndLoc);
+    break;
+  case OMPD_target_parallel:
+    Res =
+        S.ActOnOpenMPTargetParallelDirective(Clauses, AStmt, StartLoc, EndLoc);
+    break;
+  case OMPD_target_parallel_for:
+    Res = S.ActOnOpenMPTargetParallelForDirective(Clauses, AStmt, StartLoc,
+                                                  EndLoc, VarsWithInheritedDSA);
+    break;
+  case OMPD_cancellation_point:
+    assert(Clauses.empty() &&
+           "No clauses are allowed for 'omp cancellation point' directive");
+    assert(AStmt == nullptr && "No associated statement allowed for 'omp "
+                               "cancellation point' directive");
+    Res =
+        S.ActOnOpenMPCancellationPointDirective(StartLoc, EndLoc, CancelRegion);
+    break;
+  case OMPD_cancel:
+    assert(AStmt == nullptr &&
+           "No associated statement allowed for 'omp cancel' directive");
+    Res = S.ActOnOpenMPCancelDirective(Clauses, StartLoc, EndLoc, CancelRegion);
+    break;
+  case OMPD_target_data:
+    Res = S.ActOnOpenMPTargetDataDirective(Clauses, AStmt, StartLoc, EndLoc);
+    break;
+  case OMPD_target_enter_data:
+    Res =
+        S.ActOnOpenMPTargetEnterDataDirective(Clauses, StartLoc, EndLoc, AStmt);
+    break;
+  case OMPD_target_exit_data:
+    Res =
+        S.ActOnOpenMPTargetExitDataDirective(Clauses, StartLoc, EndLoc, AStmt);
+    break;
+  case OMPD_taskloop:
+    Res = S.ActOnOpenMPTaskLoopDirective(Clauses, AStmt, StartLoc, EndLoc,
+                                         VarsWithInheritedDSA);
+    break;
+  case OMPD_taskloop_simd:
+    Res = S.ActOnOpenMPTaskLoopSimdDirective(Clauses, AStmt, StartLoc, EndLoc,
+                                             VarsWithInheritedDSA);
+    break;
+  case OMPD_master_taskloop:
+    Res = S.ActOnOpenMPMasterTaskLoopDirective(Clauses, AStmt, StartLoc, EndLoc,
+                                               VarsWithInheritedDSA);
+    break;
+  case OMPD_masked_taskloop:
+    Res = S.ActOnOpenMPMaskedTaskLoopDirective(Clauses, AStmt, StartLoc, EndLoc,
+                                               VarsWithInheritedDSA);
+    break;
+  case OMPD_master_taskloop_simd:
+    Res = S.ActOnOpenMPMasterTaskLoopSimdDirective(
+        Clauses, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
+    break;
+  case OMPD_masked_taskloop_simd:
+    Res = S.ActOnOpenMPMaskedTaskLoopSimdDirective(
+        Clauses, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
+    break;
+  case OMPD_parallel_master_taskloop:
+    Res = S.ActOnOpenMPParallelMasterTaskLoopDirective(
+        Clauses, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
+    break;
+  case OMPD_parallel_masked_taskloop:
+    Res = S.ActOnOpenMPParallelMaskedTaskLoopDirective(
+        Clauses, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
+    break;
+  case OMPD_parallel_master_taskloop_simd:
+    Res = S.ActOnOpenMPParallelMasterTaskLoopSimdDirective(
+        Clauses, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
+    break;
+  case OMPD_parallel_masked_taskloop_simd:
+    Res = S.ActOnOpenMPParallelMaskedTaskLoopSimdDirective(
+        Clauses, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
+    break;
+  case OMPD_distribute:
+    Res = S.ActOnOpenMPDistributeDirective(Clauses, AStmt, StartLoc, EndLoc,
+                                           VarsWithInheritedDSA);
+    break;
+  case OMPD_target_update:
+    Res = S.ActOnOpenMPTargetUpdateDirective(Clauses, StartLoc, EndLoc, AStmt);
+    break;
+  case OMPD_distribute_parallel_for:
+    Res = S.ActOnOpenMPDistributeParallelForDirective(
+        Clauses, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
+    break;
+  case OMPD_distribute_parallel_for_simd:
+    Res = S.ActOnOpenMPDistributeParallelForSimdDirective(
+        Clauses, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
+    break;
+  case OMPD_distribute_simd:
+    Res = S.ActOnOpenMPDistributeSimdDirective(Clauses, AStmt, StartLoc, EndLoc,
+                                               VarsWithInheritedDSA);
+    break;
+  case OMPD_target_parallel_for_simd:
+    Res = S.ActOnOpenMPTargetParallelForSimdDirective(
+        Clauses, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
+    break;
+  case OMPD_target_simd:
+    Res = S.ActOnOpenMPTargetSimdDirective(Clauses, AStmt, StartLoc, EndLoc,
+                                           VarsWithInheritedDSA);
+    break;
+  case OMPD_teams_distribute:
+    Res = S.ActOnOpenMPTeamsDistributeDirective(Clauses, AStmt, StartLoc,
+                                                EndLoc, VarsWithInheritedDSA);
+    break;
+  case OMPD_teams_distribute_simd:
+    Res = S.ActOnOpenMPTeamsDistributeSimdDirective(
+        Clauses, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
+    break;
+  case OMPD_teams_distribute_parallel_for_simd:
+    Res = S.ActOnOpenMPTeamsDistributeParallelForSimdDirective(
+        Clauses, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
+    break;
+  case OMPD_teams_distribute_parallel_for:
+    Res = S.ActOnOpenMPTeamsDistributeParallelForDirective(
+        Clauses, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
+    break;
+  case OMPD_target_teams:
+    Res = S.ActOnOpenMPTargetTeamsDirective(Clauses, AStmt, StartLoc, EndLoc);
+    break;
+  case OMPD_target_teams_distribute:
+    Res = S.ActOnOpenMPTargetTeamsDistributeDirective(
+        Clauses, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
+    break;
+  case OMPD_target_teams_distribute_parallel_for:
+    Res = S.ActOnOpenMPTargetTeamsDistributeParallelForDirective(
+        Clauses, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
+    break;
+  case OMPD_target_teams_distribute_parallel_for_simd:
+    Res = S.ActOnOpenMPTargetTeamsDistributeParallelForSimdDirective(
+        Clauses, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
+    break;
+  case OMPD_target_teams_distribute_simd:
+    Res = S.ActOnOpenMPTargetTeamsDistributeSimdDirective(
+        Clauses, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
+    break;
+  case OMPD_interop:
+    assert(AStmt == nullptr &&
+           "No associated statement allowed for 'omp interop' directive");
+    Res = S.ActOnOpenMPInteropDirective(Clauses, StartLoc, EndLoc);
+    break;
+  case OMPD_dispatch:
+    Res = S.ActOnOpenMPDispatchDirective(Clauses, AStmt, StartLoc, EndLoc);
+    break;
+  case OMPD_loop:
+    Res = S.ActOnOpenMPGenericLoopDirective(Clauses, AStmt, StartLoc, EndLoc,
+                                            VarsWithInheritedDSA);
+    break;
+  case OMPD_teams_loop:
+    Res = S.ActOnOpenMPTeamsGenericLoopDirective(Clauses, AStmt, StartLoc,
+                                                 EndLoc, VarsWithInheritedDSA);
+    break;
+  case OMPD_target_teams_loop:
+    Res = S.ActOnOpenMPTargetTeamsGenericLoopDirective(
+        Clauses, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
+    break;
+  case OMPD_parallel_loop:
+    Res = S.ActOnOpenMPParallelGenericLoopDirective(
+        Clauses, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
+    break;
+  case OMPD_target_parallel_loop:
+    Res = S.ActOnOpenMPTargetParallelGenericLoopDirective(
+        Clauses, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
+    break;
+  case OMPD_declare_target:
+  case OMPD_end_declare_target:
+  case OMPD_threadprivate:
+  case OMPD_allocate:
+  case OMPD_declare_reduction:
+  case OMPD_declare_mapper:
+  case OMPD_declare_simd:
+  case OMPD_requires:
+  case OMPD_declare_variant:
+  case OMPD_begin_declare_variant:
+  case OMPD_end_declare_variant:
+    llvm_unreachable("OpenMP Directive is not allowed");
+  case OMPD_unknown:
+  default:
+    llvm_unreachable("Unknown OpenMP directive");
+  }
+  return Res;
+}
+
 StmtResult SemaOpenMP::ActOnOpenMPExecutableDirective(
     OpenMPDirectiveKind Kind, const DeclarationNameInfo &DirName,
     OpenMPDirectiveKind CancelRegion, ArrayRef<OMPClause *> Clauses,
@@ -6170,334 +6492,19 @@ StmtResult SemaOpenMP::ActOnOpenMPExecutableDirective(
                                             ClausesWithImplicit);
   }
 
-  switch (Kind) {
-  case OMPD_parallel:
-    Res = ActOnOpenMPParallelDirective(ClausesWithImplicit, AStmt, StartLoc,
-                                       EndLoc);
-    break;
-  case OMPD_simd:
-    Res = ActOnOpenMPSimdDirective(ClausesWithImplicit, AStmt, StartLoc, EndLoc,
-                                   VarsWithInheritedDSA);
-    break;
-  case OMPD_tile:
-    Res =
-        ActOnOpenMPTileDirective(ClausesWithImplicit, AStmt, StartLoc, EndLoc);
-    break;
-  case OMPD_unroll:
-    Res = ActOnOpenMPUnrollDirective(ClausesWithImplicit, AStmt, StartLoc,
-                                     EndLoc);
-    break;
-  case OMPD_reverse:
-    assert(ClausesWithImplicit.empty() &&
-           "reverse directive does not support any clauses");
-    Res = ActOnOpenMPReverseDirective(AStmt, StartLoc, EndLoc);
-    break;
-  case OMPD_interchange:
-    Res = ActOnOpenMPInterchangeDirective(ClausesWithImplicit, AStmt, StartLoc,
-                                          EndLoc);
-    break;
-  case OMPD_for:
-    Res = ActOnOpenMPForDirective(ClausesWithImplicit, AStmt, StartLoc, EndLoc,
-                                  VarsWithInheritedDSA);
-    break;
-  case OMPD_for_simd:
-    Res = ActOnOpenMPForSimdDirective(ClausesWithImplicit, AStmt, StartLoc,
-                                      EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_sections:
-    Res = ActOnOpenMPSectionsDirective(ClausesWithImplicit, AStmt, StartLoc,
-                                       EndLoc);
-    break;
-  case OMPD_section:
-    assert(ClausesWithImplicit.empty() &&
-           "No clauses are allowed for 'omp section' directive");
-    Res = ActOnOpenMPSectionDirective(AStmt, StartLoc, EndLoc);
-    break;
-  case OMPD_single:
-    Res = ActOnOpenMPSingleDirective(ClausesWithImplicit, AStmt, StartLoc,
-                                     EndLoc);
-    break;
-  case OMPD_master:
-    assert(ClausesWithImplicit.empty() &&
-           "No clauses are allowed for 'omp master' directive");
-    Res = ActOnOpenMPMasterDirective(AStmt, StartLoc, EndLoc);
-    break;
-  case OMPD_masked:
-    Res = ActOnOpenMPMaskedDirective(ClausesWithImplicit, AStmt, StartLoc,
-                                     EndLoc);
-    break;
-  case OMPD_critical:
-    Res = ActOnOpenMPCriticalDirective(DirName, ClausesWithImplicit, AStmt,
-                                       StartLoc, EndLoc);
-    break;
-  case OMPD_parallel_for:
-    Res = ActOnOpenMPParallelForDirective(ClausesWithImplicit, AStmt, StartLoc,
-                                          EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_parallel_for_simd:
-    Res = ActOnOpenMPParallelForSimdDirective(
-        ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_scope:
-    Res =
-        ActOnOpenMPScopeDirective(ClausesWithImplicit, AStmt, StartLoc, EndLoc);
-    break;
-  case OMPD_parallel_master:
-    Res = ActOnOpenMPParallelMasterDirective(ClausesWithImplicit, AStmt,
-                                             StartLoc, EndLoc);
-    break;
-  case OMPD_parallel_masked:
-    Res = ActOnOpenMPParallelMaskedDirective(ClausesWithImplicit, AStmt,
-                                             StartLoc, EndLoc);
-    break;
-  case OMPD_parallel_sections:
-    Res = ActOnOpenMPParallelSectionsDirective(ClausesWithImplicit, AStmt,
-                                               StartLoc, EndLoc);
-    break;
-  case OMPD_task:
-    Res =
-        ActOnOpenMPTaskDirective(ClausesWithImplicit, AStmt, StartLoc, EndLoc);
-    break;
-  case OMPD_taskyield:
-    assert(ClausesWithImplicit.empty() &&
-           "No clauses are allowed for 'omp taskyield' directive");
-    assert(AStmt == nullptr &&
-           "No associated statement allowed for 'omp taskyield' directive");
-    Res = ActOnOpenMPTaskyieldDirective(StartLoc, EndLoc);
-    break;
-  case OMPD_error:
-    assert(AStmt == nullptr &&
-           "No associated statement allowed for 'omp error' directive");
-    Res = ActOnOpenMPErrorDirective(ClausesWithImplicit, StartLoc, EndLoc);
-    break;
-  case OMPD_barrier:
-    assert(ClausesWithImplicit.empty() &&
-           "No clauses are allowed for 'omp barrier' directive");
-    assert(AStmt == nullptr &&
-           "No associated statement allowed for 'omp barrier' directive");
-    Res = ActOnOpenMPBarrierDirective(StartLoc, EndLoc);
-    break;
-  case OMPD_taskwait:
-    assert(AStmt == nullptr &&
-           "No associated statement allowed for 'omp taskwait' directive");
-    Res = ActOnOpenMPTaskwaitDirective(ClausesWithImplicit, StartLoc, EndLoc);
-    break;
-  case OMPD_taskgroup:
-    Res = ActOnOpenMPTaskgroupDirective(ClausesWithImplicit, AStmt, StartLoc,
-                                        EndLoc);
-    break;
-  case OMPD_flush:
-    assert(AStmt == nullptr &&
-           "No associated statement allowed for 'omp flush' directive");
-    Res = ActOnOpenMPFlushDirective(ClausesWithImplicit, StartLoc, EndLoc);
-    break;
-  case OMPD_depobj:
-    assert(AStmt == nullptr &&
-           "No associated statement allowed for 'omp depobj' directive");
-    Res = ActOnOpenMPDepobjDirective(ClausesWithImplicit, StartLoc, EndLoc);
-    break;
-  case OMPD_scan:
-    assert(AStmt == nullptr &&
-           "No associated statement allowed for 'omp scan' directive");
-    Res = ActOnOpenMPScanDirective(ClausesWithImplicit, StartLoc, EndLoc);
-    break;
-  case OMPD_ordered:
-    Res = ActOnOpenMPOrderedDirective(ClausesWithImplicit, AStmt, StartLoc,
-                                      EndLoc);
-    break;
-  case OMPD_atomic:
-    Res = ActOnOpenMPAtomicDirective(ClausesWithImplicit, AStmt, StartLoc,
-                                     EndLoc);
-    break;
-  case OMPD_teams:
-    Res =
-        ActOnOpenMPTeamsDirective(ClausesWithImplicit, AStmt, StartLoc, EndLoc);
-    break;
-  case OMPD_target:
-    Res = ActOnOpenMPTargetDirective(ClausesWithImplicit, AStmt, StartLoc,
-                                     EndLoc);
-    break;
-  case OMPD_target_parallel:
-    Res = ActOnOpenMPTargetParallelDirective(ClausesWithImplicit, AStmt,
-                                             StartLoc, EndLoc);
-    break;
-  case OMPD_target_parallel_for:
-    Res = ActOnOpenMPTargetParallelForDirective(
-        ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_cancellation_point:
-    assert(ClausesWithImplicit.empty() &&
-           "No clauses are allowed for 'omp cancellation point' directive");
-    assert(AStmt == nullptr && "No associated statement allowed for 'omp "
-                               "cancellation point' directive");
-    Res = ActOnOpenMPCancellationPointDirective(StartLoc, EndLoc, CancelRegion);
-    break;
-  case OMPD_cancel:
-    assert(AStmt == nullptr &&
-           "No associated statement allowed for 'omp cancel' directive");
-    Res = ActOnOpenMPCancelDirective(ClausesWithImplicit, StartLoc, EndLoc,
-                                     CancelRegion);
-    break;
-  case OMPD_target_data:
-    Res = ActOnOpenMPTargetDataDirective(ClausesWithImplicit, AStmt, StartLoc,
-                                         EndLoc);
-    break;
-  case OMPD_target_enter_data:
-    Res = ActOnOpenMPTargetEnterDataDirective(ClausesWithImplicit, StartLoc,
-                                              EndLoc, AStmt);
-    break;
-  case OMPD_target_exit_data:
-    Res = ActOnOpenMPTargetExitDataDirective(ClausesWithImplicit, StartLoc,
-                                             EndLoc, AStmt);
-    break;
-  case OMPD_taskloop:
-    Res = ActOnOpenMPTaskLoopDirective(ClausesWithImplicit, AStmt, StartLoc,
-                                       EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_taskloop_simd:
-    Res = ActOnOpenMPTaskLoopSimdDirective(ClausesWithImplicit, AStmt, StartLoc,
-                                           EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_master_taskloop:
-    Res = ActOnOpenMPMasterTaskLoopDirective(
-        ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_masked_taskloop:
-    Res = ActOnOpenMPMaskedTaskLoopDirective(
-        ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_master_taskloop_simd:
-    Res = ActOnOpenMPMasterTaskLoopSimdDirective(
-        ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_masked_taskloop_simd:
-    Res = ActOnOpenMPMaskedTaskLoopSimdDirective(
-        ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_parallel_master_taskloop:
-    Res = ActOnOpenMPParallelMasterTaskLoopDirective(
-        ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_parallel_masked_taskloop:
-    Res = ActOnOpenMPParallelMaskedTaskLoopDirective(
-        ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_parallel_master_taskloop_simd:
-    Res = ActOnOpenMPParallelMasterTaskLoopSimdDirective(
-        ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_parallel_masked_taskloop_simd:
-    Res = ActOnOpenMPParallelMaskedTaskLoopSimdDirective(
-        ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_distribute:
-    Res = ActOnOpenMPDistributeDirective(ClausesWithImplicit, AStmt, StartLoc,
-                                         EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_target_update:
-    Res = ActOnOpenMPTargetUpdateDirective(ClausesWithImplicit, StartLoc,
-                                           EndLoc, AStmt);
-    break;
-  case OMPD_distribute_parallel_for:
-    Res = ActOnOpenMPDistributeParallelForDirective(
-        ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_distribute_parallel_for_simd:
-    Res = ActOnOpenMPDistributeParallelForSimdDirective(
-        ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_distribute_simd:
-    Res = ActOnOpenMPDistributeSimdDirective(
-        ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_target_parallel_for_simd:
-    Res = ActOnOpenMPTargetParallelForSimdDirective(
-        ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_target_simd:
-    Res = ActOnOpenMPTargetSimdDirective(ClausesWithImplicit, AStmt, StartLoc,
-                                         EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_teams_distribute:
-    Res = ActOnOpenMPTeamsDistributeDirective(
-        ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_teams_distribute_simd:
-    Res = ActOnOpenMPTeamsDistributeSimdDirective(
-        ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_teams_distribute_parallel_for_simd:
-    Res = ActOnOpenMPTeamsDistributeParallelForSimdDirective(
-        ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_teams_distribute_parallel_for:
-    Res = ActOnOpenMPTeamsDistributeParallelForDirective(
-        ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_target_teams:
-    Res = ActOnOpenMPTargetTeamsDirective(ClausesWithImplicit, AStmt, StartLoc,
-                                          EndLoc);
-    break;
-  case OMPD_target_teams_distribute:
-    Res = ActOnOpenMPTargetTeamsDistributeDirective(
-        ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_target_teams_distribute_parallel_for:
-    Res = ActOnOpenMPTargetTeamsDistributeParallelForDirective(
-        ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_target_teams_distribute_parallel_for_simd:
-    Res = ActOnOpenMPTargetTeamsDistributeParallelForSimdDirective(
-        ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_target_teams_distribute_simd:
-    Res = ActOnOpenMPTargetTeamsDistributeSimdDirective(
-        ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_interop:
-    assert(AStmt == nullptr &&
-           "No associated statement allowed for 'omp interop' directive");
-    Res = ActOnOpenMPInteropDirective(ClausesWithImplicit, StartLoc, EndLoc);
-    break;
-  case OMPD_dispatch:
-    Res = ActOnOpenMPDispatchDirective(ClausesWithImplicit, AStmt, StartLoc,
-                                       EndLoc);
-    break;
-  case OMPD_loop:
-    Res = ActOnOpenMPGenericLoopDirective(ClausesWithImplicit, AStmt, StartLoc,
-                                          EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_teams_loop:
-    Res = ActOnOpenMPTeamsGenericLoopDirective(
-        ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_target_teams_loop:
-    Res = ActOnOpenMPTargetTeamsGenericLoopDirective(
-        ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_parallel_loop:
-    Res = ActOnOpenMPParallelGenericLoopDirective(
-        ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_target_parallel_loop:
-    Res = ActOnOpenMPTargetParallelGenericLoopDirective(
-        ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA);
-    break;
-  case OMPD_declare_target:
-  case OMPD_end_declare_target:
-  case OMPD_threadprivate:
-  case OMPD_allocate:
-  case OMPD_declare_reduction:
-  case OMPD_declare_mapper:
-  case OMPD_declare_simd:
-  case OMPD_requires:
-  case OMPD_declare_variant:
-  case OMPD_begin_declare_variant:
-  case OMPD_end_declare_variant:
-    llvm_unreachable("OpenMP Directive is not allowed");
-  case OMPD_unknown:
-  default:
-    llvm_unreachable("Unknown OpenMP directive");
+  if (!SemaRef.CurContext->isDependentContext()) {
+    Res = createASTForDirective(SemaRef, Kind, ClausesWithImplicit, AStmt,
+                                StartLoc, EndLoc, DirName, CancelRegion,
+                                VarsWithInheritedDSA);
+  } else {
+    if (getDirectiveAssociation(Kind) == Association::Loop)
+      Res = ActOnOpenMPOpaqueLoopDirective(Kind, ClausesWithImplicit, AStmt,
+                                           StartLoc, EndLoc,
+                                           VarsWithInheritedDSA);
+    else
+      Res = ActOnOpenMPOpaqueBlockDirective(Kind, ClausesWithImplicit, AStmt,
+                                            CancelRegion, DirName, StartLoc,
+                                            EndLoc);
   }
 
   ErrorFound = Res.isInvalid() || ErrorFound;
@@ -11148,6 +11155,21 @@ StmtResult SemaOpenMP::ActOnOpenMPDepobjDirective(ArrayRef<OMPClause *> Clauses,
   return OMPDepobjDirective::Create(getASTContext(), StartLoc, EndLoc, Clauses);
 }
 
+static bool checkScanScope(Sema &S, Scope *CurrentS, SourceLocation Loc) {
+  bool ErrorFound = false;
+  // Check that scan directive is used in the scope of the OpenMP loop body.
+  if (CurrentS) {
+    Scope *ParentS = CurrentS->getParent();
+    if (!ParentS || ParentS->getParent() != ParentS->getBreakParent() ||
+        !ParentS->getBreakParent()->isOpenMPLoopScope()) {
+      S.Diag(Loc, diag::err_omp_orphaned_device_directive)
+          << getOpenMPDirectiveName(OMPD_scan) << 5;
+      ErrorFound = true;
+    }
+  }
+  return ErrorFound;
+}
+
 StmtResult SemaOpenMP::ActOnOpenMPScanDirective(ArrayRef<OMPClause *> Clauses,
                                                 SourceLocation StartLoc,
                                                 SourceLocation EndLoc) {
@@ -11158,13 +11180,9 @@ StmtResult SemaOpenMP::ActOnOpenMPScanDirective(ArrayRef<OMPClause *> Clauses,
     return StmtError();
   }
   // Check that scan directive is used in the scope of the OpenMP loop body.
-  if (Scope *S = DSAStack->getCurScope()) {
-    Scope *ParentS = S->getParent();
-    if (!ParentS || ParentS->getParent() != ParentS->getBreakParent() ||
-        !ParentS->getBreakParent()->isOpenMPLoopScope())
-      return StmtError(Diag(StartLoc, diag::err_omp_orphaned_device_directive)
-                       << getOpenMPDirectiveName(OMPD_scan) << 5);
-  }
+  if (checkScanScope(SemaRef, DSAStack->getCurScope(), StartLoc))
+    return StmtError();
+
   // Check that only one instance of scan directives is used in the same outer
   // region.
   if (DSAStack->doesParentHasScanDirective()) {
@@ -23289,6 +23307,143 @@ StmtResult SemaOpenMP::ActOnOpenMPScopeDirective(ArrayRef<OMPClause *> Clauses,
                                    AStmt);
 }
 
+static bool checkScanScope(Sema &S, Scope *CurrentS, SourceLocation Loc);
+
+StmtResult SemaOpenMP::ActOnOpenMPOpaqueBlockDirective(
+    OpenMPDirectiveKind DKind, ArrayRef<OMPClause *> Clauses, Stmt *AStmt,
+    OpenMPDirectiveKind CancelRegion, const DeclarationNameInfo &DirName,
+    SourceLocation StartLoc, SourceLocation EndLoc) {
+  bool NeedsStmt = false;
+  if (DKind == OMPD_section || DKind == OMPD_target_enter_data ||
+      DKind == OMPD_target_exit_data || DKind == OMPD_target_update) {
+    // The association of these in the spec is either "none" or "separating",
+    // but they do have an associated statement in clang.
+    NeedsStmt = true;
+  } else if (DKind != OMPD_ordered) {
+    // "ordered" has two versions, one with and one without a statement.
+    Association Assoc = getDirectiveAssociation(DKind);
+    NeedsStmt = Assoc != Association::None && Assoc != Association::Separating;
+  }
+
+  if (!AStmt && NeedsStmt)
+    return StmtError();
+
+  if (AStmt && isOpenMPCapturingDirective(DKind))
+    setBranchProtectedScope(SemaRef, DKind, AStmt);
+
+  // "scan" is the only executable, non-loop-associated directive so far
+  // that relies on DSAStack->getCurScope for diagnostic checks. The scope is
+  // not available (i.e. nullptr) when instantiating a template, so perform
+  // the check early.
+  if (DKind == OMPD_scan) {
+    if (checkScanScope(SemaRef, DSAStack->getCurScope(), StartLoc))
+      return StmtError();
+  }
+
+  Expr *ReductionRef = nullptr;
+  assert(!isOpenMPSimdDirective(DKind) && "Unexpected loop directive");
+  if (DKind == OMPD_taskgroup || isOpenMPParallelDirective(DKind) ||
+      isOpenMPWorksharingDirective(DKind))
+    ReductionRef = DSAStack->getTaskgroupReductionRef();
+
+  return OMPOpaqueBlockDirective::Create(
+      getASTContext(), StartLoc, EndLoc, DKind, Clauses, AStmt, ReductionRef,
+      DSAStack->isCancelRegion(), CancelRegion, DirName);
+}
+
+StmtResult SemaOpenMP::ActOnOpenMPOpaqueLoopDirective(
+    OpenMPDirectiveKind DKind, ArrayRef<OMPClause *> Clauses, Stmt *AStmt,
+    SourceLocation StartLoc, SourceLocation EndLoc,
+    VarsWithInheritedDSAType &VarsWithImplicitDSA) {
+  if (!AStmt)
+    return StmtError();
+
+  if (isOpenMPLoopTransformationDirective(DKind)) {
+    switch (DKind) {
+    case OMPD_tile:
+      return ActOnOpenMPTileDirective(Clauses, AStmt, StartLoc, EndLoc);
+    case OMPD_unroll:
+      return ActOnOpenMPUnrollDirective(Clauses, AStmt, StartLoc, EndLoc);
+    case OMPD_reverse:
+      return ActOnOpenMPReverseDirective(AStmt, StartLoc, EndLoc);
+    case OMPD_interchange:
+      return ActOnOpenMPInterchangeDirective(Clauses, AStmt, StartLoc, EndLoc);
+    default:
+      llvm_unreachable("Unexpected loop-transformatinal directive");
+    }
+  }
+
+  ArrayRef<OpenMPDirectiveKind> Leafs = getLeafConstructsOrSelf(DKind);
+
+  if (isOpenMPGenericLoopDirective(DKind)) {
+    // OpenMP 5.1 [2.11.7, loop construct, Restrictions]
+    // A list item may not appear in a lastprivate clause unless it is the
+    // loop iteration variable of a loop that is associated with the construct.
+    if (checkGenericLoopLastprivate(SemaRef, Clauses, DKind, DSAStack))
+      return StmtError();
+  }
+
+  CapturedStmt *CS = [&]() {
+    switch (DKind) {
+    case OMPD_distribute:
+    case OMPD_for:
+    case OMPD_taskloop:
+      return cast<CapturedStmt>(AStmt);
+    default:
+      return setBranchProtectedScope(SemaRef, DKind, AStmt);
+    }
+  }();
+
+  // assert(isa<CapturedStmt>(AStmt) && "Captured statement expected");
+  OMPLoopBasedDirective::HelperExprs B;
+  // In presence of clause 'collapse' or 'ordered' with number of loops, it will
+  // define the nested loops number.
+  unsigned NestedLoopCount = checkOpenMPLoop(
+      DKind, getCollapseNumberExpr(Clauses), getOrderedNumberExpr(Clauses), CS,
+      SemaRef, *DSAStack, VarsWithImplicitDSA, B);
+  if (NestedLoopCount == 0)
+    return StmtError();
+
+  assert((SemaRef.CurContext->isDependentContext() || B.builtAll()) &&
+         "omp for loop exprs were not built");
+
+  if (finishLinearClauses(SemaRef, Clauses, B, DSAStack))
+    return StmtError();
+
+  if (isOpenMPSimdDirective(DKind)) {
+    if (checkSimdlenSafelenSpecified(SemaRef, Clauses))
+      return StmtError();
+  }
+
+  if (llvm::is_contained(Leafs, OMPD_taskloop)) {
+    // OpenMP, [2.9.2 taskloop Construct, Restrictions]
+    // The grainsize clause and num_tasks clause are mutually exclusive and may
+    // not appear on the same taskloop directive.
+    if (checkMutuallyExclusiveClauses(SemaRef, Clauses,
+                                      {OMPC_grainsize, OMPC_num_tasks}))
+      return StmtError();
+    // OpenMP, [2.9.2 taskloop Construct, Restrictions]
+    // If a reduction clause is present on the taskloop directive, the nogroup
+    // clause must not be specified.
+    if (checkReductionClauseWithNogroup(SemaRef, Clauses))
+      return StmtError();
+  }
+
+  if (llvm::is_contained(Leafs, OMPD_teams))
+    DSAStack->setParentTeamsRegionLoc(StartLoc);
+
+  Expr *ReductionRef = nullptr;
+  assert(DKind != OMPD_taskgroup && "Unexpected block directive");
+  if ((isOpenMPParallelDirective(DKind) ||
+       isOpenMPWorksharingDirective(DKind)) &&
+      !isOpenMPSimdDirective(DKind))
+    ReductionRef = DSAStack->getTaskgroupReductionRef();
+
+  return OMPOpaqueLoopDirective::Create(
+      getASTContext(), StartLoc, EndLoc, DKind, NestedLoopCount, Clauses, AStmt,
+      B, ReductionRef, DSAStack->isCancelRegion());
+}
+
 OMPClause *SemaOpenMP::ActOnOpenMPInclusiveClause(ArrayRef<Expr *> VarList,
                                                   SourceLocation StartLoc,
                                                   SourceLocation LParenLoc,
diff --git a/clang/lib/Sema/TreeTransform.h b/clang/lib/Sema/TreeTransform.h
index 81e515e..3a49681 100644
--- a/clang/lib/Sema/TreeTransform.h
+++ b/clang/lib/Sema/TreeTransform.h
@@ -9271,6 +9271,8 @@ TreeTransform<Derived>::TransformOMPCanonicalLoop(OMPCanonicalLoop *L) {
 template <typename Derived>
 StmtResult TreeTransform<Derived>::TransformOMPExecutableDirective(
     OMPExecutableDirective *D) {
+  // If D is OMPOpaqueBlockDirective or OMPOpaqueLoopDirective,
+  // then D->getDirectiveKind() will return the actual directive kind.
 
   // Transform the clauses
   llvm::SmallVector<OMPClause *, 16> TClauses;
@@ -9322,14 +9324,29 @@ StmtResult TreeTransform<Derived>::TransformOMPExecutableDirective(
   // Transform directive name for 'omp critical' directive.
   DeclarationNameInfo DirName;
   if (D->getDirectiveKind() == OMPD_critical) {
-    DirName = cast<OMPCriticalDirective>(D)->getDirectiveName();
+    if (auto *C = dyn_cast<OMPCriticalDirective>(D))
+      DirName = C->getDirectiveName();
+    else if (auto *C = dyn_cast<OMPOpaqueBlockDirective>(D))
+      DirName = C->getDirectiveName();
+    else
+      llvm_unreachable("Unexpected directive class");
     DirName = getDerived().TransformDeclarationNameInfo(DirName);
   }
   OpenMPDirectiveKind CancelRegion = OMPD_unknown;
   if (D->getDirectiveKind() == OMPD_cancellation_point) {
-    CancelRegion = cast<OMPCancellationPointDirective>(D)->getCancelRegion();
+    if (auto *C = dyn_cast<OMPCancellationPointDirective>(D))
+      CancelRegion = C->getCancelRegion();
+    else if (auto *C = dyn_cast<OMPOpaqueBlockDirective>(D))
+      CancelRegion = C->getCancelRegion();
+    else
+      llvm_unreachable("Unexpected directive class");
   } else if (D->getDirectiveKind() == OMPD_cancel) {
-    CancelRegion = cast<OMPCancelDirective>(D)->getCancelRegion();
+    if (auto *C = dyn_cast<OMPCancelDirective>(D))
+      CancelRegion = C->getCancelRegion();
+    else if (auto *C = dyn_cast<OMPOpaqueBlockDirective>(D))
+      CancelRegion = C->getCancelRegion();
+    else
+      llvm_unreachable("Unexpected directive class");
   }
 
   return getDerived().RebuildOMPExecutableDirective(
@@ -9390,6 +9407,28 @@ StmtResult TreeTransform<Derived>::TransformOMPInformationalDirective(
 }
 
 template <typename Derived>
+StmtResult TreeTransform<Derived>::TransformOMPOpaqueBlockDirective(
+    OMPOpaqueBlockDirective *D) {
+  DeclarationNameInfo DirName;
+  getDerived().getSema().OpenMP().StartOpenMPDSABlock(
+      D->getDirectiveKind(), DirName, nullptr, D->getBeginLoc());
+  StmtResult Res = getDerived().TransformOMPExecutableDirective(D);
+  getDerived().getSema().OpenMP().EndOpenMPDSABlock(Res.get());
+  return Res;
+}
+
+template <typename Derived>
+StmtResult TreeTransform<Derived>::TransformOMPOpaqueLoopDirective(
+    OMPOpaqueLoopDirective *D) {
+  DeclarationNameInfo DirName;
+  getDerived().getSema().OpenMP().StartOpenMPDSABlock(
+      D->getDirectiveKind(), DirName, nullptr, D->getBeginLoc());
+  StmtResult Res = getDerived().TransformOMPExecutableDirective(D);
+  getDerived().getSema().OpenMP().EndOpenMPDSABlock(Res.get());
+  return Res;
+}
+
+template <typename Derived>
 StmtResult
 TreeTransform<Derived>::TransformOMPMetaDirective(OMPMetaDirective *D) {
   // TODO: Fix This
diff --git a/clang/lib/Serialization/ASTReaderStmt.cpp b/clang/lib/Serialization/ASTReaderStmt.cpp
index 731ad0b..4585693 100644
--- a/clang/lib/Serialization/ASTReaderStmt.cpp
+++ b/clang/lib/Serialization/ASTReaderStmt.cpp
@@ -2400,7 +2400,6 @@ void ASTStmtReader::VisitOMPExecutableDirective(OMPExecutableDirective *E) {
 }
 
 void ASTStmtReader::VisitOMPLoopBasedDirective(OMPLoopBasedDirective *D) {
-  VisitStmt(D);
   // Field CollapsedNum was read in ReadStmtFromStream.
   Record.skipInts(1);
   VisitOMPExecutableDirective(D);
@@ -2410,6 +2409,24 @@ void ASTStmtReader::VisitOMPLoopDirective(OMPLoopDirective *D) {
   VisitOMPLoopBasedDirective(D);
 }
 
+void ASTStmtReader::VisitOMPOpaqueBlockDirective(OMPOpaqueBlockDirective *D) {
+  VisitStmt(D);
+  // The DKind was read in ReadStmtFromStream.
+  Record.skipInts(1);
+  VisitOMPExecutableDirective(D);
+  D->setHasCancel(Record.readBool());
+  D->setCancelRegion(Record.readEnum<OpenMPDirectiveKind>());
+  D->setDirectiveName(Record.readDeclarationNameInfo());
+}
+
+void ASTStmtReader::VisitOMPOpaqueLoopDirective(OMPOpaqueLoopDirective *D) {
+  VisitStmt(D);
+  // The DKind was read in ReadStmtFromStream.
+  Record.skipInts(1);
+  VisitOMPLoopDirective(D);
+  D->setHasCancel(Record.readBool());
+}
+
 void ASTStmtReader::VisitOMPMetaDirective(OMPMetaDirective *D) {
   VisitStmt(D);
   // The NumClauses field was read in ReadStmtFromStream.
@@ -2424,11 +2441,13 @@ void ASTStmtReader::VisitOMPParallelDirective(OMPParallelDirective *D) {
 }
 
 void ASTStmtReader::VisitOMPSimdDirective(OMPSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
 }
 
 void ASTStmtReader::VisitOMPLoopTransformationDirective(
     OMPLoopTransformationDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopBasedDirective(D);
   D->setNumGeneratedLoops(Record.readUInt32());
 }
@@ -2450,11 +2469,13 @@ void ASTStmtReader::VisitOMPInterchangeDirective(OMPInterchangeDirective *D) {
 }
 
 void ASTStmtReader::VisitOMPForDirective(OMPForDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   D->setHasCancel(Record.readBool());
 }
 
 void ASTStmtReader::VisitOMPForSimdDirective(OMPForSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
 }
 
@@ -2492,12 +2513,14 @@ void ASTStmtReader::VisitOMPCriticalDirective(OMPCriticalDirective *D) {
 }
 
 void ASTStmtReader::VisitOMPParallelForDirective(OMPParallelForDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   D->setHasCancel(Record.readBool());
 }
 
 void ASTStmtReader::VisitOMPParallelForSimdDirective(
     OMPParallelForSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
 }
 
@@ -2619,6 +2642,7 @@ void ASTStmtReader::VisitOMPTargetParallelDirective(
 
 void ASTStmtReader::VisitOMPTargetParallelForDirective(
     OMPTargetParallelForDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   D->setHasCancel(Record.readBool());
 }
@@ -2642,59 +2666,70 @@ void ASTStmtReader::VisitOMPCancelDirective(OMPCancelDirective *D) {
 }
 
 void ASTStmtReader::VisitOMPTaskLoopDirective(OMPTaskLoopDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   D->setHasCancel(Record.readBool());
 }
 
 void ASTStmtReader::VisitOMPTaskLoopSimdDirective(OMPTaskLoopSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
 }
 
 void ASTStmtReader::VisitOMPMasterTaskLoopDirective(
     OMPMasterTaskLoopDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   D->setHasCancel(Record.readBool());
 }
 
 void ASTStmtReader::VisitOMPMaskedTaskLoopDirective(
     OMPMaskedTaskLoopDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   D->setHasCancel(Record.readBool());
 }
 
 void ASTStmtReader::VisitOMPMasterTaskLoopSimdDirective(
     OMPMasterTaskLoopSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
 }
 
 void ASTStmtReader::VisitOMPMaskedTaskLoopSimdDirective(
     OMPMaskedTaskLoopSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
 }
 
 void ASTStmtReader::VisitOMPParallelMasterTaskLoopDirective(
     OMPParallelMasterTaskLoopDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   D->setHasCancel(Record.readBool());
 }
 
 void ASTStmtReader::VisitOMPParallelMaskedTaskLoopDirective(
     OMPParallelMaskedTaskLoopDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   D->setHasCancel(Record.readBool());
 }
 
 void ASTStmtReader::VisitOMPParallelMasterTaskLoopSimdDirective(
     OMPParallelMasterTaskLoopSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
 }
 
 void ASTStmtReader::VisitOMPParallelMaskedTaskLoopSimdDirective(
     OMPParallelMaskedTaskLoopSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
 }
 
 void ASTStmtReader::VisitOMPDistributeDirective(OMPDistributeDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
 }
 
@@ -2705,46 +2740,55 @@ void ASTStmtReader::VisitOMPTargetUpdateDirective(OMPTargetUpdateDirective *D) {
 
 void ASTStmtReader::VisitOMPDistributeParallelForDirective(
     OMPDistributeParallelForDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   D->setHasCancel(Record.readBool());
 }
 
 void ASTStmtReader::VisitOMPDistributeParallelForSimdDirective(
     OMPDistributeParallelForSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
 }
 
 void ASTStmtReader::VisitOMPDistributeSimdDirective(
     OMPDistributeSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
 }
 
 void ASTStmtReader::VisitOMPTargetParallelForSimdDirective(
     OMPTargetParallelForSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
 }
 
 void ASTStmtReader::VisitOMPTargetSimdDirective(OMPTargetSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
 }
 
 void ASTStmtReader::VisitOMPTeamsDistributeDirective(
     OMPTeamsDistributeDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
 }
 
 void ASTStmtReader::VisitOMPTeamsDistributeSimdDirective(
     OMPTeamsDistributeSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
 }
 
 void ASTStmtReader::VisitOMPTeamsDistributeParallelForSimdDirective(
     OMPTeamsDistributeParallelForSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
 }
 
 void ASTStmtReader::VisitOMPTeamsDistributeParallelForDirective(
     OMPTeamsDistributeParallelForDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   D->setHasCancel(Record.readBool());
 }
@@ -2756,22 +2800,26 @@ void ASTStmtReader::VisitOMPTargetTeamsDirective(OMPTargetTeamsDirective *D) {
 
 void ASTStmtReader::VisitOMPTargetTeamsDistributeDirective(
     OMPTargetTeamsDistributeDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
 }
 
 void ASTStmtReader::VisitOMPTargetTeamsDistributeParallelForDirective(
     OMPTargetTeamsDistributeParallelForDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   D->setHasCancel(Record.readBool());
 }
 
 void ASTStmtReader::VisitOMPTargetTeamsDistributeParallelForSimdDirective(
     OMPTargetTeamsDistributeParallelForSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
 }
 
 void ASTStmtReader::VisitOMPTargetTeamsDistributeSimdDirective(
     OMPTargetTeamsDistributeSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
 }
 
@@ -2792,27 +2840,32 @@ void ASTStmtReader::VisitOMPMaskedDirective(OMPMaskedDirective *D) {
 }
 
 void ASTStmtReader::VisitOMPGenericLoopDirective(OMPGenericLoopDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
 }
 
 void ASTStmtReader::VisitOMPTeamsGenericLoopDirective(
     OMPTeamsGenericLoopDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
 }
 
 void ASTStmtReader::VisitOMPTargetTeamsGenericLoopDirective(
     OMPTargetTeamsGenericLoopDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   D->setCanBeParallelFor(Record.readBool());
 }
 
 void ASTStmtReader::VisitOMPParallelGenericLoopDirective(
     OMPParallelGenericLoopDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
 }
 
 void ASTStmtReader::VisitOMPTargetParallelGenericLoopDirective(
     OMPTargetParallelGenericLoopDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
 }
 
@@ -3460,6 +3513,27 @@ Stmt *ASTReader::ReadStmtFromStream(ModuleFile &F) {
       S = OMPCanonicalLoop::createEmpty(Context);
       break;
 
+    case STMT_OMP_OPAQUE_BLOCK_DIRECTIVE: {
+      unsigned DKind = Record[ASTStmtReader::NumStmtFields];
+      unsigned NumClauses = Record[ASTStmtReader::NumStmtFields + 1];
+      // NumChildren is at index + 2.
+      bool HasAssociatedStmt = Record[ASTStmtReader::NumStmtFields + 3];
+      S = OMPOpaqueBlockDirective::CreateEmpty(
+          Context, static_cast<OpenMPDirectiveKind>(DKind), NumClauses,
+          HasAssociatedStmt, Empty);
+      break;
+    }
+
+    case STMT_OMP_OPAQUE_LOOP_DIRECTIVE: {
+      unsigned DKind = Record[ASTStmtReader::NumStmtFields];
+      unsigned CollapsedNum = Record[ASTStmtReader::NumStmtFields + 1];
+      unsigned NumClauses = Record[ASTStmtReader::NumStmtFields + 2];
+      S = OMPOpaqueLoopDirective::CreateEmpty(
+          Context, static_cast<OpenMPDirectiveKind>(DKind), NumClauses,
+          CollapsedNum, Empty);
+      break;
+    }
+
     case STMT_OMP_META_DIRECTIVE:
       S = OMPMetaDirective::CreateEmpty(
           Context, Record[ASTStmtReader::NumStmtFields], Empty);
diff --git a/clang/lib/Serialization/ASTWriterStmt.cpp b/clang/lib/Serialization/ASTWriterStmt.cpp
index 4994047..47968dd 100644
--- a/clang/lib/Serialization/ASTWriterStmt.cpp
+++ b/clang/lib/Serialization/ASTWriterStmt.cpp
@@ -2401,7 +2401,6 @@ void ASTStmtWriter::VisitOMPExecutableDirective(OMPExecutableDirective *E) {
 }
 
 void ASTStmtWriter::VisitOMPLoopBasedDirective(OMPLoopBasedDirective *D) {
-  VisitStmt(D);
   Record.writeUInt32(D->getLoopsNumber());
   VisitOMPExecutableDirective(D);
 }
@@ -2410,6 +2409,24 @@ void ASTStmtWriter::VisitOMPLoopDirective(OMPLoopDirective *D) {
   VisitOMPLoopBasedDirective(D);
 }
 
+void ASTStmtWriter::VisitOMPOpaqueBlockDirective(OMPOpaqueBlockDirective *D) {
+  VisitStmt(D);
+  Record.writeUInt32(static_cast<unsigned>(D->getDirectiveKind()));
+  VisitOMPExecutableDirective(D);
+  Record.writeBool(D->hasCancel());
+  Record.writeEnum(D->getCancelRegion());
+  Record.AddDeclarationNameInfo(D->getDirectiveName());
+  Code = serialization::STMT_OMP_OPAQUE_BLOCK_DIRECTIVE;
+}
+
+void ASTStmtWriter::VisitOMPOpaqueLoopDirective(OMPOpaqueLoopDirective *D) {
+  VisitStmt(D);
+  Record.writeUInt32(static_cast<unsigned>(D->getDirectiveKind()));
+  VisitOMPLoopDirective(D);
+  Record.writeBool(D->hasCancel());
+  Code = serialization::STMT_OMP_OPAQUE_LOOP_DIRECTIVE;
+}
+
 void ASTStmtWriter::VisitOMPMetaDirective(OMPMetaDirective *D) {
   VisitStmt(D);
   Record.push_back(D->getNumClauses());
@@ -2425,12 +2442,14 @@ void ASTStmtWriter::VisitOMPParallelDirective(OMPParallelDirective *D) {
 }
 
 void ASTStmtWriter::VisitOMPSimdDirective(OMPSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Code = serialization::STMT_OMP_SIMD_DIRECTIVE;
 }
 
 void ASTStmtWriter::VisitOMPLoopTransformationDirective(
     OMPLoopTransformationDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopBasedDirective(D);
   Record.writeUInt32(D->getNumGeneratedLoops());
 }
@@ -2456,12 +2475,14 @@ void ASTStmtWriter::VisitOMPInterchangeDirective(OMPInterchangeDirective *D) {
 }
 
 void ASTStmtWriter::VisitOMPForDirective(OMPForDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Record.writeBool(D->hasCancel());
   Code = serialization::STMT_OMP_FOR_DIRECTIVE;
 }
 
 void ASTStmtWriter::VisitOMPForSimdDirective(OMPForSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Code = serialization::STMT_OMP_FOR_SIMD_DIRECTIVE;
 }
@@ -2506,6 +2527,7 @@ void ASTStmtWriter::VisitOMPCriticalDirective(OMPCriticalDirective *D) {
 }
 
 void ASTStmtWriter::VisitOMPParallelForDirective(OMPParallelForDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Record.writeBool(D->hasCancel());
   Code = serialization::STMT_OMP_PARALLEL_FOR_DIRECTIVE;
@@ -2513,6 +2535,7 @@ void ASTStmtWriter::VisitOMPParallelForDirective(OMPParallelForDirective *D) {
 
 void ASTStmtWriter::VisitOMPParallelForSimdDirective(
     OMPParallelForSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Code = serialization::STMT_OMP_PARALLEL_FOR_SIMD_DIRECTIVE;
 }
@@ -2591,6 +2614,7 @@ void ASTStmtWriter::VisitOMPTargetParallelDirective(
 
 void ASTStmtWriter::VisitOMPTargetParallelForDirective(
     OMPTargetParallelForDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Record.writeBool(D->hasCancel());
   Code = serialization::STMT_OMP_TARGET_PARALLEL_FOR_DIRECTIVE;
@@ -2680,18 +2704,21 @@ void ASTStmtWriter::VisitOMPCancelDirective(OMPCancelDirective *D) {
 }
 
 void ASTStmtWriter::VisitOMPTaskLoopDirective(OMPTaskLoopDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Record.writeBool(D->hasCancel());
   Code = serialization::STMT_OMP_TASKLOOP_DIRECTIVE;
 }
 
 void ASTStmtWriter::VisitOMPTaskLoopSimdDirective(OMPTaskLoopSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Code = serialization::STMT_OMP_TASKLOOP_SIMD_DIRECTIVE;
 }
 
 void ASTStmtWriter::VisitOMPMasterTaskLoopDirective(
     OMPMasterTaskLoopDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Record.writeBool(D->hasCancel());
   Code = serialization::STMT_OMP_MASTER_TASKLOOP_DIRECTIVE;
@@ -2699,6 +2726,7 @@ void ASTStmtWriter::VisitOMPMasterTaskLoopDirective(
 
 void ASTStmtWriter::VisitOMPMaskedTaskLoopDirective(
     OMPMaskedTaskLoopDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Record.writeBool(D->hasCancel());
   Code = serialization::STMT_OMP_MASKED_TASKLOOP_DIRECTIVE;
@@ -2706,18 +2734,21 @@ void ASTStmtWriter::VisitOMPMaskedTaskLoopDirective(
 
 void ASTStmtWriter::VisitOMPMasterTaskLoopSimdDirective(
     OMPMasterTaskLoopSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Code = serialization::STMT_OMP_MASTER_TASKLOOP_SIMD_DIRECTIVE;
 }
 
 void ASTStmtWriter::VisitOMPMaskedTaskLoopSimdDirective(
     OMPMaskedTaskLoopSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Code = serialization::STMT_OMP_MASKED_TASKLOOP_SIMD_DIRECTIVE;
 }
 
 void ASTStmtWriter::VisitOMPParallelMasterTaskLoopDirective(
     OMPParallelMasterTaskLoopDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Record.writeBool(D->hasCancel());
   Code = serialization::STMT_OMP_PARALLEL_MASTER_TASKLOOP_DIRECTIVE;
@@ -2725,6 +2756,7 @@ void ASTStmtWriter::VisitOMPParallelMasterTaskLoopDirective(
 
 void ASTStmtWriter::VisitOMPParallelMaskedTaskLoopDirective(
     OMPParallelMaskedTaskLoopDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Record.writeBool(D->hasCancel());
   Code = serialization::STMT_OMP_PARALLEL_MASKED_TASKLOOP_DIRECTIVE;
@@ -2732,17 +2764,20 @@ void ASTStmtWriter::VisitOMPParallelMaskedTaskLoopDirective(
 
 void ASTStmtWriter::VisitOMPParallelMasterTaskLoopSimdDirective(
     OMPParallelMasterTaskLoopSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Code = serialization::STMT_OMP_PARALLEL_MASTER_TASKLOOP_SIMD_DIRECTIVE;
 }
 
 void ASTStmtWriter::VisitOMPParallelMaskedTaskLoopSimdDirective(
     OMPParallelMaskedTaskLoopSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Code = serialization::STMT_OMP_PARALLEL_MASKED_TASKLOOP_SIMD_DIRECTIVE;
 }
 
 void ASTStmtWriter::VisitOMPDistributeDirective(OMPDistributeDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Code = serialization::STMT_OMP_DISTRIBUTE_DIRECTIVE;
 }
@@ -2755,6 +2790,7 @@ void ASTStmtWriter::VisitOMPTargetUpdateDirective(OMPTargetUpdateDirective *D) {
 
 void ASTStmtWriter::VisitOMPDistributeParallelForDirective(
     OMPDistributeParallelForDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Record.writeBool(D->hasCancel());
   Code = serialization::STMT_OMP_DISTRIBUTE_PARALLEL_FOR_DIRECTIVE;
@@ -2762,47 +2798,55 @@ void ASTStmtWriter::VisitOMPDistributeParallelForDirective(
 
 void ASTStmtWriter::VisitOMPDistributeParallelForSimdDirective(
     OMPDistributeParallelForSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Code = serialization::STMT_OMP_DISTRIBUTE_PARALLEL_FOR_SIMD_DIRECTIVE;
 }
 
 void ASTStmtWriter::VisitOMPDistributeSimdDirective(
     OMPDistributeSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Code = serialization::STMT_OMP_DISTRIBUTE_SIMD_DIRECTIVE;
 }
 
 void ASTStmtWriter::VisitOMPTargetParallelForSimdDirective(
     OMPTargetParallelForSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Code = serialization::STMT_OMP_TARGET_PARALLEL_FOR_SIMD_DIRECTIVE;
 }
 
 void ASTStmtWriter::VisitOMPTargetSimdDirective(OMPTargetSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Code = serialization::STMT_OMP_TARGET_SIMD_DIRECTIVE;
 }
 
 void ASTStmtWriter::VisitOMPTeamsDistributeDirective(
     OMPTeamsDistributeDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Code = serialization::STMT_OMP_TEAMS_DISTRIBUTE_DIRECTIVE;
 }
 
 void ASTStmtWriter::VisitOMPTeamsDistributeSimdDirective(
     OMPTeamsDistributeSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Code = serialization::STMT_OMP_TEAMS_DISTRIBUTE_SIMD_DIRECTIVE;
 }
 
 void ASTStmtWriter::VisitOMPTeamsDistributeParallelForSimdDirective(
     OMPTeamsDistributeParallelForSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Code = serialization::STMT_OMP_TEAMS_DISTRIBUTE_PARALLEL_FOR_SIMD_DIRECTIVE;
 }
 
 void ASTStmtWriter::VisitOMPTeamsDistributeParallelForDirective(
     OMPTeamsDistributeParallelForDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Record.writeBool(D->hasCancel());
   Code = serialization::STMT_OMP_TEAMS_DISTRIBUTE_PARALLEL_FOR_DIRECTIVE;
@@ -2816,12 +2860,14 @@ void ASTStmtWriter::VisitOMPTargetTeamsDirective(OMPTargetTeamsDirective *D) {
 
 void ASTStmtWriter::VisitOMPTargetTeamsDistributeDirective(
     OMPTargetTeamsDistributeDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Code = serialization::STMT_OMP_TARGET_TEAMS_DISTRIBUTE_DIRECTIVE;
 }
 
 void ASTStmtWriter::VisitOMPTargetTeamsDistributeParallelForDirective(
     OMPTargetTeamsDistributeParallelForDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Record.writeBool(D->hasCancel());
   Code = serialization::STMT_OMP_TARGET_TEAMS_DISTRIBUTE_PARALLEL_FOR_DIRECTIVE;
@@ -2829,6 +2875,7 @@ void ASTStmtWriter::VisitOMPTargetTeamsDistributeParallelForDirective(
 
 void ASTStmtWriter::VisitOMPTargetTeamsDistributeParallelForSimdDirective(
     OMPTargetTeamsDistributeParallelForSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Code = serialization::
       STMT_OMP_TARGET_TEAMS_DISTRIBUTE_PARALLEL_FOR_SIMD_DIRECTIVE;
@@ -2836,6 +2883,7 @@ void ASTStmtWriter::VisitOMPTargetTeamsDistributeParallelForSimdDirective(
 
 void ASTStmtWriter::VisitOMPTargetTeamsDistributeSimdDirective(
     OMPTargetTeamsDistributeSimdDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Code = serialization::STMT_OMP_TARGET_TEAMS_DISTRIBUTE_SIMD_DIRECTIVE;
 }
@@ -2860,18 +2908,21 @@ void ASTStmtWriter::VisitOMPMaskedDirective(OMPMaskedDirective *D) {
 }
 
 void ASTStmtWriter::VisitOMPGenericLoopDirective(OMPGenericLoopDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Code = serialization::STMT_OMP_GENERIC_LOOP_DIRECTIVE;
 }
 
 void ASTStmtWriter::VisitOMPTeamsGenericLoopDirective(
     OMPTeamsGenericLoopDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Code = serialization::STMT_OMP_TEAMS_GENERIC_LOOP_DIRECTIVE;
 }
 
 void ASTStmtWriter::VisitOMPTargetTeamsGenericLoopDirective(
     OMPTargetTeamsGenericLoopDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Record.writeBool(D->canBeParallelFor());
   Code = serialization::STMT_OMP_TARGET_TEAMS_GENERIC_LOOP_DIRECTIVE;
@@ -2879,12 +2930,14 @@ void ASTStmtWriter::VisitOMPTargetTeamsGenericLoopDirective(
 
 void ASTStmtWriter::VisitOMPParallelGenericLoopDirective(
     OMPParallelGenericLoopDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Code = serialization::STMT_OMP_PARALLEL_GENERIC_LOOP_DIRECTIVE;
 }
 
 void ASTStmtWriter::VisitOMPTargetParallelGenericLoopDirective(
     OMPTargetParallelGenericLoopDirective *D) {
+  VisitStmt(D);
   VisitOMPLoopDirective(D);
   Code = serialization::STMT_OMP_TARGET_PARALLEL_GENERIC_LOOP_DIRECTIVE;
 }
diff --git a/clang/lib/StaticAnalyzer/Core/ExprEngine.cpp b/clang/lib/StaticAnalyzer/Core/ExprEngine.cpp
index b46cd9f..b980d6d 100644
--- a/clang/lib/StaticAnalyzer/Core/ExprEngine.cpp
+++ b/clang/lib/StaticAnalyzer/Core/ExprEngine.cpp
@@ -1748,6 +1748,8 @@ void ExprEngine::Visit(const Stmt *S, ExplodedNode *Pred,
     case Stmt::SEHLeaveStmtClass:
     case Stmt::SEHFinallyStmtClass:
     case Stmt::OMPCanonicalLoopClass:
+    case Stmt::OMPOpaqueBlockDirectiveClass:
+    case Stmt::OMPOpaqueLoopDirectiveClass:
     case Stmt::OMPParallelDirectiveClass:
     case Stmt::OMPSimdDirectiveClass:
     case Stmt::OMPForDirectiveClass:
diff --git a/clang/test/OpenMP/Inputs/nesting_of_regions.cpp b/clang/test/OpenMP/Inputs/nesting_of_regions.cpp
index 985cdc0..f17c5e5 100644
--- a/clang/test/OpenMP/Inputs/nesting_of_regions.cpp
+++ b/clang/test/OpenMP/Inputs/nesting_of_regions.cpp
@@ -19565,5 +19565,5 @@ void foo() {
     bar();
   }
 
-  return foo<int>();
+  return foo<int>(); // expected-note {{in instantiation of function template specialization 'foo<int>' requested here}}
 }
diff --git a/clang/test/OpenMP/atomic_messages.cpp b/clang/test/OpenMP/atomic_messages.cpp
index d492f6e..ab8da28 100644
--- a/clang/test/OpenMP/atomic_messages.cpp
+++ b/clang/test/OpenMP/atomic_messages.cpp
@@ -46,20 +46,21 @@ T read() {
   T a = T(), b = T();
 // Test for atomic read
 #pragma omp atomic read
-  // expected-error@+2 {{the statement for 'atomic read' must be an expression statement of form 'v = x;', where v and x are both lvalue expressions with scalar type}}
-  // expected-note@+1 {{expected an expression statement}}
+  // expected-error@+2 2 {{the statement for 'atomic read' must be an expression statement of form 'v = x;', where v and x are both lvalue expressions with scalar type}}
+  // expected-note@+1 2 {{expected an expression statement}}
   ;
 #pragma omp atomic read
-  // expected-error@+2 {{the statement for 'atomic read' must be an expression statement of form 'v = x;', where v and x are both lvalue expressions with scalar type}}
-  // expected-note@+1 {{expected built-in assignment operator}}
+  // expected-error@+2 2 {{the statement for 'atomic read' must be an expression statement of form 'v = x;', where v and x are both lvalue expressions with scalar type}}
+  // expected-note@+1 2 {{expected built-in assignment operator}}
   foo();
 #pragma omp atomic read
   // expected-error@+2 2 {{the statement for 'atomic read' must be an expression statement of form 'v = x;', where v and x are both lvalue expressions with scalar type}}
   // expected-note@+1 2 {{expected built-in assignment operator}}
   a += b;
 #pragma omp atomic read
-  // expected-error@+2 {{the statement for 'atomic read' must be an expression statement of form 'v = x;', where v and x are both lvalue expressions with scalar type}}
-  // expected-note@+1 {{expected lvalue expression}}
+  // expected-error@+3 2 {{the statement for 'atomic read' must be an expression statement of form 'v = x;', where v and x are both lvalue expressions with scalar type}}
+  // expected-note@+2 {{expected lvalue expression}}
+  // expected-note@+1 {{expected built-in assignment operator}}
   a = 0;
 #pragma omp atomic read
   // expected-error@+2 {{the statement for 'atomic read' must be an expression statement of form 'v = x;', where v and x are both lvalue expressions with scalar type}}
@@ -335,6 +336,7 @@ int update() {
   // expected-note@+1 {{expected an expression statement}}
   ;
 
+  // expected-note@+1 {{in instantiation of function template specialization 'update<int>' requested here}}
   return update<int>();
 }
 
@@ -724,6 +726,7 @@ int seq_cst() {
   // expected-note@+1 {{expected an expression statement}}
   ;
 
+  // expected-note@+1 {{in instantiation of function template specialization 'seq_cst<int>' requested here}}
  return seq_cst<int>();
 }
 
@@ -735,7 +738,7 @@ T acq_rel() {
   // expected-error@+2 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an lvalue expression with scalar type}}
   // expected-note@+1 {{expected an expression statement}}
   ;
-// omp50-error@+1 2 {{directive '#pragma omp atomic' cannot contain more than one 'seq_cst', 'relaxed', 'acq_rel', 'acquire' or 'release' clause}} omp50-note@+1 2 {{'acq_rel' clause used here}} omp45-error@+1 {{unexpected OpenMP clause 'acq_rel' in directive '#pragma omp atomic'}} omp50-error@+1 2 {{directive '#pragma omp atomic read' cannot be used with 'acq_rel' clause}} omp50-note@+1 2 {{'acq_rel' clause used here}}
+// omp50-error@+1 {{directive '#pragma omp atomic' cannot contain more than one 'seq_cst', 'relaxed', 'acq_rel', 'acquire' or 'release' clause}} omp50-note@+1 {{'acq_rel' clause used here}} omp45-error@+1 {{unexpected OpenMP clause 'acq_rel' in directive '#pragma omp atomic'}} omp50-error@+1 {{directive '#pragma omp atomic read' cannot be used with 'acq_rel' clause}} omp50-note@+1 {{'acq_rel' clause used here}}
 #pragma omp atomic read acq_rel seq_cst
   a = b;
 
@@ -766,7 +769,7 @@ int acq_rel() {
   // expected-note@+1 {{expected an expression statement}}
   ;
 
- return acq_rel<int>(); // omp50-note {{in instantiation of function template specialization 'acq_rel<int>' requested here}}
+ return acq_rel<int>(); // expected-note {{in instantiation of function template specialization 'acq_rel<int>' requested here}}
 }
 
 template <class T>
@@ -777,7 +780,7 @@ T acquire() {
   // expected-error@+2 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an lvalue expression with scalar type}}
   // expected-note@+1 {{expected an expression statement}}
   ;
-// omp50-error@+1 2 {{directive '#pragma omp atomic' cannot contain more than one 'seq_cst', 'relaxed', 'acq_rel', 'acquire' or 'release' clause}} omp50-note@+1 2 {{'acquire' clause used here}} omp45-error@+1 {{unexpected OpenMP clause 'acquire' in directive '#pragma omp atomic'}} omp50-error@+1 2 {{directive '#pragma omp atomic' cannot be used with 'acquire' clause}} omp50-note@+1 2 {{'acquire' clause used here}}
+// omp50-error@+1 {{directive '#pragma omp atomic' cannot contain more than one 'seq_cst', 'relaxed', 'acq_rel', 'acquire' or 'release' clause}} omp50-note@+1 {{'acquire' clause used here}} omp45-error@+1 {{unexpected OpenMP clause 'acquire' in directive '#pragma omp atomic'}} omp50-error@+1 {{directive '#pragma omp atomic' cannot be used with 'acquire' clause}} omp50-note@+1 {{'acquire' clause used here}}
 #pragma omp atomic acquire seq_cst
   a += b;
 
@@ -808,7 +811,7 @@ int acquire() {
   // expected-note@+1 {{expected an expression statement}}
   ;
 
- return acquire<int>(); // omp50-note {{in instantiation of function template specialization 'acquire<int>' requested here}}
+ return acquire<int>(); // expected-note {{in instantiation of function template specialization 'acquire<int>' requested here}}
 }
 
 template <class T>
@@ -819,7 +822,7 @@ T release() {
   // expected-error@+2 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an lvalue expression with scalar type}}
   // expected-note@+1 {{expected an expression statement}}
   ;
-// omp50-error@+1 2 {{directive '#pragma omp atomic' cannot contain more than one 'seq_cst', 'relaxed', 'acq_rel', 'acquire' or 'release' clause}} omp50-note@+1 2 {{'release' clause used here}} omp45-error@+1 {{unexpected OpenMP clause 'release' in directive '#pragma omp atomic'}}
+// omp50-error@+1 {{directive '#pragma omp atomic' cannot contain more than one 'seq_cst', 'relaxed', 'acq_rel', 'acquire' or 'release' clause}} omp50-note@+1 {{'release' clause used here}} omp45-error@+1 {{unexpected OpenMP clause 'release' in directive '#pragma omp atomic'}}
 #pragma omp atomic release seq_cst
   a += b;
 
@@ -850,7 +853,7 @@ int release() {
   // expected-note@+1 {{expected an expression statement}}
   ;
 
- return release<int>(); // omp50-note {{in instantiation of function template specialization 'release<int>' requested here}}
+ return release<int>(); // expected-note {{in instantiation of function template specialization 'release<int>' requested here}}
 }
 
 template <class T>
@@ -861,7 +864,7 @@ T relaxed() {
   // expected-error@+2 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an lvalue expression with scalar type}}
   // expected-note@+1 {{expected an expression statement}}
   ;
-// omp50-error@+1 2 {{directive '#pragma omp atomic' cannot contain more than one 'seq_cst', 'relaxed', 'acq_rel', 'acquire' or 'release' clause}} omp50-note@+1 2 {{'relaxed' clause used here}} omp45-error@+1 {{unexpected OpenMP clause 'relaxed' in directive '#pragma omp atomic'}}
+// omp50-error@+1 {{directive '#pragma omp atomic' cannot contain more than one 'seq_cst', 'relaxed', 'acq_rel', 'acquire' or 'release' clause}} omp50-note@+1 {{'relaxed' clause used here}} omp45-error@+1 {{unexpected OpenMP clause 'relaxed' in directive '#pragma omp atomic'}}
 #pragma omp atomic relaxed seq_cst
   a += b;
 
@@ -892,35 +895,35 @@ int relaxed() {
   // expected-note@+1 {{expected an expression statement}}
   ;
 
- return relaxed<int>(); // omp50-note {{in instantiation of function template specialization 'relaxed<int>' requested here}}
+ return relaxed<int>(); // expected-note {{in instantiation of function template specialization 'relaxed<int>' requested here}}
 }
 
 template <class T>
 T mixed() {
   T a, b = T();
-// expected-error@+2 2 {{directive '#pragma omp atomic' cannot contain more than one 'read', 'write', 'update', 'capture', or 'compare' clause}}
-// expected-note@+1 2 {{'read' clause used here}}
+// expected-error@+2 {{directive '#pragma omp atomic' cannot contain more than one 'read', 'write', 'update', 'capture', or 'compare' clause}}
+// expected-note@+1 {{'read' clause used here}}
 #pragma omp atomic read write
   a = b;
-// expected-error@+2 2 {{directive '#pragma omp atomic' cannot contain more than one 'read', 'write', 'update', 'capture', or 'compare' clause}}
-// expected-note@+1 2 {{'write' clause used here}}
+// expected-error@+2 {{directive '#pragma omp atomic' cannot contain more than one 'read', 'write', 'update', 'capture', or 'compare' clause}}
+// expected-note@+1 {{'write' clause used here}}
 #pragma omp atomic write read
   a = b;
-// expected-error@+2 2 {{directive '#pragma omp atomic' cannot contain more than one 'read', 'write', 'update', 'capture', or 'compare' clause}}
-// expected-note@+1 2 {{'update' clause used here}}
+// expected-error@+2 {{directive '#pragma omp atomic' cannot contain more than one 'read', 'write', 'update', 'capture', or 'compare' clause}}
+// expected-note@+1 {{'update' clause used here}}
 #pragma omp atomic update read
   a += b;
-// expected-error@+2 2 {{directive '#pragma omp atomic' cannot contain more than one 'read', 'write', 'update', 'capture', or 'compare' clause}}
-// expected-note@+1 2 {{'capture' clause used here}}
+// expected-error@+2 {{directive '#pragma omp atomic' cannot contain more than one 'read', 'write', 'update', 'capture', or 'compare' clause}}
+// expected-note@+1 {{'capture' clause used here}}
 #pragma omp atomic capture read
   a = ++b;
 #ifdef OMP51
-// expected-error@+2 2 {{directive '#pragma omp atomic' cannot contain more than one 'read', 'write', 'update', 'capture', or 'compare' clause}}
-// expected-note@+1 2 {{'write' clause used here}}
+// expected-error@+2 {{directive '#pragma omp atomic' cannot contain more than one 'read', 'write', 'update', 'capture', or 'compare' clause}}
+// expected-note@+1 {{'write' clause used here}}
 #pragma omp atomic write compare
   a = b;
-// expected-error@+2 2 {{directive '#pragma omp atomic' cannot contain more than one 'read', 'write', 'update', 'capture', or 'compare' clause}}
-// expected-note@+1 2 {{'read' clause used here}}
+// expected-error@+2 {{directive '#pragma omp atomic' cannot contain more than one 'read', 'write', 'update', 'capture', or 'compare' clause}}
+// expected-note@+1 {{'read' clause used here}}
 #pragma omp atomic read compare
   a = b;
 #endif
diff --git a/clang/test/OpenMP/critical_messages.cpp b/clang/test/OpenMP/critical_messages.cpp
index 9a7267c..bb41268 100644
--- a/clang/test/OpenMP/critical_messages.cpp
+++ b/clang/test/OpenMP/critical_messages.cpp
@@ -69,7 +69,9 @@ int tmain(int argc, char **argv) { // expected-note {{declared here}}
   foo();
   #pragma omp critical (name) hint(N) // expected-error {{argument to 'hint' clause must be a non-negative integer value}} expected-error {{constructs with the same name must have a 'hint' clause with the same value}} expected-note {{'hint' clause with value '4'}}
   foo();
-  #pragma omp critical hint(N) // expected-error {{the name of the construct must be specified in presence of 'hint' clause}}
+  // expected-error@+2 {{the name of the construct must be specified in presence of 'hint' clause}}
+  // expected-error@+1 {{argument to 'hint' clause must be a non-negative integer value}}
+  #pragma omp critical hint(N)
   foo();
 
   const int omp_lock_hint_none = 0;
diff --git a/clang/test/OpenMP/default_firstprivate_ast_print.cpp b/clang/test/OpenMP/default_firstprivate_ast_print.cpp
index 4bf9fc6..aa70339 100644
--- a/clang/test/OpenMP/default_firstprivate_ast_print.cpp
+++ b/clang/test/OpenMP/default_firstprivate_ast_print.cpp
@@ -48,7 +48,7 @@ struct SomeKernel {
     // PRINT-NEXT:     return this->targetDev++;
     // PRINT-NEXT:  }();
     // PRINT-NEXT: }
-    // DUMP: -OMPParallelDirective
+    // DUMP: -OMPOpaqueBlockDirective{{.*}}'parallel'
     // DUMP-NEXT: -OMPDefaultClause
     // DUMP-NOT:   -OMPFirstprivateClause
   }
diff --git a/clang/test/OpenMP/default_private_ast_print.cpp b/clang/test/OpenMP/default_private_ast_print.cpp
index b15a764..f1f8369 100644
--- a/clang/test/OpenMP/default_private_ast_print.cpp
+++ b/clang/test/OpenMP/default_private_ast_print.cpp
@@ -44,7 +44,7 @@ struct SomeKernel {
     // PRINT-NEXT: {
     // PRINT-NEXT:  this->targetDev++;
     // CHECK-NEXT: }
-    // DUMP: -OMPParallelDirective
+    // DUMP: -OMPOpaqueBlockDirective{{.*}}'parallel'
     // DUMP->NEXT: -OMPDefaultClause
   }
   // PRINT: template<> void apply<32U>()
diff --git a/clang/test/OpenMP/error_message.cpp b/clang/test/OpenMP/error_message.cpp
index 227cbf6..e434125 100644
--- a/clang/test/OpenMP/error_message.cpp
+++ b/clang/test/OpenMP/error_message.cpp
@@ -81,15 +81,12 @@ if (1)
 // expected-error@+1 {{ERROR}}
 #pragma omp error severity(fatal) severity(fatal) // expected-error {{directive '#pragma omp error' cannot contain more than one 'severity' clause}}
 
-// expected-warning@+2 {{WARNING}}
 // expected-warning@+1 {{WARNING}}
 #pragma omp error severity(warning) severity(warning) // expected-error {{directive '#pragma omp error' cannot contain more than one 'severity' clause}}
 
-// expected-warning@+1 {{WARNING}}
 #pragma omp error severity(warning) // expected-warning {{WARNING}}
 #pragma omp error severity(fatal) // expected-error {{ERROR}}
 
-// expected-warning@+1 {{WARNING}}
 #pragma omp error at(compilation) severity(warning) // expected-warning {{WARNING}}
 #pragma omp error at(execution) severity(warning) // no error, diagnosic at runtime
 #pragma omp error at(compilation) severity(fatal) // expected-error {{ERROR}}
@@ -98,10 +95,8 @@ if (1)
 #pragma omp error message("GPU compiler is needed.") // expected-error {{GPU compiler is needed}}
 #pragma omp error at(compilation) message("GPU compiler is needed.") // expected-error {{GPU compiler is needed}}
 #pragma omp error at(execution) message("GPU compiler is needed.") // no error
-// expected-warning@+1 {{GPU compiler is needed.}}
 #pragma omp error severity(warning) message("GPU compiler is needed.") // expected-warning {{GPU compiler is needed.}}
 #pragma omp error severity(fatal) message("GPU compiler is needed.") // expected-error {{GPU compiler is needed}}
-// expected-warning@+1 {{GPU compiler is needed.}}
 #pragma omp error at(compilation) severity(warning) message("GPU compiler is needed.") // expected-warning {{GPU compiler is needed.}}
 #pragma omp error at(compilation) severity(fatal) message("GPU compiler is needed.") // expected-error {{GPU compiler is needed.}}
 #pragma omp error at(execution) severity(warning) message("GPU compiler is needed.") // no warning warning will emit at runtime.
diff --git a/clang/test/OpenMP/generic_loop_ast_print.cpp b/clang/test/OpenMP/generic_loop_ast_print.cpp
index b361724..46eb98d 100644
--- a/clang/test/OpenMP/generic_loop_ast_print.cpp
+++ b/clang/test/OpenMP/generic_loop_ast_print.cpp
@@ -31,7 +31,7 @@
 //DUMP: FunctionTemplateDecl{{.*}}templ_foo
 //DUMP: TemplateTypeParmDecl{{.*}}T
 //DUMP: NonTypeTemplateParmDecl{{.*}}C
-//DUMP: OMPGenericLoopDirective
+//DUMP: OMPOpaqueLoopDirective{{.*}}'loop'
 //DUMP: OMPCollapseClause
 //DUMP: DeclRefExpr{{.*}}'C' 'int'
 //DUMP: OMPReductionClause
diff --git a/clang/test/OpenMP/interchange_ast_print.cpp b/clang/test/OpenMP/interchange_ast_print.cpp
index 4b0818f..2e3f1e4 100644
--- a/clang/test/OpenMP/interchange_ast_print.cpp
+++ b/clang/test/OpenMP/interchange_ast_print.cpp
@@ -135,7 +135,7 @@ void foo4(int start, int end, int step) {
 template<typename T, T P>
 void foo5(T start, T end) {
   // PRINT: #pragma omp for
-  // DUMP:  OMPForDirective
+  // DUMP:  OMPOpaqueLoopDirective
   #pragma omp for
     // PRINT:     #pragma omp interchange permutation(P + 1, 2 - P)
     // DUMP:      OMPInterchangeDirective
diff --git a/clang/test/OpenMP/interop_ast_print.cpp b/clang/test/OpenMP/interop_ast_print.cpp
index fed6feb..b0bbf27 100644
--- a/clang/test/OpenMP/interop_ast_print.cpp
+++ b/clang/test/OpenMP/interop_ast_print.cpp
@@ -259,7 +259,7 @@ void fooTemp() {
   omp_interop_t interop_var;
   //PRINT: #pragma omp interop init(prefer_type(I,4,"level_one"), target : interop_var)
   //DUMP: FunctionDecl{{.*}}fooTemp
-  //DUMP: OMPInteropDirective
+  //DUMP: OMPOpaqueBlockDirective{{.*}}'interop'
   //DUMP: OMPInitClause
   //DUMP: DeclRefExpr{{.*}}'omp_interop_t'{{.*}}'interop_var'
   //DUMP: DeclRefExpr{{.*}}NonTypeTemplateParm{{.*}}'I' 'int'
@@ -287,7 +287,7 @@ void barTemp(T t) {
   //PRINT: #pragma omp interop init(prefer_type(4,"level_one"), target : t)
   //DUMP: FunctionDecl{{.*}}barTemp 'void (T)'
   //DUMP: ParmVarDecl{{.*}}t 'T'
-  //DUMP: OMPInteropDirective
+  //DUMP: OMPOpaqueBlockDirective{{.*}}'interop'
   //DUMP: OMPInitClause
   //DUMP: DeclRefExpr{{.*}}ParmVar{{.*}}'t' 'T'
   //DUMP: IntegerLiteral{{.*}}'int' 4
@@ -295,13 +295,13 @@ void barTemp(T t) {
   #pragma omp interop init(prefer_type(4,"level_one"), target: t)
 
   //PRINT: #pragma omp interop use(t)
-  //DUMP: OMPInteropDirective
+  //DUMP: OMPOpaqueBlockDirective{{.*}}'interop'
   //DUMP: OMPUseClause
   //DUMP: DeclRefExpr{{.*}}ParmVar{{.*}}'t' 'T'
   #pragma omp interop use(t)
 
   //PRINT: #pragma omp interop destroy(t)
-  //DUMP: OMPInteropDirective
+  //DUMP: OMPOpaqueBlockDirective{{.*}}'interop'
   //DUMP: OMPDestroyClause
   //DUMP: DeclRefExpr{{.*}}ParmVar{{.*}}'t' 'T'
   #pragma omp interop destroy(t)
diff --git a/clang/test/OpenMP/scope_ast_print.cpp b/clang/test/OpenMP/scope_ast_print.cpp
index c5c3a29..73abb1c 100644
--- a/clang/test/OpenMP/scope_ast_print.cpp
+++ b/clang/test/OpenMP/scope_ast_print.cpp
@@ -71,7 +71,7 @@ int template_test() {
 //DUMP: FunctionTemplateDecl {{.*}}run
 //DUMP: TemplateTypeParmDecl {{.*}}referenced typename depth 0 index 0 T
 //DUMP: FunctionDecl {{.*}}run 'T ()'
-//DUMP: OMPScopeDirective
+//DUMP: OMPOpaqueBlockDirective{{.*}}'scope'
 //DUMP: OMPPrivateClause
 //DUMP: DeclRefExpr {{.*}}'T' lvalue Var {{.*}} 'a' 'T'
 //DUMP: OMPReductionClause
diff --git a/clang/test/OpenMP/target_update_from_messages.cpp b/clang/test/OpenMP/target_update_from_messages.cpp
index 02cd293..d9126c2 100644
--- a/clang/test/OpenMP/target_update_from_messages.cpp
+++ b/clang/test/OpenMP/target_update_from_messages.cpp
@@ -1,7 +1,7 @@
-// RUN: %clang_cc1 -verify=expected,le51 -fopenmp -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
-// RUN: %clang_cc1 -verify=expected,le45 -fopenmp -fopenmp-version=40 -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
-// RUN: %clang_cc1 -verify=expected,le45 -fopenmp -fopenmp-version=45 -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
-// RUN: %clang_cc1 -verify=expected,le51 -fopenmp -fopenmp-version=50 -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le51 -fopenmp -ferror-limit 150 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le45 -fopenmp -fopenmp-version=40 -ferror-limit 150 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le45 -fopenmp -fopenmp-version=45 -ferror-limit 150 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le51 -fopenmp -fopenmp-version=50 -ferror-limit 150 %s -Wno-openmp-mapping -Wuninitialized
 
 // RUN: %clang_cc1 -verify=expected,le51 -fopenmp-simd -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
 
@@ -99,14 +99,14 @@ T tmain(T argc) {
   T *m;
   S7 s7;
 
-#pragma omp target update from // expected-error {{expected '(' after 'from'}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
-#pragma omp target update from( // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected expression}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
-#pragma omp target update from() // expected-error {{expected expression}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
-#pragma omp target update() // expected-warning {{extra tokens at the end of '#pragma omp target update' are ignored}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
-#pragma omp target update from(alloc) // expected-error {{use of undeclared identifier 'alloc'}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from // expected-error {{expected '(' after 'from'}} expected-error 2 {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from( // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected expression}} expected-error 2 {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from() // expected-error {{expected expression}} expected-error 2 {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update() // expected-warning {{extra tokens at the end of '#pragma omp target update' are ignored}} expected-error 2 {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(alloc) // expected-error {{use of undeclared identifier 'alloc'}} expected-error 2 {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
 #pragma omp target update from(x)
 #pragma omp target update from(t[:I])
-#pragma omp target update from(T) // expected-error {{'T' does not refer to a value}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(T) // expected-error {{'T' does not refer to a value}} expected-error 2 {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
 #pragma omp target update from(I) // le51-error 2 {{expected addressable lvalue in 'from' clause}} le45-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
 #pragma omp target update from(S2::S2s)
 #pragma omp target update from(S2::S2sc)
@@ -114,10 +114,10 @@ T tmain(T argc) {
 #pragma omp target update from(y x) // expected-error {{expected ',' or ')' in 'from' clause}}
 #pragma omp target update from(argc > 0 ? x : y) // le51-error 2 {{expected addressable lvalue in 'from' clause}} le45-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
 
-#pragma omp target update from(S1) // expected-error {{'S1' does not refer to a value}}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(S1) // expected-error {{'S1' does not refer to a value}}} expected-error 2 {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
 #pragma omp target update from(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}}
 #pragma omp target update from(ba)
-#pragma omp target update from(h) // expected-error {{threadprivate variables are not allowed in 'from' clause}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(h) // expected-error {{threadprivate variables are not allowed in 'from' clause}} expected-error 2 {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
 #pragma omp target update from(k), to(k) // expected-error 2 {{variable can appear only once in OpenMP 'target update' construct}} expected-note 2 {{used here}}
 #pragma omp target update from(t), from(t[:5]) // expected-error 2 {{variable can appear only once in OpenMP 'target update' construct}} expected-note 2 {{used here}}
 #pragma omp target update from(da)
diff --git a/clang/test/OpenMP/target_update_to_messages.cpp b/clang/test/OpenMP/target_update_to_messages.cpp
index 33aaf3b..8b6ffbd 100644
--- a/clang/test/OpenMP/target_update_to_messages.cpp
+++ b/clang/test/OpenMP/target_update_to_messages.cpp
@@ -1,10 +1,10 @@
-// RUN: %clang_cc1 -verify=expected,le51 -fopenmp -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
-// RUN: %clang_cc1 -verify=expected,le45 -fopenmp -fopenmp-version=40 -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
-// RUN: %clang_cc1 -verify=expected,le45 -fopenmp -fopenmp-version=45 -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le51 -fopenmp -ferror-limit 150 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le45 -fopenmp -fopenmp-version=40 -ferror-limit 150 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le45 -fopenmp -fopenmp-version=45 -ferror-limit 150 %s -Wno-openmp-mapping -Wuninitialized
 
-// RUN: %clang_cc1 -verify=expected,le51 -fopenmp-simd -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
-// RUN: %clang_cc1 -verify=expected,le45 -fopenmp-simd -fopenmp-version=40 -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
-// RUN: %clang_cc1 -verify=expected,le45 -fopenmp-simd -fopenmp-version=45 -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le51 -fopenmp-simd -ferror-limit 150 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le45 -fopenmp-simd -fopenmp-version=40 -ferror-limit 150 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le45 -fopenmp-simd -fopenmp-version=45 -ferror-limit 150 %s -Wno-openmp-mapping -Wuninitialized
 
 void foo() {
 }
@@ -105,24 +105,24 @@ T tmain(T argc) {
   const T (&l)[5] = da;
   S7 s7;
 
-#pragma omp target update to // expected-error {{expected '(' after 'to'}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
-#pragma omp target update to( // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected expression}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
-#pragma omp target update to() // expected-error {{expected expression}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
-#pragma omp target update() // expected-warning {{extra tokens at the end of '#pragma omp target update' are ignored}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
-#pragma omp target update to(alloc) // expected-error {{use of undeclared identifier 'alloc'}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to // expected-error {{expected '(' after 'to'}} expected-error 2 {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to( // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected expression}} expected-error 2 {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to() // expected-error {{expected expression}} expected-error 2 {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update() // expected-warning {{extra tokens at the end of '#pragma omp target update' are ignored}} expected-error 2 {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(alloc) // expected-error {{use of undeclared identifier 'alloc'}} expected-error 2 {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
 #pragma omp target update to(x)
 #pragma omp target update to(t[:I])
-#pragma omp target update to(T) // expected-error {{'T' does not refer to a value}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(T) // expected-error {{'T' does not refer to a value}} expected-error 2 {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
 #pragma omp target update to(I) // le51-error 2 {{expected addressable lvalue in 'to' clause}} le45-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
 #pragma omp target update to(S2::S2s)
 #pragma omp target update to(S2::S2sc)
 #pragma omp target update to(to)
 #pragma omp target update to(y x) // expected-error {{expected ',' or ')' in 'to' clause}}
 #pragma omp target update to(argc > 0 ? x : y) // le51-error 2 {{expected addressable lvalue in 'to' clause}} le45-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
-#pragma omp target update to(S1) // expected-error {{'S1' does not refer to a value}}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(S1) // expected-error {{'S1' does not refer to a value}}} expected-error 2 {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
 #pragma omp target update to(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}}
 #pragma omp target update to(ba)
-#pragma omp target update to(h) // expected-error {{threadprivate variables are not allowed in 'to' clause}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(h) // expected-error {{threadprivate variables are not allowed in 'to' clause}} expected-error 2 {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
 #pragma omp target update to(k), from(k) // expected-error 2 {{variable can appear only once in OpenMP 'target update' construct}} expected-note 2 {{used here}}
 #pragma omp target update to(t), to(t[:5]) // expected-error 2 {{variable can appear only once in OpenMP 'target update' construct}} expected-note 2 {{used here}}
 #pragma omp target update to(da)
diff --git a/clang/test/OpenMP/task_messages.cpp b/clang/test/OpenMP/task_messages.cpp
index cd91ba6..e30379d 100644
--- a/clang/test/OpenMP/task_messages.cpp
+++ b/clang/test/OpenMP/task_messages.cpp
@@ -157,9 +157,11 @@ int foo() {
   ;
 #pragma omp task detach(evt) detach(evt) // omp45-error 2 {{unexpected OpenMP clause 'detach' in directive '#pragma omp task'}} expected-error {{directive '#pragma omp task' cannot contain more than one 'detach' clause}}
 #pragma omp task detach(cevt) detach(revt) // omp45-error 2 {{unexpected OpenMP clause 'detach' in directive '#pragma omp task'}} expected-error {{directive '#pragma omp task' cannot contain more than one 'detach' clause}} omp5-error {{expected variable of the 'omp_event_handle_t' type, not 'const omp_event_handle_t' (aka 'const unsigned long')}} omp5-error {{expected variable of the 'omp_event_handle_t' type, not 'omp_event_handle_t &' (aka 'unsigned long &')}}
-#pragma omp task detach(evt) mergeable // omp45-error {{unexpected OpenMP clause 'detach' in directive '#pragma omp task'}} omp5-error {{'mergeable' and 'detach' clause are mutually exclusive and may not appear on the same directive}} omp5-note {{'detach' clause is specified here}}
+// omp5-note@+1 {{'detach' clause is specified here}}
+#pragma omp task detach(evt) mergeable // omp45-error {{unexpected OpenMP clause 'detach' in directive '#pragma omp task'}} omp5-error 2 {{'mergeable' and 'detach' clause are mutually exclusive and may not appear on the same directive}} omp5-note {{'detach' clause is specified here}}
   ;
-#pragma omp task mergeable detach(evt) // omp45-error {{unexpected OpenMP clause 'detach' in directive '#pragma omp task'}} omp5-error {{'detach' and 'mergeable' clause are mutually exclusive and may not appear on the same directive}} omp5-note {{'mergeable' clause is specified here}}
+// omp5-note@+1 {{'mergeable' clause is specified here}}
+#pragma omp task mergeable detach(evt) // omp45-error {{unexpected OpenMP clause 'detach' in directive '#pragma omp task'}} omp5-error 2 {{'detach' and 'mergeable' clause are mutually exclusive and may not appear on the same directive}} omp5-note {{'mergeable' clause is specified here}}
 #pragma omp task detach(-evt) // omp45-error {{unexpected OpenMP clause 'detach' in directive '#pragma omp task'}} omp5-error {{expected variable of the 'omp_event_handle_t' type}}
   ;
 #pragma omp task detach(evt) shared(evt) // omp45-error {{unexpected OpenMP clause 'detach' in directive '#pragma omp task'}}
diff --git a/clang/tools/libclang/CIndex.cpp b/clang/tools/libclang/CIndex.cpp
index def4524..6a6d7fd 100644
--- a/clang/tools/libclang/CIndex.cpp
+++ b/clang/tools/libclang/CIndex.cpp
@@ -2188,6 +2188,8 @@ public:
   void VisitOMPExecutableDirective(const OMPExecutableDirective *D);
   void VisitOMPLoopBasedDirective(const OMPLoopBasedDirective *D);
   void VisitOMPLoopDirective(const OMPLoopDirective *D);
+  void VisitOMPOpaqueBlockDirective(const OMPOpaqueBlockDirective *D);
+  void VisitOMPOpaqueLoopDirective(const OMPOpaqueLoopDirective *D);
   void VisitOMPParallelDirective(const OMPParallelDirective *D);
   void VisitOMPSimdDirective(const OMPSimdDirective *D);
   void
@@ -3266,6 +3268,16 @@ void EnqueueVisitor::VisitOMPLoopDirective(const OMPLoopDirective *D) {
   VisitOMPLoopBasedDirective(D);
 }
 
+void EnqueueVisitor::VisitOMPOpaqueBlockDirective(
+    const OMPOpaqueBlockDirective *D) {
+  VisitOMPExecutableDirective(D);
+}
+
+void EnqueueVisitor::VisitOMPOpaqueLoopDirective(
+    const OMPOpaqueLoopDirective *D) {
+  VisitOMPLoopDirective(D);
+}
+
 void EnqueueVisitor::VisitOMPParallelDirective(const OMPParallelDirective *D) {
   VisitOMPExecutableDirective(D);
 }
@@ -6318,6 +6330,10 @@ CXString clang_getCursorKindSpelling(enum CXCursorKind Kind) {
     return cxstring::createRef("OMPParallelGenericLoopDirective");
   case CXCursor_OMPTargetParallelGenericLoopDirective:
     return cxstring::createRef("OMPTargetParallelGenericLoopDirective");
+  case CXCursor_OMPOpaqueBlockDirective:
+    return cxstring::createRef("OMPOpaqueBlockDirective");
+  case CXCursor_OMPOpaqueLoopDirective:
+    return cxstring::createRef("OMPOpaqueLoopDirective");
   case CXCursor_OverloadCandidate:
     return cxstring::createRef("OverloadCandidate");
   case CXCursor_TypeAliasTemplateDecl:
diff --git a/clang/tools/libclang/CXCursor.cpp b/clang/tools/libclang/CXCursor.cpp
index c8cf51d..c0806fd 100644
--- a/clang/tools/libclang/CXCursor.cpp
+++ b/clang/tools/libclang/CXCursor.cpp
@@ -891,6 +891,12 @@ CXCursor cxcursor::MakeCXCursor(const Stmt *S, const Decl *Parent,
   case Stmt::OMPTargetParallelGenericLoopDirectiveClass:
     K = CXCursor_OMPTargetParallelGenericLoopDirective;
     break;
+  case Stmt::OMPOpaqueBlockDirectiveClass:
+    K = CXCursor_OMPOpaqueBlockDirective;
+    break;
+  case Stmt::OMPOpaqueLoopDirectiveClass:
+    K = CXCursor_OMPOpaqueLoopDirective;
+    break;
   case Stmt::BuiltinBitCastExprClass:
     K = CXCursor_BuiltinBitCastExpr;
     break;