Repository navigation
[OpenMP] Add 'replayable' clause (for 'taskgraph' support) - #194053
Draft
Conversation
Created using spr 1.3.5
llvmbot
added
clang:frontend
Member
|
@llvm/pr-subscribers-clang-codegen @llvm/pr-subscribers-flang-openmp Author: Julian Brown (jtb20) ChangesThis patch adds support for the 'replayable' clause to Clang. This is commit-id:4a9a97fd Patch is 27.09 KiB, truncated to 20.00 KiB below, full version: https://github.com/llvm/llvm-project/pull/194053.diff 15 Files Affected:
diff --git a/clang/include/clang/AST/OpenMPClause.h b/clang/include/clang/AST/OpenMPClause.h
index bee7bbaf6f952..36e5cdb4edbac 100644
--- a/clang/include/clang/AST/OpenMPClause.h
+++ b/clang/include/clang/AST/OpenMPClause.h
@@ -2040,6 +2040,75 @@ class OMPSelfMapsClause final : public OMPClause {
}
};
+/// This represents a 'replayable' clause in the '#pragma omp target',
+// '#pragma omp target enter data', '#pragma omp target exit data',
+// '#pragma omp target update', '#pragma omp task', '#pragma omp taskloop' or
+// '#pragma omp taskwait' directive.
+///
+/// \code
+/// #pragma omp task replayable(1)
+/// \endcode
+/// In this example directive '#pragma omp task' has the 'replayable' clause.
+class OMPReplayableClause final : public OMPClause {
+ friend class OMPClauseReader;
+
+ /// Location of '('.
+ SourceLocation LParenLoc;
+
+ /// Condition of the 'replayable' clause.
+ Stmt *Condition = nullptr;
+
+public:
+ /// Build 'replayable' clause.
+ ///
+ /// \param Cond Condition of the clause.
+ /// \param StartLoc Starting location of the clause.
+ /// \param LParenLoc Location of '('.
+ /// \param EndLoc Ending location of the clause.
+ OMPReplayableClause(Expr *Cond, SourceLocation StartLoc,
+ SourceLocation LParenLoc, SourceLocation EndLoc)
+ : OMPClause(llvm::omp::OMPC_replayable, StartLoc, EndLoc),
+ LParenLoc(LParenLoc), Condition(Cond) {}
+
+ /// Build an empty clause.
+ OMPReplayableClause()
+ : OMPClause(llvm::omp::OMPC_replayable, SourceLocation(),
+ SourceLocation()) {}
+
+ /// Sets the location of '('.
+ void setLParenLoc(SourceLocation Loc) { LParenLoc = Loc; }
+
+ /// Returns the location of '('.
+ SourceLocation getLParenLoc() const { return LParenLoc; }
+
+ /// Set condition.
+ void setCondition(Expr *Cond) { Condition = Cond; }
+
+ /// Returns condition.
+ Expr *getCondition() const { return cast_or_null<Expr>(Condition); }
+
+ child_range children() {
+ if (Condition)
+ return child_range(&Condition, &Condition + 1);
+ return child_range(child_iterator(), child_iterator());
+ }
+
+ const_child_range children() const {
+ if (Condition)
+ return const_child_range(&Condition, &Condition + 1);
+ return const_child_range(const_child_iterator(), const_child_iterator());
+ }
+
+ child_range used_children();
+ const_child_range used_children() const {
+ return const_cast<OMPReplayableClause *>(this)->used_children();
+ }
+
+ static bool classof(const OMPClause *T) {
+ return T->getClauseKind() == llvm::omp::OMPC_replayable;
+ }
+};
+
/// This represents 'at' clause in the '#pragma omp error' directive
///
/// \code
diff --git a/clang/include/clang/AST/RecursiveASTVisitor.h b/clang/include/clang/AST/RecursiveASTVisitor.h
index 8f1d4c2a23034..827d0840dfe21 100644
--- a/clang/include/clang/AST/RecursiveASTVisitor.h
+++ b/clang/include/clang/AST/RecursiveASTVisitor.h
@@ -3657,6 +3657,12 @@ bool RecursiveASTVisitor<Derived>::VisitOMPNowaitClause(OMPNowaitClause *C) {
return true;
}
+template <typename Derived>
+bool RecursiveASTVisitor<Derived>::VisitOMPReplayableClause(OMPReplayableClause *C) {
+ TRY_TO(TraverseStmt(C->getCondition()));
+ return true;
+}
+
template <typename Derived>
bool RecursiveASTVisitor<Derived>::VisitOMPUntiedClause(OMPUntiedClause *) {
return true;
diff --git a/clang/include/clang/Sema/SemaOpenMP.h b/clang/include/clang/Sema/SemaOpenMP.h
index 7c500847881f0..98405b871641a 100644
--- a/clang/include/clang/Sema/SemaOpenMP.h
+++ b/clang/include/clang/Sema/SemaOpenMP.h
@@ -1173,6 +1173,12 @@ class SemaOpenMP : public SemaBase {
OMPClause *ActOnOpenMPSelfMapsClause(SourceLocation StartLoc,
SourceLocation EndLoc);
+ /// Called on well-formed 'replayable' clause.
+ OMPClause *ActOnOpenMPReplayableClause(SourceLocation StartLoc,
+ SourceLocation EndLoc,
+ SourceLocation LParenLoc,
+ Expr *Condition);
+
/// Called on well-formed 'at' clause.
OMPClause *ActOnOpenMPAtClause(OpenMPAtClauseKind Kind,
SourceLocation KindLoc,
diff --git a/clang/lib/AST/OpenMPClause.cpp b/clang/lib/AST/OpenMPClause.cpp
index db4d5519acb38..7d2b4aa64a1df 100644
--- a/clang/lib/AST/OpenMPClause.cpp
+++ b/clang/lib/AST/OpenMPClause.cpp
@@ -324,6 +324,12 @@ OMPClause::child_range OMPNowaitClause::used_children() {
return children();
}
+OMPClause::child_range OMPReplayableClause::used_children() {
+ if (Condition)
+ return child_range(&Condition, &Condition + 1);
+ return children();
+}
+
OMPClause::child_range OMPGrainsizeClause::used_children() {
if (Stmt **C = getAddrOfExprAsWritten(getPreInitStmt()))
return child_range(C, C + 1);
@@ -2201,6 +2207,15 @@ void OMPClausePrinter::VisitOMPNowaitClause(OMPNowaitClause *Node) {
}
}
+void OMPClausePrinter::VisitOMPReplayableClause(OMPReplayableClause *Node) {
+ OS << "replayable";
+ if (auto *Cond = Node->getCondition()) {
+ OS << "(";
+ Cond->printPretty(OS, nullptr, Policy, 0);
+ OS << ")";
+ }
+}
+
void OMPClausePrinter::VisitOMPUntiedClause(OMPUntiedClause *) {
OS << "untied";
}
diff --git a/clang/lib/AST/StmtProfile.cpp b/clang/lib/AST/StmtProfile.cpp
index 39b4377ec1577..d8caf515d27d7 100644
--- a/clang/lib/AST/StmtProfile.cpp
+++ b/clang/lib/AST/StmtProfile.cpp
@@ -606,6 +606,11 @@ void OMPClauseProfiler::VisitOMPNowaitClause(const OMPNowaitClause *C) {
Profiler->VisitStmt(C->getCondition());
}
+void OMPClauseProfiler::VisitOMPReplayableClause(const OMPReplayableClause *C) {
+ if (C->getCondition())
+ Profiler->VisitStmt(C->getCondition());
+}
+
void OMPClauseProfiler::VisitOMPUntiedClause(const OMPUntiedClause *) {}
void OMPClauseProfiler::VisitOMPMergeableClause(const OMPMergeableClause *) {}
diff --git a/clang/lib/CodeGen/CGOpenMPRuntime.cpp b/clang/lib/CodeGen/CGOpenMPRuntime.cpp
index 5361e830dcab3..63e85f9e8d2a4 100644
--- a/clang/lib/CodeGen/CGOpenMPRuntime.cpp
+++ b/clang/lib/CodeGen/CGOpenMPRuntime.cpp
@@ -4789,6 +4789,7 @@ void CGOpenMPRuntime::emitTaskCall(CodeGenFunction &CGF, SourceLocation Loc,
llvm::Function *TaskFunction,
QualType SharedsTy, Address Shareds,
const Expr *IfCond,
+ const Expr *ReplayableCond,
const OMPTaskDataTy &Data) {
if (!CGF.HaveInsertPoint())
return;
@@ -4953,8 +4954,16 @@ void CGOpenMPRuntime::emitTaskCall(CodeGenFunction &CGF, SourceLocation Loc,
RegionCodeGenTy TaskgraphRCG(TaskgraphTaskCodeGen);
TaskgraphRCG(CGF);
} else {
- RegionCodeGenTy NonTaskgraphRCG(NonTaskgraphTaskCodeGen);
- NonTaskgraphRCG(CGF);
+ if (ReplayableCond) {
+ // We have a replayable clause. Task is replayable if its argument is
+ // omitted or evaluates to TRUE.
+ emitIfClause(CGF, ReplayableCond, TaskgraphTaskCodeGen,
+ NonTaskgraphTaskCodeGen);
+ } else {
+ // Not taskgraph, not replayable.
+ RegionCodeGenTy NonTaskgraphRCG(NonTaskgraphTaskCodeGen);
+ NonTaskgraphRCG(CGF);
+ }
}
}
@@ -4963,6 +4972,7 @@ void CGOpenMPRuntime::emitTaskLoopCall(CodeGenFunction &CGF, SourceLocation Loc,
llvm::Function *TaskFunction,
QualType SharedsTy, Address Shareds,
const Expr *IfCond,
+ const Expr *ReplayableCond,
const OMPTaskDataTy &Data) {
if (!CGF.HaveInsertPoint())
return;
@@ -5136,8 +5146,16 @@ void CGOpenMPRuntime::emitTaskLoopCall(CodeGenFunction &CGF, SourceLocation Loc,
RegionCodeGenTy TaskgraphRCG(TaskgraphTaskloopCodeGen);
TaskgraphRCG(CGF);
} else {
- RegionCodeGenTy NonTaskgraphRCG(NonTaskgraphTaskloopCodeGen);
- NonTaskgraphRCG(CGF);
+ if (ReplayableCond) {
+ // We have a replayable clause. Taskloop is replayable if its argument
+ // is omitted or evaluates to TRUE.
+ emitIfClause(CGF, ReplayableCond, TaskgraphTaskloopCodeGen,
+ NonTaskgraphTaskloopCodeGen);
+ } else {
+ // Not taskgraph, not replayable.
+ RegionCodeGenTy NonTaskgraphRCG(NonTaskgraphTaskloopCodeGen);
+ NonTaskgraphRCG(CGF);
+ }
}
}
@@ -6335,6 +6353,7 @@ Address CGOpenMPRuntime::getTaskReductionItem(CodeGenFunction &CGF,
}
void CGOpenMPRuntime::emitTaskwaitCall(CodeGenFunction &CGF, SourceLocation Loc,
+ const Expr *ReplayableCond,
const OMPTaskDataTy &Data) {
if (!CGF.HaveInsertPoint())
return;
@@ -6409,8 +6428,16 @@ void CGOpenMPRuntime::emitTaskwaitCall(CodeGenFunction &CGF, SourceLocation Loc,
RegionCodeGenTy TaskgraphRCG(TaskgraphTaskwaitCodeGen);
TaskgraphRCG(CGF);
} else {
- RegionCodeGenTy NonTaskgraphRCG(NonTaskgraphTaskwaitCodeGen);
- NonTaskgraphRCG(CGF);
+ if (ReplayableCond) {
+ // We have a replayable clause. Taskwait is replayable if its argument
+ // is omitted or evaluates to TRUE.
+ emitIfClause(CGF, ReplayableCond, TaskgraphTaskwaitCodeGen,
+ NonTaskgraphTaskwaitCodeGen);
+ } else {
+ // Not taskgraph, not replayable.
+ RegionCodeGenTy NonTaskgraphRCG(NonTaskgraphTaskwaitCodeGen);
+ NonTaskgraphRCG(CGF);
+ }
}
}
@@ -13568,6 +13595,7 @@ void CGOpenMPSIMDRuntime::emitTaskCall(CodeGenFunction &CGF, SourceLocation Loc,
llvm::Function *TaskFunction,
QualType SharedsTy, Address Shareds,
const Expr *IfCond,
+ const Expr *ReplayableCond,
const OMPTaskDataTy &Data) {
llvm_unreachable("Not supported in SIMD-only mode");
}
@@ -13575,7 +13603,7 @@ void CGOpenMPSIMDRuntime::emitTaskCall(CodeGenFunction &CGF, SourceLocation Loc,
void CGOpenMPSIMDRuntime::emitTaskLoopCall(
CodeGenFunction &CGF, SourceLocation Loc, const OMPLoopDirective &D,
llvm::Function *TaskFunction, QualType SharedsTy, Address Shareds,
- const Expr *IfCond, const OMPTaskDataTy &Data) {
+ const Expr *IfCond, const Expr *ReplayableCond, const OMPTaskDataTy &Data) {
llvm_unreachable("Not supported in SIMD-only mode");
}
@@ -13616,6 +13644,7 @@ Address CGOpenMPSIMDRuntime::getTaskReductionItem(CodeGenFunction &CGF,
void CGOpenMPSIMDRuntime::emitTaskwaitCall(CodeGenFunction &CGF,
SourceLocation Loc,
+ const Expr *ReplayableCond,
const OMPTaskDataTy &Data) {
llvm_unreachable("Not supported in SIMD-only mode");
}
diff --git a/clang/lib/CodeGen/CGOpenMPRuntime.h b/clang/lib/CodeGen/CGOpenMPRuntime.h
index 9a4de26418e45..7ac06547a5409 100644
--- a/clang/lib/CodeGen/CGOpenMPRuntime.h
+++ b/clang/lib/CodeGen/CGOpenMPRuntime.h
@@ -123,6 +123,7 @@ struct OMPTaskDataTy final {
bool IsWorksharingReduction = false;
bool HasNowaitClause = false;
bool HasModifier = false;
+ const Expr *ReplayableCond = nullptr;
};
/// Class intended to support codegen of all kind of the reduction clauses.
@@ -1177,6 +1178,7 @@ class CGOpenMPRuntime {
const OMPExecutableDirective &D,
llvm::Function *TaskFunction, QualType SharedsTy,
Address Shareds, const Expr *IfCond,
+ const Expr *ReplayableCond,
const OMPTaskDataTy &Data);
/// Emit task region for the taskloop directive. The taskloop region is
@@ -1212,7 +1214,8 @@ class CGOpenMPRuntime {
const OMPLoopDirective &D,
llvm::Function *TaskFunction,
QualType SharedsTy, Address Shareds,
- const Expr *IfCond, const OMPTaskDataTy &Data);
+ const Expr *IfCond, const Expr *ReplayableCond,
+ const OMPTaskDataTy &Data);
/// Emit code for the directive that does not require outlining.
///
@@ -1380,6 +1383,7 @@ class CGOpenMPRuntime {
/// Emit code for 'taskwait' directive.
virtual void emitTaskwaitCall(CodeGenFunction &CGF, SourceLocation Loc,
+ const Expr *ReplayableCond,
const OMPTaskDataTy &Data);
/// Emit code for 'taskgraph' directive.
@@ -2058,6 +2062,7 @@ class CGOpenMPSIMDRuntime final : public CGOpenMPRuntime {
const OMPExecutableDirective &D,
llvm::Function *TaskFunction, QualType SharedsTy,
Address Shareds, const Expr *IfCond,
+ const Expr *ReplayableCond,
const OMPTaskDataTy &Data) override;
/// Emit task region for the taskloop directive. The taskloop region is
@@ -2092,7 +2097,8 @@ class CGOpenMPSIMDRuntime final : public CGOpenMPRuntime {
void emitTaskLoopCall(CodeGenFunction &CGF, SourceLocation Loc,
const OMPLoopDirective &D, llvm::Function *TaskFunction,
QualType SharedsTy, Address Shareds, const Expr *IfCond,
- const OMPTaskDataTy &Data) override;
+ const Expr *ReplayableCond, const OMPTaskDataTy &Data)
+ override;
/// Emit a code for reduction clause. Next code should be emitted for
/// reduction:
@@ -2212,7 +2218,8 @@ class CGOpenMPSIMDRuntime final : public CGOpenMPRuntime {
/// Emit code for 'taskwait' directive.
void emitTaskwaitCall(CodeGenFunction &CGF, SourceLocation Loc,
- const OMPTaskDataTy &Data) override;
+ const Expr *ReplayableCond, const OMPTaskDataTy &Data)
+ override;
/// Emit code for 'taskgraph' directive.
/// \param IfCond Expression evaluated in if clause associated with the target
diff --git a/clang/lib/CodeGen/CGStmtOpenMP.cpp b/clang/lib/CodeGen/CGStmtOpenMP.cpp
index 65eb8177652a8..afcdc95237874 100644
--- a/clang/lib/CodeGen/CGStmtOpenMP.cpp
+++ b/clang/lib/CodeGen/CGStmtOpenMP.cpp
@@ -5721,8 +5721,20 @@ void CodeGenFunction::EmitOMPTargetTaskBasedDirective(
IntegerLiteral IfCond(getContext(), TrueOrFalse,
getContext().getIntTypeForBitwidth(32, /*Signed=*/0),
SourceLocation());
+ const Expr *ReplayableCond = nullptr;
+ if (auto *RC = S.getSingleClause<OMPReplayableClause>()) {
+ ReplayableCond = RC->getCondition();
+ if (!ReplayableCond) {
+ ReplayableCond =
+ IntegerLiteral::Create(
+ getContext(), llvm::APInt(32, 1),
+ getContext().getIntTypeForBitwidth(32, /*Signed=*/0),
+ SourceLocation());
+ }
+ }
CGM.getOpenMPRuntime().emitTaskCall(*this, S.getBeginLoc(), S, OutlinedFn,
- SharedsTy, CapturedStruct, &IfCond, Data);
+ SharedsTy, CapturedStruct, &IfCond,
+ ReplayableCond, Data);
}
void CodeGenFunction::processInReduction(const OMPExecutableDirective &S,
@@ -5831,15 +5843,27 @@ void CodeGenFunction::EmitOMPTaskDirective(const OMPTaskDirective &S) {
OMPTaskDataTy Data;
// Check if we should emit tied or untied task.
Data.Tied = !S.getSingleClause<OMPUntiedClause>();
+ const Expr *ReplayableCond = nullptr;
+ if (auto *RC = S.getSingleClause<OMPReplayableClause>()) {
+ ReplayableCond = RC->getCondition();
+ if (!ReplayableCond) {
+ ReplayableCond =
+ IntegerLiteral::Create(
+ getContext(), llvm::APInt(32, 1),
+ getContext().getIntTypeForBitwidth(32, /*Signed=*/0),
+ SourceLocation());
+ }
+ }
auto &&BodyGen = [CS](CodeGenFunction &CGF, PrePostActionTy &) {
CGF.EmitStmt(CS->getCapturedStmt());
};
auto &&TaskGen = [&S, SharedsTy, CapturedStruct,
- IfCond](CodeGenFunction &CGF, llvm::Function *OutlinedFn,
+ IfCond, ReplayableCond](CodeGenFunction &CGF,
+ llvm::Function *OutlinedFn,
const OMPTaskDataTy &Data) {
CGF.CGM.getOpenMPRuntime().emitTaskCall(CGF, S.getBeginLoc(), S, OutlinedFn,
SharedsTy, CapturedStruct, IfCond,
- Data);
+ ReplayableCond, Data);
};
auto LPCRegion =
CGOpenMPRuntime::LastprivateConditionalRAII::disable(*this, S);
@@ -5870,7 +5894,19 @@ void CodeGenFunction::EmitOMPTaskwaitDirective(const OMPTaskwaitDirective &S) {
// Build list of dependences
buildDependences(S, Data);
Data.HasNowaitClause = S.hasClausesOfKind<OMPNowaitClause>();
- CGM.getOpenMPRuntime().emitTaskwaitCall(*this, S.getBeginLoc(), Data);
+ const Expr *ReplayableCond = nullptr;
+ if (auto *RC = S.getSingleClause<OMPReplayableClause>()) {
+ ReplayableCond = RC->getCondition();
+ if (!ReplayableCond) {
+ ReplayableCond =
+ IntegerLiteral::Create(
+ getContext(), llvm::APInt(32, 1),
+ getContext().getIntTypeForBitwidth(32, /*Signed=*/0),
+ SourceLocation());
+ }
+ }
+ CGM.getOpenMPRuntime().emitTaskwaitCall(*this, S.getBeginLoc(),
+ ReplayableCond, Data);
}
void CodeGenFunction::EmitOMPTaskgraphDirective(
@@ -8225,6 +8261,18 @@ void CodeGenFunction::EmitOMPTaskLoopBasedDirective(const OMPLoopDirective &S) {
}
}
+ const Expr *ReplayableCond = nullptr;
+ if (auto *RC = S.getSingleClause<OMPReplayableClause>()) {
+ ReplayableCond = RC->getCondition();
+ if (!ReplayableCond) {
+ ReplayableCond =
+ IntegerLiteral::Create(
+ getContext(), llvm::APInt(32, 1),
+ getContext().getIntTypeForBitwidth(32, /*Signed=*/0),
+ SourceLocation());
+ }
+ }
+
OMPTaskDataTy Data;
// Check if taskloop must be emitted without taskgroup.
Data.Nogroup = S.getSingleClause<OMPNogroupClause>();
@@ -8345,16 +8393,17 @@ void CodeGenFunction::EmitOMPTaskLoopBasedDirective(const OMPLoopDirective &S) {
});
};
auto &&TaskGen =
- [&S, SharedsTy, CapturedStruct, IfCond]
+ [&S, SharedsTy, CapturedStruct, IfCond, ReplayableCond]
(CodeGenFunction &CGF, llvm::Function *OutlinedFn,
const OMPTaskDataTy &Data) {
auto &&CodeGen =
- [&S, OutlinedFn, SharedsTy, CapturedStruct, IfCond, &Data]
- (CodeGenFunction &CGF, PrePostActionTy &) {
+ [&S, OutlinedFn, SharedsTy, CapturedStruct, IfCond, ReplayableCond,
+ &Data](CodeGenFunction &CGF, PrePostActionTy &) {
OMPLoopScope PreInitScope(CGF, S);
CGF.CGM.getOpenMPRuntime().emitTaskLoopCall(CGF, S.getBeginLoc(), S,
OutlinedFn, SharedsTy,
- CapturedStruct, IfCond, Data);
+ CapturedStruct, IfCond,
+ ReplayableCond, Data);
};
CGF.CGM.getOpenMPRuntime().emitInlinedDirective(CGF, OMPD_taskloop,
CodeGen);
diff --git a/clang/lib/Parse/ParseOpenMP.cpp b/clang/lib/Parse/ParseOpenMP.cpp
index 5be99550f1ef6..d1394e1c62674 100644
--- a/clang/lib/Parse/ParseOpenMP.cpp
+++ b/clang/lib/Parse/ParseOpenMP.cpp
@@ -3386,13 +3386,15 @@ OMPClause *Parser::ParseOpenMPClause(OpenMPDirectiveKind DKind,
ErrorFound = true;
}
- if (CKind == OMPC_nowait && PP.LookAhead(/*N=*/0).is(tok::l_paren) &&
+ if (CKind == OMPC_nowait &&
+ PP.LookAhead(/*N=*/0).is(tok::l_paren) &&
getLangOpts().OpenMP >= 60)
Clause = ParseOpenMPSingleExprClause(CKind, WrongDirective);
else
Clause = ParseOpenMPClause(CKind, WrongDirective);
...
[truncated]
|
Member
|
@llvm/pr-subscribers-clang-modules Author: Julian Brown (jtb20) ChangesThis patch adds support for the 'replayable' clause to Clang. This is commit-id:4a9a97fd Patch is 27.09 KiB, truncated to 20.00 KiB below, full version: https://github.com/llvm/llvm-project/pull/194053.diff 15 Files Affected:
diff --git a/clang/include/clang/AST/OpenMPClause.h b/clang/include/clang/AST/OpenMPClause.h
index bee7bbaf6f952..36e5cdb4edbac 100644
--- a/clang/include/clang/AST/OpenMPClause.h
+++ b/clang/include/clang/AST/OpenMPClause.h
@@ -2040,6 +2040,75 @@ class OMPSelfMapsClause final : public OMPClause {
}
};
+/// This represents a 'replayable' clause in the '#pragma omp target',
+// '#pragma omp target enter data', '#pragma omp target exit data',
+// '#pragma omp target update', '#pragma omp task', '#pragma omp taskloop' or
+// '#pragma omp taskwait' directive.
+///
+/// \code
+/// #pragma omp task replayable(1)
+/// \endcode
+/// In this example directive '#pragma omp task' has the 'replayable' clause.
+class OMPReplayableClause final : public OMPClause {
+ friend class OMPClauseReader;
+
+ /// Location of '('.
+ SourceLocation LParenLoc;
+
+ /// Condition of the 'replayable' clause.
+ Stmt *Condition = nullptr;
+
+public:
+ /// Build 'replayable' clause.
+ ///
+ /// \param Cond Condition of the clause.
+ /// \param StartLoc Starting location of the clause.
+ /// \param LParenLoc Location of '('.
+ /// \param EndLoc Ending location of the clause.
+ OMPReplayableClause(Expr *Cond, SourceLocation StartLoc,
+ SourceLocation LParenLoc, SourceLocation EndLoc)
+ : OMPClause(llvm::omp::OMPC_replayable, StartLoc, EndLoc),
+ LParenLoc(LParenLoc), Condition(Cond) {}
+
+ /// Build an empty clause.
+ OMPReplayableClause()
+ : OMPClause(llvm::omp::OMPC_replayable, SourceLocation(),
+ SourceLocation()) {}
+
+ /// Sets the location of '('.
+ void setLParenLoc(SourceLocation Loc) { LParenLoc = Loc; }
+
+ /// Returns the location of '('.
+ SourceLocation getLParenLoc() const { return LParenLoc; }
+
+ /// Set condition.
+ void setCondition(Expr *Cond) { Condition = Cond; }
+
+ /// Returns condition.
+ Expr *getCondition() const { return cast_or_null<Expr>(Condition); }
+
+ child_range children() {
+ if (Condition)
+ return child_range(&Condition, &Condition + 1);
+ return child_range(child_iterator(), child_iterator());
+ }
+
+ const_child_range children() const {
+ if (Condition)
+ return const_child_range(&Condition, &Condition + 1);
+ return const_child_range(const_child_iterator(), const_child_iterator());
+ }
+
+ child_range used_children();
+ const_child_range used_children() const {
+ return const_cast<OMPReplayableClause *>(this)->used_children();
+ }
+
+ static bool classof(const OMPClause *T) {
+ return T->getClauseKind() == llvm::omp::OMPC_replayable;
+ }
+};
+
/// This represents 'at' clause in the '#pragma omp error' directive
///
/// \code
diff --git a/clang/include/clang/AST/RecursiveASTVisitor.h b/clang/include/clang/AST/RecursiveASTVisitor.h
index 8f1d4c2a23034..827d0840dfe21 100644
--- a/clang/include/clang/AST/RecursiveASTVisitor.h
+++ b/clang/include/clang/AST/RecursiveASTVisitor.h
@@ -3657,6 +3657,12 @@ bool RecursiveASTVisitor<Derived>::VisitOMPNowaitClause(OMPNowaitClause *C) {
return true;
}
+template <typename Derived>
+bool RecursiveASTVisitor<Derived>::VisitOMPReplayableClause(OMPReplayableClause *C) {
+ TRY_TO(TraverseStmt(C->getCondition()));
+ return true;
+}
+
template <typename Derived>
bool RecursiveASTVisitor<Derived>::VisitOMPUntiedClause(OMPUntiedClause *) {
return true;
diff --git a/clang/include/clang/Sema/SemaOpenMP.h b/clang/include/clang/Sema/SemaOpenMP.h
index 7c500847881f0..98405b871641a 100644
--- a/clang/include/clang/Sema/SemaOpenMP.h
+++ b/clang/include/clang/Sema/SemaOpenMP.h
@@ -1173,6 +1173,12 @@ class SemaOpenMP : public SemaBase {
OMPClause *ActOnOpenMPSelfMapsClause(SourceLocation StartLoc,
SourceLocation EndLoc);
+ /// Called on well-formed 'replayable' clause.
+ OMPClause *ActOnOpenMPReplayableClause(SourceLocation StartLoc,
+ SourceLocation EndLoc,
+ SourceLocation LParenLoc,
+ Expr *Condition);
+
/// Called on well-formed 'at' clause.
OMPClause *ActOnOpenMPAtClause(OpenMPAtClauseKind Kind,
SourceLocation KindLoc,
diff --git a/clang/lib/AST/OpenMPClause.cpp b/clang/lib/AST/OpenMPClause.cpp
index db4d5519acb38..7d2b4aa64a1df 100644
--- a/clang/lib/AST/OpenMPClause.cpp
+++ b/clang/lib/AST/OpenMPClause.cpp
@@ -324,6 +324,12 @@ OMPClause::child_range OMPNowaitClause::used_children() {
return children();
}
+OMPClause::child_range OMPReplayableClause::used_children() {
+ if (Condition)
+ return child_range(&Condition, &Condition + 1);
+ return children();
+}
+
OMPClause::child_range OMPGrainsizeClause::used_children() {
if (Stmt **C = getAddrOfExprAsWritten(getPreInitStmt()))
return child_range(C, C + 1);
@@ -2201,6 +2207,15 @@ void OMPClausePrinter::VisitOMPNowaitClause(OMPNowaitClause *Node) {
}
}
+void OMPClausePrinter::VisitOMPReplayableClause(OMPReplayableClause *Node) {
+ OS << "replayable";
+ if (auto *Cond = Node->getCondition()) {
+ OS << "(";
+ Cond->printPretty(OS, nullptr, Policy, 0);
+ OS << ")";
+ }
+}
+
void OMPClausePrinter::VisitOMPUntiedClause(OMPUntiedClause *) {
OS << "untied";
}
diff --git a/clang/lib/AST/StmtProfile.cpp b/clang/lib/AST/StmtProfile.cpp
index 39b4377ec1577..d8caf515d27d7 100644
--- a/clang/lib/AST/StmtProfile.cpp
+++ b/clang/lib/AST/StmtProfile.cpp
@@ -606,6 +606,11 @@ void OMPClauseProfiler::VisitOMPNowaitClause(const OMPNowaitClause *C) {
Profiler->VisitStmt(C->getCondition());
}
+void OMPClauseProfiler::VisitOMPReplayableClause(const OMPReplayableClause *C) {
+ if (C->getCondition())
+ Profiler->VisitStmt(C->getCondition());
+}
+
void OMPClauseProfiler::VisitOMPUntiedClause(const OMPUntiedClause *) {}
void OMPClauseProfiler::VisitOMPMergeableClause(const OMPMergeableClause *) {}
diff --git a/clang/lib/CodeGen/CGOpenMPRuntime.cpp b/clang/lib/CodeGen/CGOpenMPRuntime.cpp
index 5361e830dcab3..63e85f9e8d2a4 100644
--- a/clang/lib/CodeGen/CGOpenMPRuntime.cpp
+++ b/clang/lib/CodeGen/CGOpenMPRuntime.cpp
@@ -4789,6 +4789,7 @@ void CGOpenMPRuntime::emitTaskCall(CodeGenFunction &CGF, SourceLocation Loc,
llvm::Function *TaskFunction,
QualType SharedsTy, Address Shareds,
const Expr *IfCond,
+ const Expr *ReplayableCond,
const OMPTaskDataTy &Data) {
if (!CGF.HaveInsertPoint())
return;
@@ -4953,8 +4954,16 @@ void CGOpenMPRuntime::emitTaskCall(CodeGenFunction &CGF, SourceLocation Loc,
RegionCodeGenTy TaskgraphRCG(TaskgraphTaskCodeGen);
TaskgraphRCG(CGF);
} else {
- RegionCodeGenTy NonTaskgraphRCG(NonTaskgraphTaskCodeGen);
- NonTaskgraphRCG(CGF);
+ if (ReplayableCond) {
+ // We have a replayable clause. Task is replayable if its argument is
+ // omitted or evaluates to TRUE.
+ emitIfClause(CGF, ReplayableCond, TaskgraphTaskCodeGen,
+ NonTaskgraphTaskCodeGen);
+ } else {
+ // Not taskgraph, not replayable.
+ RegionCodeGenTy NonTaskgraphRCG(NonTaskgraphTaskCodeGen);
+ NonTaskgraphRCG(CGF);
+ }
}
}
@@ -4963,6 +4972,7 @@ void CGOpenMPRuntime::emitTaskLoopCall(CodeGenFunction &CGF, SourceLocation Loc,
llvm::Function *TaskFunction,
QualType SharedsTy, Address Shareds,
const Expr *IfCond,
+ const Expr *ReplayableCond,
const OMPTaskDataTy &Data) {
if (!CGF.HaveInsertPoint())
return;
@@ -5136,8 +5146,16 @@ void CGOpenMPRuntime::emitTaskLoopCall(CodeGenFunction &CGF, SourceLocation Loc,
RegionCodeGenTy TaskgraphRCG(TaskgraphTaskloopCodeGen);
TaskgraphRCG(CGF);
} else {
- RegionCodeGenTy NonTaskgraphRCG(NonTaskgraphTaskloopCodeGen);
- NonTaskgraphRCG(CGF);
+ if (ReplayableCond) {
+ // We have a replayable clause. Taskloop is replayable if its argument
+ // is omitted or evaluates to TRUE.
+ emitIfClause(CGF, ReplayableCond, TaskgraphTaskloopCodeGen,
+ NonTaskgraphTaskloopCodeGen);
+ } else {
+ // Not taskgraph, not replayable.
+ RegionCodeGenTy NonTaskgraphRCG(NonTaskgraphTaskloopCodeGen);
+ NonTaskgraphRCG(CGF);
+ }
}
}
@@ -6335,6 +6353,7 @@ Address CGOpenMPRuntime::getTaskReductionItem(CodeGenFunction &CGF,
}
void CGOpenMPRuntime::emitTaskwaitCall(CodeGenFunction &CGF, SourceLocation Loc,
+ const Expr *ReplayableCond,
const OMPTaskDataTy &Data) {
if (!CGF.HaveInsertPoint())
return;
@@ -6409,8 +6428,16 @@ void CGOpenMPRuntime::emitTaskwaitCall(CodeGenFunction &CGF, SourceLocation Loc,
RegionCodeGenTy TaskgraphRCG(TaskgraphTaskwaitCodeGen);
TaskgraphRCG(CGF);
} else {
- RegionCodeGenTy NonTaskgraphRCG(NonTaskgraphTaskwaitCodeGen);
- NonTaskgraphRCG(CGF);
+ if (ReplayableCond) {
+ // We have a replayable clause. Taskwait is replayable if its argument
+ // is omitted or evaluates to TRUE.
+ emitIfClause(CGF, ReplayableCond, TaskgraphTaskwaitCodeGen,
+ NonTaskgraphTaskwaitCodeGen);
+ } else {
+ // Not taskgraph, not replayable.
+ RegionCodeGenTy NonTaskgraphRCG(NonTaskgraphTaskwaitCodeGen);
+ NonTaskgraphRCG(CGF);
+ }
}
}
@@ -13568,6 +13595,7 @@ void CGOpenMPSIMDRuntime::emitTaskCall(CodeGenFunction &CGF, SourceLocation Loc,
llvm::Function *TaskFunction,
QualType SharedsTy, Address Shareds,
const Expr *IfCond,
+ const Expr *ReplayableCond,
const OMPTaskDataTy &Data) {
llvm_unreachable("Not supported in SIMD-only mode");
}
@@ -13575,7 +13603,7 @@ void CGOpenMPSIMDRuntime::emitTaskCall(CodeGenFunction &CGF, SourceLocation Loc,
void CGOpenMPSIMDRuntime::emitTaskLoopCall(
CodeGenFunction &CGF, SourceLocation Loc, const OMPLoopDirective &D,
llvm::Function *TaskFunction, QualType SharedsTy, Address Shareds,
- const Expr *IfCond, const OMPTaskDataTy &Data) {
+ const Expr *IfCond, const Expr *ReplayableCond, const OMPTaskDataTy &Data) {
llvm_unreachable("Not supported in SIMD-only mode");
}
@@ -13616,6 +13644,7 @@ Address CGOpenMPSIMDRuntime::getTaskReductionItem(CodeGenFunction &CGF,
void CGOpenMPSIMDRuntime::emitTaskwaitCall(CodeGenFunction &CGF,
SourceLocation Loc,
+ const Expr *ReplayableCond,
const OMPTaskDataTy &Data) {
llvm_unreachable("Not supported in SIMD-only mode");
}
diff --git a/clang/lib/CodeGen/CGOpenMPRuntime.h b/clang/lib/CodeGen/CGOpenMPRuntime.h
index 9a4de26418e45..7ac06547a5409 100644
--- a/clang/lib/CodeGen/CGOpenMPRuntime.h
+++ b/clang/lib/CodeGen/CGOpenMPRuntime.h
@@ -123,6 +123,7 @@ struct OMPTaskDataTy final {
bool IsWorksharingReduction = false;
bool HasNowaitClause = false;
bool HasModifier = false;
+ const Expr *ReplayableCond = nullptr;
};
/// Class intended to support codegen of all kind of the reduction clauses.
@@ -1177,6 +1178,7 @@ class CGOpenMPRuntime {
const OMPExecutableDirective &D,
llvm::Function *TaskFunction, QualType SharedsTy,
Address Shareds, const Expr *IfCond,
+ const Expr *ReplayableCond,
const OMPTaskDataTy &Data);
/// Emit task region for the taskloop directive. The taskloop region is
@@ -1212,7 +1214,8 @@ class CGOpenMPRuntime {
const OMPLoopDirective &D,
llvm::Function *TaskFunction,
QualType SharedsTy, Address Shareds,
- const Expr *IfCond, const OMPTaskDataTy &Data);
+ const Expr *IfCond, const Expr *ReplayableCond,
+ const OMPTaskDataTy &Data);
/// Emit code for the directive that does not require outlining.
///
@@ -1380,6 +1383,7 @@ class CGOpenMPRuntime {
/// Emit code for 'taskwait' directive.
virtual void emitTaskwaitCall(CodeGenFunction &CGF, SourceLocation Loc,
+ const Expr *ReplayableCond,
const OMPTaskDataTy &Data);
/// Emit code for 'taskgraph' directive.
@@ -2058,6 +2062,7 @@ class CGOpenMPSIMDRuntime final : public CGOpenMPRuntime {
const OMPExecutableDirective &D,
llvm::Function *TaskFunction, QualType SharedsTy,
Address Shareds, const Expr *IfCond,
+ const Expr *ReplayableCond,
const OMPTaskDataTy &Data) override;
/// Emit task region for the taskloop directive. The taskloop region is
@@ -2092,7 +2097,8 @@ class CGOpenMPSIMDRuntime final : public CGOpenMPRuntime {
void emitTaskLoopCall(CodeGenFunction &CGF, SourceLocation Loc,
const OMPLoopDirective &D, llvm::Function *TaskFunction,
QualType SharedsTy, Address Shareds, const Expr *IfCond,
- const OMPTaskDataTy &Data) override;
+ const Expr *ReplayableCond, const OMPTaskDataTy &Data)
+ override;
/// Emit a code for reduction clause. Next code should be emitted for
/// reduction:
@@ -2212,7 +2218,8 @@ class CGOpenMPSIMDRuntime final : public CGOpenMPRuntime {
/// Emit code for 'taskwait' directive.
void emitTaskwaitCall(CodeGenFunction &CGF, SourceLocation Loc,
- const OMPTaskDataTy &Data) override;
+ const Expr *ReplayableCond, const OMPTaskDataTy &Data)
+ override;
/// Emit code for 'taskgraph' directive.
/// \param IfCond Expression evaluated in if clause associated with the target
diff --git a/clang/lib/CodeGen/CGStmtOpenMP.cpp b/clang/lib/CodeGen/CGStmtOpenMP.cpp
index 65eb8177652a8..afcdc95237874 100644
--- a/clang/lib/CodeGen/CGStmtOpenMP.cpp
+++ b/clang/lib/CodeGen/CGStmtOpenMP.cpp
@@ -5721,8 +5721,20 @@ void CodeGenFunction::EmitOMPTargetTaskBasedDirective(
IntegerLiteral IfCond(getContext(), TrueOrFalse,
getContext().getIntTypeForBitwidth(32, /*Signed=*/0),
SourceLocation());
+ const Expr *ReplayableCond = nullptr;
+ if (auto *RC = S.getSingleClause<OMPReplayableClause>()) {
+ ReplayableCond = RC->getCondition();
+ if (!ReplayableCond) {
+ ReplayableCond =
+ IntegerLiteral::Create(
+ getContext(), llvm::APInt(32, 1),
+ getContext().getIntTypeForBitwidth(32, /*Signed=*/0),
+ SourceLocation());
+ }
+ }
CGM.getOpenMPRuntime().emitTaskCall(*this, S.getBeginLoc(), S, OutlinedFn,
- SharedsTy, CapturedStruct, &IfCond, Data);
+ SharedsTy, CapturedStruct, &IfCond,
+ ReplayableCond, Data);
}
void CodeGenFunction::processInReduction(const OMPExecutableDirective &S,
@@ -5831,15 +5843,27 @@ void CodeGenFunction::EmitOMPTaskDirective(const OMPTaskDirective &S) {
OMPTaskDataTy Data;
// Check if we should emit tied or untied task.
Data.Tied = !S.getSingleClause<OMPUntiedClause>();
+ const Expr *ReplayableCond = nullptr;
+ if (auto *RC = S.getSingleClause<OMPReplayableClause>()) {
+ ReplayableCond = RC->getCondition();
+ if (!ReplayableCond) {
+ ReplayableCond =
+ IntegerLiteral::Create(
+ getContext(), llvm::APInt(32, 1),
+ getContext().getIntTypeForBitwidth(32, /*Signed=*/0),
+ SourceLocation());
+ }
+ }
auto &&BodyGen = [CS](CodeGenFunction &CGF, PrePostActionTy &) {
CGF.EmitStmt(CS->getCapturedStmt());
};
auto &&TaskGen = [&S, SharedsTy, CapturedStruct,
- IfCond](CodeGenFunction &CGF, llvm::Function *OutlinedFn,
+ IfCond, ReplayableCond](CodeGenFunction &CGF,
+ llvm::Function *OutlinedFn,
const OMPTaskDataTy &Data) {
CGF.CGM.getOpenMPRuntime().emitTaskCall(CGF, S.getBeginLoc(), S, OutlinedFn,
SharedsTy, CapturedStruct, IfCond,
- Data);
+ ReplayableCond, Data);
};
auto LPCRegion =
CGOpenMPRuntime::LastprivateConditionalRAII::disable(*this, S);
@@ -5870,7 +5894,19 @@ void CodeGenFunction::EmitOMPTaskwaitDirective(const OMPTaskwaitDirective &S) {
// Build list of dependences
buildDependences(S, Data);
Data.HasNowaitClause = S.hasClausesOfKind<OMPNowaitClause>();
- CGM.getOpenMPRuntime().emitTaskwaitCall(*this, S.getBeginLoc(), Data);
+ const Expr *ReplayableCond = nullptr;
+ if (auto *RC = S.getSingleClause<OMPReplayableClause>()) {
+ ReplayableCond = RC->getCondition();
+ if (!ReplayableCond) {
+ ReplayableCond =
+ IntegerLiteral::Create(
+ getContext(), llvm::APInt(32, 1),
+ getContext().getIntTypeForBitwidth(32, /*Signed=*/0),
+ SourceLocation());
+ }
+ }
+ CGM.getOpenMPRuntime().emitTaskwaitCall(*this, S.getBeginLoc(),
+ ReplayableCond, Data);
}
void CodeGenFunction::EmitOMPTaskgraphDirective(
@@ -8225,6 +8261,18 @@ void CodeGenFunction::EmitOMPTaskLoopBasedDirective(const OMPLoopDirective &S) {
}
}
+ const Expr *ReplayableCond = nullptr;
+ if (auto *RC = S.getSingleClause<OMPReplayableClause>()) {
+ ReplayableCond = RC->getCondition();
+ if (!ReplayableCond) {
+ ReplayableCond =
+ IntegerLiteral::Create(
+ getContext(), llvm::APInt(32, 1),
+ getContext().getIntTypeForBitwidth(32, /*Signed=*/0),
+ SourceLocation());
+ }
+ }
+
OMPTaskDataTy Data;
// Check if taskloop must be emitted without taskgroup.
Data.Nogroup = S.getSingleClause<OMPNogroupClause>();
@@ -8345,16 +8393,17 @@ void CodeGenFunction::EmitOMPTaskLoopBasedDirective(const OMPLoopDirective &S) {
});
};
auto &&TaskGen =
- [&S, SharedsTy, CapturedStruct, IfCond]
+ [&S, SharedsTy, CapturedStruct, IfCond, ReplayableCond]
(CodeGenFunction &CGF, llvm::Function *OutlinedFn,
const OMPTaskDataTy &Data) {
auto &&CodeGen =
- [&S, OutlinedFn, SharedsTy, CapturedStruct, IfCond, &Data]
- (CodeGenFunction &CGF, PrePostActionTy &) {
+ [&S, OutlinedFn, SharedsTy, CapturedStruct, IfCond, ReplayableCond,
+ &Data](CodeGenFunction &CGF, PrePostActionTy &) {
OMPLoopScope PreInitScope(CGF, S);
CGF.CGM.getOpenMPRuntime().emitTaskLoopCall(CGF, S.getBeginLoc(), S,
OutlinedFn, SharedsTy,
- CapturedStruct, IfCond, Data);
+ CapturedStruct, IfCond,
+ ReplayableCond, Data);
};
CGF.CGM.getOpenMPRuntime().emitInlinedDirective(CGF, OMPD_taskloop,
CodeGen);
diff --git a/clang/lib/Parse/ParseOpenMP.cpp b/clang/lib/Parse/ParseOpenMP.cpp
index 5be99550f1ef6..d1394e1c62674 100644
--- a/clang/lib/Parse/ParseOpenMP.cpp
+++ b/clang/lib/Parse/ParseOpenMP.cpp
@@ -3386,13 +3386,15 @@ OMPClause *Parser::ParseOpenMPClause(OpenMPDirectiveKind DKind,
ErrorFound = true;
}
- if (CKind == OMPC_nowait && PP.LookAhead(/*N=*/0).is(tok::l_paren) &&
+ if (CKind == OMPC_nowait &&
+ PP.LookAhead(/*N=*/0).is(tok::l_paren) &&
getLangOpts().OpenMP >= 60)
Clause = ParseOpenMPSingleExprClause(CKind, WrongDirective);
else
Clause = ParseOpenMPClause(CKind, WrongDirective);
...
[truncated]
|
|
✅ With the latest revision this PR passed the C/C++ code formatter. |
🐧 Linux x64 Test Results
✅ The build succeeded and all tests passed. |
Created using spr 1.3.5
alexey-bataev
left a comment
alexey-bataev
left a comment
Member
There was a problem hiding this comment.
Tests? Docs updates (OpenMPSupport.rst and ReleaseNotes.rst)?
This was referenced May 5, 2026
Contributor
Author
Tests are in the next PR (#194054). I'll look at adjusting the docs. Thanks! |
Contributor
Author
|
Actually I think I'd say this falls under taskgraph support (already marked as "in progress" in OpenMPSupport.rst), and I don't think it is quite ready to be announced as a new feature under ReleaseNotes.rst, especially not specifically pertaining to this patch. As of the current state of the PR stack, taskgraph runtime support (and thus taskgraph support in general) is still hidden behind the LIBOMP_TASKGRAPH_EXPERIMENTAL cmake option -- and it should remain so at least until some outstanding problems with shared/firstprivate data in recorded tasks are resolved, IMO. (That's currently in progress.) |
Created using spr 1.3.5
Created using spr 1.3.5
jtb20
added a commit
that referenced
this pull request
Jun 23, 2026
This patch adds support for the 'replayable' clause to Clang. This is used to mark that the directive the clause is used on participates in taskgraph recording and replay: thus, when the clause is present, the new API routines introduced in a previous patch are emitted. Pull Request: #194053
jtb20
added a commit
that referenced
this pull request
Aug 14, 2026
This patch adds support for the 'replayable' clause to Clang. This is used to mark that the directive the clause is used on participates in taskgraph recording and replay: thus, when the clause is present, the new API routines introduced in a previous patch are emitted, even when the directive is not lexically nested within a taskgraph construct. Pull Request: #194053
Created using spr 1.3.7
jtb20
added a commit
that referenced
this pull request
Sep 23, 2026
This patch adds support for the 'replayable' clause to Clang. This is used to mark that the directive the clause is used on participates in taskgraph recording and replay: thus, when the clause is present, the new API routines introduced in a previous patch are emitted, even when the directive is not lexically nested within a taskgraph construct. Pull Request: #194053
Contributor
Author
|
Update (2026-10-29): an updated version of this patch is in the current series on the overview PR, #188765, and that's the version to read; this PR has not been updated and stays in draft. Please hold off reviewing the rest of the series until #216324 (New taskgraph runtime implementation) is approved. This PR is one patch of a stacked series adding OpenMP 6.0 The full stack, grouped by topic and in the order it should be reviewed and landed, is listed in the parent PR: #188765 That list is the single source of truth; any older stack listings in these PRs have been removed. |
jtb20
marked this pull request as draft
October 9, 2026 10:22
jtb20
added a commit
that referenced
this pull request
Oct 29, 2026
This patch adds support for the 'replayable' clause to Clang. This is
used to indicate whether the directive the clause is used on participates
in taskgraph recording and replay: thus, when the clause is present (and
its argument is true), the new API routines introduced in a previous
patch are emitted, even when the directive is not lexically nested
within a taskgraph construct. Conversely, a construct inside a taskgraph
region is replayable unless its 'replayable' clause says otherwise, so
'replayable(false)' there keeps it out of the recording. The clause is
accepted on the target-family compound directives as well ('target teams',
'target parallel for' and the rest), as OpenMP 6.0 allows.
Tests have been added to cover parsing and diagnostics, including
interaction with reductions.
Assisted-By: Codex
Assisted-By: Claude
Pull Request: #194053
This branch has not been deployed
This file contains hidden or bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
Sign up for free
to join this conversation on GitHub.
Already have an account?
Sign in to comment
Add this suggestion to a batch that can be applied as a single commit.This suggestion is invalid because no changes were made to the code.Suggestions cannot be applied while the pull request is closed.Suggestions cannot be applied while viewing a subset of changes.Only one suggestion per line can be applied in a batch.Add this suggestion to a batch that can be applied as a single commit.Applying suggestions on deleted lines is not supported.You must change the existing code in this line in order to create a valid suggestion.Outdated suggestions cannot be applied.This suggestion has been applied or marked resolved.Suggestions cannot be applied from pending reviews.Suggestions cannot be applied on multi-line comments.Suggestions cannot be applied while the pull request is queued to merge.Suggestion cannot be applied right now. Please check back later.
This patch adds support for the 'replayable' clause to Clang. This is
used to mark that the directive the clause is used on participates in
taskgraph recording and replay: thus, when the clause is present, the
new API routines introduced in a previous patch are emitted, even when
the directive is not lexically nested within a taskgraph construct.