[clang] [clang][OpenMP] Move "loop" directive mapping from sema to codegen (PR #99905)
via cfe-commits
cfe-commits at lists.llvm.org
Mon Jul 22 10:29:01 PDT 2024
llvmbot wrote:
<!--LLVM PR SUMMARY COMMENT-->
@llvm/pr-subscribers-clang-codegen
Author: Krzysztof Parzyszek (kparzysz)
<details>
<summary>Changes</summary>
Given "loop" construct, clang will try to treat it as "for", "distribute" or "simd", depending on either the implied binding, or the bind clause if present. This patch moves the code that performs this construct remapping from sema to codegen.
For a "loop" construct without a bind clause, this patch will create an implicit bind clause based on implied binding to simplify further analysis.
During codegen the function `EmitOMPGenericLoopDirective` (i.e. "loop") will invoke the "emit" functions for "for", "distribute" or "simd", depending on the bind clause.
---
Patch is 91.90 KiB, truncated to 20.00 KiB below, full version: https://github.com/llvm/llvm-project/pull/99905.diff
11 Files Affected:
- (modified) clang/include/clang/AST/StmtOpenMP.h (+3-21)
- (modified) clang/include/clang/Sema/SemaOpenMP.h (+1-22)
- (modified) clang/lib/AST/StmtOpenMP.cpp (+3-7)
- (modified) clang/lib/CodeGen/CGStmtOpenMP.cpp (+213-135)
- (modified) clang/lib/Sema/SemaOpenMP.cpp (+48-174)
- (modified) clang/lib/Sema/TreeTransform.h (+3-6)
- (modified) clang/lib/Serialization/ASTReaderStmt.cpp (-1)
- (modified) clang/lib/Serialization/ASTWriterStmt.cpp (-1)
- (modified) clang/test/OpenMP/generic_loop_ast_print.cpp (+19-14)
- (modified) clang/test/OpenMP/generic_loop_codegen.cpp (+80-70)
- (modified) clang/test/PCH/pragma-loop.cpp (+3-3)
``````````diff
diff --git a/clang/include/clang/AST/StmtOpenMP.h b/clang/include/clang/AST/StmtOpenMP.h
index 194eb2d10dcb3..f313c480f9a08 100644
--- a/clang/include/clang/AST/StmtOpenMP.h
+++ b/clang/include/clang/AST/StmtOpenMP.h
@@ -281,15 +281,6 @@ class OMPExecutableDirective : public Stmt {
return Data->getClauses();
}
- /// Was this directive mapped from an another directive?
- /// e.g. 1) omp loop bind(parallel) is mapped to OMPD_for
- /// 2) omp loop bind(teams) is mapped to OMPD_distribute
- /// 3) omp loop bind(thread) is mapped to OMPD_simd
- /// It was necessary to note it down in the Directive because of
- /// clang::TreeTransform::TransformOMPExecutableDirective() pass in
- /// the frontend.
- OpenMPDirectiveKind PrevMappedDirective = llvm::omp::OMPD_unknown;
-
protected:
/// Data, associated with the directive.
OMPChildren *Data = nullptr;
@@ -354,10 +345,6 @@ class OMPExecutableDirective : public Stmt {
return Inst;
}
- void setMappedDirective(OpenMPDirectiveKind MappedDirective) {
- PrevMappedDirective = MappedDirective;
- }
-
public:
/// Iterates over expressions/statements used in the construct.
class used_clauses_child_iterator
@@ -611,8 +598,6 @@ class OMPExecutableDirective : public Stmt {
"Expected directive with the associated statement.");
return Data->getRawStmt();
}
-
- OpenMPDirectiveKind getMappedDirective() const { return PrevMappedDirective; }
};
/// This represents '#pragma omp parallel' directive.
@@ -1620,8 +1605,7 @@ class OMPSimdDirective : public OMPLoopDirective {
SourceLocation EndLoc, unsigned CollapsedNum,
ArrayRef<OMPClause *> Clauses,
Stmt *AssociatedStmt,
- const HelperExprs &Exprs,
- OpenMPDirectiveKind ParamPrevMappedDirective);
+ const HelperExprs &Exprs);
/// Creates an empty directive with the place
/// for \a NumClauses clauses.
@@ -1699,8 +1683,7 @@ class OMPForDirective : public OMPLoopDirective {
SourceLocation EndLoc, unsigned CollapsedNum,
ArrayRef<OMPClause *> Clauses,
Stmt *AssociatedStmt, const HelperExprs &Exprs,
- Expr *TaskRedRef, bool HasCancel,
- OpenMPDirectiveKind ParamPrevMappedDirective);
+ Expr *TaskRedRef, bool HasCancel);
/// Creates an empty directive with the place
/// for \a NumClauses clauses.
@@ -4478,8 +4461,7 @@ class OMPDistributeDirective : public OMPLoopDirective {
static OMPDistributeDirective *
Create(const ASTContext &C, SourceLocation StartLoc, SourceLocation EndLoc,
unsigned CollapsedNum, ArrayRef<OMPClause *> Clauses,
- Stmt *AssociatedStmt, const HelperExprs &Exprs,
- OpenMPDirectiveKind ParamPrevMappedDirective);
+ Stmt *AssociatedStmt, const HelperExprs &Exprs);
/// Creates an empty directive with the place
/// for \a NumClauses clauses.
diff --git a/clang/include/clang/Sema/SemaOpenMP.h b/clang/include/clang/Sema/SemaOpenMP.h
index 54d81f91ffebc..aa61dae9415e2 100644
--- a/clang/include/clang/Sema/SemaOpenMP.h
+++ b/clang/include/clang/Sema/SemaOpenMP.h
@@ -398,8 +398,7 @@ class SemaOpenMP : public SemaBase {
StmtResult ActOnOpenMPExecutableDirective(
OpenMPDirectiveKind Kind, const DeclarationNameInfo &DirName,
OpenMPDirectiveKind CancelRegion, ArrayRef<OMPClause *> Clauses,
- Stmt *AStmt, SourceLocation StartLoc, SourceLocation EndLoc,
- OpenMPDirectiveKind PrevMappedDirective = llvm::omp::OMPD_unknown);
+ Stmt *AStmt, SourceLocation StartLoc, SourceLocation EndLoc);
/// Called on well-formed '\#pragma omp parallel' after parsing
/// of the associated statement.
StmtResult ActOnOpenMPParallelDirective(ArrayRef<OMPClause *> Clauses,
@@ -1430,26 +1429,6 @@ class SemaOpenMP : public SemaBase {
/// All `omp assumes` we encountered so far.
SmallVector<OMPAssumeAttr *, 4> OMPAssumeGlobal;
-
- /// OMPD_loop is mapped to OMPD_for, OMPD_distribute or OMPD_simd depending
- /// on the parameter of the bind clause. In the methods for the
- /// mapped directives, check the parameters of the lastprivate clause.
- bool checkLastPrivateForMappedDirectives(ArrayRef<OMPClause *> Clauses);
- /// Depending on the bind clause of OMPD_loop map the directive to new
- /// directives.
- /// 1) loop bind(parallel) --> OMPD_for
- /// 2) loop bind(teams) --> OMPD_distribute
- /// 3) loop bind(thread) --> OMPD_simd
- /// This is being handled in Sema instead of Codegen because of the need for
- /// rigorous semantic checking in the new mapped directives.
- bool mapLoopConstruct(llvm::SmallVector<OMPClause *> &ClausesWithoutBind,
- ArrayRef<OMPClause *> Clauses,
- OpenMPBindClauseKind &BindKind,
- OpenMPDirectiveKind &Kind,
- OpenMPDirectiveKind &PrevMappedDirective,
- SourceLocation StartLoc, SourceLocation EndLoc,
- const DeclarationNameInfo &DirName,
- OpenMPDirectiveKind CancelRegion);
};
} // namespace clang
diff --git a/clang/lib/AST/StmtOpenMP.cpp b/clang/lib/AST/StmtOpenMP.cpp
index a2325b177d41e..525d079da2670 100644
--- a/clang/lib/AST/StmtOpenMP.cpp
+++ b/clang/lib/AST/StmtOpenMP.cpp
@@ -300,7 +300,7 @@ OMPParallelDirective *OMPParallelDirective::CreateEmpty(const ASTContext &C,
OMPSimdDirective *OMPSimdDirective::Create(
const ASTContext &C, SourceLocation StartLoc, SourceLocation EndLoc,
unsigned CollapsedNum, ArrayRef<OMPClause *> Clauses, Stmt *AssociatedStmt,
- const HelperExprs &Exprs, OpenMPDirectiveKind ParamPrevMappedDirective) {
+ const HelperExprs &Exprs) {
auto *Dir = createDirective<OMPSimdDirective>(
C, Clauses, AssociatedStmt, numLoopChildren(CollapsedNum, OMPD_simd),
StartLoc, EndLoc, CollapsedNum);
@@ -320,7 +320,6 @@ OMPSimdDirective *OMPSimdDirective::Create(
Dir->setDependentInits(Exprs.DependentInits);
Dir->setFinalsConditions(Exprs.FinalsConditions);
Dir->setPreInits(Exprs.PreInits);
- Dir->setMappedDirective(ParamPrevMappedDirective);
return Dir;
}
@@ -336,8 +335,7 @@ OMPSimdDirective *OMPSimdDirective::CreateEmpty(const ASTContext &C,
OMPForDirective *OMPForDirective::Create(
const ASTContext &C, SourceLocation StartLoc, SourceLocation EndLoc,
unsigned CollapsedNum, ArrayRef<OMPClause *> Clauses, Stmt *AssociatedStmt,
- const HelperExprs &Exprs, Expr *TaskRedRef, bool HasCancel,
- OpenMPDirectiveKind ParamPrevMappedDirective) {
+ const HelperExprs &Exprs, Expr *TaskRedRef, bool HasCancel) {
auto *Dir = createDirective<OMPForDirective>(
C, Clauses, AssociatedStmt, numLoopChildren(CollapsedNum, OMPD_for) + 1,
StartLoc, EndLoc, CollapsedNum);
@@ -367,7 +365,6 @@ OMPForDirective *OMPForDirective::Create(
Dir->setPreInits(Exprs.PreInits);
Dir->setTaskReductionRefExpr(TaskRedRef);
Dir->setHasCancel(HasCancel);
- Dir->setMappedDirective(ParamPrevMappedDirective);
return Dir;
}
@@ -1572,7 +1569,7 @@ OMPParallelMaskedTaskLoopSimdDirective::CreateEmpty(const ASTContext &C,
OMPDistributeDirective *OMPDistributeDirective::Create(
const ASTContext &C, SourceLocation StartLoc, SourceLocation EndLoc,
unsigned CollapsedNum, ArrayRef<OMPClause *> Clauses, Stmt *AssociatedStmt,
- const HelperExprs &Exprs, OpenMPDirectiveKind ParamPrevMappedDirective) {
+ const HelperExprs &Exprs) {
auto *Dir = createDirective<OMPDistributeDirective>(
C, Clauses, AssociatedStmt,
numLoopChildren(CollapsedNum, OMPD_distribute), StartLoc, EndLoc,
@@ -1601,7 +1598,6 @@ OMPDistributeDirective *OMPDistributeDirective::Create(
Dir->setDependentInits(Exprs.DependentInits);
Dir->setFinalsConditions(Exprs.FinalsConditions);
Dir->setPreInits(Exprs.PreInits);
- Dir->setMappedDirective(ParamPrevMappedDirective);
return Dir;
}
diff --git a/clang/lib/CodeGen/CGStmtOpenMP.cpp b/clang/lib/CodeGen/CGStmtOpenMP.cpp
index adf74ea16c895..2e83213fa03e1 100644
--- a/clang/lib/CodeGen/CGStmtOpenMP.cpp
+++ b/clang/lib/CodeGen/CGStmtOpenMP.cpp
@@ -44,6 +44,8 @@ using namespace llvm::omp;
#define TTL_CODEGEN_TYPE "target-teams-loop-codegen"
static const VarDecl *getBaseDecl(const Expr *Ref);
+static OpenMPDirectiveKind
+getEffectiveDirectiveKind(const OMPExecutableDirective &S);
namespace {
/// Lexical scope for OpenMP executable constructs, that handles correct codegen
@@ -111,10 +113,10 @@ class OMPLexicalScope : public CodeGenFunction::LexicalScope {
/// for captured expressions.
class OMPParallelScope final : public OMPLexicalScope {
bool EmitPreInitStmt(const OMPExecutableDirective &S) {
- OpenMPDirectiveKind Kind = S.getDirectiveKind();
- return !(isOpenMPTargetExecutionDirective(Kind) ||
- isOpenMPLoopBoundSharingDirective(Kind)) &&
- isOpenMPParallelDirective(Kind);
+ OpenMPDirectiveKind EKind = getEffectiveDirectiveKind(S);
+ return !(isOpenMPTargetExecutionDirective(EKind) ||
+ isOpenMPLoopBoundSharingDirective(EKind)) &&
+ isOpenMPParallelDirective(EKind);
}
public:
@@ -127,9 +129,9 @@ class OMPParallelScope final : public OMPLexicalScope {
/// for captured expressions.
class OMPTeamsScope final : public OMPLexicalScope {
bool EmitPreInitStmt(const OMPExecutableDirective &S) {
- OpenMPDirectiveKind Kind = S.getDirectiveKind();
- return !isOpenMPTargetExecutionDirective(Kind) &&
- isOpenMPTeamsDirective(Kind);
+ OpenMPDirectiveKind EKind = getEffectiveDirectiveKind(S);
+ return !isOpenMPTargetExecutionDirective(EKind) &&
+ isOpenMPTeamsDirective(EKind);
}
public:
@@ -268,7 +270,7 @@ class OMPSimdLexicalScope : public CodeGenFunction::LexicalScope {
}
}
}
- if (!isOpenMPSimdDirective(S.getDirectiveKind()))
+ if (!isOpenMPSimdDirective(getEffectiveDirectiveKind(S)))
CGF.EmitOMPPrivateClause(S, InlinedShareds);
if (const auto *TG = dyn_cast<OMPTaskgroupDirective>(&S)) {
if (const Expr *E = TG->getReductionRef())
@@ -309,6 +311,30 @@ class OMPSimdLexicalScope : public CodeGenFunction::LexicalScope {
} // namespace
+// The loop directive with a bind clause will be mapped to a different
+// directive with corresponding semantics.
+static OpenMPDirectiveKind
+getEffectiveDirectiveKind(const OMPExecutableDirective &S) {
+ OpenMPDirectiveKind Kind = S.getDirectiveKind();
+ if (Kind != OMPD_loop)
+ return Kind;
+
+ OpenMPBindClauseKind BindKind = OMPC_BIND_unknown;
+ if (const auto *C = S.getSingleClause<OMPBindClause>())
+ BindKind = C->getBindKind();
+
+ switch (BindKind) {
+ case OMPC_BIND_parallel:
+ return OMPD_for;
+ case OMPC_BIND_teams:
+ return OMPD_distribute;
+ case OMPC_BIND_thread:
+ return OMPD_simd;
+ default:
+ return OMPD_loop;
+ }
+}
+
static void emitCommonOMPTargetDirective(CodeGenFunction &CGF,
const OMPExecutableDirective &S,
const RegionCodeGenTy &CodeGen);
@@ -825,9 +851,9 @@ bool CodeGenFunction::EmitOMPFirstprivateClause(const OMPExecutableDirective &D,
OMPPrivateScope &PrivateScope) {
if (!HaveInsertPoint())
return false;
- bool DeviceConstTarget =
- getLangOpts().OpenMPIsTargetDevice &&
- isOpenMPTargetExecutionDirective(D.getDirectiveKind());
+ OpenMPDirectiveKind EKind = getEffectiveDirectiveKind(D);
+ bool DeviceConstTarget = getLangOpts().OpenMPIsTargetDevice &&
+ isOpenMPTargetExecutionDirective(EKind);
bool FirstprivateIsLastprivate = false;
llvm::DenseMap<const VarDecl *, OpenMPLastprivateModifier> Lastprivates;
for (const auto *C : D.getClausesOfKind<OMPLastprivateClause>()) {
@@ -838,7 +864,7 @@ bool CodeGenFunction::EmitOMPFirstprivateClause(const OMPExecutableDirective &D,
}
llvm::DenseSet<const VarDecl *> EmittedAsFirstprivate;
llvm::SmallVector<OpenMPDirectiveKind, 4> CaptureRegions;
- getOpenMPCaptureRegions(CaptureRegions, D.getDirectiveKind());
+ getOpenMPCaptureRegions(CaptureRegions, EKind);
// Force emission of the firstprivate copy if the directive does not emit
// outlined function, like omp for, omp simd, omp distribute etc.
bool MustEmitFirstprivateCopy =
@@ -1067,8 +1093,9 @@ bool CodeGenFunction::EmitOMPLastprivateClauseInit(
if (!HaveInsertPoint())
return false;
bool HasAtLeastOneLastprivate = false;
+ OpenMPDirectiveKind EKind = getEffectiveDirectiveKind(D);
llvm::DenseSet<const VarDecl *> SIMDLCVs;
- if (isOpenMPSimdDirective(D.getDirectiveKind())) {
+ if (isOpenMPSimdDirective(EKind)) {
const auto *LoopDirective = cast<OMPLoopDirective>(&D);
for (const Expr *C : LoopDirective->counters()) {
SIMDLCVs.insert(
@@ -1078,7 +1105,7 @@ bool CodeGenFunction::EmitOMPLastprivateClauseInit(
llvm::DenseSet<const VarDecl *> AlreadyEmittedVars;
for (const auto *C : D.getClausesOfKind<OMPLastprivateClause>()) {
HasAtLeastOneLastprivate = true;
- if (isOpenMPTaskLoopDirective(D.getDirectiveKind()) &&
+ if (isOpenMPTaskLoopDirective(EKind) &&
!getLangOpts().OpenMPSimd)
break;
const auto *IRef = C->varlist_begin();
@@ -1312,13 +1339,13 @@ void CodeGenFunction::EmitOMPReductionClauseInit(
++Count;
}
if (!Data.ReductionVars.empty()) {
+ OpenMPDirectiveKind EKind = getEffectiveDirectiveKind(D);
Data.IsReductionWithTaskMod = true;
- Data.IsWorksharingReduction =
- isOpenMPWorksharingDirective(D.getDirectiveKind());
+ Data.IsWorksharingReduction = isOpenMPWorksharingDirective(EKind);
llvm::Value *ReductionDesc = CGM.getOpenMPRuntime().emitTaskReductionInit(
*this, D.getBeginLoc(), TaskLHSs, TaskRHSs, Data);
const Expr *TaskRedRef = nullptr;
- switch (D.getDirectiveKind()) {
+ switch (EKind) {
case OMPD_parallel:
TaskRedRef = cast<OMPParallelDirective>(D).getTaskReductionRefExpr();
break;
@@ -1449,16 +1476,16 @@ void CodeGenFunction::EmitOMPReductionClauseFinal(
IsReductionWithTaskMod || C->getModifier() == OMPC_REDUCTION_task;
}
if (HasAtLeastOneReduction) {
+ OpenMPDirectiveKind EKind = getEffectiveDirectiveKind(D);
if (IsReductionWithTaskMod) {
CGM.getOpenMPRuntime().emitTaskReductionFini(
- *this, D.getBeginLoc(),
- isOpenMPWorksharingDirective(D.getDirectiveKind()));
+ *this, D.getBeginLoc(), isOpenMPWorksharingDirective(EKind));
}
bool TeamsLoopCanBeParallel = false;
if (auto *TTLD = dyn_cast<OMPTargetTeamsGenericLoopDirective>(&D))
TeamsLoopCanBeParallel = TTLD->canBeParallelFor();
bool WithNowait = D.getSingleClause<OMPNowaitClause>() ||
- isOpenMPParallelDirective(D.getDirectiveKind()) ||
+ isOpenMPParallelDirective(EKind) ||
TeamsLoopCanBeParallel || ReductionKind == OMPD_simd;
bool SimpleReduction = ReductionKind == OMPD_simd;
// Emit nowait reduction if nowait clause is present or directive is a
@@ -1915,7 +1942,8 @@ void CodeGenFunction::EmitOMPLoopBody(const OMPLoopDirective &D,
// Update the linear variables.
// In distribute directives only loop counters may be marked as linear, no
// need to generate the code for them.
- if (!isOpenMPDistributeDirective(D.getDirectiveKind())) {
+ OpenMPDirectiveKind EKind = getEffectiveDirectiveKind(D);
+ if (!isOpenMPDistributeDirective(EKind)) {
for (const auto *C : D.getClausesOfKind<OMPLinearClause>()) {
for (const Expr *UE : C->updates())
EmitIgnoredExpr(UE);
@@ -1949,7 +1977,7 @@ void CodeGenFunction::EmitOMPLoopBody(const OMPLoopDirective &D,
OMPAfterScanBlock = createBasicBlock("omp.after.scan.bb");
// No need to allocate inscan exit block, in simd mode it is selected in the
// codegen for the scan directive.
- if (D.getDirectiveKind() != OMPD_simd && !getLangOpts().OpenMPSimd)
+ if (EKind != OMPD_simd && !getLangOpts().OpenMPSimd)
OMPScanExitBlock = createBasicBlock("omp.exit.inscan.bb");
OMPScanDispatch = createBasicBlock("omp.inscan.dispatch");
EmitBranch(OMPScanDispatch);
@@ -2362,7 +2390,8 @@ void CodeGenFunction::EmitOMPLinearClause(
if (!HaveInsertPoint())
return;
llvm::DenseSet<const VarDecl *> SIMDLCVs;
- if (isOpenMPSimdDirective(D.getDirectiveKind())) {
+ OpenMPDirectiveKind EKind = getEffectiveDirectiveKind(D);
+ if (isOpenMPSimdDirective(EKind)) {
const auto *LoopDirective = cast<OMPLoopDirective>(&D);
for (const Expr *C : LoopDirective->counters()) {
SIMDLCVs.insert(
@@ -2424,9 +2453,9 @@ void CodeGenFunction::EmitOMPSimdInit(const OMPLoopDirective &D) {
if (const auto *C = D.getSingleClause<OMPOrderClause>())
if (C->getKind() == OMPC_ORDER_concurrent)
LoopStack.setParallel(/*Enable=*/true);
- if ((D.getDirectiveKind() == OMPD_simd ||
- (getLangOpts().OpenMPSimd &&
- isOpenMPSimdDirective(D.getDirectiveKind()))) &&
+ OpenMPDirectiveKind EKind = getEffectiveDirectiveKind(D);
+ if ((EKind == OMPD_simd ||
+ (getLangOpts().OpenMPSimd && isOpenMPSimdDirective(EKind))) &&
llvm::any_of(D.getClausesOfKind<OMPReductionClause>(),
[](const OMPReductionClause *C) {
return C->getModifier() == OMPC_REDUCTION_inscan;
@@ -2513,7 +2542,8 @@ static void emitCommonSimdLoop(CodeGenFunction &CGF, const OMPLoopDirective &S,
BodyCodeGen(CGF);
};
const Expr *IfCond = nullptr;
- if (isOpenMPSimdDirective(S.getDirectiveKind())) {
+ OpenMPDirectiveKind EKind = getEffectiveDirectiveKind(S);
+ if (isOpenMPSimdDirective(EKind)) {
for (const auto *C : S.getClausesOfKind<OMPIfClause>()) {
if (CGF.getLangOpts().OpenMP >= 50 &&
(C->getNameModifier() == OMPD_unknown ||
@@ -2534,21 +2564,24 @@ static void emitCommonSimdLoop(CodeGenFunction &CGF, const OMPLoopDirective &S,
static void emitOMPSimdRegion(CodeGenFunction &CGF, const OMPLoopDirective &S,
PrePostActionTy &Action) {
Action.Enter(CGF);
- assert(isOpenMPSimdDirective(S.getDirectiveKind()) &&
- "Expected simd directive");
OMPLoopScope PreInitScope(CGF, S);
// if (PreCond) {
// for (IV in 0..LastIteration) BODY;
// <Final counter/linear vars updates>;
// }
- //
- if (isOpenMPDistributeDirective(S.getDirectiveKind()) ||
- isOpenMPWorksharingDirective(S.getDirectiveKind()) ||
- isOpenMPTaskLoopDirective(S.getDirectiveKind())) {
- (void)EmitOMPHelperVar(CGF, cast<DeclRefExpr>(S.getLowerBoundVariable()));
- (void)EmitOMPHelperVar(CGF, cast<DeclRefExpr>(S.getUpperBoundVariable()));
+
+ // The presence of lower/upper bound variable depends on the actual directive
+ // kind in the AST node. The variables must be emitted because some of the
+ // expressions associated with the loop will use them.
+ OpenMPDirectiveKind DKind = S.getDirectiveKind();
+ if (isOpenMPDistributeDirective(DKind) ||
+ isOpenMPWorksharingDirective(DKind) || isOpenMPTaskLoopDirective(DKind) ||
+ isOpenMPGenericLoopDirective(DKind)) {
+ EmitOMPHelperVar(CGF, cast<DeclRefExpr>(S.getLowerBoundVariable()));
+ EmitOMPHelperVar(CGF, cast<DeclRefExpr>(S.getUpperBoundVariable()));
}
+ OpenMPDirectiveKind EKind = getEffectiveDirectiveKind(S);
// Emit: if (PreCond) - begin.
// If the condition constant folds and can be elided, avoid emitting the
// whole loop.
@@ -2593,7 +2626,7 @@ static void emitOMPSimdRegion(CodeGenFunction &CGF, const OMPLoopDirective &S,
CGF, S, CGF.EmitLValue(S.getIterationVariable()));
bool HasLastprivateClause = CGF.EmitOMPLastprivateClauseInit(S, LoopScope);
(void)LoopScope.Privatize();
- if (isOpenMPTargetExecutionDirective(S.getDirectiveKind()))
+ if (isOpenMPTargetExecutionDirective(EKind))
CGF.CGM.getOpenMPRuntime().adjustTargetSpecificDataForLambdas(CGF, S);
emitCommonSimdLoop(
@@ -2627,7 +2660,9 @@ static void emitOMPSimdRegion(CodeGenFunction &CGF, const OMPLo...
[truncated]
``````````
</details>
https://github.com/llvm/llvm-project/pull/99905
More information about the cfe-commits
mailing list