[clang] 9afb651 - Adding support for iterator in motion clauses. (#159112)
via cfe-commits
cfe-commits at lists.llvm.org
Mon Dec 1 00:33:36 PST 2025
Author: ShashwathiNavada
Date: 2025-12-01T14:03:32+05:30
New Revision: 9afb651613a9383923b0f52885fb2221a5ec134f
URL: https://github.com/llvm/llvm-project/commit/9afb651613a9383923b0f52885fb2221a5ec134f
DIFF: https://github.com/llvm/llvm-project/commit/9afb651613a9383923b0f52885fb2221a5ec134f.diff
LOG: Adding support for iterator in motion clauses. (#159112)
As described in section 2.14.6 of openmp spec, the patch implements
support for iterator in motion clauses.
---------
Co-authored-by: Shashwathi N <nshashwa at pe31.hpc.amslabs.hpecorp.net>
Added:
clang/test/OpenMP/target_update_iterator_ast_print.cpp
clang/test/OpenMP/target_update_iterator_serialization.cpp
Modified:
clang/docs/OpenMPSupport.rst
clang/include/clang/AST/OpenMPClause.h
clang/include/clang/Basic/OpenMPKinds.def
clang/include/clang/Sema/SemaOpenMP.h
clang/lib/AST/OpenMPClause.cpp
clang/lib/CodeGen/CGOpenMPRuntime.cpp
clang/lib/Parse/ParseOpenMP.cpp
clang/lib/Sema/SemaOpenMP.cpp
clang/lib/Sema/TreeTransform.h
clang/lib/Serialization/ASTReader.cpp
clang/lib/Serialization/ASTWriter.cpp
clang/test/OpenMP/target_update_codegen.cpp
Removed:
################################################################################
diff --git a/clang/docs/OpenMPSupport.rst b/clang/docs/OpenMPSupport.rst
index e7ca7b0bd0792..ab3f2c48983ca 100644
--- a/clang/docs/OpenMPSupport.rst
+++ b/clang/docs/OpenMPSupport.rst
@@ -266,7 +266,7 @@ implementation.
+------------------------------+--------------------------------------------------------------+--------------------------+-----------------------------------------------------------------------+
| device | has_device_addr clause on target construct | :none:`unclaimed` | |
+------------------------------+--------------------------------------------------------------+--------------------------+-----------------------------------------------------------------------+
-| device | iterators in map clause or motion clauses | :none:`unclaimed` | |
+| device | iterators in map clause or motion clauses | :none:`done` | https://github.com/llvm/llvm-project/pull/159112 |
+------------------------------+--------------------------------------------------------------+--------------------------+-----------------------------------------------------------------------+
| device | indirect clause on declare target directive | :part:`In Progress` | |
+------------------------------+--------------------------------------------------------------+--------------------------+-----------------------------------------------------------------------+
diff --git a/clang/include/clang/AST/OpenMPClause.h b/clang/include/clang/AST/OpenMPClause.h
index 72c5efde7449b..d9c3cf239451e 100644
--- a/clang/include/clang/AST/OpenMPClause.h
+++ b/clang/include/clang/AST/OpenMPClause.h
@@ -7582,7 +7582,8 @@ class OMPToClause final : public OMPMappableExprListClause<OMPToClause>,
/// Motion-modifiers for the 'to' clause.
OpenMPMotionModifierKind MotionModifiers[NumberOfOMPMotionModifiers] = {
- OMPC_MOTION_MODIFIER_unknown, OMPC_MOTION_MODIFIER_unknown};
+ OMPC_MOTION_MODIFIER_unknown, OMPC_MOTION_MODIFIER_unknown,
+ OMPC_MOTION_MODIFIER_unknown};
/// Location of motion-modifiers for the 'to' clause.
SourceLocation MotionModifiersLoc[NumberOfOMPMotionModifiers];
@@ -7654,6 +7655,9 @@ class OMPToClause final : public OMPMappableExprListClause<OMPToClause>,
MotionModifiersLoc[I] = TLoc;
}
+ void setIteratorModifier(Expr *IteratorModifier) {
+ getTrailingObjects<Expr *>()[2 * varlist_size()] = IteratorModifier;
+ }
/// Set colon location.
void setColonLoc(SourceLocation Loc) { ColonLoc = Loc; }
@@ -7662,7 +7666,7 @@ class OMPToClause final : public OMPMappableExprListClause<OMPToClause>,
size_t numTrailingObjects(OverloadToken<Expr *>) const {
// There are varlist_size() of expressions, and varlist_size() of
// user-defined mappers.
- return 2 * varlist_size();
+ return 2 * varlist_size() + 1;
}
size_t numTrailingObjects(OverloadToken<ValueDecl *>) const {
return getUniqueDeclarationsNum();
@@ -7688,15 +7692,14 @@ class OMPToClause final : public OMPMappableExprListClause<OMPToClause>,
/// \param UDMQualifierLoc C++ nested name specifier for the associated
/// user-defined mapper.
/// \param MapperId The identifier of associated user-defined mapper.
- static OMPToClause *Create(const ASTContext &C, const OMPVarListLocTy &Locs,
- ArrayRef<Expr *> Vars,
- ArrayRef<ValueDecl *> Declarations,
- MappableExprComponentListsRef ComponentLists,
- ArrayRef<Expr *> UDMapperRefs,
- ArrayRef<OpenMPMotionModifierKind> MotionModifiers,
- ArrayRef<SourceLocation> MotionModifiersLoc,
- NestedNameSpecifierLoc UDMQualifierLoc,
- DeclarationNameInfo MapperId);
+ static OMPToClause *
+ Create(const ASTContext &C, const OMPVarListLocTy &Locs,
+ ArrayRef<Expr *> Vars, ArrayRef<ValueDecl *> Declarations,
+ MappableExprComponentListsRef ComponentLists,
+ ArrayRef<Expr *> UDMapperRefs, Expr *IteratorModifier,
+ ArrayRef<OpenMPMotionModifierKind> MotionModifiers,
+ ArrayRef<SourceLocation> MotionModifiersLoc,
+ NestedNameSpecifierLoc UDMQualifierLoc, DeclarationNameInfo MapperId);
/// Creates an empty clause with the place for \a NumVars variables.
///
@@ -7717,7 +7720,9 @@ class OMPToClause final : public OMPMappableExprListClause<OMPToClause>,
"Requested modifier exceeds the total number of modifiers.");
return MotionModifiers[Cnt];
}
-
+ Expr *getIteratorModifier() const {
+ return getTrailingObjects<Expr *>()[2 * varlist_size()];
+ }
/// Fetches the motion-modifier location at 'Cnt' index of array of modifiers'
/// locations.
///
@@ -7782,7 +7787,8 @@ class OMPFromClause final
/// Motion-modifiers for the 'from' clause.
OpenMPMotionModifierKind MotionModifiers[NumberOfOMPMotionModifiers] = {
- OMPC_MOTION_MODIFIER_unknown, OMPC_MOTION_MODIFIER_unknown};
+ OMPC_MOTION_MODIFIER_unknown, OMPC_MOTION_MODIFIER_unknown,
+ OMPC_MOTION_MODIFIER_unknown};
/// Location of motion-modifiers for the 'from' clause.
SourceLocation MotionModifiersLoc[NumberOfOMPMotionModifiers];
@@ -7843,7 +7849,9 @@ class OMPFromClause final
"Unexpected index to store motion modifier, exceeds array size.");
MotionModifiers[I] = T;
}
-
+ void setIteratorModifier(Expr *IteratorModifier) {
+ getTrailingObjects<Expr *>()[2 * varlist_size()] = IteratorModifier;
+ }
/// Set location for the motion-modifier.
///
/// \param I index for motion-modifier location.
@@ -7862,7 +7870,7 @@ class OMPFromClause final
size_t numTrailingObjects(OverloadToken<Expr *>) const {
// There are varlist_size() of expressions, and varlist_size() of
// user-defined mappers.
- return 2 * varlist_size();
+ return 2 * varlist_size() + 1;
}
size_t numTrailingObjects(OverloadToken<ValueDecl *>) const {
return getUniqueDeclarationsNum();
@@ -7892,7 +7900,7 @@ class OMPFromClause final
Create(const ASTContext &C, const OMPVarListLocTy &Locs,
ArrayRef<Expr *> Vars, ArrayRef<ValueDecl *> Declarations,
MappableExprComponentListsRef ComponentLists,
- ArrayRef<Expr *> UDMapperRefs,
+ ArrayRef<Expr *> UDMapperRefs, Expr *IteratorExpr,
ArrayRef<OpenMPMotionModifierKind> MotionModifiers,
ArrayRef<SourceLocation> MotionModifiersLoc,
NestedNameSpecifierLoc UDMQualifierLoc, DeclarationNameInfo MapperId);
@@ -7916,7 +7924,9 @@ class OMPFromClause final
"Requested modifier exceeds the total number of modifiers.");
return MotionModifiers[Cnt];
}
-
+ Expr *getIteratorModifier() const {
+ return getTrailingObjects<Expr *>()[2 * varlist_size()];
+ }
/// Fetches the motion-modifier location at 'Cnt' index of array of modifiers'
/// locations.
///
diff --git a/clang/include/clang/Basic/OpenMPKinds.def b/clang/include/clang/Basic/OpenMPKinds.def
index b98b946cad75a..ceac89d3aba6d 100644
--- a/clang/include/clang/Basic/OpenMPKinds.def
+++ b/clang/include/clang/Basic/OpenMPKinds.def
@@ -207,6 +207,7 @@ OPENMP_MAP_MODIFIER_KIND(ompx_hold)
// Modifiers for 'to' or 'from' clause.
OPENMP_MOTION_MODIFIER_KIND(mapper)
+OPENMP_MOTION_MODIFIER_KIND(iterator)
OPENMP_MOTION_MODIFIER_KIND(present)
// Static attributes for 'dist_schedule' clause.
diff --git a/clang/include/clang/Sema/SemaOpenMP.h b/clang/include/clang/Sema/SemaOpenMP.h
index 686e51ee92a08..2d05b4423140b 100644
--- a/clang/include/clang/Sema/SemaOpenMP.h
+++ b/clang/include/clang/Sema/SemaOpenMP.h
@@ -1351,7 +1351,7 @@ class SemaOpenMP : public SemaBase {
OMPClause *
ActOnOpenMPToClause(ArrayRef<OpenMPMotionModifierKind> MotionModifiers,
ArrayRef<SourceLocation> MotionModifiersLoc,
- CXXScopeSpec &MapperIdScopeSpec,
+ Expr *IteratorModifier, CXXScopeSpec &MapperIdScopeSpec,
DeclarationNameInfo &MapperId, SourceLocation ColonLoc,
ArrayRef<Expr *> VarList, const OMPVarListLocTy &Locs,
ArrayRef<Expr *> UnresolvedMappers = {});
@@ -1359,7 +1359,7 @@ class SemaOpenMP : public SemaBase {
OMPClause *
ActOnOpenMPFromClause(ArrayRef<OpenMPMotionModifierKind> MotionModifiers,
ArrayRef<SourceLocation> MotionModifiersLoc,
- CXXScopeSpec &MapperIdScopeSpec,
+ Expr *IteratorModifier, CXXScopeSpec &MapperIdScopeSpec,
DeclarationNameInfo &MapperId, SourceLocation ColonLoc,
ArrayRef<Expr *> VarList, const OMPVarListLocTy &Locs,
ArrayRef<Expr *> UnresolvedMappers = {});
diff --git a/clang/lib/AST/OpenMPClause.cpp b/clang/lib/AST/OpenMPClause.cpp
index 0640fed823771..2183d77de8fa7 100644
--- a/clang/lib/AST/OpenMPClause.cpp
+++ b/clang/lib/AST/OpenMPClause.cpp
@@ -1321,7 +1321,7 @@ OMPToClause *OMPToClause::Create(
const ASTContext &C, const OMPVarListLocTy &Locs, ArrayRef<Expr *> Vars,
ArrayRef<ValueDecl *> Declarations,
MappableExprComponentListsRef ComponentLists, ArrayRef<Expr *> UDMapperRefs,
- ArrayRef<OpenMPMotionModifierKind> MotionModifiers,
+ Expr *IteratorModifier, ArrayRef<OpenMPMotionModifierKind> MotionModifiers,
ArrayRef<SourceLocation> MotionModifiersLoc,
NestedNameSpecifierLoc UDMQualifierLoc, DeclarationNameInfo MapperId) {
OMPMappableExprListSizeTy Sizes;
@@ -1343,7 +1343,7 @@ OMPToClause *OMPToClause::Create(
void *Mem = C.Allocate(
totalSizeToAlloc<Expr *, ValueDecl *, unsigned,
OMPClauseMappableExprCommon::MappableComponent>(
- 2 * Sizes.NumVars, Sizes.NumUniqueDeclarations,
+ 2 * Sizes.NumVars + 1, Sizes.NumUniqueDeclarations,
Sizes.NumUniqueDeclarations + Sizes.NumComponentLists,
Sizes.NumComponents));
@@ -1353,6 +1353,7 @@ OMPToClause *OMPToClause::Create(
Clause->setVarRefs(Vars);
Clause->setUDMapperRefs(UDMapperRefs);
Clause->setClauseInfo(Declarations, ComponentLists);
+ Clause->setIteratorModifier(IteratorModifier);
return Clause;
}
@@ -1361,17 +1362,19 @@ OMPToClause *OMPToClause::CreateEmpty(const ASTContext &C,
void *Mem = C.Allocate(
totalSizeToAlloc<Expr *, ValueDecl *, unsigned,
OMPClauseMappableExprCommon::MappableComponent>(
- 2 * Sizes.NumVars, Sizes.NumUniqueDeclarations,
+ 2 * Sizes.NumVars + 1, Sizes.NumUniqueDeclarations,
Sizes.NumUniqueDeclarations + Sizes.NumComponentLists,
Sizes.NumComponents));
- return new (Mem) OMPToClause(Sizes);
+ OMPToClause *Clause = new (Mem) OMPToClause(Sizes);
+ Clause->setIteratorModifier(nullptr);
+ return Clause;
}
OMPFromClause *OMPFromClause::Create(
const ASTContext &C, const OMPVarListLocTy &Locs, ArrayRef<Expr *> Vars,
ArrayRef<ValueDecl *> Declarations,
MappableExprComponentListsRef ComponentLists, ArrayRef<Expr *> UDMapperRefs,
- ArrayRef<OpenMPMotionModifierKind> MotionModifiers,
+ Expr *IteratorModifier, ArrayRef<OpenMPMotionModifierKind> MotionModifiers,
ArrayRef<SourceLocation> MotionModifiersLoc,
NestedNameSpecifierLoc UDMQualifierLoc, DeclarationNameInfo MapperId) {
OMPMappableExprListSizeTy Sizes;
@@ -1393,7 +1396,7 @@ OMPFromClause *OMPFromClause::Create(
void *Mem = C.Allocate(
totalSizeToAlloc<Expr *, ValueDecl *, unsigned,
OMPClauseMappableExprCommon::MappableComponent>(
- 2 * Sizes.NumVars, Sizes.NumUniqueDeclarations,
+ 2 * Sizes.NumVars + 1, Sizes.NumUniqueDeclarations,
Sizes.NumUniqueDeclarations + Sizes.NumComponentLists,
Sizes.NumComponents));
@@ -1404,6 +1407,7 @@ OMPFromClause *OMPFromClause::Create(
Clause->setVarRefs(Vars);
Clause->setUDMapperRefs(UDMapperRefs);
Clause->setClauseInfo(Declarations, ComponentLists);
+ Clause->setIteratorModifier(IteratorModifier);
return Clause;
}
@@ -1413,10 +1417,12 @@ OMPFromClause::CreateEmpty(const ASTContext &C,
void *Mem = C.Allocate(
totalSizeToAlloc<Expr *, ValueDecl *, unsigned,
OMPClauseMappableExprCommon::MappableComponent>(
- 2 * Sizes.NumVars, Sizes.NumUniqueDeclarations,
+ 2 * Sizes.NumVars + 1, Sizes.NumUniqueDeclarations,
Sizes.NumUniqueDeclarations + Sizes.NumComponentLists,
Sizes.NumComponents));
- return new (Mem) OMPFromClause(Sizes);
+ OMPFromClause *Clause = new (Mem) OMPFromClause(Sizes);
+ Clause->setIteratorModifier(nullptr);
+ return Clause;
}
void OMPUseDevicePtrClause::setPrivateCopies(ArrayRef<Expr *> VL) {
@@ -2694,12 +2700,16 @@ template <typename T> void OMPClausePrinter::VisitOMPMotionClause(T *Node) {
OS << '(';
for (unsigned I = 0; I < NumberOfOMPMotionModifiers; ++I) {
if (Node->getMotionModifier(I) != OMPC_MOTION_MODIFIER_unknown) {
- OS << getOpenMPSimpleClauseTypeName(Node->getClauseKind(),
- Node->getMotionModifier(I));
- if (Node->getMotionModifier(I) == OMPC_MOTION_MODIFIER_mapper)
- PrintMapper(OS, Node, Policy);
- if (I < ModifierCount - 1)
- OS << ", ";
+ if (Node->getMotionModifier(I) == OMPC_MOTION_MODIFIER_iterator) {
+ PrintIterator(OS, Node, Policy);
+ } else {
+ OS << getOpenMPSimpleClauseTypeName(Node->getClauseKind(),
+ Node->getMotionModifier(I));
+ if (Node->getMotionModifier(I) == OMPC_MOTION_MODIFIER_mapper)
+ PrintMapper(OS, Node, Policy);
+ if (I < ModifierCount - 1)
+ OS << ", ";
+ }
}
}
OS << ':';
diff --git a/clang/lib/CodeGen/CGOpenMPRuntime.cpp b/clang/lib/CodeGen/CGOpenMPRuntime.cpp
index a8255ac74cfcf..9bd6da4a38df8 100644
--- a/clang/lib/CodeGen/CGOpenMPRuntime.cpp
+++ b/clang/lib/CodeGen/CGOpenMPRuntime.cpp
@@ -8634,6 +8634,15 @@ class MappableExprsHandler {
if (llvm::is_contained(C->getMotionModifiers(),
OMPC_MOTION_MODIFIER_present))
Kind = Present;
+ if (llvm::is_contained(C->getMotionModifiers(),
+ OMPC_MOTION_MODIFIER_iterator)) {
+ if (auto *IteratorExpr = dyn_cast<OMPIteratorExpr>(
+ C->getIteratorModifier()->IgnoreParenImpCasts())) {
+ const auto *VD = cast<VarDecl>(IteratorExpr->getIteratorDecl(0));
+ CGF.EmitVarDecl(*VD);
+ }
+ }
+
const auto *EI = C->getVarRefs().begin();
for (const auto L : C->component_lists()) {
InfoGen(std::get<0>(L), Kind, std::get<1>(L), OMPC_MAP_to, {},
@@ -8650,6 +8659,15 @@ class MappableExprsHandler {
if (llvm::is_contained(C->getMotionModifiers(),
OMPC_MOTION_MODIFIER_present))
Kind = Present;
+ if (llvm::is_contained(C->getMotionModifiers(),
+ OMPC_MOTION_MODIFIER_iterator)) {
+ if (auto *IteratorExpr = dyn_cast<OMPIteratorExpr>(
+ C->getIteratorModifier()->IgnoreParenImpCasts())) {
+ const auto *VD = cast<VarDecl>(IteratorExpr->getIteratorDecl(0));
+ CGF.EmitVarDecl(*VD);
+ }
+ }
+
const auto *EI = C->getVarRefs().begin();
for (const auto L : C->component_lists()) {
InfoGen(std::get<0>(L), Kind, std::get<1>(L), OMPC_MAP_from, {},
diff --git a/clang/lib/Parse/ParseOpenMP.cpp b/clang/lib/Parse/ParseOpenMP.cpp
index 3b69c286634bb..15c3f7594bf44 100644
--- a/clang/lib/Parse/ParseOpenMP.cpp
+++ b/clang/lib/Parse/ParseOpenMP.cpp
@@ -4925,19 +4925,28 @@ bool Parser::ParseOpenMPVarList(OpenMPDirectiveKind DKind,
break;
Data.MotionModifiers.push_back(Modifier);
Data.MotionModifiersLoc.push_back(Tok.getLocation());
- ConsumeToken();
- if (Modifier == OMPC_MOTION_MODIFIER_mapper) {
- IsInvalidMapperModifier = parseMapperModifier(Data);
- if (IsInvalidMapperModifier)
+ if (PP.getSpelling(Tok) == "iterator" && getLangOpts().OpenMP >= 51) {
+ ExprResult Tail;
+ Tail = ParseOpenMPIteratorsExpr();
+ Tail = Actions.ActOnFinishFullExpr(Tail.get(), T.getOpenLocation(),
+ /*DiscardedValue=*/false);
+ if (Tail.isUsable())
+ Data.IteratorExpr = Tail.get();
+ } else {
+ ConsumeToken();
+ if (Modifier == OMPC_MOTION_MODIFIER_mapper) {
+ IsInvalidMapperModifier = parseMapperModifier(Data);
+ if (IsInvalidMapperModifier)
+ break;
+ }
+ // OpenMP < 5.1 doesn't permit a ',' or additional modifiers.
+ if (getLangOpts().OpenMP < 51)
break;
+ // OpenMP 5.1 accepts an optional ',' even if the next character is ':'.
+ // TODO: Is that intentional?
+ if (Tok.is(tok::comma))
+ ConsumeToken();
}
- // OpenMP < 5.1 doesn't permit a ',' or additional modifiers.
- if (getLangOpts().OpenMP < 51)
- break;
- // OpenMP 5.1 accepts an optional ',' even if the next character is ':'.
- // TODO: Is that intentional?
- if (Tok.is(tok::comma))
- ConsumeToken();
}
if (!Data.MotionModifiers.empty() && Tok.isNot(tok::colon)) {
if (!IsInvalidMapperModifier) {
diff --git a/clang/lib/Sema/SemaOpenMP.cpp b/clang/lib/Sema/SemaOpenMP.cpp
index 31c8f0cd30c56..431c545c07e47 100644
--- a/clang/lib/Sema/SemaOpenMP.cpp
+++ b/clang/lib/Sema/SemaOpenMP.cpp
@@ -18712,16 +18712,16 @@ OMPClause *SemaOpenMP::ActOnOpenMPVarListClause(OpenMPClauseKind Kind,
ExtraModifierLoc, ColonLoc, VarList, Locs);
break;
case OMPC_to:
- Res =
- ActOnOpenMPToClause(Data.MotionModifiers, Data.MotionModifiersLoc,
- Data.ReductionOrMapperIdScopeSpec,
- Data.ReductionOrMapperId, ColonLoc, VarList, Locs);
+ Res = ActOnOpenMPToClause(
+ Data.MotionModifiers, Data.MotionModifiersLoc, Data.IteratorExpr,
+ Data.ReductionOrMapperIdScopeSpec, Data.ReductionOrMapperId, ColonLoc,
+ VarList, Locs);
break;
case OMPC_from:
- Res = ActOnOpenMPFromClause(Data.MotionModifiers, Data.MotionModifiersLoc,
- Data.ReductionOrMapperIdScopeSpec,
- Data.ReductionOrMapperId, ColonLoc, VarList,
- Locs);
+ Res = ActOnOpenMPFromClause(
+ Data.MotionModifiers, Data.MotionModifiersLoc, Data.IteratorExpr,
+ Data.ReductionOrMapperIdScopeSpec, Data.ReductionOrMapperId, ColonLoc,
+ VarList, Locs);
break;
case OMPC_use_device_ptr:
Res = ActOnOpenMPUseDevicePtrClause(VarList, Locs);
@@ -24457,11 +24457,12 @@ void SemaOpenMP::ActOnOpenMPDeclareTargetInitializer(Decl *TargetDecl) {
OMPClause *SemaOpenMP::ActOnOpenMPToClause(
ArrayRef<OpenMPMotionModifierKind> MotionModifiers,
- ArrayRef<SourceLocation> MotionModifiersLoc,
+ ArrayRef<SourceLocation> MotionModifiersLoc, Expr *IteratorExpr,
CXXScopeSpec &MapperIdScopeSpec, DeclarationNameInfo &MapperId,
SourceLocation ColonLoc, ArrayRef<Expr *> VarList,
const OMPVarListLocTy &Locs, ArrayRef<Expr *> UnresolvedMappers) {
OpenMPMotionModifierKind Modifiers[] = {OMPC_MOTION_MODIFIER_unknown,
+ OMPC_MOTION_MODIFIER_unknown,
OMPC_MOTION_MODIFIER_unknown};
SourceLocation ModifiersLoc[NumberOfOMPMotionModifiers];
@@ -24485,20 +24486,25 @@ OMPClause *SemaOpenMP::ActOnOpenMPToClause(
MapperIdScopeSpec, MapperId, UnresolvedMappers);
if (MVLI.ProcessedVarList.empty())
return nullptr;
-
+ if (IteratorExpr)
+ if (auto *DRE = dyn_cast<DeclRefExpr>(IteratorExpr))
+ if (auto *VD = dyn_cast<VarDecl>(DRE->getDecl()))
+ DSAStack->addIteratorVarDecl(VD);
return OMPToClause::Create(
getASTContext(), Locs, MVLI.ProcessedVarList, MVLI.VarBaseDeclarations,
- MVLI.VarComponents, MVLI.UDMapperList, Modifiers, ModifiersLoc,
- MapperIdScopeSpec.getWithLocInContext(getASTContext()), MapperId);
+ MVLI.VarComponents, MVLI.UDMapperList, IteratorExpr, Modifiers,
+ ModifiersLoc, MapperIdScopeSpec.getWithLocInContext(getASTContext()),
+ MapperId);
}
OMPClause *SemaOpenMP::ActOnOpenMPFromClause(
ArrayRef<OpenMPMotionModifierKind> MotionModifiers,
- ArrayRef<SourceLocation> MotionModifiersLoc,
+ ArrayRef<SourceLocation> MotionModifiersLoc, Expr *IteratorExpr,
CXXScopeSpec &MapperIdScopeSpec, DeclarationNameInfo &MapperId,
SourceLocation ColonLoc, ArrayRef<Expr *> VarList,
const OMPVarListLocTy &Locs, ArrayRef<Expr *> UnresolvedMappers) {
OpenMPMotionModifierKind Modifiers[] = {OMPC_MOTION_MODIFIER_unknown,
+ OMPC_MOTION_MODIFIER_unknown,
OMPC_MOTION_MODIFIER_unknown};
SourceLocation ModifiersLoc[NumberOfOMPMotionModifiers];
@@ -24522,11 +24528,15 @@ OMPClause *SemaOpenMP::ActOnOpenMPFromClause(
MapperIdScopeSpec, MapperId, UnresolvedMappers);
if (MVLI.ProcessedVarList.empty())
return nullptr;
-
+ if (IteratorExpr)
+ if (auto *DRE = dyn_cast<DeclRefExpr>(IteratorExpr))
+ if (auto *VD = dyn_cast<VarDecl>(DRE->getDecl()))
+ DSAStack->addIteratorVarDecl(VD);
return OMPFromClause::Create(
getASTContext(), Locs, MVLI.ProcessedVarList, MVLI.VarBaseDeclarations,
- MVLI.VarComponents, MVLI.UDMapperList, Modifiers, ModifiersLoc,
- MapperIdScopeSpec.getWithLocInContext(getASTContext()), MapperId);
+ MVLI.VarComponents, MVLI.UDMapperList, IteratorExpr, Modifiers,
+ ModifiersLoc, MapperIdScopeSpec.getWithLocInContext(getASTContext()),
+ MapperId);
}
OMPClause *
diff --git a/clang/lib/Sema/TreeTransform.h b/clang/lib/Sema/TreeTransform.h
index 0e8b674a006d0..8e5dbeb792348 100644
--- a/clang/lib/Sema/TreeTransform.h
+++ b/clang/lib/Sema/TreeTransform.h
@@ -2221,13 +2221,14 @@ class TreeTransform {
OMPClause *
RebuildOMPToClause(ArrayRef<OpenMPMotionModifierKind> MotionModifiers,
ArrayRef<SourceLocation> MotionModifiersLoc,
- CXXScopeSpec &MapperIdScopeSpec,
+ Expr *IteratorModifier, CXXScopeSpec &MapperIdScopeSpec,
DeclarationNameInfo &MapperId, SourceLocation ColonLoc,
ArrayRef<Expr *> VarList, const OMPVarListLocTy &Locs,
ArrayRef<Expr *> UnresolvedMappers) {
return getSema().OpenMP().ActOnOpenMPToClause(
- MotionModifiers, MotionModifiersLoc, MapperIdScopeSpec, MapperId,
- ColonLoc, VarList, Locs, UnresolvedMappers);
+ MotionModifiers, MotionModifiersLoc, IteratorModifier,
+ MapperIdScopeSpec, MapperId, ColonLoc, VarList, Locs,
+ UnresolvedMappers);
}
/// Build a new OpenMP 'from' clause.
@@ -2237,13 +2238,14 @@ class TreeTransform {
OMPClause *
RebuildOMPFromClause(ArrayRef<OpenMPMotionModifierKind> MotionModifiers,
ArrayRef<SourceLocation> MotionModifiersLoc,
- CXXScopeSpec &MapperIdScopeSpec,
+ Expr *IteratorModifier, CXXScopeSpec &MapperIdScopeSpec,
DeclarationNameInfo &MapperId, SourceLocation ColonLoc,
ArrayRef<Expr *> VarList, const OMPVarListLocTy &Locs,
ArrayRef<Expr *> UnresolvedMappers) {
return getSema().OpenMP().ActOnOpenMPFromClause(
- MotionModifiers, MotionModifiersLoc, MapperIdScopeSpec, MapperId,
- ColonLoc, VarList, Locs, UnresolvedMappers);
+ MotionModifiers, MotionModifiersLoc, IteratorModifier,
+ MapperIdScopeSpec, MapperId, ColonLoc, VarList, Locs,
+ UnresolvedMappers);
}
/// Build a new OpenMP 'use_device_ptr' clause.
@@ -11535,6 +11537,13 @@ template <typename Derived>
OMPClause *TreeTransform<Derived>::TransformOMPToClause(OMPToClause *C) {
OMPVarListLocTy Locs(C->getBeginLoc(), C->getLParenLoc(), C->getEndLoc());
llvm::SmallVector<Expr *, 16> Vars;
+ Expr *IteratorModifier = C->getIteratorModifier();
+ if (IteratorModifier) {
+ ExprResult MapModRes = getDerived().TransformExpr(IteratorModifier);
+ if (MapModRes.isInvalid())
+ return nullptr;
+ IteratorModifier = MapModRes.get();
+ }
CXXScopeSpec MapperIdScopeSpec;
DeclarationNameInfo MapperIdInfo;
llvm::SmallVector<Expr *, 16> UnresolvedMappers;
@@ -11542,14 +11551,22 @@ OMPClause *TreeTransform<Derived>::TransformOMPToClause(OMPToClause *C) {
*this, C, Vars, MapperIdScopeSpec, MapperIdInfo, UnresolvedMappers))
return nullptr;
return getDerived().RebuildOMPToClause(
- C->getMotionModifiers(), C->getMotionModifiersLoc(), MapperIdScopeSpec,
- MapperIdInfo, C->getColonLoc(), Vars, Locs, UnresolvedMappers);
+ C->getMotionModifiers(), C->getMotionModifiersLoc(), IteratorModifier,
+ MapperIdScopeSpec, MapperIdInfo, C->getColonLoc(), Vars, Locs,
+ UnresolvedMappers);
}
template <typename Derived>
OMPClause *TreeTransform<Derived>::TransformOMPFromClause(OMPFromClause *C) {
OMPVarListLocTy Locs(C->getBeginLoc(), C->getLParenLoc(), C->getEndLoc());
llvm::SmallVector<Expr *, 16> Vars;
+ Expr *IteratorModifier = C->getIteratorModifier();
+ if (IteratorModifier) {
+ ExprResult MapModRes = getDerived().TransformExpr(IteratorModifier);
+ if (MapModRes.isInvalid())
+ return nullptr;
+ IteratorModifier = MapModRes.get();
+ }
CXXScopeSpec MapperIdScopeSpec;
DeclarationNameInfo MapperIdInfo;
llvm::SmallVector<Expr *, 16> UnresolvedMappers;
@@ -11557,8 +11574,9 @@ OMPClause *TreeTransform<Derived>::TransformOMPFromClause(OMPFromClause *C) {
*this, C, Vars, MapperIdScopeSpec, MapperIdInfo, UnresolvedMappers))
return nullptr;
return getDerived().RebuildOMPFromClause(
- C->getMotionModifiers(), C->getMotionModifiersLoc(), MapperIdScopeSpec,
- MapperIdInfo, C->getColonLoc(), Vars, Locs, UnresolvedMappers);
+ C->getMotionModifiers(), C->getMotionModifiersLoc(), IteratorModifier,
+ MapperIdScopeSpec, MapperIdInfo, C->getColonLoc(), Vars, Locs,
+ UnresolvedMappers);
}
template <typename Derived>
diff --git a/clang/lib/Serialization/ASTReader.cpp b/clang/lib/Serialization/ASTReader.cpp
index 55c52154c4113..67ba1fd70dff7 100644
--- a/clang/lib/Serialization/ASTReader.cpp
+++ b/clang/lib/Serialization/ASTReader.cpp
@@ -12387,6 +12387,8 @@ void OMPClauseReader::VisitOMPToClause(OMPToClause *C) {
C->setMotionModifier(
I, static_cast<OpenMPMotionModifierKind>(Record.readInt()));
C->setMotionModifierLoc(I, Record.readSourceLocation());
+ if (C->getMotionModifier(I) == OMPC_MOTION_MODIFIER_iterator)
+ C->setIteratorModifier(Record.readExpr());
}
C->setMapperQualifierLoc(Record.readNestedNameSpecifierLoc());
C->setMapperIdInfo(Record.readDeclarationNameInfo());
@@ -12443,6 +12445,8 @@ void OMPClauseReader::VisitOMPFromClause(OMPFromClause *C) {
C->setMotionModifier(
I, static_cast<OpenMPMotionModifierKind>(Record.readInt()));
C->setMotionModifierLoc(I, Record.readSourceLocation());
+ if (C->getMotionModifier(I) == OMPC_MOTION_MODIFIER_iterator)
+ C->setIteratorModifier(Record.readExpr());
}
C->setMapperQualifierLoc(Record.readNestedNameSpecifierLoc());
C->setMapperIdInfo(Record.readDeclarationNameInfo());
diff --git a/clang/lib/Serialization/ASTWriter.cpp b/clang/lib/Serialization/ASTWriter.cpp
index e8c0d3f2b4ee9..fcee93c0ebbd3 100644
--- a/clang/lib/Serialization/ASTWriter.cpp
+++ b/clang/lib/Serialization/ASTWriter.cpp
@@ -8417,6 +8417,8 @@ void OMPClauseWriter::VisitOMPToClause(OMPToClause *C) {
for (unsigned I = 0; I < NumberOfOMPMotionModifiers; ++I) {
Record.push_back(C->getMotionModifier(I));
Record.AddSourceLocation(C->getMotionModifierLoc(I));
+ if (C->getMotionModifier(I) == OMPC_MOTION_MODIFIER_iterator)
+ Record.AddStmt(C->getIteratorModifier());
}
Record.AddNestedNameSpecifierLoc(C->getMapperQualifierLoc());
Record.AddDeclarationNameInfo(C->getMapperIdInfo());
@@ -8447,6 +8449,8 @@ void OMPClauseWriter::VisitOMPFromClause(OMPFromClause *C) {
for (unsigned I = 0; I < NumberOfOMPMotionModifiers; ++I) {
Record.push_back(C->getMotionModifier(I));
Record.AddSourceLocation(C->getMotionModifierLoc(I));
+ if (C->getMotionModifier(I) == OMPC_MOTION_MODIFIER_iterator)
+ Record.AddStmt(C->getIteratorModifier());
}
Record.AddNestedNameSpecifierLoc(C->getMapperQualifierLoc());
Record.AddDeclarationNameInfo(C->getMapperIdInfo());
diff --git a/clang/test/OpenMP/target_update_codegen.cpp b/clang/test/OpenMP/target_update_codegen.cpp
index c8211f475c7fc..6c754c1c953ea 100644
--- a/clang/test/OpenMP/target_update_codegen.cpp
+++ b/clang/test/OpenMP/target_update_codegen.cpp
@@ -1560,5 +1560,37 @@ void foo(int arg) {
{ ++arg; }
}
+#endif
+// RUN: %clang_cc1 -DCK26 -verify -Wno-vla -fopenmp -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck %s --check-prefix CK26 --check-prefix CK26-64
+// RUN: %clang_cc1 -DCK26 -fopenmp -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify -Wno-vla %s -emit-llvm -o - | FileCheck %s --check-prefix CK26 --check-prefix CK26-64
+// RUN: %clang_cc1 -DCK26 -fopenmp-version=51 -verify -Wno-vla -fopenmp -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck %s --check-prefix CK26 --check-prefix CK26-32
+// RUN: %clang_cc1 -DCK26 -fopenmp -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify -Wno-vla %s -emit-llvm -o - | FileCheck %s --check-prefix CK26 --check-prefix CK26-32
+
+// RUN: %clang_cc1 -DCK26 -verify -Wno-vla -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY19 %s
+// RUN: %clang_cc1 -DCK26 -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify -Wno-vla %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY19 %s
+// RUN: %clang_cc1 -DCK26 -verify -Wno-vla -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY19 %s
+// RUN: %clang_cc1 -DCK26 -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify -Wno-vla %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY19 %s
+// SIMD-ONLY19-NOT: {{__kmpc|__tgt}}
+#ifdef CK26
+void foo() {
+int a[10];
+#pragma omp target update to(iterator(int it = 0:10) : a[it])
+// CK26-LABEL: define {{.+}}foo
+// CK26: %[[ITER:[a-zA-Z0-9_]+]] = alloca i32, align 4
+// CK26: %[[LOAD2:.*]] = load i32, ptr %[[ITER]], align 4
+}
+
+void foo1() {
+int a[10];
+#pragma omp target update from(iterator(int it = 0:10) : a[it])
+// CK26-LABEL: define {{.+}}foo1
+// CK26: %[[ITER:[a-zA-Z0-9_]+]] = alloca i32, align 4
+// CK26: %[[LOAD2:.*]] = load i32, ptr %[[ITER]], align 4
+}
+
#endif
#endif
diff --git a/clang/test/OpenMP/target_update_iterator_ast_print.cpp b/clang/test/OpenMP/target_update_iterator_ast_print.cpp
new file mode 100644
index 0000000000000..322f565c9c732
--- /dev/null
+++ b/clang/test/OpenMP/target_update_iterator_ast_print.cpp
@@ -0,0 +1,16 @@
+// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=51 -ast-print %s | FileCheck %s
+// expected-no-diagnostics
+
+#ifndef HEADER
+#define HEADER
+
+void test() {
+ int a[10];
+ #pragma omp target update to(iterator(int it = 0:10): a[it])
+ // CHECK: int a[10];
+ // CHECK: #pragma omp target update to(iterator(int it = 0:10): a[it])
+ #pragma omp target update from(iterator(int it = 0:10): a[it])
+ // CHECK: #pragma omp target update from(iterator(int it = 0:10): a[it])
+}
+
+#endif
diff --git a/clang/test/OpenMP/target_update_iterator_serialization.cpp b/clang/test/OpenMP/target_update_iterator_serialization.cpp
new file mode 100644
index 0000000000000..c1ad380f7c9a5
--- /dev/null
+++ b/clang/test/OpenMP/target_update_iterator_serialization.cpp
@@ -0,0 +1,35 @@
+// Test without serialization:
+// RUN: %clang_cc1 -std=c++20 -fopenmp %s -ast-dump | FileCheck %s
+
+// Test with serialization:
+// RUN: %clang_cc1 -std=c++20 -fopenmp -emit-pch -o %t %s
+// RUN: %clang_cc1 -x c++ -std=c++20 -fopenmp -include-pch %t -ast-dump-all /dev/null \
+// RUN: | sed -e "s/ <undeserialized declarations>//" -e "s/ imported//" \
+// RUN: | FileCheck %s
+
+// CHECK: OMPTargetUpdateDirective
+// CHECK-NEXT: OMPFromClause
+// CHECK-NEXT: ArraySubscriptExpr
+// CHECK: DeclRefExpr {{.*}} 'a'
+// CHECK: DeclRefExpr {{.*}} 'it'
+
+
+void foo1() {
+ int a[10];
+
+#pragma omp target update from(iterator(int it = 0:10) : a[it])
+ ;
+}
+
+// CHECK: OMPTargetUpdateDirective
+// CHECK-NEXT: OMPToClause
+// CHECK-NEXT: ArraySubscriptExpr
+// CHECK: DeclRefExpr {{.*}} 'a'
+// CHECK: DeclRefExpr {{.*}} 'it'
+
+void foo2() {
+ int a[10];
+
+#pragma omp target update to(iterator(int it = 0:10) : a[it])
+ ;
+}
More information about the cfe-commits
mailing list