[OPENMP52] Initial support for doacross clause.

[jyu2]

Summary

Initial support for doacross clause in ordered directive.
This is intend to replace depend clause in ordered directive.

[ABataev]

clang/lib/Parse/ParseOpenMP.cpp

4424–4448

Can it be unified with depenbd clause parsing? (Maybe in a separate template function)

clang/lib/Sema/SemaOpenMP.cpp

91

DoacrossClauseMapTy? Since it is intended to handle also doacross clauses.

11296–11332

Try to avoid copy-paste. Maybe introduce templated function?

11346–11359

Same, try to avoid copy-paste

24033

Same, if possible try to unify handling with the depend clause, if possible

[jyu2]

Thanks Alex for the review. This is to address his comments.

[jyu2]

Thanks Alexey for the review.

clang/lib/Parse/ParseOpenMP.cpp

4424–4448

I don't really has an idea on how to combine this two with template function. Since depend clause in ordered is deprecated in 52, and will be removed, should we leave as this?

clang/lib/Sema/SemaOpenMP.cpp

91

Changed. Thanks.

11296–11332

Not sure how to do this part.

11346–11359

The code is for both Depend and Doacross with is really try to avoid copy-paste.

24033

ActOnOpenMPDoacrossClauseCommon is added which called in
ActOnOpenMPDependClause
and
ActOnOpenMPDoacrossClause

[ABataev]

clang/lib/Parse/ParseOpenMP.cpp

4424–4448

Even ff it will be removed in 52, it will still stay for OpenMP < 52. Would be good to try to unify it.

clang/lib/Sema/SemaOpenMP.cpp

20693–20699

Better to create clauses in ActOnDoAcross and ActOnDepend, this function better to return required data as a struct/class/bolean, etc.

20700

No need for else

[jyu2]

Thanks Alexey's review. This is address his comments.

[jyu2]

Thanks.

clang/lib/Parse/ParseOpenMP.cpp

4424–4448

OKay thanks. Changed

clang/lib/Sema/SemaOpenMP.cpp

11296–11332

I just merged code into OMPC_depend.

20693–20699

Okay I create static function instead. Thanks.

[ABataev]

Upload full context

[ABataev]

clang/lib/Sema/SemaOpenMP.cpp

20871

Remove this new line

[jyu2]

Sorry, full diff.

[ABataev]

LG

[jyu2]

Checked in with rG085845a2acbe

clang/include/clang/AST/OpenMPClause.h

public:
 
   /// Return the size expression.
   Expr *getSize() { return getStmtAs<Expr>(); }
 
   /// Return the size expression.
   Expr *getSize() const { return getStmtAs<Expr>(); }
 };
 
+/// This represents the 'doacross' clause for the '#pragma omp ordered'
+/// directive.
+///
+/// \code
+/// #pragma omp ordered doacross(sink: i-1, j-1)
+/// \endcode
+/// In this example directive '#pragma omp ordered' with clause 'doacross' with
+/// a dependence-type 'sink' and loop-iteration vector expressions i-1 and j-1.
+class OMPDoacrossClause final
+    : public OMPVarListClause<OMPDoacrossClause>,
+      private llvm::TrailingObjects<OMPDoacrossClause, Expr *> {
+  friend class OMPClauseReader;
+  friend OMPVarListClause;
+  friend TrailingObjects;
+
+  /// Dependence type (sink or source).
+  OpenMPDoacrossClauseModifier DepType = OMPC_DOACROSS_unknown;
+
+  /// Dependence type location.
+  SourceLocation DepLoc;
+
+  /// Colon location.
+  SourceLocation ColonLoc;
+
+  /// Number of loops, associated with the doacross clause.
+  unsigned NumLoops = 0;
+
+  /// Build clause with number of expressions \a N.
+  ///
+  /// \param StartLoc Starting location of the clause.
+  /// \param LParenLoc Location of '('.
+  /// \param EndLoc Ending location of the clause.
+  /// \param N Number of expressions in the clause.
+  /// \param NumLoops Number of loops associated with the clause.
+  OMPDoacrossClause(SourceLocation StartLoc, SourceLocation LParenLoc,
+                    SourceLocation EndLoc, unsigned N, unsigned NumLoops)
+      : OMPVarListClause<OMPDoacrossClause>(llvm::omp::OMPC_doacross, StartLoc,
+                                            LParenLoc, EndLoc, N),
+        NumLoops(NumLoops) {}
+
+  /// Build an empty clause.
+  ///
+  /// \param N Number of expressions in the clause.
+  /// \param NumLoops Number of loops associated with the clause.
+  explicit OMPDoacrossClause(unsigned N, unsigned NumLoops)
+      : OMPVarListClause<OMPDoacrossClause>(llvm::omp::OMPC_doacross,
+                                            SourceLocation(), SourceLocation(),
+                                            SourceLocation(), N),
+        NumLoops(NumLoops) {}
+
+  /// Set dependence type.
+  void setDependenceType(OpenMPDoacrossClauseModifier M) { DepType = M; }
+
+  /// Set dependence type location.
+  void setDependenceLoc(SourceLocation Loc) { DepLoc = Loc; }
+
+  /// Set colon location.
+  void setColonLoc(SourceLocation Loc) { ColonLoc = Loc; }
+
+public:
+  /// Creates clause with a list of expressions \a VL.
+  ///
+  /// \param C AST context.
+  /// \param StartLoc Starting location of the clause.
+  /// \param LParenLoc Location of '('.
+  /// \param EndLoc Ending location of the clause.
+  /// \param DepType The dependence type.
+  /// \param DepLoc Location of the dependence type.
+  /// \param ColonLoc Location of ':'.
+  /// \param VL List of references to the expressions.
+  /// \param NumLoops Number of loops that associated with the clause.
+  static OMPDoacrossClause *
+  Create(const ASTContext &C, SourceLocation StartLoc, SourceLocation LParenLoc,
+         SourceLocation EndLoc, OpenMPDoacrossClauseModifier DepType,
+         SourceLocation DepLoc, SourceLocation ColonLoc, ArrayRef<Expr *> VL,
+         unsigned NumLoops);
+
+  /// Creates an empty clause with \a N expressions.
+  ///
+  /// \param C AST context.
+  /// \param N The number of expressions.
+  /// \param NumLoops Number of loops that is associated with this clause.
+  static OMPDoacrossClause *CreateEmpty(const ASTContext &C, unsigned N,
+                                        unsigned NumLoops);
+
+  /// Get dependence type.
+  OpenMPDoacrossClauseModifier getDependenceType() const { return DepType; }
+
+  /// Get dependence type location.
+  SourceLocation getDependenceLoc() const { return DepLoc; }
+
+  /// Get colon location.
+  SourceLocation getColonLoc() const { return ColonLoc; }
+
+  /// Get number of loops associated with the clause.
+  unsigned getNumLoops() const { return NumLoops; }
+
+  /// Set the loop data.
+  void setLoopData(unsigned NumLoop, Expr *Cnt);
+
+  /// Get the loop data.
+  Expr *getLoopData(unsigned NumLoop);
+  const Expr *getLoopData(unsigned NumLoop) const;
+
+  child_range children() {
+    return child_range(reinterpret_cast<Stmt **>(varlist_begin()),
+                       reinterpret_cast<Stmt **>(varlist_end()));
+  }
+
+  const_child_range children() const {
+    auto Children = const_cast<OMPDoacrossClause *>(this)->children();
+    return const_child_range(Children.begin(), Children.end());
+  }
+
+  child_range used_children() {
+    return child_range(child_iterator(), child_iterator());
+  }
+  const_child_range used_children() const {
+    return const_child_range(const_child_iterator(), const_child_iterator());
+  }
+
+  static bool classof(const OMPClause *T) {
+    return T->getClauseKind() == llvm::omp::OMPC_doacross;
+  }
+};
+
 } // namespace clang
 
 #endif // LLVM_CLANG_AST_OPENMPCLAUSE_H

clang/include/clang/AST/RecursiveASTVisitor.h

 template <typename Derived>
 bool RecursiveASTVisitor<Derived>::VisitOMPXDynCGroupMemClause(
     OMPXDynCGroupMemClause *C) {
   TRY_TO(VisitOMPClauseWithPreInit(C));
   TRY_TO(TraverseStmt(C->getSize()));
   return true;
 }
 
+template <typename Derived>
+bool RecursiveASTVisitor<Derived>::VisitOMPDoacrossClause(
+    OMPDoacrossClause *C) {
+  TRY_TO(VisitOMPClauseList(C));
+  return true;
+}
+
 // FIXME: look at the following tricky-seeming exprs to see if we
 // need to recurse on anything.  These are ones that have methods
 // returning decls or qualtypes or nestednamespecifier -- though I'm
 // not sure if they own them -- or just seemed very complicated, or
 // had lots of sub-types to explore.
 //
 // VisitOverloadExpr and its children: recurse on template args? etc?
 

clang/include/clang/Basic/DiagnosticSemaKinds.td

def err_omp_no_more_if_clause : Error<
   "no more 'if' clause is allowed">;
 def err_omp_unnamed_if_clause : Error<
   "expected%select{| one of}0 %1 directive name modifier%select{|s}0">;
 def note_omp_previous_named_if_clause : Note<
   "previous clause with directive name modifier specified here">;
 def err_omp_ordered_directive_with_param : Error<
   "'ordered' directive %select{without any clauses|with 'threads' clause}0 cannot be closely nested inside ordered region with specified parameter">;
 def err_omp_ordered_directive_without_param : Error<
-  "'ordered' directive with 'depend' clause cannot be closely nested inside ordered region without specified parameter">;
+  "'ordered' directive with '%0' clause cannot be closely nested inside ordered region without specified parameter">;
 def note_omp_ordered_param : Note<
   "'ordered' clause%select{| with specified parameter}0">;
 def err_omp_expected_base_var_name : Error<
   "expected variable name as a base of the array %select{subscript|section}0">;
 def err_omp_map_shared_storage : Error<
   "variable already marked as mapped in current construct">;
 def err_omp_invalid_map_type_for_directive : Error<
   "%select{map type '%1' is not allowed|map type must be specified}0 for '#pragma omp %2'">;
def err_omp_hint_clause_no_name : Error<
   "the name of the construct must be specified in presence of 'hint' clause">;
 def err_omp_critical_with_hint : Error<
   "constructs with the same name must have a 'hint' clause with the same value">;
 def note_omp_critical_hint_here : Note<
   "%select{|previous }0'hint' clause with value '%1'">;
 def note_omp_critical_no_hint : Note<
   "%select{|previous }0directive with no 'hint' clause specified">;
 def err_omp_depend_clause_thread_simd : Error<
-  "'depend' clauses cannot be mixed with '%0' clause">;
+  "'%0' clauses cannot be mixed with '%1' clause">;
 def err_omp_depend_sink_expected_loop_iteration : Error<
   "expected%select{| %1}0 loop iteration variable">;
 def err_omp_depend_sink_unexpected_expr : Error<
   "unexpected expression: number of expressions is larger than the number of associated loops">;
 def err_omp_depend_sink_expected_plus_minus : Error<
   "expected '+' or '-' operation">;
 def err_omp_taskwait_depend_mutexinoutset_not_allowed : Error<
   "'mutexinoutset' modifier not allowed in 'depend' clause on 'taskwait' directive">;
-def err_omp_depend_sink_source_not_allowed : Error<
-  "'depend(%select{source|sink:vec}0)' clause%select{|s}0 cannot be mixed with 'depend(%select{sink:vec|source}0)' clause%select{s|}0">;
+def err_omp_sink_and_source_not_allowed : Error<
+  "'%0(%select{source|sink:vec}1)' clause%select{|s}1 cannot be mixed with '%0(%select{sink:vec|source}1)' clause%select{s|}1">;
 def err_omp_depend_zero_length_array_section_not_allowed : Error<
   "zero-length array section is not allowed in 'depend' clause">;
 def err_omp_depend_sink_source_with_modifier : Error<
   "depend modifier cannot be used with 'sink' or 'source' depend type">;
 def err_omp_depend_modifier_not_iterator : Error<
   "expected iterator specification as depend modifier">;
 def err_omp_map_modifier_not_iterator : Error<
   "expected iterator specification as map modifier">;

clang/include/clang/Basic/OpenMPKinds.h

 };
 
 enum OpenMPNumTasksClauseModifier {
 #define OPENMP_NUMTASKS_MODIFIER(Name) OMPC_NUMTASKS_##Name,
 #include "clang/Basic/OpenMPKinds.def"
   OMPC_NUMTASKS_unknown
 };
 
+/// OpenMP dependence types for 'doacross' clause.
+enum OpenMPDoacrossClauseModifier {
+#define OPENMP_DOACROSS_MODIFIER(Name) OMPC_DOACROSS_##Name,
+#include "clang/Basic/OpenMPKinds.def"
+  OMPC_DOACROSS_unknown
+};
+
 /// Contains 'interop' data for 'append_args' and 'init' clauses.
 class Expr;
 struct OMPInteropInfo final {
   OMPInteropInfo(bool IsTarget = false, bool IsTargetSync = false)
       : IsTarget(IsTarget), IsTargetSync(IsTargetSync) {}
   bool IsTarget;
   bool IsTargetSync;
   llvm::SmallVector<Expr *, 4> PreferTypes;

clang/include/clang/Basic/OpenMPKinds.def

 #define OPENMP_BIND_KIND(Name)
 #endif
 #ifndef OPENMP_GRAINSIZE_MODIFIER
 #define OPENMP_GRAINSIZE_MODIFIER(Name)
 #endif
 #ifndef OPENMP_NUMTASKS_MODIFIER
 #define OPENMP_NUMTASKS_MODIFIER(Name)
 #endif
+#ifndef OPENMP_DOACROSS_MODIFIER
+#define OPENMP_DOACROSS_MODIFIER(Name)
+#endif
 
 // Static attributes for 'schedule' clause.
 OPENMP_SCHEDULE_KIND(static)
 OPENMP_SCHEDULE_KIND(dynamic)
 OPENMP_SCHEDULE_KIND(guided)
 OPENMP_SCHEDULE_KIND(auto)
 OPENMP_SCHEDULE_KIND(runtime)
 

 OPENMP_BIND_KIND(thread)
 
 // Modifiers for the 'grainsize' clause.
 OPENMP_GRAINSIZE_MODIFIER(strict)
 
 // Modifiers for the 'num_tasks' clause.
 OPENMP_NUMTASKS_MODIFIER(strict)
 
+// Modifiers for the 'doacross' clause.
+OPENMP_DOACROSS_MODIFIER(source)
+OPENMP_DOACROSS_MODIFIER(sink)
+
 #undef OPENMP_NUMTASKS_MODIFIER
 #undef OPENMP_GRAINSIZE_MODIFIER
 #undef OPENMP_BIND_KIND
 #undef OPENMP_ADJUST_ARGS_KIND
 #undef OPENMP_REDUCTION_MODIFIER
 #undef OPENMP_DEVICE_MODIFIER
 #undef OPENMP_ORDER_KIND
 #undef OPENMP_ORDER_MODIFIER
 #undef OPENMP_LASTPRIVATE_KIND
 #undef OPENMP_DEVICE_TYPE_KIND
 #undef OPENMP_LINEAR_KIND
 #undef OPENMP_DEPEND_KIND
 #undef OPENMP_SCHEDULE_MODIFIER
 #undef OPENMP_SCHEDULE_KIND
 #undef OPENMP_ATOMIC_DEFAULT_MEM_ORDER_KIND
 #undef OPENMP_AT_KIND
 #undef OPENMP_SEVERITY_KIND
 #undef OPENMP_MAP_KIND
 #undef OPENMP_MAP_MODIFIER_KIND
 #undef OPENMP_MOTION_MODIFIER_KIND
 #undef OPENMP_DIST_SCHEDULE_KIND
 #undef OPENMP_DEFAULTMAP_KIND
 #undef OPENMP_DEFAULTMAP_MODIFIER
+#undef OPENMP_DOACROSS_MODIFIER
 

clang/include/clang/Sema/Sema.h

private:
   };
 
   /// Number of nested '#pragma omp declare target' directives.
   SmallVector<DeclareTargetContextInfo, 4> DeclareTargetNesting;
 
   /// Initialization of data-sharing attributes stack.
   void InitDataSharingAttributesStack();
   void DestroyDataSharingAttributesStack();
-  ExprResult
-  VerifyPositiveIntegerConstantInClause(Expr *Op, OpenMPClauseKind CKind,
-                                        bool StrictlyPositive = true,
-                                        bool SuppressExprDiags = false);
+
   /// Returns OpenMP nesting level for current directive.
   unsigned getOpenMPNestingLevel() const;
 
   /// Adjusts the function scopes index for the target-based regions.
   void adjustOpenMPTargetScopeIndex(unsigned &FunctionScopesIndex,
                                     unsigned Level) const;
 
   /// Returns the number of scopes associated with the construct on the given
public:
   /// Act on \p D, a function definition inside of an `omp [begin/end] assumes`.
   void ActOnFinishedFunctionDefinitionInOpenMPAssumeScope(Decl *D);
 
   /// Can we exit an OpenMP declare variant scope at the moment.
   bool isInOpenMPDeclareVariantScope() const {
     return !OMPDeclareVariantScopes.empty();
   }
 
+  ExprResult
+  VerifyPositiveIntegerConstantInClause(Expr *Op, OpenMPClauseKind CKind,
+                                        bool StrictlyPositive = true,
+                                        bool SuppressExprDiags = false);
+
   /// Given the potential call expression \p Call, determine if there is a
   /// specialization via the OpenMP declare variant mechanism available. If
   /// there is, return the specialized call expression, otherwise return the
   /// original \p Call.
   ExprResult ActOnOpenMPCall(ExprResult Call, Scope *Scope,
                              SourceLocation LParenLoc, MultiExprArg ArgExprs,
                              SourceLocation RParenLoc, Expr *ExecConfig);
 
OMPClause *ActOnOpenMPBindClause(OpenMPBindClauseKind Kind,
                                    SourceLocation LParenLoc,
                                    SourceLocation EndLoc);
 
   /// Called on a well-formed 'ompx_dyn_cgroup_mem' clause.
   OMPClause *ActOnOpenMPXDynCGroupMemClause(Expr *Size, SourceLocation StartLoc,
                                             SourceLocation LParenLoc,
                                             SourceLocation EndLoc);
 
+  /// Called on well-formed 'doacross' clause.
+  OMPClause *
+  ActOnOpenMPDoacrossClause(OpenMPDoacrossClauseModifier DepType,
+                            SourceLocation DepLoc, SourceLocation ColonLoc,
+                            ArrayRef<Expr *> VarList, SourceLocation StartLoc,
+                            SourceLocation LParenLoc, SourceLocation EndLoc);
+
   /// The kind of conversion being performed.
   enum CheckedConversionKind {
     /// An implicit conversion.
     CCK_ImplicitConversion,
     /// A C-style cast.
     CCK_CStyleCast,
     /// A functional-style cast.
     CCK_FunctionalCast,

clang/lib/AST/OpenMPClause.cpp

OMPBindClause::Create(const ASTContext &C, OpenMPBindClauseKind K,
                       SourceLocation KLoc, SourceLocation StartLoc,
                       SourceLocation LParenLoc, SourceLocation EndLoc) {
   return new (C) OMPBindClause(K, KLoc, StartLoc, LParenLoc, EndLoc);
 }
 
 OMPBindClause *OMPBindClause::CreateEmpty(const ASTContext &C) {
   return new (C) OMPBindClause();
 }
+
+OMPDoacrossClause *
+OMPDoacrossClause::Create(const ASTContext &C, SourceLocation StartLoc,
+                          SourceLocation LParenLoc, SourceLocation EndLoc,
+                          OpenMPDoacrossClauseModifier DepType,
+                          SourceLocation DepLoc, SourceLocation ColonLoc,
+                          ArrayRef<Expr *> VL, unsigned NumLoops) {
+  void *Mem = C.Allocate(totalSizeToAlloc<Expr *>(VL.size() + NumLoops),
+                         alignof(OMPDoacrossClause));
+  OMPDoacrossClause *Clause = new (Mem)
+      OMPDoacrossClause(StartLoc, LParenLoc, EndLoc, VL.size(), NumLoops);
+  Clause->setDependenceType(DepType);
+  Clause->setDependenceLoc(DepLoc);
+  Clause->setColonLoc(ColonLoc);
+  Clause->setVarRefs(VL);
+  for (unsigned I = 0; I < NumLoops; ++I)
+    Clause->setLoopData(I, nullptr);
+  return Clause;
+}
+
+OMPDoacrossClause *OMPDoacrossClause::CreateEmpty(const ASTContext &C,
+                                                  unsigned N,
+                                                  unsigned NumLoops) {
+  void *Mem = C.Allocate(totalSizeToAlloc<Expr *>(N + NumLoops),
+                         alignof(OMPDoacrossClause));
+  return new (Mem) OMPDoacrossClause(N, NumLoops);
+}
+
+void OMPDoacrossClause::setLoopData(unsigned NumLoop, Expr *Cnt) {
+  assert(NumLoop < NumLoops && "Loop index must be less number of loops.");
+  auto *It = std::next(getVarRefs().end(), NumLoop);
+  *It = Cnt;
+}
+
+Expr *OMPDoacrossClause::getLoopData(unsigned NumLoop) {
+  assert(NumLoop < NumLoops && "Loop index must be less number of loops.");
+  auto *It = std::next(getVarRefs().end(), NumLoop);
+  return *It;
+}
+
+const Expr *OMPDoacrossClause::getLoopData(unsigned NumLoop) const {
+  assert(NumLoop < NumLoops && "Loop index must be less number of loops.");
+  const auto *It = std::next(getVarRefs().end(), NumLoop);
+  return *It;
+}
+
 //===----------------------------------------------------------------------===//
 //  OpenMP clauses printing methods
 //===----------------------------------------------------------------------===//
 
 void OMPClausePrinter::VisitOMPIfClause(OMPIfClause *Node) {
   OS << "if(";
   if (Node->getNameModifier() != OMPD_unknown)
     OS << getOpenMPDirectiveName(Node->getNameModifier()) << ": ";

 
 void OMPClausePrinter::VisitOMPXDynCGroupMemClause(
     OMPXDynCGroupMemClause *Node) {
   OS << "ompx_dyn_cgroup_mem(";
   Node->getSize()->printPretty(OS, nullptr, Policy, 0);
   OS << ")";
 }
 
+void OMPClausePrinter::VisitOMPDoacrossClause(OMPDoacrossClause *Node) {
+  OS << "doacross(";
+  OpenMPDoacrossClauseModifier DepType = Node->getDependenceType();
+  OS << (DepType == OMPC_DOACROSS_source ? "source:" : "sink:");
+  VisitOMPClauseList(Node, ' ');
+  OS << ")";
+}
+
 void OMPTraitInfo::getAsVariantMatchInfo(ASTContext &ASTCtx,
                                          VariantMatchInfo &VMI) const {
   for (const OMPTraitSet &Set : Sets) {
     for (const OMPTraitSelector &Selector : Set.Selectors) {
 
       // User conditions are special as we evaluate the condition here.
       if (Selector.Kind == TraitSelector::user_condition) {
         assert(Selector.ScoreOrCondition &&

clang/lib/AST/StmtProfile.cpp

 void OMPClauseProfiler::VisitOMPOrderClause(const OMPOrderClause *C) {}
 void OMPClauseProfiler::VisitOMPBindClause(const OMPBindClause *C) {}
 void OMPClauseProfiler::VisitOMPXDynCGroupMemClause(
     const OMPXDynCGroupMemClause *C) {
   VistOMPClauseWithPreInit(C);
   if (Expr *Size = C->getSize())
     Profiler->VisitStmt(Size);
 }
+void OMPClauseProfiler::VisitOMPDoacrossClause(const OMPDoacrossClause *C) {
+  VisitOMPClauseList(C);
+}
 } // namespace
 
 void
 StmtProfiler::VisitOMPExecutableDirective(const OMPExecutableDirective *S) {
   VisitStmt(S);
   OMPClauseProfiler P(this);
   ArrayRef<OMPClause *> Clauses = S->clauses();
   for (ArrayRef<OMPClause *>::iterator I = Clauses.begin(), E = Clauses.end();

clang/lib/Basic/OpenMPKinds.cpp

case OMPC_depend: {
     unsigned Type = llvm::StringSwitch<unsigned>(Str)
 #define OPENMP_DEPEND_KIND(Name) .Case(#Name, OMPC_DEPEND_##Name)
 #include "clang/Basic/OpenMPKinds.def"
                         .Default(OMPC_DEPEND_unknown);
     if (LangOpts.OpenMP < 51 && Type == OMPC_DEPEND_inoutset)
       return OMPC_DEPEND_unknown;
     return Type;
   }
+  case OMPC_doacross:
+    return llvm::StringSwitch<OpenMPDoacrossClauseModifier>(Str)
+#define OPENMP_DOACROSS_MODIFIER(Name) .Case(#Name, OMPC_DOACROSS_##Name)
+#include "clang/Basic/OpenMPKinds.def"
+        .Default(OMPC_DOACROSS_unknown);
   case OMPC_linear:
     return llvm::StringSwitch<OpenMPLinearClauseKind>(Str)
 #define OPENMP_LINEAR_KIND(Name) .Case(#Name, OMPC_LINEAR_##Name)
 #include "clang/Basic/OpenMPKinds.def"
         .Default(OMPC_LINEAR_unknown);
   case OMPC_map: {
     unsigned Type = llvm::StringSwitch<unsigned>(Str)
 #define OPENMP_MAP_KIND(Name)                                                  \
case OMPC_depend:
     case OMPC_DEPEND_unknown:
       return "unknown";
 #define OPENMP_DEPEND_KIND(Name)                                             \
   case OMPC_DEPEND_##Name:                                                   \
     return #Name;
 #include "clang/Basic/OpenMPKinds.def"
     }
     llvm_unreachable("Invalid OpenMP 'depend' clause type");
+  case OMPC_doacross:
+    switch (Type) {
+    case OMPC_DOACROSS_unknown:
+      return "unknown";
+#define OPENMP_DOACROSS_MODIFIER(Name)                                         \
+  case OMPC_DOACROSS_##Name:                                                   \
+    return #Name;
+#include "clang/Basic/OpenMPKinds.def"
+    }
+    llvm_unreachable("Invalid OpenMP 'doacross' clause type");
   case OMPC_linear:
     switch (Type) {
     case OMPC_LINEAR_unknown:
       return "unknown";
 #define OPENMP_LINEAR_KIND(Name)                                             \
   case OMPC_LINEAR_##Name:                                                   \
     return #Name;
 #include "clang/Basic/OpenMPKinds.def"

clang/lib/Parse/ParseOpenMP.cpp

while (Tok.isNot(tok::annot_pragma_openmp_end)) {
         ConsumeToken();
       Actions.EndOpenMPClause();
     }
     // End location of the directive.
     EndLoc = Tok.getLocation();
     // Consume final annot_pragma_openmp_end.
     ConsumeAnnotationToken();
 
-    // OpenMP [2.13.8, ordered Construct, Syntax]
-    // If the depend clause is specified, the ordered construct is a stand-alone
-    // directive.
-    if (DKind == OMPD_ordered && FirstClauses[unsigned(OMPC_depend)].getInt()) {
+    if (DKind == OMPD_ordered) {
+      // If the depend or doacross clause is specified, the ordered construct
+      // is a stand-alone directive.
+      for (auto CK : {OMPC_depend, OMPC_doacross}) {
+        if (FirstClauses[unsigned(CK)].getInt()) {
           if ((StmtCtx & ParsedStmtContext::AllowStandaloneOpenMPDirectives) ==
               ParsedStmtContext()) {
             Diag(Loc, diag::err_omp_immediate_directive)
                 << getOpenMPDirectiveName(DKind) << 1
-            << getOpenMPClauseName(OMPC_depend);
+                << getOpenMPClauseName(CK);
           }
           HasAssociatedStatement = false;
         }
+      }
+    }
 
     if (DKind == OMPD_tile && !FirstClauses[unsigned(OMPC_sizes)].getInt()) {
       Diag(Loc, diag::err_omp_required_clause)
           << getOpenMPDirectiveName(OMPD_tile) << "sizes";
     }
 
     StmtResult AssociatedStmt;
     if (HasAssociatedStatement) {
OMPClause *Parser::ParseOpenMPClause(OpenMPDirectiveKind DKind,
   case OMPC_use_device_addr:
   case OMPC_is_device_ptr:
   case OMPC_has_device_addr:
   case OMPC_allocate:
   case OMPC_nontemporal:
   case OMPC_inclusive:
   case OMPC_exclusive:
   case OMPC_affinity:
+  case OMPC_doacross:
     Clause = ParseOpenMPVarListClause(DKind, CKind, WrongDirective);
     break;
   case OMPC_sizes:
     if (!FirstClause) {
       Diag(Tok, diag::err_omp_more_one_clause)
           << getOpenMPDirectiveName(DKind) << getOpenMPClauseName(CKind) << 0;
       ErrorFound = true;
     }
if (Kind == OMPC_reduction || Kind == OMPC_task_reduction ||
     }
     if (Tok.is(tok::colon))
       Data.ColonLoc = ConsumeToken();
     else
       Diag(Tok, diag::warn_pragma_expected_colon) << "reduction identifier";
     if (!InvalidReductionId)
       Data.ReductionOrMapperId =
           Actions.GetNameFromUnqualifiedId(UnqualifiedReductionId);
-  } else if (Kind == OMPC_depend) {
+  } else if (Kind == OMPC_depend || Kind == OMPC_doacross) {
     if (getLangOpts().OpenMP >= 50) {
       if (Tok.is(tok::identifier) && PP.getSpelling(Tok) == "iterator") {
         // Handle optional dependence modifier.
         // iterator(iterators-definition)
         // where iterators-definition is iterator-specifier [,
         // iterators-definition ]
         // where iterator-specifier is [ iterator-type ] identifier =
         // range-specification
         HasIterator = true;
         EnterScope(Scope::OpenMPDirectiveScope | Scope::DeclScope);
         ExprResult IteratorRes = ParseOpenMPIteratorsExpr();
         Data.DepModOrTailExpr = IteratorRes.get();
         // Parse ','
         ExpectAndConsume(tok::comma);
       }
     }
     // Handle dependency type for depend clause.
     ColonProtectionRAIIObject ColonRAII(*this);
     Data.ExtraModifier = getOpenMPSimpleClauseType(
         Kind, Tok.is(tok::identifier) ? PP.getSpelling(Tok) : "",
         getLangOpts());
     Data.ExtraModifierLoc = Tok.getLocation();
-    if (Data.ExtraModifier == OMPC_DEPEND_unknown) {
+    if ((Kind == OMPC_depend && Data.ExtraModifier == OMPC_DEPEND_unknown) ||
+        (Kind == OMPC_doacross &&
+         Data.ExtraModifier == OMPC_DOACROSS_unknown)) {
       SkipUntil(tok::colon, tok::r_paren, tok::annot_pragma_openmp_end,
                 StopBeforeMatch);
     } else {
       ConsumeToken();
       // Special processing for depend(source) clause.
-      if (DKind == OMPD_ordered && Data.ExtraModifier == OMPC_DEPEND_source) {
+      if (DKind == OMPD_ordered && Kind == OMPC_depend &&
+          Data.ExtraModifier == OMPC_DEPEND_source) {
         // Parse ')'.
         T.consumeClose();
         return false;
       }
     }
     if (Tok.is(tok::colon)) {
       Data.ColonLoc = ConsumeToken();
     } else {
       Diag(Tok, DKind == OMPD_ordered ? diag::warn_pragma_expected_colon_r_paren
                                       : diag::warn_pragma_expected_colon)
-          << "dependency type";
+          << (Kind == OMPC_depend ? "dependency type" : "dependence-type");
+    }
+    // Special processing for doacross(source) clause.
+    if (Kind == OMPC_doacross && Data.ExtraModifier == OMPC_DOACROSS_source) {
+      // Parse ')'.
+      T.consumeClose();
+      return false;
     }
   } else if (Kind == OMPC_linear) {
     // Try to parse modifier if any.
     Data.ExtraModifier = OMPC_LINEAR_val;
     if (Tok.is(tok::identifier) && PP.LookAhead(0).is(tok::l_paren)) {
       Data.ExtraModifier =
           getOpenMPSimpleClauseType(Kind, PP.getSpelling(Tok), getLangOpts());
       Data.ExtraModifierLoc = ConsumeToken();
       LinearT.consumeOpen();
       NeedRParenForLinear = true;
     }
   } else if (Kind == OMPC_lastprivate) {
     // Try to parse modifier if any.
     Data.ExtraModifier = OMPC_LASTPRIVATE_unknown;
     // Conditional modifier allowed only in OpenMP 5.0 and not supported in
     // distribute and taskloop based directives.
     if ((getLangOpts().OpenMP >= 50 && !isOpenMPDistributeDirective(DKind) &&
          !isOpenMPTaskLoopDirective(DKind)) &&
         Tok.is(tok::identifier) && PP.LookAhead(0).is(tok::colon)) {
       Data.ExtraModifier =
           getOpenMPSimpleClauseType(Kind, PP.getSpelling(Tok), getLangOpts());
       Data.ExtraModifierLoc = Tok.getLocation();
       ConsumeToken();
       assert(Tok.is(tok::colon) && "Expected colon.");
       Data.ColonLoc = ConsumeToken();
     }
   } else if (Kind == OMPC_map) {
     // Handle optional iterator map modifier.
     if (Tok.is(tok::identifier) && PP.getSpelling(Tok) == "iterator") {
       HasIterator = true;
       EnterScope(Scope::OpenMPDirectiveScope | Scope::DeclScope);
       Data.MapTypeModifiers.push_back(OMPC_MAP_MODIFIER_iterator);
       Data.MapTypeModifiersLoc.push_back(Tok.getLocation());
       ExprResult IteratorRes = ParseOpenMPIteratorsExpr();
if (Data.ExtraModifier == OMPC_ADJUST_ARGS_unknown) {
         Data.ColonLoc = Tok.getLocation();
       ExpectAndConsume(tok::colon, diag::warn_pragma_expected_colon,
                        "adjust-op");
     }
   }
 
   bool IsComma =
       (Kind != OMPC_reduction && Kind != OMPC_task_reduction &&
-       Kind != OMPC_in_reduction && Kind != OMPC_depend && Kind != OMPC_map) ||
+       Kind != OMPC_in_reduction && Kind != OMPC_depend &&
+       Kind != OMPC_doacross && Kind != OMPC_map) ||
       (Kind == OMPC_reduction && !InvalidReductionId) ||
       (Kind == OMPC_map && Data.ExtraModifier != OMPC_MAP_unknown) ||
       (Kind == OMPC_depend && Data.ExtraModifier != OMPC_DEPEND_unknown) ||
+      (Kind == OMPC_doacross && Data.ExtraModifier != OMPC_DOACROSS_unknown) ||
       (Kind == OMPC_adjust_args &&
        Data.ExtraModifier != OMPC_ADJUST_ARGS_unknown);
   const bool MayHaveTail = (Kind == OMPC_linear || Kind == OMPC_aligned);
   while (IsComma || (Tok.isNot(tok::r_paren) && Tok.isNot(tok::colon) &&
                      Tok.isNot(tok::annot_pragma_openmp_end))) {
     ParseScope OMPListScope(this, Scope::OpenMPDirectiveScope);
     ColonProtectionRAIIObject ColonRAII(*this, MayHaveTail);
     if (!ParseOpenMPReservedLocator(Kind, Data, getLangOpts())) {
bool Parser::ParseOpenMPVarList(OpenMPDirectiveKind DKind,
 
   // Parse ')'.
   Data.RLoc = Tok.getLocation();
   if (!T.consumeClose())
     Data.RLoc = T.getCloseLocation();
   // Exit from scope when the iterator is used in depend clause.
   if (HasIterator)
     ExitScope();
-  return (Kind != OMPC_depend && Kind != OMPC_map && Vars.empty()) ||
+  return (Kind != OMPC_depend && Kind != OMPC_doacross && Kind != OMPC_map &&
+          Vars.empty()) ||
          (MustHaveTail && !Data.DepModOrTailExpr) || InvalidReductionId ||
          IsInvalidMapperModifier || InvalidIterator;
 }
 
 /// Parsing of OpenMP clause 'private', 'firstprivate', 'lastprivate',
 /// 'shared', 'copyin', 'copyprivate', 'flush', 'reduction', 'task_reduction',
 /// 'in_reduction', 'nontemporal', 'exclusive' or 'inclusive'.
 ///

clang/lib/Sema/SemaOpenMP.cpp

DSAVarData(OpenMPDirectiveKind DKind, OpenMPClauseKind CKind,
                SourceLocation ImplicitDSALoc, unsigned Modifier,
                bool AppliedToPointee)
         : DKind(DKind), CKind(CKind), Modifier(Modifier), RefExpr(RefExpr),
           PrivateCopy(PrivateCopy), ImplicitDSALoc(ImplicitDSALoc),
           AppliedToPointee(AppliedToPointee) {}
   };
   using OperatorOffsetTy =
       llvm::SmallVector<std::pair<Expr *, OverloadedOperatorKind>, 4>;
-  using DoacrossDependMapTy =
-      llvm::DenseMap<OMPDependClause *, OperatorOffsetTy>;
+  using DoacrossClauseMapTy = llvm::DenseMap<OMPClause *, OperatorOffsetTy>;
   /// Kind of the declaration used in the uses_allocators clauses.
   enum class UsesAllocatorsDeclKind {
     /// Predefined allocator
     PredefinedAllocator,
     /// User-defined allocator
     UserDefinedAllocator,
     /// The declaration that represent allocator trait
     AllocatorTrait,
struct SharingMapTy {
     OpenMPDirectiveKind Directive = OMPD_unknown;
     DeclarationNameInfo DirectiveName;
     Scope *CurScope = nullptr;
     DeclContext *Context = nullptr;
     SourceLocation ConstructLoc;
     /// Set of 'depend' clauses with 'sink|source' dependence kind. Required to
     /// get the data (loop counters etc.) about enclosing loop-based construct.
     /// This data is required during codegen.
-    DoacrossDependMapTy DoacrossDepends;
+    DoacrossClauseMapTy DoacrossDepends;
     /// First argument (Expr *) contains optional argument of the
     /// 'ordered' clause, the second one is true if the regions has 'ordered'
     /// clause, false otherwise.
     std::optional<std::pair<const Expr *, OMPOrderedClause *>> OrderedRegion;
     bool RegionHasOrderConcurrent = false;
     unsigned AssociatedLoops = 1;
     bool HasMutipleLoops = false;
     const Decl *PossiblyLoopCounter = nullptr;
void addMappableExpressionComponents(
     MEC.Components.back().append(Components.begin(), Components.end());
     MEC.Kind = WhereFoundClauseKind;
   }
 
   unsigned getNestingLevel() const {
     assert(!isStackEmpty());
     return getStackSize() - 1;
   }
-  void addDoacrossDependClause(OMPDependClause *C,
-                               const OperatorOffsetTy &OpsOffs) {
+  void addDoacrossDependClause(OMPClause *C, const OperatorOffsetTy &OpsOffs) {
     SharingMapTy *Parent = getSecondOnStackOrNull();
     assert(Parent && isOpenMPWorksharingDirective(Parent->Directive));
     Parent->DoacrossDepends.try_emplace(C, OpsOffs);
   }
-  llvm::iterator_range<DoacrossDependMapTy::const_iterator>
+  llvm::iterator_range<DoacrossClauseMapTy::const_iterator>
   getDoacrossDependClauses() const {
     const SharingMapTy &StackElem = getTopOfStack();
     if (isOpenMPWorksharingDirective(StackElem.Directive)) {
-      const DoacrossDependMapTy &Ref = StackElem.DoacrossDepends;
+      const DoacrossClauseMapTy &Ref = StackElem.DoacrossDepends;
       return llvm::make_range(Ref.begin(), Ref.end());
     }
     return llvm::make_range(StackElem.DoacrossDepends.end(),
                             StackElem.DoacrossDepends.end());
   }
 
   // Store types of classes which have been explicitly mapped
   void addMappedClassesQualTypes(QualType QT) {
if (DSA.getOrderedRegionParam().second->getNumForLoops()) {
             CurrentNestedLoopCount,
             ResultIterSpaces[CurrentNestedLoopCount].NumIterations);
         DSA.getOrderedRegionParam().second->setLoopCounter(
             CurrentNestedLoopCount,
             ResultIterSpaces[CurrentNestedLoopCount].CounterVar);
       }
     }
     for (auto &Pair : DSA.getDoacrossDependClauses()) {
-      if (CurrentNestedLoopCount >= Pair.first->getNumLoops()) {
+      auto *DependC = dyn_cast<OMPDependClause>(Pair.first);
+      auto *DoacrossC = dyn_cast<OMPDoacrossClause>(Pair.first);
+      unsigned NumLoops =
+          DependC ? DependC->getNumLoops() : DoacrossC->getNumLoops();
+      if (CurrentNestedLoopCount >= NumLoops) {
         // Erroneous case - clause has some problems.
         continue;
       }
-      if (Pair.first->getDependencyKind() == OMPC_DEPEND_sink &&
+      if (DependC && DependC->getDependencyKind() == OMPC_DEPEND_sink &&
           Pair.second.size() <= CurrentNestedLoopCount) {
         // Erroneous case - clause has some problems.
-        Pair.first->setLoopData(CurrentNestedLoopCount, nullptr);
+        DependC->setLoopData(CurrentNestedLoopCount, nullptr);
+        continue;
+      }
+      if (DoacrossC && DoacrossC->getDependenceType() == OMPC_DOACROSS_sink &&
+          Pair.second.size() <= CurrentNestedLoopCount) {
+        // Erroneous case - clause has some problems.
+        DoacrossC->setLoopData(CurrentNestedLoopCount, nullptr);
         continue;
       }
       Expr *CntValue;
-      if (Pair.first->getDependencyKind() == OMPC_DEPEND_source)
+      SourceLocation DepLoc =
+          DependC ? DependC->getDependencyLoc() : DoacrossC->getDependenceLoc();
+      if ((DependC && DependC->getDependencyKind() == OMPC_DEPEND_source) ||
+          (DoacrossC && DoacrossC->getDependenceType() == OMPC_DOACROSS_source))
         CntValue = ISC.buildOrderedLoopData(
             DSA.getCurScope(),
             ResultIterSpaces[CurrentNestedLoopCount].CounterVar, Captures,
-            Pair.first->getDependencyLoc());
+            DepLoc);
       else
         CntValue = ISC.buildOrderedLoopData(
             DSA.getCurScope(),
             ResultIterSpaces[CurrentNestedLoopCount].CounterVar, Captures,
-            Pair.first->getDependencyLoc(),
-            Pair.second[CurrentNestedLoopCount].first,
+            DepLoc, Pair.second[CurrentNestedLoopCount].first,
             Pair.second[CurrentNestedLoopCount].second);
-      Pair.first->setLoopData(CurrentNestedLoopCount, CntValue);
+      if (DependC)
+        DependC->setLoopData(CurrentNestedLoopCount, CntValue);
+      else
+        DoacrossC->setLoopData(CurrentNestedLoopCount, CntValue);
     }
   }
 
   return HasErrors;
 }
 
 /// Build 'VarRef = Start.
 static ExprResult

 
 StmtResult Sema::ActOnOpenMPOrderedDirective(ArrayRef<OMPClause *> Clauses,
                                              Stmt *AStmt,
                                              SourceLocation StartLoc,
                                              SourceLocation EndLoc) {
   const OMPClause *DependFound = nullptr;
   const OMPClause *DependSourceClause = nullptr;
   const OMPClause *DependSinkClause = nullptr;
+  const OMPClause *DoacrossFound = nullptr;
+  const OMPClause *DoacrossSourceClause = nullptr;
+  const OMPClause *DoacrossSinkClause = nullptr;
   bool ErrorFound = false;
   const OMPThreadsClause *TC = nullptr;
   const OMPSIMDClause *SC = nullptr;
   for (const OMPClause *C : Clauses) {
-    if (auto *DC = dyn_cast<OMPDependClause>(C)) {
-      DependFound = C;
-      if (DC->getDependencyKind() == OMPC_DEPEND_source) {
-        if (DependSourceClause) {
+    auto DOC = dyn_cast<OMPDoacrossClause>(C);
+    auto DC = dyn_cast<OMPDependClause>(C);
+    if (DC || DOC) {
+      DependFound = DC ? C : nullptr;
+      DoacrossFound = DOC ? C : nullptr;
+      if ((DC && DC->getDependencyKind() == OMPC_DEPEND_source) ||
+          (DOC && DOC->getDependenceType() == OMPC_DOACROSS_source)) {
+        if ((DC && DependSourceClause) || (DOC && DoacrossSourceClause)) {
           Diag(C->getBeginLoc(), diag::err_omp_more_one_clause)
               << getOpenMPDirectiveName(OMPD_ordered)
-              << getOpenMPClauseName(OMPC_depend) << 2;
+              << getOpenMPClauseName(DC ? OMPC_depend : OMPC_doacross) << 2;
           ErrorFound = true;
         } else {
+          if (DC)
             DependSourceClause = C;
+          else
+            DoacrossSourceClause = C;
         }
-        if (DependSinkClause) {
-          Diag(C->getBeginLoc(), diag::err_omp_depend_sink_source_not_allowed)
-              << 0;
+        if ((DC && DependSinkClause) || (DOC && DoacrossSinkClause)) {
+          Diag(C->getBeginLoc(), diag::err_omp_sink_and_source_not_allowed)
+              << (DC ? "depend" : "doacross") << 0;
           ErrorFound = true;
         }
-      } else if (DC->getDependencyKind() == OMPC_DEPEND_sink) {
-        if (DependSourceClause) {
-          Diag(C->getBeginLoc(), diag::err_omp_depend_sink_source_not_allowed)
-              << 1;
+      } else if ((DC && DC->getDependencyKind() == OMPC_DEPEND_sink) ||
+                 (DOC && DOC->getDependenceType() == OMPC_DOACROSS_sink)) {
+        if (DependSourceClause || DoacrossSourceClause) {
+          Diag(C->getBeginLoc(), diag::err_omp_sink_and_source_not_allowed)
+              << (DC ? "depend" : "doacross") << 1;
           ErrorFound = true;
         }
+        if (DC)
           DependSinkClause = C;
+        else
+          DoacrossSinkClause = C;
       }
     } else if (C->getClauseKind() == OMPC_threads) {
       TC = cast<OMPThreadsClause>(C);
     } else if (C->getClauseKind() == OMPC_simd) {
       SC = cast<OMPSIMDClause>(C);
     }
   }
   if (!ErrorFound && !SC &&
       isOpenMPSimdDirective(DSAStack->getParentDirective())) {
     // OpenMP [2.8.1,simd Construct, Restrictions]
     // An ordered construct with the simd clause is the only OpenMP construct
     // that can appear in the simd region.
     Diag(StartLoc, diag::err_omp_prohibited_region_simd)
         << (LangOpts.OpenMP >= 50 ? 1 : 0);
     ErrorFound = true;
-  } else if (DependFound && (TC || SC)) {
-    Diag(DependFound->getBeginLoc(), diag::err_omp_depend_clause_thread_simd)
+  } else if ((DependFound || DoacrossFound) && (TC || SC)) {
+    SourceLocation Loc =
+        DependFound ? DependFound->getBeginLoc() : DoacrossFound->getBeginLoc();
+    Diag(Loc, diag::err_omp_depend_clause_thread_simd)
+        << getOpenMPClauseName(DependFound ? OMPC_depend : OMPC_doacross)
         << getOpenMPClauseName(TC ? TC->getClauseKind() : SC->getClauseKind());
     ErrorFound = true;
-  } else if (DependFound && !DSAStack->getParentOrderedRegionParam().first) {
-    Diag(DependFound->getBeginLoc(),
-         diag::err_omp_ordered_directive_without_param);
+  } else if ((DependFound || DoacrossFound) &&
+             !DSAStack->getParentOrderedRegionParam().first) {
+    SourceLocation Loc =
+        DependFound ? DependFound->getBeginLoc() : DoacrossFound->getBeginLoc();
+    Diag(Loc, diag::err_omp_ordered_directive_without_param)
+        << getOpenMPClauseName(DependFound ? OMPC_depend : OMPC_doacross);
     ErrorFound = true;
   } else if (TC || Clauses.empty()) {
     if (const Expr *Param = DSAStack->getParentOrderedRegionParam().first) {
       SourceLocation ErrLoc = TC ? TC->getBeginLoc() : StartLoc;
       Diag(ErrLoc, diag::err_omp_ordered_directive_with_param)
           << (TC != nullptr);
       Diag(Param->getBeginLoc(), diag::note_omp_ordered_param) << 1;
       ErrorFound = true;
     }
   }
-  if ((!AStmt && !DependFound) || ErrorFound)
+  if ((!AStmt && !DependFound && !DoacrossFound) || ErrorFound)
     return StmtError();
 
   // OpenMP 5.0, 2.17.9, ordered Construct, Restrictions.
   // During execution of an iteration of a worksharing-loop or a loop nest
   // within a worksharing-loop, simd, or worksharing-loop SIMD region, a thread
   // must not execute more than one ordered region corresponding to an ordered
   // construct without a depend clause.
-  if (!DependFound) {
+  if (!DependFound && !DoacrossFound) {
     if (DSAStack->doesParentHasOrderedDirective()) {
       Diag(StartLoc, diag::err_omp_several_directives_in_region) << "ordered";
       Diag(DSAStack->getParentOrderedDirectiveLoc(),
            diag::note_omp_previous_directive)
           << "ordered";
       return StmtError();
     }
     DSAStack->setParentHasOrderedDirective(StartLoc);
case OMPC_inclusive:
     break;
   case OMPC_exclusive:
     Res = ActOnOpenMPExclusiveClause(VarList, StartLoc, LParenLoc, EndLoc);
     break;
   case OMPC_affinity:
     Res = ActOnOpenMPAffinityClause(StartLoc, LParenLoc, ColonLoc, EndLoc,
                                     Data.DepModOrTailExpr, VarList);
     break;
+  case OMPC_doacross:
+    Res = ActOnOpenMPDoacrossClause(
+        static_cast<OpenMPDoacrossClauseModifier>(ExtraModifier),
+        ExtraModifierLoc, ColonLoc, VarList, StartLoc, LParenLoc, EndLoc);
+    break;
   case OMPC_if:
   case OMPC_depobj:
   case OMPC_final:
   case OMPC_num_threads:
   case OMPC_safelen:
   case OMPC_simdlen:
   case OMPC_sizes:
   case OMPC_allocator:
OMPClause *Sema::ActOnOpenMPDepobjClause(Expr *Depobj, SourceLocation StartLoc,
   if (!Depobj->isLValue()) {
     Diag(Depobj->getExprLoc(), diag::err_omp_expected_omp_depend_t_lvalue)
         << 1 << Depobj->getSourceRange();
   }
 
   return OMPDepobjClause::Create(Context, StartLoc, LParenLoc, EndLoc, Depobj);
 }
 
-OMPClause *
-Sema::ActOnOpenMPDependClause(const OMPDependClause::DependDataTy &Data,
-                              Expr *DepModifier, ArrayRef<Expr *> VarList,
-                              SourceLocation StartLoc, SourceLocation LParenLoc,
+namespace {
+// Utility struct that gathers the related info for doacross clause.
+struct DoacrossDataInfoTy {
+  // The list of expressions.
+  SmallVector<Expr *, 8> Vars;
+  // The OperatorOffset for doacross loop.
+  DSAStackTy::OperatorOffsetTy OpsOffs;
+  // The depended loop count.
+  llvm::APSInt TotalDepCount;
+};
+} // namespace
+static DoacrossDataInfoTy
+ProcessOpenMPDoacrossClauseCommon(Sema &SemaRef, bool IsSource,
+                                  ArrayRef<Expr *> VarList, DSAStackTy *Stack,
                                   SourceLocation EndLoc) {
-  OpenMPDependClauseKind DepKind = Data.DepKind;
-  SourceLocation DepLoc = Data.DepLoc;
-  if (DSAStack->getCurrentDirective() == OMPD_ordered &&
-      DepKind != OMPC_DEPEND_source && DepKind != OMPC_DEPEND_sink) {
-    Diag(DepLoc, diag::err_omp_unexpected_clause_value)
-        << "'source' or 'sink'" << getOpenMPClauseName(OMPC_depend);
-    return nullptr;
-  }
-  if (DSAStack->getCurrentDirective() == OMPD_taskwait &&
-      DepKind == OMPC_DEPEND_mutexinoutset) {
-    Diag(DepLoc, diag::err_omp_taskwait_depend_mutexinoutset_not_allowed);
-    return nullptr;
-  }
-  if ((DSAStack->getCurrentDirective() != OMPD_ordered ||
-       DSAStack->getCurrentDirective() == OMPD_depobj) &&
-      (DepKind == OMPC_DEPEND_unknown || DepKind == OMPC_DEPEND_source ||
-       DepKind == OMPC_DEPEND_sink ||
-       ((LangOpts.OpenMP < 50 ||
-         DSAStack->getCurrentDirective() == OMPD_depobj) &&
-        DepKind == OMPC_DEPEND_depobj))) {
-    SmallVector<unsigned, 6> Except = {OMPC_DEPEND_source, OMPC_DEPEND_sink,
-                                       OMPC_DEPEND_outallmemory,
-                                       OMPC_DEPEND_inoutallmemory};
-    if (LangOpts.OpenMP < 50 || DSAStack->getCurrentDirective() == OMPD_depobj)
-      Except.push_back(OMPC_DEPEND_depobj);
-    if (LangOpts.OpenMP < 51)
-      Except.push_back(OMPC_DEPEND_inoutset);
-    std::string Expected = (LangOpts.OpenMP >= 50 && !DepModifier)
-                               ? "depend modifier(iterator) or "
-                               : "";
-    Diag(DepLoc, diag::err_omp_unexpected_clause_value)
-        << Expected + getListOfPossibleValues(OMPC_depend, /*First=*/0,
-                                              /*Last=*/OMPC_DEPEND_unknown,
-                                              Except)
-        << getOpenMPClauseName(OMPC_depend);
-    return nullptr;
-  }
-  if (DepModifier &&
-      (DepKind == OMPC_DEPEND_source || DepKind == OMPC_DEPEND_sink)) {
-    Diag(DepModifier->getExprLoc(),
-         diag::err_omp_depend_sink_source_with_modifier);
-    return nullptr;
-  }
-  if (DepModifier &&
-      !DepModifier->getType()->isSpecificBuiltinType(BuiltinType::OMPIterator))
-    Diag(DepModifier->getExprLoc(), diag::err_omp_depend_modifier_not_iterator);
 
   SmallVector<Expr *, 8> Vars;
   DSAStackTy::OperatorOffsetTy OpsOffs;
   llvm::APSInt DepCounter(/*BitWidth=*/32);
   llvm::APSInt TotalDepCount(/*BitWidth=*/32);
-  if (DepKind == OMPC_DEPEND_sink || DepKind == OMPC_DEPEND_source) {
-    if (const Expr *OrderedCountExpr =
-            DSAStack->getParentOrderedRegionParam().first) {
-      TotalDepCount = OrderedCountExpr->EvaluateKnownConstInt(Context);
-      TotalDepCount.setIsUnsigned(/*Val=*/true);
-    }
-  }
+
+          Stack->getParentOrderedRegionParam().first) {
+    TotalDepCount = OrderedCountExpr->EvaluateKnownConstInt(SemaRef.Context);
+
   for (Expr *RefExpr : VarList) {
-    assert(RefExpr && "NULL expr in OpenMP shared clause.");
+    assert(RefExpr && "NULL expr in OpenMP doacross clause.");
     if (isa<DependentScopeDeclRefExpr>(RefExpr)) {
       // It will be analyzed later.
       Vars.push_back(RefExpr);
       continue;
     }
 
     SourceLocation ELoc = RefExpr->getExprLoc();
     Expr *SimpleExpr = RefExpr->IgnoreParenCasts();
-    if (DepKind == OMPC_DEPEND_sink) {
-      if (DSAStack->getParentOrderedRegionParam().first &&
+    if (!IsSource) {
+      if (Stack->getParentOrderedRegionParam().first &&
           DepCounter >= TotalDepCount) {
-        Diag(ELoc, diag::err_omp_depend_sink_unexpected_expr);
+        SemaRef.Diag(ELoc, diag::err_omp_depend_sink_unexpected_expr);
         continue;
       }
       ++DepCounter;
       // OpenMP  [2.13.9, Summary]
       // depend(dependence-type : vec), where dependence-type is:
       // 'sink' and where vec is the iteration vector, which has the form:
       //  x1 [+- d1], x2 [+- d2 ], . . . , xn [+- dn]
       // where n is the value specified by the ordered clause in the loop
       // directive, xi denotes the loop iteration variable of the i-th nested
       // loop associated with the loop directive, and di is a constant
       // non-negative integer.
-      if (CurContext->isDependentContext()) {
+      if (SemaRef.CurContext->isDependentContext()) {
         // It will be analyzed later.
         Vars.push_back(RefExpr);
         continue;
       }
       SimpleExpr = SimpleExpr->IgnoreImplicit();
       OverloadedOperatorKind OOK = OO_None;
       SourceLocation OOLoc;
       Expr *LHS = SimpleExpr;
if (!IsSource) {
                   .getName()
                   .getCXXOverloadedOperator();
         OOLoc = MCE->getCallee()->getExprLoc();
         LHS = MCE->getImplicitObjectArgument()->IgnoreParenImpCasts();
         RHS = MCE->getArg(/*Arg=*/0)->IgnoreParenImpCasts();
       }
       SourceLocation ELoc;
       SourceRange ERange;
-      auto Res = getPrivateItem(*this, LHS, ELoc, ERange);
+      auto Res = getPrivateItem(SemaRef, LHS, ELoc, ERange);
       if (Res.second) {
         // It will be analyzed later.
         Vars.push_back(RefExpr);
       }
       ValueDecl *D = Res.first;
       if (!D)
         continue;
 
       if (OOK != OO_Plus && OOK != OO_Minus && (RHS || OOK != OO_None)) {
-        Diag(OOLoc, diag::err_omp_depend_sink_expected_plus_minus);
+        SemaRef.Diag(OOLoc, diag::err_omp_depend_sink_expected_plus_minus);
         continue;
       }
       if (RHS) {
-        ExprResult RHSRes = VerifyPositiveIntegerConstantInClause(
+        ExprResult RHSRes = SemaRef.VerifyPositiveIntegerConstantInClause(
             RHS, OMPC_depend, /*StrictlyPositive=*/false);
         if (RHSRes.isInvalid())
           continue;
       }
-      if (!CurContext->isDependentContext() &&
-          DSAStack->getParentOrderedRegionParam().first &&
-          DepCounter != DSAStack->isParentLoopControlVariable(D).first) {
+      if (!SemaRef.CurContext->isDependentContext() &&
+          Stack->getParentOrderedRegionParam().first &&
+          DepCounter != Stack->isParentLoopControlVariable(D).first) {
         const ValueDecl *VD =
-            DSAStack->getParentLoopControlVariable(DepCounter.getZExtValue());
+            Stack->getParentLoopControlVariable(DepCounter.getZExtValue());
         if (VD)
-          Diag(ELoc, diag::err_omp_depend_sink_expected_loop_iteration)
+          SemaRef.Diag(ELoc, diag::err_omp_depend_sink_expected_loop_iteration)
               << 1 << VD;
         else
-          Diag(ELoc, diag::err_omp_depend_sink_expected_loop_iteration) << 0;
+          SemaRef.Diag(ELoc, diag::err_omp_depend_sink_expected_loop_iteration)
+              << 0;
         continue;
       }
       OpsOffs.emplace_back(RHS, OOK);
-    } else {
+    }
+    Vars.push_back(RefExpr->IgnoreParenImpCasts());
+  }
+  if (!SemaRef.CurContext->isDependentContext() && !IsSource &&
+      TotalDepCount > VarList.size() &&
+      Stack->getParentOrderedRegionParam().first &&
+      Stack->getParentLoopControlVariable(VarList.size() + 1)) {
+    SemaRef.Diag(EndLoc, diag::err_omp_depend_sink_expected_loop_iteration)
+        << 1 << Stack->getParentLoopControlVariable(VarList.size() + 1);
+  }
+  return {Vars, OpsOffs, TotalDepCount};
+}
+
+OMPClause *
+Sema::ActOnOpenMPDependClause(const OMPDependClause::DependDataTy &Data,
+                              Expr *DepModifier, ArrayRef<Expr *> VarList,
+                              SourceLocation StartLoc, SourceLocation LParenLoc,
+                              SourceLocation EndLoc) {
+  OpenMPDependClauseKind DepKind = Data.DepKind;
+  SourceLocation DepLoc = Data.DepLoc;
+  if (DSAStack->getCurrentDirective() == OMPD_ordered &&
+      DepKind != OMPC_DEPEND_source && DepKind != OMPC_DEPEND_sink) {
+    Diag(DepLoc, diag::err_omp_unexpected_clause_value)
+        << "'source' or 'sink'" << getOpenMPClauseName(OMPC_depend);
+    return nullptr;
+  }
+  if (DSAStack->getCurrentDirective() == OMPD_taskwait &&
+      DepKind == OMPC_DEPEND_mutexinoutset) {
+    Diag(DepLoc, diag::err_omp_taskwait_depend_mutexinoutset_not_allowed);
+    return nullptr;
+  }
+  if ((DSAStack->getCurrentDirective() != OMPD_ordered ||
+       DSAStack->getCurrentDirective() == OMPD_depobj) &&
+      (DepKind == OMPC_DEPEND_unknown || DepKind == OMPC_DEPEND_source ||
+       DepKind == OMPC_DEPEND_sink ||
+       ((LangOpts.OpenMP < 50 ||
+         DSAStack->getCurrentDirective() == OMPD_depobj) &&
+        DepKind == OMPC_DEPEND_depobj))) {
+    SmallVector<unsigned, 6> Except = {OMPC_DEPEND_source, OMPC_DEPEND_sink,
+                                       OMPC_DEPEND_outallmemory,
+                                       OMPC_DEPEND_inoutallmemory};
+    if (LangOpts.OpenMP < 50 || DSAStack->getCurrentDirective() == OMPD_depobj)
+      Except.push_back(OMPC_DEPEND_depobj);
+    if (LangOpts.OpenMP < 51)
+      Except.push_back(OMPC_DEPEND_inoutset);
+    std::string Expected = (LangOpts.OpenMP >= 50 && !DepModifier)
+                               ? "depend modifier(iterator) or "
+                               : "";
+    Diag(DepLoc, diag::err_omp_unexpected_clause_value)
+        << Expected + getListOfPossibleValues(OMPC_depend, /*First=*/0,
+                                              /*Last=*/OMPC_DEPEND_unknown,
+                                              Except)
+        << getOpenMPClauseName(OMPC_depend);
+    return nullptr;
+  }
+  if (DepModifier &&
+      (DepKind == OMPC_DEPEND_source || DepKind == OMPC_DEPEND_sink)) {
+    Diag(DepModifier->getExprLoc(),
+         diag::err_omp_depend_sink_source_with_modifier);
+    return nullptr;
+  }
+  if (DepModifier &&
+      !DepModifier->getType()->isSpecificBuiltinType(BuiltinType::OMPIterator))
+    Diag(DepModifier->getExprLoc(), diag::err_omp_depend_modifier_not_iterator);
+
+  SmallVector<Expr *, 8> Vars;
+  DSAStackTy::OperatorOffsetTy OpsOffs;
+  llvm::APSInt TotalDepCount(/*BitWidth=*/32);
+
+  if (DepKind == OMPC_DEPEND_sink || DepKind == OMPC_DEPEND_source) {
+    DoacrossDataInfoTy VarOffset = ProcessOpenMPDoacrossClauseCommon(
+        *this, DepKind == OMPC_DEPEND_source, VarList, DSAStack, EndLoc);
+    Vars = VarOffset.Vars;
+    OpsOffs = VarOffset.OpsOffs;
+    TotalDepCount = VarOffset.TotalDepCount;
+  } else
+  for (Expr *RefExpr : VarList) {
+    assert(RefExpr && "NULL expr in OpenMP shared clause.");
+    if (isa<DependentScopeDeclRefExpr>(RefExpr)) {
+      // It will be analyzed later.
+      Vars.push_back(RefExpr);
+      continue;
+    }
+
+    SourceLocation ELoc = RefExpr->getExprLoc();
+    Expr *SimpleExpr = RefExpr->IgnoreParenCasts();
+    if (DepKind != OMPC_DEPEND_sink && DepKind != OMPC_DEPEND_source) {
       bool OMPDependTFound = LangOpts.OpenMP >= 50;
       if (OMPDependTFound)
         OMPDependTFound = findOMPDependT(*this, StartLoc, DSAStack,
                                          DepKind == OMPC_DEPEND_depobj);
       if (DepKind == OMPC_DEPEND_depobj) {
         // OpenMP 5.0, 2.17.11 depend Clause, Restrictions, C/C++
         // List items used in depend clauses with the depobj dependence type
         // must be expressions of the omp_depend_t type.
if (DepKind != OMPC_DEPEND_sink && DepKind != OMPC_DEPEND_source) {
               << (LangOpts.OpenMP >= 50 ? 1 : 0) << RefExpr->getSourceRange();
           continue;
         }
       }
     }
     Vars.push_back(RefExpr->IgnoreParenImpCasts());
   }
 
-  if (!CurContext->isDependentContext() && DepKind == OMPC_DEPEND_sink &&
-      TotalDepCount > VarList.size() &&
-      DSAStack->getParentOrderedRegionParam().first &&
-      DSAStack->getParentLoopControlVariable(VarList.size() + 1)) {
-    Diag(EndLoc, diag::err_omp_depend_sink_expected_loop_iteration)
-        << 1 << DSAStack->getParentLoopControlVariable(VarList.size() + 1);
-  }
   if (DepKind != OMPC_DEPEND_source && DepKind != OMPC_DEPEND_sink &&
       DepKind != OMPC_DEPEND_outallmemory &&
       DepKind != OMPC_DEPEND_inoutallmemory && Vars.empty())
     return nullptr;
 
   auto *C = OMPDependClause::Create(
       Context, StartLoc, LParenLoc, EndLoc,
       {DepKind, DepLoc, Data.ColonLoc, Data.OmpAllMemoryLoc}, DepModifier, Vars,
       TotalDepCount.getZExtValue());
   if ((DepKind == OMPC_DEPEND_sink || DepKind == OMPC_DEPEND_source) &&
       DSAStack->isParentOrderedRegion())
     DSAStack->addDoacrossDependClause(C, OpsOffs);
   return C;
 }
 
 OMPClause *Sema::ActOnOpenMPDeviceClause(OpenMPDeviceClauseModifier Modifier,
                                          Expr *Device, SourceLocation StartLoc,
                                          SourceLocation LParenLoc,
if (CaptureRegion != OMPD_unknown && !CurContext->isDependentContext()) {
     llvm::MapVector<const Expr *, DeclRefExpr *> Captures;
     ValExpr = tryBuildCapture(*this, ValExpr, Captures).get();
     HelperValStmt = buildPreInits(Context, Captures);
   }
 
   return new (Context) OMPXDynCGroupMemClause(
       ValExpr, HelperValStmt, CaptureRegion, StartLoc, LParenLoc, EndLoc);
 }
+
+OMPClause *Sema::ActOnOpenMPDoacrossClause(
+    OpenMPDoacrossClauseModifier DepType, SourceLocation DepLoc,
+    SourceLocation ColonLoc, ArrayRef<Expr *> VarList, SourceLocation StartLoc,
+    SourceLocation LParenLoc, SourceLocation EndLoc) {
+
+  if (DSAStack->getCurrentDirective() == OMPD_ordered &&
+      DepType != OMPC_DOACROSS_source && DepType != OMPC_DOACROSS_sink) {
+    Diag(DepLoc, diag::err_omp_unexpected_clause_value)
+        << "'source' or 'sink'" << getOpenMPClauseName(OMPC_doacross);
+    return nullptr;
+  }
+
+  SmallVector<Expr *, 8> Vars;
+  DSAStackTy::OperatorOffsetTy OpsOffs;
+  llvm::APSInt TotalDepCount(/*BitWidth=*/32);
+  DoacrossDataInfoTy VarOffset = ProcessOpenMPDoacrossClauseCommon(
+      *this, DepType == OMPC_DOACROSS_source, VarList, DSAStack, EndLoc);
+  Vars = VarOffset.Vars;
+  OpsOffs = VarOffset.OpsOffs;
+  TotalDepCount = VarOffset.TotalDepCount;
+  auto *C = OMPDoacrossClause::Create(Context, StartLoc, LParenLoc, EndLoc,
+                                      DepType, DepLoc, ColonLoc, Vars,
+                                      TotalDepCount.getZExtValue());
+  if (DSAStack->isParentOrderedRegion())
+    DSAStack->addDoacrossDependClause(C, OpsOffs);
+  return C;
+}

clang/lib/Sema/TreeTransform.h

   /// By default, performs semantic analysis to build the new OpenMP clause.
   /// Subclasses may override this routine to provide different behavior.
   OMPClause *RebuildOMPMessageClause(Expr *MS, SourceLocation StartLoc,
                                      SourceLocation LParenLoc,
                                      SourceLocation EndLoc) {
     return getSema().ActOnOpenMPMessageClause(MS, StartLoc, LParenLoc, EndLoc);
   }
 
+  /// Build a new OpenMP 'doacross' clause.
+  ///
+  /// By default, performs semantic analysis to build the new OpenMP clause.
+  /// Subclasses may override this routine to provide different behavior.
+  OMPClause *
+  RebuildOMPDoacrossClause(OpenMPDoacrossClauseModifier DepType,
+                           SourceLocation DepLoc, SourceLocation ColonLoc,
+                           ArrayRef<Expr *> VarList, SourceLocation StartLoc,
+                           SourceLocation LParenLoc, SourceLocation EndLoc) {
+    return getSema().ActOnOpenMPDoacrossClause(
+        DepType, DepLoc, ColonLoc, VarList, StartLoc, LParenLoc, EndLoc);
+  }
+
   /// Rebuild the operand to an Objective-C \@synchronized statement.
   ///
   /// By default, performs semantic analysis to build the new statement.
   /// Subclasses may override this routine to provide different behavior.
   ExprResult RebuildObjCAtSynchronizedOperand(SourceLocation atLoc,
                                               Expr *object) {
     return getSema().ActOnObjCAtSynchronizedOperand(atLoc, object);
   }
OMPClause *TreeTransform<Derived>::TransformOMPXDynCGroupMemClause(
     OMPXDynCGroupMemClause *C) {
   ExprResult Size = getDerived().TransformExpr(C->getSize());
   if (Size.isInvalid())
     return nullptr;
   return getDerived().RebuildOMPXDynCGroupMemClause(
       Size.get(), C->getBeginLoc(), C->getLParenLoc(), C->getEndLoc());
 }
 
+template <typename Derived>
+OMPClause *
+TreeTransform<Derived>::TransformOMPDoacrossClause(OMPDoacrossClause *C) {
+  llvm::SmallVector<Expr *, 16> Vars;
+  Vars.reserve(C->varlist_size());
+  for (auto *VE : C->varlists()) {
+    ExprResult EVar = getDerived().TransformExpr(cast<Expr>(VE));
+    if (EVar.isInvalid())
+      return nullptr;
+    Vars.push_back(EVar.get());
+  }
+  return getDerived().RebuildOMPDoacrossClause(
+      C->getDependenceType(), C->getDependenceLoc(), C->getColonLoc(), Vars,
+      C->getBeginLoc(), C->getLParenLoc(), C->getEndLoc());
+}
+
 //===----------------------------------------------------------------------===//
 // Expression transformation
 //===----------------------------------------------------------------------===//
 template<typename Derived>
 ExprResult
 TreeTransform<Derived>::TransformConstantExpr(ConstantExpr *E) {
   return TransformExpr(E->getSubExpr());
 }

clang/lib/Serialization/ASTReader.cpp

case llvm::omp::OMPC_bind:
     C = OMPBindClause::CreateEmpty(Context);
     break;
   case llvm::omp::OMPC_align:
     C = new (Context) OMPAlignClause();
     break;
   case llvm::omp::OMPC_ompx_dyn_cgroup_mem:
     C = new (Context) OMPXDynCGroupMemClause();
     break;
+  case llvm::omp::OMPC_doacross: {
+    unsigned NumVars = Record.readInt();
+    unsigned NumLoops = Record.readInt();
+    C = OMPDoacrossClause::CreateEmpty(Context, NumVars, NumLoops);
+    break;
+  }
 #define OMP_CLAUSE_NO_CLASS(Enum, Str)                                         \
   case llvm::omp::Enum:                                                        \
     break;
 #include "llvm/Frontend/OpenMP/OMPKinds.def"
   default:
     break;
   }
   assert(C && "Unknown OMPClause type");

 }
 
 void OMPClauseReader::VisitOMPXDynCGroupMemClause(OMPXDynCGroupMemClause *C) {
   VisitOMPClauseWithPreInit(C);
   C->setSize(Record.readSubExpr());
   C->setLParenLoc(Record.readSourceLocation());
 }
 
+void OMPClauseReader::VisitOMPDoacrossClause(OMPDoacrossClause *C) {
+  C->setLParenLoc(Record.readSourceLocation());
+  C->setDependenceType(
+      static_cast<OpenMPDoacrossClauseModifier>(Record.readInt()));
+  C->setDependenceLoc(Record.readSourceLocation());
+  C->setColonLoc(Record.readSourceLocation());
+  unsigned NumVars = C->varlist_size();
+  SmallVector<Expr *, 16> Vars;
+  Vars.reserve(NumVars);
+  for (unsigned I = 0; I != NumVars; ++I)
+    Vars.push_back(Record.readSubExpr());
+  C->setVarRefs(Vars);
+  for (unsigned I = 0, E = C->getNumLoops(); I < E; ++I)
+    C->setLoopData(I, Record.readSubExpr());
+}
+
 OMPTraitInfo *ASTRecordReader::readOMPTraitInfo() {
   OMPTraitInfo &TI = getContext().getNewOMPTraitInfo();
   TI.Sets.resize(readUInt32());
   for (auto &Set : TI.Sets) {
     Set.Kind = readEnum<llvm::omp::TraitSet>();
     Set.Selectors.resize(readUInt32());
     for (auto &Selector : Set.Selectors) {
       Selector.Kind = readEnum<llvm::omp::TraitSelector>();

clang/lib/Serialization/ASTWriter.cpp

 }
 
 void OMPClauseWriter::VisitOMPXDynCGroupMemClause(OMPXDynCGroupMemClause *C) {
   VisitOMPClauseWithPreInit(C);
   Record.AddStmt(C->getSize());
   Record.AddSourceLocation(C->getLParenLoc());
 }
 
+void OMPClauseWriter::VisitOMPDoacrossClause(OMPDoacrossClause *C) {
+  Record.push_back(C->varlist_size());
+  Record.push_back(C->getNumLoops());
+  Record.AddSourceLocation(C->getLParenLoc());
+  Record.push_back(C->getDependenceType());
+  Record.AddSourceLocation(C->getDependenceLoc());
+  Record.AddSourceLocation(C->getColonLoc());
+  for (auto *VE : C->varlists())
+    Record.AddStmt(VE);
+  for (unsigned I = 0, E = C->getNumLoops(); I < E; ++I)
+    Record.AddStmt(C->getLoopData(I));
+}
+
 void ASTRecordWriter::writeOMPTraitInfo(const OMPTraitInfo *TI) {
   writeUInt32(TI->Sets.size());
   for (const auto &Set : TI->Sets) {
     writeEnum(Set.Kind);
     writeUInt32(Set.Selectors.size());
     for (const auto &Selector : Set.Selectors) {
       writeEnum(Selector.Kind);
       writeBool(Selector.ScoreOrCondition);

clang/test/OpenMP/ordered_ast_print.cpp

-// RUN: %clang_cc1 -verify -fopenmp -ast-print %s | FileCheck %s
+// RUN: %clang_cc1 -verify -fopenmp -ast-print %s | FileCheck -check-prefixes CHECK,OMP51 %s
 // RUN: %clang_cc1 -fopenmp -x c++ -std=c++11 -emit-pch -o %t %s
-// RUN: %clang_cc1 -fopenmp -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print | FileCheck %s
+// RUN: %clang_cc1 -fopenmp -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print | FileCheck -check-prefixes CHECK,OMP51 %s
 
-// RUN: %clang_cc1 -verify -fopenmp-simd -ast-print %s | FileCheck %s
+// RUN: %clang_cc1 -verify -fopenmp-version=52 -fopenmp -ast-print %s | FileCheck -check-prefixes CHECK,OMP52 %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=52 -x c++ -std=c++11 -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=52 -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print | FileCheck -check-prefixes CHECK,OMP52 %s
+
+// RUN: %clang_cc1 -verify -fopenmp-simd -ast-print %s | FileCheck -check-prefixes CHECK,OMP51 %s
 // RUN: %clang_cc1 -fopenmp-simd -x c++ -std=c++11 -emit-pch -o %t %s
-// RUN: %clang_cc1 -fopenmp-simd -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print | FileCheck %s
+// RUN: %clang_cc1 -fopenmp-simd -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print | FileCheck -check-prefixes CHECK,OMP51 %s
 // expected-no-diagnostics
 
 #ifndef HEADER
 #define HEADER
 
 void foo() {}
 
 template <class T, int N>
T tmain (T argc) {
   #pragma omp parallel for simd
   for (int i =0 ; i < argc; ++i)
   #pragma omp ordered simd
   {
     a=2;
   }
   #pragma omp parallel for ordered(1)
   for (int i =0 ; i < argc; ++i) {
+#if _OPENMP >= 202111
+  #pragma omp ordered doacross(source:)
+  #pragma omp ordered doacross(sink:i+N)
+#else
   #pragma omp ordered depend(source)
   #pragma omp ordered depend(sink:i+N)
+#endif
     a = 2;
   }
   return (0);
 }
 
 // CHECK: static T a;
 // CHECK-NEXT: #pragma omp for ordered
 // CHECK-NEXT: for (int i = 0; i < argc; ++i)

 // CHECK-NEXT: #pragma omp parallel for simd
 // CHECK-NEXT: for (int i = 0; i < argc; ++i)
 // CHECK-NEXT: #pragma omp ordered simd
 // CHECK-NEXT: {
 // CHECK-NEXT: a = 2;
 // CHECK-NEXT: }
 // CHECK-NEXT: #pragma omp parallel for ordered(1)
 // CHECK-NEXT: for (int i = 0; i < argc; ++i) {
-// CHECK-NEXT: #pragma omp ordered depend(source)
-// CHECK-NEXT: #pragma omp ordered depend(sink : i + N)
+#if _OPENMP >= 202111
+// OMP52: #pragma omp ordered doacross(source:)
+// OMP52-NEXT: #pragma omp ordered doacross(sink: i + N)
+#else
+// OMP51: #pragma omp ordered depend(source)
+// OMP51-NEXT: #pragma omp ordered depend(sink : i + N)
+#endif
 // CHECK-NEXT: a = 2;
 // CHECK-NEXT: }
 // CHECK: static int a;
 // CHECK-NEXT: #pragma omp for ordered
 // CHECK-NEXT: for (int i = 0; i < argc; ++i)
 // CHECK-NEXT: #pragma omp ordered
 // CHECK-NEXT: {
 // CHECK-NEXT: a = 2;

 // CHECK-NEXT: #pragma omp parallel for simd
 // CHECK-NEXT: for (int i = 0; i < argc; ++i)
 // CHECK-NEXT: #pragma omp ordered simd
 // CHECK-NEXT: {
 // CHECK-NEXT: a = 2;
 // CHECK-NEXT: }
 // CHECK-NEXT: #pragma omp parallel for ordered(1)
 // CHECK-NEXT: for (int i = 0; i < argc; ++i) {
-// CHECK-NEXT: #pragma omp ordered depend(source)
-// CHECK-NEXT: #pragma omp ordered depend(sink : i + 3)
+#if _OPENMP >= 202111
+// OMP52: #pragma omp ordered doacross(source:)
+// OMP52-NEXT: #pragma omp ordered doacross(sink: i + 3)
+#else
+// OMP51: #pragma omp ordered depend(source)
+// OMP51-NEXT: #pragma omp ordered depend(sink : i + 3)
+#endif
 // CHECK-NEXT: a = 2;
 // CHECK-NEXT: }
 
 // CHECK-LABEL: int main(
 int main (int argc, char **argv) {
   int b = argc, c, d, e, f, g;
   static int a;
 // CHECK: static int a;
// CHECK: static int a;
   #pragma omp parallel for simd
   for (int i =0 ; i < argc; ++i)
   #pragma omp ordered simd
   {
     a=2;
   }
   #pragma omp parallel for ordered(1)
   for (int i =0 ; i < argc; ++i) {
+#if _OPENMP >= 202111
+  #pragma omp ordered doacross(source:)
+  #pragma omp ordered doacross(sink: i - 5)
+#else
   #pragma omp ordered depend(source)
   #pragma omp ordered depend(sink: i - 5)
+#endif
     a = 2;
   }
 // CHECK-NEXT: #pragma omp for ordered
 // CHECK-NEXT: for (int i = 0; i < argc; ++i)
 // CHECK-NEXT: #pragma omp ordered
 // CHECK-NEXT: {
 // CHECK-NEXT: a = 2;
 // CHECK-NEXT: }

 // CHECK-NEXT: #pragma omp parallel for simd
 // CHECK-NEXT: for (int i = 0; i < argc; ++i)
 // CHECK-NEXT: #pragma omp ordered simd
 // CHECK-NEXT: {
 // CHECK-NEXT: a = 2;
 // CHECK-NEXT: }
 // CHECK-NEXT: #pragma omp parallel for ordered(1)
 // CHECK-NEXT: for (int i = 0; i < argc; ++i) {
-// CHECK-NEXT: #pragma omp ordered depend(source)
-// CHECK-NEXT: #pragma omp ordered depend(sink : i - 5)
+#if _OPENMP >= 202111
+// OMP52: #pragma omp ordered doacross(source:)
+// OMP52-NEXT: #pragma omp ordered doacross(sink: i - 5)
+#else
+// OMP51: #pragma omp ordered depend(source)
+// OMP51-NEXT: #pragma omp ordered depend(sink : i - 5)
+#endif
 // CHECK-NEXT: a = 2;
 // CHECK-NEXT: }
   return tmain<int, 3>(argc);
 }
 
 #endif

clang/test/OpenMP/ordered_messages.cpp

 // RUN: %clang_cc1 -verify -fopenmp -ferror-limit 100 -o - %s -Wuninitialized
 // RUN: %clang_cc1 -verify -fopenmp -ferror-limit 100 -std=c++98 -o - %s -Wuninitialized
 // RUN: %clang_cc1 -verify -fopenmp -ferror-limit 100 -std=c++11 -o - %s -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,omp52 -fopenmp -fopenmp-version=52 -ferror-limit 100 -o - %s -Wuninitialized
 
 // RUN: %clang_cc1 -verify -fopenmp-simd -ferror-limit 100 -o - %s -Wuninitialized
 // RUN: %clang_cc1 -verify -fopenmp-simd -ferror-limit 100 -std=c++98 -o - %s -Wuninitialized
 // RUN: %clang_cc1 -verify -fopenmp-simd -ferror-limit 100 -std=c++11 -o - %s -Wuninitialized
 
 void xxx(int argc) {
   int x; // expected-note {{initialize the variable 'x' to silence this warning}}
 #pragma omp for ordered
for (int i = 0; i < 10; ++i) {
     #pragma omp ordered threads // expected-error {{OpenMP constructs may not be nested inside a simd region}}
     {
       foo();
     }
     #pragma omp ordered depend(source) // expected-error {{OpenMP constructs may not be nested inside a simd region}}
   }
 #pragma omp parallel for ordered
   for (int i = 0; i < 10; ++i) {
+#if _OPENMP >= 202111
+    #pragma omp ordered doacross(source:) // omp52-error {{'ordered' directive with 'doacross' clause cannot be closely nested inside ordered region without specified parameter}}
+    #pragma omp ordered doacross(sink : i) // omp52-error {{'ordered' directive with 'doacross' clause cannot be closely nested inside ordered region without specified parameter}}
+#else
     #pragma omp ordered depend(source) // expected-error {{'ordered' directive with 'depend' clause cannot be closely nested inside ordered region without specified parameter}}
     #pragma omp ordered depend(sink : i) // expected-error {{'ordered' directive with 'depend' clause cannot be closely nested inside ordered region without specified parameter}}
+#endif
   }
 #pragma omp parallel for ordered(2) // expected-note 3 {{'ordered' clause with specified parameter}}
   for (int i = 0; i < 10; ++i) {
     for (int j = 0; j < 10; ++j) {
+#if _OPENMP >= 202111
+#pragma omp ordered doacross // omp52-error {{expected '(' after 'doacross'}} omp52-error {{'ordered' directive without any clauses cannot be closely nested inside ordered region with specified parameter}}
+#pragma omp ordered doacross( // omp52-error {{expected ')'}} omp52-error {{expected 'source' or 'sink' in OpenMP clause 'doacross'}} omp52-error {{'ordered' directive without any clauses cannot be closely nested inside ordered region with specified parameter}} omp52-warning {{missing ':' or ')' after dependence-type - ignoring}} omp52-note {{to match this '('}}
+#pragma omp ordered doacross(source // omp52-warning {{missing ':' or ')' after dependence-type - ignoring}} omp52-error {{expected ')'}} omp52-note {{to match this '('}}
+#pragma omp ordered doacross(sink // omp52-error {{expected expression}} omp52-warning {{missing ':' or ')' after dependence-type - ignoring}} omp52-error {{expected ')'}} omp52-note {{to match this '('}} omp52-error {{expected 'i' loop iteration variable}}
+#pragma omp ordered doacross(sink : // omp52-error {{expected ')'}} omp52-note {{to match this '('}} omp52-error {{expected expression}} omp52-error {{expected 'i' loop iteration variable}}
+#pragma omp ordered doacross(sink : i // omp52-error {{expected ')'}} omp52-note {{to match this '('}} omp52-error {{expected 'j' loop iteration variable}}
+#pragma omp ordered doacross(sink : i) // omp52-error {{expected 'j' loop iteration variable}}
+#pragma omp ordered doacross(source:)
+                           if (i == j)
+#pragma omp ordered doacross(source:) // omp52-error {{'#pragma omp ordered' with 'doacross' clause cannot be an immediate substatement}}
+                             ;
+                           if (i == j)
+#pragma omp ordered doacross(sink : i, j) // omp52-error {{'#pragma omp ordered' with 'doacross' clause cannot be an immediate substatement}}
+                             ;
+#pragma omp ordered doacross(source:) threads // omp52-error {{'doacross' clauses cannot be mixed with 'threads' clause}}
+#pragma omp ordered simd doacross(source:) // omp52-error {{'doacross' clauses cannot be mixed with 'simd' clause}}
+#pragma omp ordered doacross(source:) doacross(source:) // omp52-error {{directive '#pragma omp ordered' cannot contain more than one 'doacross' clause with 'source' dependence}}
+#pragma omp ordered doacross(in : i) // omp52-error {{expected 'source' or 'sink' in OpenMP clause 'doacross'}} omp52-error {{'ordered' directive without any clauses cannot be closely nested inside ordered region with specified parameter}}
+#pragma omp ordered doacross(sink : i, j)
+#pragma omp ordered doacross(sink : j, i) // omp52-error {{expected 'i' loop iteration variable}} omp52-error {{expected 'j' loop iteration variable}}
+#pragma omp ordered doacross(sink : i, j, k) // omp52-error {{unexpected expression: number of expressions is larger than the number of associated loops}}
+#pragma omp ordered doacross(sink : i+foo(), j/4) // omp52-error {{integral constant expression}} omp52-error {{expected '+' or '-' operation}}
+#if __cplusplus >= 201103L
+// omp52-note@-2 {{non-constexpr function 'foo' cannot be used in a constant expression}}
+#endif
+#pragma omp ordered doacross(sink : i*0, j-4)// omp52-error {{expected '+' or '-' operation}}
+#pragma omp ordered doacross(sink : i-0, j+sizeof(T)) doacross(sink : i-0, j+sizeof(T))
+#pragma omp ordered doacross(sink : i-0, j+sizeof(T)) doacross(source:) // omp52-error {{'doacross(source)' clause cannot be mixed with 'doacross(sink:vec)' clauses}}
+#pragma omp ordered doacross(source:) doacross(sink : i-0, j+sizeof(T)) // omp52-error {{'doacross(sink:vec)' clauses cannot be mixed with 'doacross(source)' clause}}
+#else
 #pragma omp ordered depend // expected-error {{expected '(' after 'depend'}} expected-error {{'ordered' directive without any clauses cannot be closely nested inside ordered region with specified parameter}}
 #pragma omp ordered depend( // expected-error {{expected ')'}} expected-error {{expected 'source' or 'sink' in OpenMP clause 'depend'}} expected-error {{'ordered' directive without any clauses cannot be closely nested inside ordered region with specified parameter}} expected-warning {{missing ':' or ')' after dependency type - ignoring}} expected-note {{to match this '('}}
 #pragma omp ordered depend(source // expected-error {{expected ')'}} expected-note {{to match this '('}}
 #pragma omp ordered depend(sink // expected-error {{expected expression}} expected-warning {{missing ':' or ')' after dependency type - ignoring}} expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected 'i' loop iteration variable}}
 #pragma omp ordered depend(sink : // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected expression}} expected-error {{expected 'i' loop iteration variable}}
 #pragma omp ordered depend(sink : i // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected 'j' loop iteration variable}}
 #pragma omp ordered depend(sink : i) // expected-error {{expected 'j' loop iteration variable}}
 #pragma omp ordered depend(source)

 #pragma omp ordered depend(sink : i+foo(), j/4) // expected-error {{integral constant expression}} expected-error {{expected '+' or '-' operation}}
 #if __cplusplus >= 201103L
 // expected-note@-2 {{non-constexpr function 'foo' cannot be used in a constant expression}}
 #endif
 #pragma omp ordered depend(sink : i*0, j-4)// expected-error {{expected '+' or '-' operation}}
 #pragma omp ordered depend(sink : i-0, j+sizeof(T)) depend(sink : i-0, j+sizeof(T))
 #pragma omp ordered depend(sink : i-0, j+sizeof(T)) depend(source) // expected-error {{'depend(source)' clause cannot be mixed with 'depend(sink:vec)' clauses}}
 #pragma omp ordered depend(source) depend(sink : i-0, j+sizeof(T)) // expected-error {{'depend(sink:vec)' clauses cannot be mixed with 'depend(source)' clause}}
+#endif
     }
   }
+#if _OPENMP >= 202111
+#else
 #pragma omp ordered depend(source) // expected-error {{'ordered' directive with 'depend' clause cannot be closely nested inside ordered region without specified parameter}}
 #pragma omp ordered depend(sink:k) // expected-error {{'ordered' directive with 'depend' clause cannot be closely nested inside ordered region without specified parameter}}
+#endif
   return T();
 }
 
 int foo() {
 #if __cplusplus >= 201103L
 // expected-note@-2 {{declared here}}
 #endif
 int k;
for (int i = 0; i < 10; ++i) {
     #pragma omp ordered threads // expected-error {{OpenMP constructs may not be nested inside a simd region}}
     {
       foo();
     }
     #pragma omp ordered depend(source) // expected-error {{OpenMP constructs may not be nested inside a simd region}}
   }
 #pragma omp parallel for ordered
   for (int i = 0; i < 10; ++i) {
+#if _OPENMP >= 202111
+#else
     #pragma omp ordered depend(source) // expected-error {{'ordered' directive with 'depend' clause cannot be closely nested inside ordered region without specified parameter}}
     #pragma omp ordered depend(sink : i) // expected-error {{'ordered' directive with 'depend' clause cannot be closely nested inside ordered region without specified parameter}}
+#endif
   }
 #pragma omp parallel for ordered(2) // expected-note 3 {{'ordered' clause with specified parameter}}
   for (int i = 0; i < 10; ++i) {
     for (int j = 0; j < 10; ++j) {
+#if _OPENMP >= 202111
+#pragma omp ordered doacross // omp52-error {{expected '(' after 'doacross'}} omp52-error {{'ordered' directive without any clauses cannot be closely nested inside ordered region with specified parameter}}
+#pragma omp ordered doacross( // omp52-error {{expected ')'}} omp52-error {{expected 'source' or 'sink' in OpenMP clause 'doacross'}} omp52-error {{'ordered' directive without any clauses cannot be closely nested inside ordered region with specified parameter}} omp52-warning {{missing ':' or ')' after dependence-type - ignoring}} omp52-note {{to match this '('}}
+#pragma omp ordered doacross(source // omp52-warning {{missing ':' or ')' after dependence-type - ignoring}} omp52-error {{expected ')'}} omp52-note {{to match this '('}}
+#pragma omp ordered doacross(sink // omp52-error {{expected expression}} omp52-warning {{missing ':' or ')' after dependence-type - ignoring}} omp52-error {{expected ')'}} omp52-note {{to match this '('}} omp52-error {{expected 'i' loop iteration variable}}
+#pragma omp ordered doacross(sink : // omp52-error {{expected ')'}} omp52-note {{to match this '('}} omp52-error {{expected expression}} omp52-error {{expected 'i' loop iteration variable}}
+#pragma omp ordered doacross(sink : i // omp52-error {{expected ')'}} omp52-note {{to match this '('}} omp52-error {{expected 'j' loop iteration variable}}
+#pragma omp ordered doacross(sink : i) // omp52-error {{expected 'j' loop iteration variable}}
+#pragma omp ordered doacross(source:)
+                           if (i == j)
+#pragma omp ordered doacross(source:) // omp52-error {{'#pragma omp ordered' with 'doacross' clause cannot be an immediate substatement}}
+                             ;
+                           if (i == j)
+#pragma omp ordered doacross(sink : i, j) // omp52-error {{'#pragma omp ordered' with 'doacross' clause cannot be an immediate substatement}}
+                             ;
+#pragma omp ordered doacross(source:) threads // omp52-error {{'doacross' clauses cannot be mixed with 'threads' clause}}
+#pragma omp ordered simd doacross(source:) // omp52-error {{'doacross' clauses cannot be mixed with 'simd' clause}}
+#pragma omp ordered doacross(source:) doacross(source:) // omp52-error {{directive '#pragma omp ordered' cannot contain more than one 'doacross' clause with 'source' dependence}}
+#pragma omp ordered doacross(in : i) // omp52-error {{expected 'source' or 'sink' in OpenMP clause 'doacross'}} omp52-error {{'ordered' directive without any clauses cannot be closely nested inside ordered region with specified parameter}}
+#pragma omp ordered doacross(sink : i, j) allocate(i) // omp52-error {{unexpected OpenMP clause 'allocate' in directive '#pragma omp ordered'}}
+#pragma omp ordered doacross(sink : j, i) // omp52-error {{expected 'i' loop iteration variable}} omp52-error {{expected 'j' loop iteration variable}}
+#pragma omp ordered doacross(sink : i, j, k) // omp52-error {{unexpected expression: number of expressions is larger than the number of associated loops}}
+#pragma omp ordered doacross(sink : i+foo(), j/4) // omp52-error {{integral constant expression}} omp52-error {{expected '+' or '-' operation}}
+#if __cplusplus >= 201103L
+// omp52-note@-2 {{non-constexpr function 'foo' cannot be used in a constant expression}}
+#endif
+#pragma omp ordered doacross(sink : i*0, j-4)// omp52-error {{expected '+' or '-' operation}}
+#pragma omp ordered doacross(sink : i-0, j+sizeof(int)) doacross(sink : i-0, j+sizeof(int))
+#pragma omp ordered doacross(sink : i-0, j+sizeof(int)) doacross(source:) // omp52-error {{'doacross(source)' clause cannot be mixed with 'doacross(sink:vec)' clauses}}
+#pragma omp ordered doacross(source:) doacross(sink : i-0, j+sizeof(int)) // omp52-error {{'doacross(sink:vec)' clauses cannot be mixed with 'doacross(source)' clause}}
+#else
 #pragma omp ordered depend // expected-error {{expected '(' after 'depend'}} expected-error {{'ordered' directive without any clauses cannot be closely nested inside ordered region with specified parameter}}
 #pragma omp ordered depend( // expected-error {{expected ')'}} expected-error {{expected 'source' or 'sink' in OpenMP clause 'depend'}} expected-error {{'ordered' directive without any clauses cannot be closely nested inside ordered region with specified parameter}} expected-warning {{missing ':' or ')' after dependency type - ignoring}} expected-note {{to match this '('}}
 #pragma omp ordered depend(source // expected-error {{expected ')'}} expected-note {{to match this '('}}
 #pragma omp ordered depend(sink // expected-error {{expected expression}} expected-warning {{missing ':' or ')' after dependency type - ignoring}} expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected 'i' loop iteration variable}}
 #pragma omp ordered depend(sink : // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected expression}} expected-error {{expected 'i' loop iteration variable}}
 #pragma omp ordered depend(sink : i // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected 'j' loop iteration variable}}
 #pragma omp ordered depend(sink : i) // expected-error {{expected 'j' loop iteration variable}}
 #pragma omp ordered depend(source)

 #pragma omp ordered depend(sink : i+foo(), j/4) // expected-error {{integral constant expression}} expected-error {{expected '+' or '-' operation}}
 #if __cplusplus >= 201103L
 // expected-note@-2 {{non-constexpr function 'foo' cannot be used in a constant expression}}
 #endif
 #pragma omp ordered depend(sink : i*0, j-4)// expected-error {{expected '+' or '-' operation}}
 #pragma omp ordered depend(sink : i-0, j+sizeof(int)) depend(sink : i-0, j+sizeof(int))
 #pragma omp ordered depend(sink : i-0, j+sizeof(int)) depend(source) // expected-error {{'depend(source)' clause cannot be mixed with 'depend(sink:vec)' clauses}}
 #pragma omp ordered depend(source) depend(sink : i-0, j+sizeof(int)) // expected-error {{'depend(sink:vec)' clauses cannot be mixed with 'depend(source)' clause}}
+#endif
     }
   }
 
 #pragma omp for ordered(2) // expected-note {{as specified in 'ordered' clause}}
   for (int i = 0; i < 10; ++i) { // expected-error {{expected 2 for loops after '#pragma omp for', but found only 1}}
+#if _OPENMP >= 202111
+#pragma omp ordered doacross(sink : i)
+    int j;
+#pragma omp ordered doacross(sink : i, j) // omp52-error {{expected loop iteration variable}}
+#else
 #pragma omp ordered depend(sink : i)
     int j;
 #pragma omp ordered depend(sink : i, j) // expected-error {{expected loop iteration variable}}
+#endif
     foo();
   }
 
   return foo<int>(); // expected-note {{in instantiation of function template specialization 'foo<int>' requested here}}
 }

clang/tools/libclang/CIndex.cpp

for (const Expr *E : C->varlists())
     Visitor->AddStmt(E);
 }
 void OMPClauseEnqueue::VisitOMPBindClause(const OMPBindClause *C) {}
 void OMPClauseEnqueue::VisitOMPXDynCGroupMemClause(
     const OMPXDynCGroupMemClause *C) {
   VisitOMPClauseWithPreInit(C);
   Visitor->AddStmt(C->getSize());
 }
+void OMPClauseEnqueue::VisitOMPDoacrossClause(const OMPDoacrossClause *C) {
+  VisitOMPClauseList(C);
+}
 
 } // namespace
 
 void EnqueueVisitor::EnqueueChildren(const OMPClause *S) {
   unsigned size = WL.size();
   OMPClauseEnqueue Visitor(this);
   Visitor.Visit(S);
   if (size == WL.size())

flang/lib/Semantics/check-omp-structure.cpp

 CHECK_SIMPLE_CLAUSE(When, OMPC_when)
 CHECK_SIMPLE_CLAUSE(AdjustArgs, OMPC_adjust_args)
 CHECK_SIMPLE_CLAUSE(AppendArgs, OMPC_append_args)
 CHECK_SIMPLE_CLAUSE(MemoryOrder, OMPC_memory_order)
 CHECK_SIMPLE_CLAUSE(Bind, OMPC_bind)
 CHECK_SIMPLE_CLAUSE(Align, OMPC_align)
 CHECK_SIMPLE_CLAUSE(Compare, OMPC_compare)
 CHECK_SIMPLE_CLAUSE(CancellationConstructType, OMPC_cancellation_construct_type)
+CHECK_SIMPLE_CLAUSE(Doacross, OMPC_doacross)
 
 CHECK_REQ_SCALAR_INT_CLAUSE(Grainsize, OMPC_grainsize)
 CHECK_REQ_SCALAR_INT_CLAUSE(NumTasks, OMPC_num_tasks)
 CHECK_REQ_SCALAR_INT_CLAUSE(NumTeams, OMPC_num_teams)
 CHECK_REQ_SCALAR_INT_CLAUSE(NumThreads, OMPC_num_threads)
 CHECK_REQ_SCALAR_INT_CLAUSE(OmpxDynCgroupMem, OMPC_ompx_dyn_cgroup_mem)
 CHECK_REQ_SCALAR_INT_CLAUSE(Priority, OMPC_priority)
 CHECK_REQ_SCALAR_INT_CLAUSE(ThreadLimit, OMPC_thread_limit)

llvm/include/llvm/Frontend/OpenMP/OMP.td

def OMPC_Bind : Clause<"bind"> {
   let clangClass = "OMPBindClause";
 }
 
 def OMPC_OMPX_DynCGroupMem : Clause<"ompx_dyn_cgroup_mem"> {
   let clangClass = "OMPXDynCGroupMemClause";
   let flangClass = "ScalarIntExpr";
 }
 
+def OMPC_Doacross : Clause<"doacross"> {
+  let clangClass = "OMPDoacrossClause";
+}
+
 //===----------------------------------------------------------------------===//
 // Definition of OpenMP directives
 //===----------------------------------------------------------------------===//
 
 def OMP_ThreadPrivate : Directive<"threadprivate"> {}
 def OMP_Parallel : Directive<"parallel"> {
   let allowedClauses = [
     VersionedClause<OMPC_Private>,
let allowedOnceClauses = [
     VersionedClause<OMPC_Release, 50>,
     // TODO This should ne `none` instead. Comment carried over from
     // OMPKinds.def.
     VersionedClause<OMPC_Flush>
   ];
 }
 def OMP_Ordered : Directive<"ordered"> {
   let allowedClauses = [
-    VersionedClause<OMPC_Depend>
+    VersionedClause<OMPC_Depend>,
+    VersionedClause<OMPC_Doacross, 52>
   ];
   let allowedOnceClauses = [
     VersionedClause<OMPC_Threads>,
     VersionedClause<OMPC_Simd>
   ];
 }
 def OMP_Atomic : Directive<"atomic"> {
   let allowedClauses = [