https://github.com/ShashwathiNavada updated https://github.com/llvm/llvm-project/pull/159112
>From 7bec4d102da2e3c99a8a8d83167471637d94f897 Mon Sep 17 00:00:00 2001 From: Shashwathi N <[email protected]> Date: Tue, 16 Sep 2025 10:23:50 -0500 Subject: [PATCH 1/6] Adding support for iterator in motion clauses --- clang/docs/OpenMPSupport.rst | 2 +- clang/include/clang/AST/OpenMPClause.h | 27 ++++++++----- clang/include/clang/Basic/OpenMPKinds.def | 1 + clang/include/clang/Sema/SemaOpenMP.h | 4 +- clang/lib/AST/OpenMPClause.cpp | 26 +++++++++---- clang/lib/Parse/ParseOpenMP.cpp | 14 +++++++ clang/lib/Sema/SemaOpenMP.cpp | 38 ++++++++++++------- clang/lib/Sema/TreeTransform.h | 36 ++++++++++++++---- clang/lib/Serialization/ASTReader.cpp | 4 ++ clang/lib/Serialization/ASTWriter.cpp | 4 ++ .../target_update_iterator_ast_print.cpp | 16 ++++++++ 11 files changed, 130 insertions(+), 42 deletions(-) create mode 100644 clang/test/OpenMP/target_update_iterator_ast_print.cpp diff --git a/clang/docs/OpenMPSupport.rst b/clang/docs/OpenMPSupport.rst index 0bc8590815220..a3e4272ceaf2f 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 | :good:`done` | | +------------------------------+--------------------------------------------------------------+--------------------------+-----------------------------------------------------------------------+ | 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 b2a6d4b9182b0..4e5a86482fbd4 100644 --- a/clang/include/clang/AST/OpenMPClause.h +++ b/clang/include/clang/AST/OpenMPClause.h @@ -7476,7 +7476,7 @@ 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]; @@ -7548,6 +7548,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; } @@ -7556,7 +7559,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(); @@ -7586,7 +7589,7 @@ class OMPToClause final : public OMPMappableExprListClause<OMPToClause>, ArrayRef<Expr *> Vars, ArrayRef<ValueDecl *> Declarations, MappableExprComponentListsRef ComponentLists, - ArrayRef<Expr *> UDMapperRefs, + ArrayRef<Expr *> UDMapperRefs,Expr *IteratorModifier, ArrayRef<OpenMPMotionModifierKind> MotionModifiers, ArrayRef<SourceLocation> MotionModifiersLoc, NestedNameSpecifierLoc UDMQualifierLoc, @@ -7611,7 +7614,9 @@ class OMPToClause final : public OMPMappableExprListClause<OMPToClause>, "Requested modifier exceeds the total number of modifiers."); return MotionModifiers[Cnt]; } - + Expr *getIteratorModifier() { + return getTrailingObjects<Expr *>()[2 * varlist_size()]; + } /// Fetches the motion-modifier location at 'Cnt' index of array of modifiers' /// locations. /// @@ -7677,7 +7682,7 @@ 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]; @@ -7738,7 +7743,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. @@ -7757,7 +7764,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(); @@ -7787,7 +7794,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); @@ -7811,7 +7818,9 @@ class OMPFromClause final "Requested modifier exceeds the total number of modifiers."); return MotionModifiers[Cnt]; } - + Expr *getIteratorModifier() { + 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 79c11b851c557..7944dc6916c94 100644 --- a/clang/include/clang/Basic/OpenMPKinds.def +++ b/clang/include/clang/Basic/OpenMPKinds.def @@ -194,6 +194,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 23827051ed724..ecfbe3284a202 100644 --- a/clang/include/clang/Sema/SemaOpenMP.h +++ b/clang/include/clang/Sema/SemaOpenMP.h @@ -1323,7 +1323,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 = {}); @@ -1331,7 +1331,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 69d33019c0952..de352df0f93f9 100644 --- a/clang/lib/AST/OpenMPClause.cpp +++ b/clang/lib/AST/OpenMPClause.cpp @@ -1218,7 +1218,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; @@ -1240,7 +1240,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)); @@ -1250,6 +1250,7 @@ OMPToClause *OMPToClause::Create( Clause->setVarRefs(Vars); Clause->setUDMapperRefs(UDMapperRefs); Clause->setClauseInfo(Declarations, ComponentLists); + Clause->setIteratorModifier(IteratorModifier); return Clause; } @@ -1258,17 +1259,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; @@ -1290,7 +1293,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)); @@ -1301,6 +1304,7 @@ OMPFromClause *OMPFromClause::Create( Clause->setVarRefs(Vars); Clause->setUDMapperRefs(UDMapperRefs); Clause->setClauseInfo(Declarations, ComponentLists); + Clause->setIteratorModifier(IteratorModifier); return Clause; } @@ -1310,10 +1314,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) { @@ -2564,12 +2570,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) { + 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/Parse/ParseOpenMP.cpp b/clang/lib/Parse/ParseOpenMP.cpp index 7dceb2d208352..75fd5a12ff3a1 100644 --- a/clang/lib/Parse/ParseOpenMP.cpp +++ b/clang/lib/Parse/ParseOpenMP.cpp @@ -4699,6 +4699,19 @@ bool Parser::ParseOpenMPVarList(OpenMPDirectiveKind DKind, break; Data.MotionModifiers.push_back(Modifier); Data.MotionModifiersLoc.push_back(Tok.getLocation()); + if (PP.getSpelling(Tok) == "iterator" && getLangOpts().OpenMP >= 51) { + ColonProtectionRAIIObject ColonRAII(*this); + TentativeParsingAction TPA(*this); + ExprResult Tail; + HasIterator = true; + EnterScope(Scope::OpenMPDirectiveScope | Scope::DeclScope); + 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); @@ -4712,6 +4725,7 @@ bool Parser::ParseOpenMPVarList(OpenMPDirectiveKind DKind, // 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 66bdd6e72d827..f44e6be061c5d 100644 --- a/clang/lib/Sema/SemaOpenMP.cpp +++ b/clang/lib/Sema/SemaOpenMP.cpp @@ -17723,16 +17723,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); @@ -23465,10 +23465,12 @@ void SemaOpenMP::ActOnOpenMPDeclareTargetInitializer(Decl *TargetDecl) { OMPClause *SemaOpenMP::ActOnOpenMPToClause( ArrayRef<OpenMPMotionModifierKind> MotionModifiers, 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]; @@ -23492,20 +23494,24 @@ 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, + 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]; @@ -23529,11 +23535,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 242ffb09af006..636f54c212f71 100644 --- a/clang/lib/Sema/TreeTransform.h +++ b/clang/lib/Sema/TreeTransform.h @@ -2196,13 +2196,15 @@ class TreeTransform { OMPClause * RebuildOMPToClause(ArrayRef<OpenMPMotionModifierKind> MotionModifiers, ArrayRef<SourceLocation> MotionModifiersLoc, + 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. @@ -2212,13 +2214,15 @@ class TreeTransform { OMPClause * RebuildOMPFromClause(ArrayRef<OpenMPMotionModifierKind> MotionModifiers, ArrayRef<SourceLocation> MotionModifiersLoc, + 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. @@ -11411,6 +11415,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; @@ -11418,14 +11429,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; @@ -11433,8 +11452,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 9ee8a0fb0f060..736ca0e4382bc 100644 --- a/clang/lib/Serialization/ASTReader.cpp +++ b/clang/lib/Serialization/ASTReader.cpp @@ -12321,6 +12321,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()); @@ -12377,6 +12379,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 2aa77934c08d1..205408ca6e9d2 100644 --- a/clang/lib/Serialization/ASTWriter.cpp +++ b/clang/lib/Serialization/ASTWriter.cpp @@ -8372,6 +8372,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()); @@ -8402,6 +8404,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_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 >From d7aa585f082bbe531461f968a73c2cf42eb8584e Mon Sep 17 00:00:00 2001 From: Shashwathi N <[email protected]> Date: Tue, 16 Sep 2025 10:39:52 -0500 Subject: [PATCH 2/6] Formatting --- clang/include/clang/AST/OpenMPClause.h | 23 ++++++++++++----------- clang/lib/AST/OpenMPClause.cpp | 12 ++++++------ clang/lib/Parse/ParseOpenMP.cpp | 24 ++++++++++++------------ clang/lib/Sema/SemaOpenMP.cpp | 18 +++++++++--------- clang/lib/Sema/TreeTransform.h | 8 +++----- 5 files changed, 42 insertions(+), 43 deletions(-) diff --git a/clang/include/clang/AST/OpenMPClause.h b/clang/include/clang/AST/OpenMPClause.h index 4e5a86482fbd4..9681e35078064 100644 --- a/clang/include/clang/AST/OpenMPClause.h +++ b/clang/include/clang/AST/OpenMPClause.h @@ -7476,7 +7476,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, + OMPC_MOTION_MODIFIER_unknown}; /// Location of motion-modifiers for the 'to' clause. SourceLocation MotionModifiersLoc[NumberOfOMPMotionModifiers]; @@ -7585,15 +7586,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,Expr *IteratorModifier, - 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. /// @@ -7682,7 +7682,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, + OMPC_MOTION_MODIFIER_unknown}; /// Location of motion-modifiers for the 'from' clause. SourceLocation MotionModifiersLoc[NumberOfOMPMotionModifiers]; diff --git a/clang/lib/AST/OpenMPClause.cpp b/clang/lib/AST/OpenMPClause.cpp index de352df0f93f9..8146c1fa4d8f4 100644 --- a/clang/lib/AST/OpenMPClause.cpp +++ b/clang/lib/AST/OpenMPClause.cpp @@ -2573,12 +2573,12 @@ template <typename T> void OMPClausePrinter::VisitOMPMotionClause(T *Node) { 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 << getOpenMPSimpleClauseTypeName(Node->getClauseKind(), + Node->getMotionModifier(I)); + if (Node->getMotionModifier(I) == OMPC_MOTION_MODIFIER_mapper) + PrintMapper(OS, Node, Policy); + if (I < ModifierCount - 1) + OS << ", "; } } } diff --git a/clang/lib/Parse/ParseOpenMP.cpp b/clang/lib/Parse/ParseOpenMP.cpp index 75fd5a12ff3a1..48296c50da5a7 100644 --- a/clang/lib/Parse/ParseOpenMP.cpp +++ b/clang/lib/Parse/ParseOpenMP.cpp @@ -4712,19 +4712,19 @@ bool Parser::ParseOpenMPVarList(OpenMPDirectiveKind DKind, 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(); + 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(); } } if (!Data.MotionModifiers.empty() && Tok.isNot(tok::colon)) { diff --git a/clang/lib/Sema/SemaOpenMP.cpp b/clang/lib/Sema/SemaOpenMP.cpp index f44e6be061c5d..2e37c85a9abab 100644 --- a/clang/lib/Sema/SemaOpenMP.cpp +++ b/clang/lib/Sema/SemaOpenMP.cpp @@ -23464,8 +23464,7 @@ void SemaOpenMP::ActOnOpenMPDeclareTargetInitializer(Decl *TargetDecl) { OMPClause *SemaOpenMP::ActOnOpenMPToClause( ArrayRef<OpenMPMotionModifierKind> MotionModifiers, - ArrayRef<SourceLocation> MotionModifiersLoc, - Expr *IteratorExpr, + ArrayRef<SourceLocation> MotionModifiersLoc, Expr *IteratorExpr, CXXScopeSpec &MapperIdScopeSpec, DeclarationNameInfo &MapperId, SourceLocation ColonLoc, ArrayRef<Expr *> VarList, const OMPVarListLocTy &Locs, ArrayRef<Expr *> UnresolvedMappers) { @@ -23494,14 +23493,15 @@ 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())) + 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,IteratorExpr, Modifiers, ModifiersLoc, - MapperIdScopeSpec.getWithLocInContext(getASTContext()), MapperId); + MVLI.VarComponents, MVLI.UDMapperList, IteratorExpr, Modifiers, + ModifiersLoc, MapperIdScopeSpec.getWithLocInContext(getASTContext()), + MapperId); } OMPClause *SemaOpenMP::ActOnOpenMPFromClause( @@ -23536,8 +23536,8 @@ OMPClause *SemaOpenMP::ActOnOpenMPFromClause( if (MVLI.ProcessedVarList.empty()) return nullptr; if (IteratorExpr) - if (auto *DRE = dyn_cast<DeclRefExpr>(IteratorExpr)) - if (auto *VD = dyn_cast<VarDecl>(DRE->getDecl())) + 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, diff --git a/clang/lib/Sema/TreeTransform.h b/clang/lib/Sema/TreeTransform.h index 636f54c212f71..63da489a52164 100644 --- a/clang/lib/Sema/TreeTransform.h +++ b/clang/lib/Sema/TreeTransform.h @@ -2196,8 +2196,7 @@ class TreeTransform { OMPClause * RebuildOMPToClause(ArrayRef<OpenMPMotionModifierKind> MotionModifiers, ArrayRef<SourceLocation> MotionModifiersLoc, - Expr *IteratorModifier, - CXXScopeSpec &MapperIdScopeSpec, + Expr *IteratorModifier, CXXScopeSpec &MapperIdScopeSpec, DeclarationNameInfo &MapperId, SourceLocation ColonLoc, ArrayRef<Expr *> VarList, const OMPVarListLocTy &Locs, ArrayRef<Expr *> UnresolvedMappers) { @@ -2214,8 +2213,7 @@ class TreeTransform { OMPClause * RebuildOMPFromClause(ArrayRef<OpenMPMotionModifierKind> MotionModifiers, ArrayRef<SourceLocation> MotionModifiersLoc, - Expr *IteratorModifier, - CXXScopeSpec &MapperIdScopeSpec, + Expr *IteratorModifier, CXXScopeSpec &MapperIdScopeSpec, DeclarationNameInfo &MapperId, SourceLocation ColonLoc, ArrayRef<Expr *> VarList, const OMPVarListLocTy &Locs, ArrayRef<Expr *> UnresolvedMappers) { @@ -11438,7 +11436,7 @@ 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(); + Expr *IteratorModifier = C->getIteratorModifier(); if (IteratorModifier) { ExprResult MapModRes = getDerived().TransformExpr(IteratorModifier); if (MapModRes.isInvalid()) >From 98eebb6a44b72b15a19ee2078313469ee895ac2a Mon Sep 17 00:00:00 2001 From: ShashwathiNavada <[email protected]> Date: Wed, 17 Sep 2025 11:17:01 +0530 Subject: [PATCH 3/6] Update ParseOpenMP.cpp --- clang/lib/Parse/ParseOpenMP.cpp | 4 ++++ 1 file changed, 4 insertions(+) diff --git a/clang/lib/Parse/ParseOpenMP.cpp b/clang/lib/Parse/ParseOpenMP.cpp index 48296c50da5a7..790eb23b48b2c 100644 --- a/clang/lib/Parse/ParseOpenMP.cpp +++ b/clang/lib/Parse/ParseOpenMP.cpp @@ -4710,6 +4710,10 @@ bool Parser::ParseOpenMPVarList(OpenMPDirectiveKind DKind, /*DiscardedValue=*/false); if (Tail.isUsable()) { Data.IteratorExpr = Tail.get(); + TPA.Commit(); + } + else + TPA.revert(); } } else { ConsumeToken(); >From 2ce56c225032db8275627584b08091e7af50e8f0 Mon Sep 17 00:00:00 2001 From: Shashwathi N <[email protected]> Date: Wed, 17 Sep 2025 01:35:20 -0500 Subject: [PATCH 4/6] Fixed the error --- clang/lib/Parse/ParseOpenMP.cpp | 6 ++---- 1 file changed, 2 insertions(+), 4 deletions(-) diff --git a/clang/lib/Parse/ParseOpenMP.cpp b/clang/lib/Parse/ParseOpenMP.cpp index 790eb23b48b2c..3b1b16ba39ccf 100644 --- a/clang/lib/Parse/ParseOpenMP.cpp +++ b/clang/lib/Parse/ParseOpenMP.cpp @@ -4711,10 +4711,8 @@ bool Parser::ParseOpenMPVarList(OpenMPDirectiveKind DKind, if (Tail.isUsable()) { Data.IteratorExpr = Tail.get(); TPA.Commit(); - } - else - TPA.revert(); - } + } else + TPA.Revert(); } else { ConsumeToken(); if (Modifier == OMPC_MOTION_MODIFIER_mapper) { >From f32b8c81bb7f44cb29ede02b871ecd55f65deee7 Mon Sep 17 00:00:00 2001 From: Shashwathi N <[email protected]> Date: Tue, 7 Oct 2025 02:47:07 -0500 Subject: [PATCH 5/6] modifications done --- clang/include/clang/AST/OpenMPClause.h | 4 +- clang/lib/AST/OpenMPClause.cpp | 4 +- clang/test/OpenMP/target_update_codegen.cpp | 55 +++++++++++++++++++++ 3 files changed, 59 insertions(+), 4 deletions(-) diff --git a/clang/include/clang/AST/OpenMPClause.h b/clang/include/clang/AST/OpenMPClause.h index 9681e35078064..4fbc44d33caba 100644 --- a/clang/include/clang/AST/OpenMPClause.h +++ b/clang/include/clang/AST/OpenMPClause.h @@ -7614,7 +7614,7 @@ class OMPToClause final : public OMPMappableExprListClause<OMPToClause>, "Requested modifier exceeds the total number of modifiers."); return MotionModifiers[Cnt]; } - Expr *getIteratorModifier() { + Expr *getIteratorModifier() const { return getTrailingObjects<Expr *>()[2 * varlist_size()]; } /// Fetches the motion-modifier location at 'Cnt' index of array of modifiers' @@ -7819,7 +7819,7 @@ class OMPFromClause final "Requested modifier exceeds the total number of modifiers."); return MotionModifiers[Cnt]; } - Expr *getIteratorModifier() { + Expr *getIteratorModifier() const { return getTrailingObjects<Expr *>()[2 * varlist_size()]; } /// Fetches the motion-modifier location at 'Cnt' index of array of modifiers' diff --git a/clang/lib/AST/OpenMPClause.cpp b/clang/lib/AST/OpenMPClause.cpp index 8146c1fa4d8f4..e9127c5052bf1 100644 --- a/clang/lib/AST/OpenMPClause.cpp +++ b/clang/lib/AST/OpenMPClause.cpp @@ -2570,9 +2570,9 @@ template <typename T> void OMPClausePrinter::VisitOMPMotionClause(T *Node) { OS << '('; for (unsigned I = 0; I < NumberOfOMPMotionModifiers; ++I) { if (Node->getMotionModifier(I) != OMPC_MOTION_MODIFIER_unknown) { - if (Node->getMotionModifier(I) == OMPC_MOTION_MODIFIER_iterator) + if (Node->getMotionModifier(I) == OMPC_MOTION_MODIFIER_iterator) { PrintIterator(OS, Node, Policy); - else { + } else { OS << getOpenMPSimpleClauseTypeName(Node->getClauseKind(), Node->getMotionModifier(I)); if (Node->getMotionModifier(I) == OMPC_MOTION_MODIFIER_mapper) diff --git a/clang/test/OpenMP/target_update_codegen.cpp b/clang/test/OpenMP/target_update_codegen.cpp index c8211f475c7fc..2221c78ea197b 100644 --- a/clang/test/OpenMP/target_update_codegen.cpp +++ b/clang/test/OpenMP/target_update_codegen.cpp @@ -1560,5 +1560,60 @@ 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 +// CK26: [[IDENT_T:%.+]] = type { i32, i32, i32, i32, ptr } +// CK26: @[[ITER:[a-zA-Z0-9_]+]] = internal global i32 0, align 4 +void foo(){ +int a[10]; +#pragma omp target update to(iterator(int it = 0:10) : a[it]) +// CHECK: define dso_local void @_Z4testv() +// CHECK-NEXT: %[[A1:.*]] = alloca \[10 x i32\], align 16 +// CHECK-NEXT: %[[A2:.*]] = alloca \[1 x ptr\], align 8 +// CHECK-NEXT: %[[A3:.*]] = alloca \[1 x ptr\], align 8 +// CHECK-NEXT: %[[A4:.*]] = alloca \[1 x ptr\], align 8 +// CHECK-NEXT: %[[A5:.*]] = alloca \[1 x ptr\], align 8 +// CHECK-NEXT: %[[A6:.*]] = alloca \[1 x ptr\], align 8 +// CHECK-NEXT: %[[A7:.*]] = alloca \[1 x ptr\], align 8 +// CHECK: %[[LOAD1:.*]] = load i32, ptr @_ZZ4testvE16iteratorvariable, align 4 +// CHECK: %[[SEXT1:.*]] = sext i32 %[[LOAD1]] to i64 +// CHECK: %[[GEP1:.*]] = getelementptr inbounds \[10 x i32\], ptr %[[A1]], i64 0, i64 %[[SEXT1]] +// CHECK: %[[GEP2:.*]] = getelementptr inbounds \[1 x ptr\], ptr %[[A2]], i32 0, i32 0 +// CHECK: store ptr %[[A1]], ptr %[[GEP2]], align 8 +// CHECK: %[[GEP3:.*]] = getelementptr inbounds \[1 x ptr\], ptr %[[A3]], i32 0, i32 0 +// CHECK: store ptr %[[GEP1]], ptr %[[GEP3]], align 8 +// CHECK: %[[GEP4:.*]] = getelementptr inbounds \[1 x ptr\], ptr %[[A4]], i64 0, i64 0 +// CHECK: store ptr null, ptr %[[GEP4]], align 8 +// CHECK: %[[GEP5:.*]] = getelementptr inbounds \[1 x ptr\], ptr %[[A2]], i32 0, i32 0 +// CHECK: %[[GEP6:.*]] = getelementptr inbounds \[1 x ptr\], ptr %[[A3]], i32 0, i32 0 +// CHECK: call void @__tgt_target_data_update_mapper(ptr @2, i64 -1, i32 1, ptr %[[GEP5]], ptr %[[GEP6]], ptr @.offload_sizes, ptr @.offload_maptypes, ptr @.offload_mapnames, ptr null) +// CHECK: %[[LOAD2:.*]] = load i32, ptr @[[ITER]], align 4 +// CHECK: %[[SEXT2:.*]] = sext i32 %[[LOAD2]] to i64 +// CHECK: %[[GEP7:.*]] = getelementptr inbounds \[10 x i32\], ptr %[[A1]], i64 0, i64 %[[SEXT2]] +// CHECK: %[[GEP8:.*]] = getelementptr inbounds \[1 x ptr\], ptr %[[A5]], i32 0, i32 0 +// CHECK: store ptr %[[A1]], ptr %[[GEP8]], align 8 +// CHECK: %[[GEP9:.*]] = getelementptr inbounds \[1 x ptr\], ptr %[[A6]], i32 0, i32 0 +// CHECK: store ptr %[[GEP7]], ptr %[[GEP9]], align 8 +// CHECK: %[[GEP10:.*]] = getelementptr inbounds \[1 x ptr\], ptr %[[A7]], i64 0, i64 0 +// CHECK: store ptr null, ptr %[[GEP10]], align 8 +// CHECK: %[[GEP11:.*]] = getelementptr inbounds \[1 x ptr\], ptr %[[A5]], i32 0, i32 0 +// CHECK: %[[GEP12:.*]] = getelementptr inbounds \[1 x ptr\], ptr %[[A6]], i32 0, i32 0 +// CHECK: call void @__tgt_target_data_update_mapper(ptr @4, i64 -1, i32 1, ptr %[[GEP11]], ptr %[[GEP12]], ptr @.offload_sizes.1, ptr @.offload_maptypes.2, ptr @.offload_mapnames.3, ptr null) +// CHECK: ret void +} #endif #endif >From b0cc40d7b63dc01e5e09bacfd0cb7440b7a1c7a1 Mon Sep 17 00:00:00 2001 From: Shashwathi N <[email protected]> Date: Tue, 7 Oct 2025 04:08:41 -0500 Subject: [PATCH 6/6] updated --- clang/test/OpenMP/target_update_codegen.cpp | 35 ++------------------- 1 file changed, 2 insertions(+), 33 deletions(-) diff --git a/clang/test/OpenMP/target_update_codegen.cpp b/clang/test/OpenMP/target_update_codegen.cpp index 2221c78ea197b..c39a77a0c4935 100644 --- a/clang/test/OpenMP/target_update_codegen.cpp +++ b/clang/test/OpenMP/target_update_codegen.cpp @@ -1581,39 +1581,8 @@ void foo(int arg) { void foo(){ int a[10]; #pragma omp target update to(iterator(int it = 0:10) : a[it]) -// CHECK: define dso_local void @_Z4testv() -// CHECK-NEXT: %[[A1:.*]] = alloca \[10 x i32\], align 16 -// CHECK-NEXT: %[[A2:.*]] = alloca \[1 x ptr\], align 8 -// CHECK-NEXT: %[[A3:.*]] = alloca \[1 x ptr\], align 8 -// CHECK-NEXT: %[[A4:.*]] = alloca \[1 x ptr\], align 8 -// CHECK-NEXT: %[[A5:.*]] = alloca \[1 x ptr\], align 8 -// CHECK-NEXT: %[[A6:.*]] = alloca \[1 x ptr\], align 8 -// CHECK-NEXT: %[[A7:.*]] = alloca \[1 x ptr\], align 8 -// CHECK: %[[LOAD1:.*]] = load i32, ptr @_ZZ4testvE16iteratorvariable, align 4 -// CHECK: %[[SEXT1:.*]] = sext i32 %[[LOAD1]] to i64 -// CHECK: %[[GEP1:.*]] = getelementptr inbounds \[10 x i32\], ptr %[[A1]], i64 0, i64 %[[SEXT1]] -// CHECK: %[[GEP2:.*]] = getelementptr inbounds \[1 x ptr\], ptr %[[A2]], i32 0, i32 0 -// CHECK: store ptr %[[A1]], ptr %[[GEP2]], align 8 -// CHECK: %[[GEP3:.*]] = getelementptr inbounds \[1 x ptr\], ptr %[[A3]], i32 0, i32 0 -// CHECK: store ptr %[[GEP1]], ptr %[[GEP3]], align 8 -// CHECK: %[[GEP4:.*]] = getelementptr inbounds \[1 x ptr\], ptr %[[A4]], i64 0, i64 0 -// CHECK: store ptr null, ptr %[[GEP4]], align 8 -// CHECK: %[[GEP5:.*]] = getelementptr inbounds \[1 x ptr\], ptr %[[A2]], i32 0, i32 0 -// CHECK: %[[GEP6:.*]] = getelementptr inbounds \[1 x ptr\], ptr %[[A3]], i32 0, i32 0 -// CHECK: call void @__tgt_target_data_update_mapper(ptr @2, i64 -1, i32 1, ptr %[[GEP5]], ptr %[[GEP6]], ptr @.offload_sizes, ptr @.offload_maptypes, ptr @.offload_mapnames, ptr null) -// CHECK: %[[LOAD2:.*]] = load i32, ptr @[[ITER]], align 4 -// CHECK: %[[SEXT2:.*]] = sext i32 %[[LOAD2]] to i64 -// CHECK: %[[GEP7:.*]] = getelementptr inbounds \[10 x i32\], ptr %[[A1]], i64 0, i64 %[[SEXT2]] -// CHECK: %[[GEP8:.*]] = getelementptr inbounds \[1 x ptr\], ptr %[[A5]], i32 0, i32 0 -// CHECK: store ptr %[[A1]], ptr %[[GEP8]], align 8 -// CHECK: %[[GEP9:.*]] = getelementptr inbounds \[1 x ptr\], ptr %[[A6]], i32 0, i32 0 -// CHECK: store ptr %[[GEP7]], ptr %[[GEP9]], align 8 -// CHECK: %[[GEP10:.*]] = getelementptr inbounds \[1 x ptr\], ptr %[[A7]], i64 0, i64 0 -// CHECK: store ptr null, ptr %[[GEP10]], align 8 -// CHECK: %[[GEP11:.*]] = getelementptr inbounds \[1 x ptr\], ptr %[[A5]], i32 0, i32 0 -// CHECK: %[[GEP12:.*]] = getelementptr inbounds \[1 x ptr\], ptr %[[A6]], i32 0, i32 0 -// CHECK: call void @__tgt_target_data_update_mapper(ptr @4, i64 -1, i32 1, ptr %[[GEP11]], ptr %[[GEP12]], ptr @.offload_sizes.1, ptr @.offload_maptypes.2, ptr @.offload_mapnames.3, ptr null) -// CHECK: ret void +// CK26-LABEL: define {{.+}}foo +// CK26: %[[LOAD2:.*]] = load i32, ptr @[[ITER]], align 4 } #endif #endif _______________________________________________ cfe-commits mailing list [email protected] https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
