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 <[email protected]> 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]) + ; +} _______________________________________________ cfe-commits mailing list [email protected] https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
