diff --git a/clang/include/clang/AST/OpenMPClause.h b/clang/include/clang/AST/OpenMPClause.h --- a/clang/include/clang/AST/OpenMPClause.h +++ b/clang/include/clang/AST/OpenMPClause.h @@ -7127,6 +7127,131 @@ } }; +/// This represents clause 'uses_allocators' in the '#pragma omp target'-based +/// directives. +/// +/// \code +/// #pragma omp target uses_allocators(default_allocator, my_allocator(traits)) +/// \endcode +/// In this example directive '#pragma omp target' has clause 'uses_allocators' +/// with the allocators 'default_allocator' and user-defined 'my_allocator'. +class OMPUsesAllocatorsClause final + : public OMPClause, + private llvm::TrailingObjects { +public: + /// Data for list of allocators. + struct Data { + /// Allocator. + Expr *Allocator = nullptr; + /// Allocator traits. + Expr *AllocatorTraits = nullptr; + /// Locations of '(' and ')' symbols. + SourceLocation LParenLoc, RParenLoc; + }; + +private: + friend class OMPClauseReader; + friend TrailingObjects; + + enum class ExprOffsets { + Allocator, + AllocatorTraits, + Total, + }; + + enum class ParenLocsOffsets { + LParen, + RParen, + Total, + }; + + /// Location of '('. + SourceLocation LParenLoc; + /// Total number of allocators in the clause. + unsigned NumOfAllocators = 0; + + /// Build clause. + /// + /// \param StartLoc Starting location of the clause. + /// \param LParenLoc Location of '('. + /// \param EndLoc Ending location of the clause. + /// \param N Number of allocators asssociated with the clause. + OMPUsesAllocatorsClause(SourceLocation StartLoc, SourceLocation LParenLoc, + SourceLocation EndLoc, unsigned N) + : OMPClause(llvm::omp::OMPC_uses_allocators, StartLoc, EndLoc), + LParenLoc(LParenLoc), NumOfAllocators(N) {} + + /// Build an empty clause. + /// \param N Number of allocators asssociated with the clause. + /// + explicit OMPUsesAllocatorsClause(unsigned N) + : OMPClause(llvm::omp::OMPC_uses_allocators, SourceLocation(), + SourceLocation()), + NumOfAllocators(N) {} + + unsigned numTrailingObjects(OverloadToken) const { + return NumOfAllocators * static_cast(ExprOffsets::Total); + } + + /// Sets the location of '('. + void setLParenLoc(SourceLocation Loc) { LParenLoc = Loc; } + + /// Sets the allocators data for the clause. + void setAllocatorsData(ArrayRef Data); + +public: + /// Creates clause with a list of allocators \p Data. + /// + /// \param C AST context. + /// \param StartLoc Starting location of the clause. + /// \param LParenLoc Location of '('. + /// \param EndLoc Ending location of the clause. + /// \param Data List of allocators. + static OMPUsesAllocatorsClause * + Create(const ASTContext &C, SourceLocation StartLoc, SourceLocation LParenLoc, + SourceLocation EndLoc, ArrayRef Data); + + /// Creates an empty clause with the place for \p N allocators. + /// + /// \param C AST context. + /// \param N The number of allocators. + static OMPUsesAllocatorsClause *CreateEmpty(const ASTContext &C, unsigned N); + + /// Returns the location of '('. + SourceLocation getLParenLoc() const { return LParenLoc; } + + /// Returns number of allocators associated with the clause. + unsigned getNumberOfAllocators() const { return NumOfAllocators; } + + /// Returns data for the specified allocator. + OMPUsesAllocatorsClause::Data getAllocatorData(unsigned I) const; + + // Iterators + child_range children() { + Stmt **Begin = reinterpret_cast(getTrailingObjects()); + return child_range(Begin, Begin + NumOfAllocators * + static_cast(ExprOffsets::Total)); + } + const_child_range children() const { + Stmt *const *Begin = + reinterpret_cast(getTrailingObjects()); + return const_child_range( + Begin, Begin + NumOfAllocators * static_cast(ExprOffsets::Total)); + } + + child_range used_children() { + return child_range(child_iterator(), child_iterator()); + } + const_child_range used_children() const { + return const_child_range(const_child_iterator(), const_child_iterator()); + } + + static bool classof(const OMPClause *T) { + return T->getClauseKind() == llvm::omp::OMPC_uses_allocators; + } +}; + /// This class implements a simple visitor for OMPClause /// subclasses. template class Ptr, typename RetTy> diff --git a/clang/include/clang/AST/RecursiveASTVisitor.h b/clang/include/clang/AST/RecursiveASTVisitor.h --- a/clang/include/clang/AST/RecursiveASTVisitor.h +++ b/clang/include/clang/AST/RecursiveASTVisitor.h @@ -3522,6 +3522,17 @@ return true; } +template +bool RecursiveASTVisitor::VisitOMPUsesAllocatorsClause( + OMPUsesAllocatorsClause *C) { + for (unsigned I = 0, E = C->getNumberOfAllocators(); I < E; ++I) { + const OMPUsesAllocatorsClause::Data Data = C->getAllocatorData(I); + TRY_TO(TraverseStmt(Data.Allocator)); + TRY_TO(TraverseStmt(Data.AllocatorTraits)); + } + return true; +} + // FIXME: look at the following tricky-seeming exprs to see if we // need to recurse on anything. These are ones that have methods // returning decls or qualtypes or nestednamespecifier -- though I'm diff --git a/clang/include/clang/Basic/DiagnosticSemaKinds.td b/clang/include/clang/Basic/DiagnosticSemaKinds.td --- a/clang/include/clang/Basic/DiagnosticSemaKinds.td +++ b/clang/include/clang/Basic/DiagnosticSemaKinds.td @@ -10221,8 +10221,8 @@ "memory order clause '%0' is specified here">; def err_omp_non_lvalue_in_map_or_motion_clauses: Error< "expected addressable lvalue in '%0' clause">; -def err_omp_event_var_expected : Error< - "expected variable of the 'omp_event_handle_t' type%select{|, not %1}0">; +def err_omp_var_expected : Error< + "expected variable of the '%0' type%select{|, not %2}1">; def warn_nested_declare_variant : Warning<"nesting `omp begin/end declare variant` is not supported yet; " "nested context ignored">, @@ -10233,6 +10233,19 @@ def err_omp_reduction_task_not_parallel_or_worksharing : Error< "'reduction' clause with 'task' modifier allowed only on non-simd parallel or" " worksharing constructs">; +def err_omp_expected_array_alloctraits : Error< + "expected constant sized array of 'omp_alloctrait_t' elements, not %0">; +def err_omp_predefined_allocator_with_traits : Error< + "predefined allocator cannot have traits specified">; +def note_omp_predefined_allocator : Note< + "predefined trait '%0' used here">; +def err_omp_nonpredefined_allocator_without_traits : Error< + "non-predefined allocator must have traits specified">; +def err_omp_allocator_used_in_clauses : Error< + "allocators used in 'uses_allocators' clause cannot appear in other " + "data-sharing or data-mapping attribute clauses">; +def err_omp_allocator_not_in_uses_allocators : Error< + "allocator must be specified in the 'uses_allocators' clause">; } // end of OpenMP category let CategoryName = "Related Result Type Issue" in { diff --git a/clang/include/clang/Parse/Parser.h b/clang/include/clang/Parse/Parser.h --- a/clang/include/clang/Parse/Parser.h +++ b/clang/include/clang/Parse/Parser.h @@ -3111,6 +3111,11 @@ /// }+ ')' ExprResult ParseOpenMPIteratorsExpr(); + /// Parses allocators and traits in the context of the uses_allocator clause. + /// Expected format: + /// '(' { [ '(' ')' ] }+ ')' + OMPClause *ParseOpenMPUsesAllocatorClause(OpenMPDirectiveKind DKind); + public: /// Parses simple expression in parens for single-expression clauses of OpenMP /// constructs. diff --git a/clang/include/clang/Sema/Sema.h b/clang/include/clang/Sema/Sema.h --- a/clang/include/clang/Sema/Sema.h +++ b/clang/include/clang/Sema/Sema.h @@ -10766,6 +10766,21 @@ SourceLocation LParenLoc, SourceLocation EndLoc); + /// Data for list of allocators. + struct UsesAllocatorsData { + /// Allocator. + Expr *Allocator = nullptr; + /// Allocator traits. + Expr *AllocatorTraits = nullptr; + /// Locations of '(' and ')' symbols. + SourceLocation LParenLoc, RParenLoc; + }; + /// Called on well-formed 'uses_allocators' clause. + OMPClause *ActOnOpenMPUsesAllocatorClause(SourceLocation StartLoc, + SourceLocation LParenLoc, + SourceLocation EndLoc, + ArrayRef Data); + /// The kind of conversion being performed. enum CheckedConversionKind { /// An implicit conversion. diff --git a/clang/lib/AST/OpenMPClause.cpp b/clang/lib/AST/OpenMPClause.cpp --- a/clang/lib/AST/OpenMPClause.cpp +++ b/clang/lib/AST/OpenMPClause.cpp @@ -150,6 +150,7 @@ case OMPC_detach: case OMPC_inclusive: case OMPC_exclusive: + case OMPC_uses_allocators: break; } @@ -239,6 +240,7 @@ case OMPC_detach: case OMPC_inclusive: case OMPC_exclusive: + case OMPC_uses_allocators: break; } @@ -1302,6 +1304,70 @@ return new (Mem) OMPExclusiveClause(N); } +void OMPUsesAllocatorsClause::setAllocatorsData( + ArrayRef Data) { + assert(Data.size() == NumOfAllocators && + "Size of allocators data is not the same as the preallocated buffer."); + for (unsigned I = 0, E = Data.size(); I < E; ++I) { + const OMPUsesAllocatorsClause::Data &D = Data[I]; + getTrailingObjects()[I * static_cast(ExprOffsets::Total) + + static_cast(ExprOffsets::Allocator)] = + D.Allocator; + getTrailingObjects()[I * static_cast(ExprOffsets::Total) + + static_cast( + ExprOffsets::AllocatorTraits)] = + D.AllocatorTraits; + getTrailingObjects< + SourceLocation>()[I * static_cast(ParenLocsOffsets::Total) + + static_cast(ParenLocsOffsets::LParen)] = + D.LParenLoc; + getTrailingObjects< + SourceLocation>()[I * static_cast(ParenLocsOffsets::Total) + + static_cast(ParenLocsOffsets::RParen)] = + D.RParenLoc; + } +} + +OMPUsesAllocatorsClause::Data +OMPUsesAllocatorsClause::getAllocatorData(unsigned I) const { + OMPUsesAllocatorsClause::Data Data; + Data.Allocator = + getTrailingObjects()[I * static_cast(ExprOffsets::Total) + + static_cast(ExprOffsets::Allocator)]; + Data.AllocatorTraits = + getTrailingObjects()[I * static_cast(ExprOffsets::Total) + + static_cast( + ExprOffsets::AllocatorTraits)]; + Data.LParenLoc = getTrailingObjects< + SourceLocation>()[I * static_cast(ParenLocsOffsets::Total) + + static_cast(ParenLocsOffsets::LParen)]; + Data.RParenLoc = getTrailingObjects< + SourceLocation>()[I * static_cast(ParenLocsOffsets::Total) + + static_cast(ParenLocsOffsets::RParen)]; + return Data; +} + +OMPUsesAllocatorsClause * +OMPUsesAllocatorsClause::Create(const ASTContext &C, SourceLocation StartLoc, + SourceLocation LParenLoc, SourceLocation EndLoc, + ArrayRef Data) { + void *Mem = C.Allocate(totalSizeToAlloc( + static_cast(ExprOffsets::Total) * Data.size(), + static_cast(ParenLocsOffsets::Total) * Data.size())); + auto *Clause = new (Mem) + OMPUsesAllocatorsClause(StartLoc, LParenLoc, EndLoc, Data.size()); + Clause->setAllocatorsData(Data); + return Clause; +} + +OMPUsesAllocatorsClause * +OMPUsesAllocatorsClause::CreateEmpty(const ASTContext &C, unsigned N) { + void *Mem = C.Allocate(totalSizeToAlloc( + static_cast(ExprOffsets::Total) * N, + static_cast(ParenLocsOffsets::Total) * N)); + return new (Mem) OMPUsesAllocatorsClause(N); +} + //===----------------------------------------------------------------------===// // OpenMP clauses printing methods //===----------------------------------------------------------------------===// @@ -1884,6 +1950,25 @@ } } +void OMPClausePrinter::VisitOMPUsesAllocatorsClause( + OMPUsesAllocatorsClause *Node) { + if (Node->getNumberOfAllocators() == 0) + return; + OS << "uses_allocators("; + for (unsigned I = 0, E = Node->getNumberOfAllocators(); I < E; ++I) { + OMPUsesAllocatorsClause::Data Data = Node->getAllocatorData(I); + Data.Allocator->printPretty(OS, nullptr, Policy); + if (Data.AllocatorTraits) { + OS << "("; + Data.AllocatorTraits->printPretty(OS, nullptr, Policy); + OS << ")"; + } + if (I < E - 1) + OS << ","; + } + OS << ")"; +} + void OMPTraitInfo::getAsVariantMatchInfo(ASTContext &ASTCtx, VariantMatchInfo &VMI) const { for (const OMPTraitSet &Set : Sets) { diff --git a/clang/lib/AST/StmtProfile.cpp b/clang/lib/AST/StmtProfile.cpp --- a/clang/lib/AST/StmtProfile.cpp +++ b/clang/lib/AST/StmtProfile.cpp @@ -800,6 +800,15 @@ void OMPClauseProfiler::VisitOMPExclusiveClause(const OMPExclusiveClause *C) { VisitOMPClauseList(C); } +void OMPClauseProfiler::VisitOMPUsesAllocatorsClause( + const OMPUsesAllocatorsClause *C) { + for (unsigned I = 0, E = C->getNumberOfAllocators(); I < E; ++I) { + OMPUsesAllocatorsClause::Data D = C->getAllocatorData(I); + Profiler->VisitStmt(D.Allocator); + if (D.AllocatorTraits) + Profiler->VisitStmt(D.AllocatorTraits); + } +} void OMPClauseProfiler::VisitOMPOrderClause(const OMPOrderClause *C) {} } // namespace diff --git a/clang/lib/Basic/OpenMPKinds.cpp b/clang/lib/Basic/OpenMPKinds.cpp --- a/clang/lib/Basic/OpenMPKinds.cpp +++ b/clang/lib/Basic/OpenMPKinds.cpp @@ -174,6 +174,7 @@ case OMPC_detach: case OMPC_inclusive: case OMPC_exclusive: + case OMPC_uses_allocators: break; } llvm_unreachable("Invalid OpenMP simple clause kind"); @@ -420,6 +421,7 @@ case OMPC_detach: case OMPC_inclusive: case OMPC_exclusive: + case OMPC_uses_allocators: break; } llvm_unreachable("Invalid OpenMP simple clause kind"); diff --git a/clang/lib/CodeGen/CGStmtOpenMP.cpp b/clang/lib/CodeGen/CGStmtOpenMP.cpp --- a/clang/lib/CodeGen/CGStmtOpenMP.cpp +++ b/clang/lib/CodeGen/CGStmtOpenMP.cpp @@ -4627,6 +4627,7 @@ case OMPC_detach: case OMPC_inclusive: case OMPC_exclusive: + case OMPC_uses_allocators: llvm_unreachable("Clause is not allowed in 'omp atomic'."); } } diff --git a/clang/lib/Parse/ParseOpenMP.cpp b/clang/lib/Parse/ParseOpenMP.cpp --- a/clang/lib/Parse/ParseOpenMP.cpp +++ b/clang/lib/Parse/ParseOpenMP.cpp @@ -2436,6 +2436,50 @@ return !IsCorrect; } +OMPClause *Parser::ParseOpenMPUsesAllocatorClause(OpenMPDirectiveKind DKind) { + SourceLocation Loc = Tok.getLocation(); + ConsumeAnyToken(); + + // Parse '('. + BalancedDelimiterTracker T(*this, tok::l_paren, tok::annot_pragma_openmp_end); + if (T.expectAndConsume(diag::err_expected_lparen_after, "uses_allocator")) + return nullptr; + SmallVector Data; + do { + ExprResult Allocator = ParseCXXIdExpression(); + if (Allocator.isInvalid()) { + SkipUntil(tok::comma, tok::r_paren, tok::annot_pragma_openmp_end, + StopBeforeMatch); + break; + } + Sema::UsesAllocatorsData &D = Data.emplace_back(); + D.Allocator = Allocator.get(); + if (Tok.is(tok::l_paren)) { + BalancedDelimiterTracker T(*this, tok::l_paren, + tok::annot_pragma_openmp_end); + T.consumeOpen(); + ExprResult AllocatorTraits = ParseCXXIdExpression(); + T.consumeClose(); + if (AllocatorTraits.isInvalid()) { + SkipUntil(tok::comma, tok::r_paren, tok::annot_pragma_openmp_end, + StopBeforeMatch); + break; + } + D.AllocatorTraits = AllocatorTraits.get(); + D.LParenLoc = T.getOpenLocation(); + D.RParenLoc = T.getCloseLocation(); + } + if (Tok.isNot(tok::comma) && Tok.isNot(tok::r_paren)) + Diag(Tok, diag::err_omp_expected_punc) << "uses_allocators" << 0; + // Parse ',' + if (Tok.is(tok::comma)) + ConsumeAnyToken(); + } while (Tok.isNot(tok::r_paren) && Tok.isNot(tok::annot_pragma_openmp_end)); + T.consumeClose(); + return Actions.ActOnOpenMPUsesAllocatorClause(Loc, T.getOpenLocation(), + T.getCloseLocation(), Data); +} + /// Parsing of OpenMP clauses. /// /// clause: @@ -2453,7 +2497,7 @@ /// in_reduction-clause | allocator-clause | allocate-clause | /// acq_rel-clause | acquire-clause | release-clause | relaxed-clause | /// depobj-clause | destroy-clause | detach-clause | inclusive-clause | -/// exclusive-clause +/// exclusive-clause | uses_allocators-clause /// OMPClause *Parser::ParseOpenMPClause(OpenMPDirectiveKind DKind, OpenMPClauseKind CKind, bool FirstClause) { @@ -2626,6 +2670,9 @@ case OMPC_exclusive: Clause = ParseOpenMPVarListClause(DKind, CKind, WrongDirective); break; + case OMPC_uses_allocators: + Clause = ParseOpenMPUsesAllocatorClause(DKind); + break; case OMPC_device_type: case OMPC_unknown: skipUntilPragmaOpenMPEnd(DKind); diff --git a/clang/lib/Sema/SemaOpenMP.cpp b/clang/lib/Sema/SemaOpenMP.cpp --- a/clang/lib/Sema/SemaOpenMP.cpp +++ b/clang/lib/Sema/SemaOpenMP.cpp @@ -170,6 +170,7 @@ llvm::SmallVector DeclareTargetLinkVarDecls; /// List of decls used in inclusive/exclusive clauses of the scan directive. llvm::DenseSet> UsedInScanDirective; + llvm::DenseSet> UsesAllocatorsDecls; SharingMapTy(OpenMPDirectiveKind DKind, DeclarationNameInfo Name, Scope *CurScope, SourceLocation Loc) : Directive(DKind), DirectiveName(Name), CurScope(CurScope), @@ -279,6 +280,8 @@ QualType OMPDependT; /// omp_event_handle_t type. QualType OMPEventHandleT; + /// omp_alloctrait_t type. + QualType OMPAlloctraitT; /// Expression for the predefined allocators. Expr *OMPPredefinedAllocators[OMPAllocateDeclAttr::OMPUserDefinedMemAlloc] = { nullptr}; @@ -293,6 +296,10 @@ void setOMPAllocatorHandleT(QualType Ty) { OMPAllocatorHandleT = Ty; } /// Gets omp_allocator_handle_t type. QualType getOMPAllocatorHandleT() const { return OMPAllocatorHandleT; } + /// Sets omp_alloctrait_t type. + void setOMPAlloctraitT(QualType Ty) { OMPAlloctraitT = Ty; } + /// Gets omp_alloctrait_t type. + QualType getOMPAlloctraitT() const { return OMPAlloctraitT; } /// Sets the given default allocator. void setAllocator(OMPAllocateDeclAttr::AllocatorTypeTy AllocatorKind, Expr *Allocator) { @@ -1006,6 +1013,19 @@ bool isImplicitTaskFirstprivate(Decl *D) const { return getTopOfStack().ImplicitTaskFirstprivates.count(D) > 0; } + + /// Marks decl as used in uses_allocators clause as the allocator. + void addUsesAllocatorsDecl(const Decl *D) { + getTopOfStack().UsesAllocatorsDecls.insert(D); + } + /// Checks if specified decl is used in uses allocator clause as the + /// allocator. + bool isUsesAllocatorsDecl(unsigned Level, const Decl *D) const { + return getStackElemAtLevel(Level).UsesAllocatorsDecls.count(D) > 0; + } + bool isUsesAllocatorsDecl(const Decl *D) const { + return getTopOfStack().UsesAllocatorsDecls.count(D) > 0; + } }; bool isImplicitTaskingRegion(OpenMPDirectiveKind DKind) { @@ -1990,10 +2010,13 @@ ((IsVariableUsedInMapClause && DSAStack->getCaptureRegion(Level, OpenMPCaptureLevel) == OMPD_target) || - !DSAStack->hasExplicitDSA( - D, - [](OpenMPClauseKind K) -> bool { return K == OMPC_firstprivate; }, - Level, /*NotLastprivate=*/true)) && + !(DSAStack->hasExplicitDSA( + D, + [](OpenMPClauseKind K) -> bool { + return K == OMPC_firstprivate; + }, + Level, /*NotLastprivate=*/true) || + DSAStack->isUsesAllocatorsDecl(Level, D))) && // If the variable is artificial and must be captured by value - try to // capture by value. !(isa(D) && !D->hasAttr() && @@ -2435,6 +2458,11 @@ static DeclRefExpr *buildCapture(Sema &S, ValueDecl *D, Expr *CaptureExpr, bool WithInit); +static void reportOriginalDsa(Sema &SemaRef, const DSAStackTy *Stack, + const ValueDecl *D, + const DSAStackTy::DSAVarData &DVar, + bool IsLoopIterVar = false); + void Sema::EndOpenMPDSABlock(Stmt *CurDirective) { // OpenMP [2.14.3.5, Restrictions, C/C++, p.1] // A variable of class type (or array thereof) that appears in a lastprivate @@ -2504,6 +2532,51 @@ Clause->setPrivateRefs(PrivateRefs); continue; } + if (auto *Clause = dyn_cast(C)) { + for (unsigned I = 0, E = Clause->getNumberOfAllocators(); I < E; ++I) { + OMPUsesAllocatorsClause::Data D = Clause->getAllocatorData(I); + auto *DRE = dyn_cast(D.Allocator->IgnoreParenImpCasts()); + if (!DRE) + continue; + ValueDecl *VD = DRE->getDecl(); + if (!VD) + continue; + DSAStackTy::DSAVarData DVar = + DSAStack->getTopDSA(VD, /*FromParent=*/false); + // OpenMP [2.12.5, target Construct] + // Memory allocators that appear in a uses_allocators clause cannot + // appear in other data-sharing attribute clauses or data-mapping + // attribute clauses in the same construct. + Expr *MapExpr = nullptr; + if (DVar.RefExpr || + DSAStack->checkMappableExprComponentListsForDecl( + VD, /*CurrentRegionOnly=*/true, + [VD, &MapExpr]( + OMPClauseMappableExprCommon::MappableExprComponentListRef + MapExprComponents, + OpenMPClauseKind C) { + auto MI = MapExprComponents.rbegin(); + auto ME = MapExprComponents.rend(); + if (MI != ME && + MI->getAssociatedDeclaration()->getCanonicalDecl() == + VD->getCanonicalDecl()) { + MapExpr = MI->getAssociatedExpression(); + return true; + } + return false; + })) { + Diag(D.Allocator->getExprLoc(), + diag::err_omp_allocator_used_in_clauses) + << D.Allocator->getSourceRange(); + if (DVar.RefExpr) + reportOriginalDsa(*this, DSAStack, VD, DVar); + else + Diag(MapExpr->getExprLoc(), diag::note_used_here) + << MapExpr->getSourceRange(); + } + } + continue; + } } // Check allocate clauses. if (!CurContext->isDependentContext()) @@ -3032,7 +3105,7 @@ static void reportOriginalDsa(Sema &SemaRef, const DSAStackTy *Stack, const ValueDecl *D, const DSAStackTy::DSAVarData &DVar, - bool IsLoopIterVar = false) { + bool IsLoopIterVar) { if (DVar.RefExpr) { SemaRef.Diag(DVar.RefExpr->getExprLoc(), diag::note_omp_explicit_dsa) << getOpenMPClauseName(DVar.CKind); @@ -3187,6 +3260,9 @@ if (VD->hasLocalStorage() && CS && !CS->capturesVariable(VD) && !Stack->isImplicitTaskFirstprivate(VD)) return; + // Skip allocators in uses_allocators clauses. + if (Stack->isUsesAllocatorsDecl(VD)) + return; DSAStackTy::DSAVarData DVar = Stack->getTopDSA(VD, /*FromParent=*/false); // Check if the variable has explicit DSA set and stop analysis if it so. @@ -4640,6 +4716,27 @@ getCanonicalDecl(DE ? DE->getDecl() : ME->getMemberDecl()), false); } +namespace { +/// Checks if the allocator is used in uses_allocators clause to be allowed in +/// target regions. +class AllocatorChecker final : public ConstStmtVisitor { + DSAStackTy *S = nullptr; + +public: + bool VisitDeclRefExpr(const DeclRefExpr *E) { + return !S->isUsesAllocatorsDecl(E->getDecl()); + } + bool VisitStmt(const Stmt *S) { + for (const Stmt *Child : S->children()) { + if (Child && Visit(Child)) + return true; + } + return false; + } + explicit AllocatorChecker(DSAStackTy *S) : S(S) {} +}; +} // namespace + static void checkAllocateClauses(Sema &S, DSAStackTy *Stack, ArrayRef Clauses) { assert(!S.CurContext->isDependentContext() && @@ -4708,6 +4805,22 @@ } for (OMPClause *C : AllocateRange) { auto *AC = cast(C); + if (S.getLangOpts().OpenMP >= 50 && + !Stack->hasRequiresDeclWithClause() && + isOpenMPTargetExecutionDirective(Stack->getCurrentDirective()) && + AC->getAllocator()) { + Expr *Allocator = AC->getAllocator(); + // OpenMP, 2.12.5 target Construct + // Memory allocators that do not appear in a uses_allocators clause cannot + // appear as an allocator in an allocate clause or be used in the target + // region unless a requires directive with the dynamic_allocators clause + // is present in the same compilation unit. + AllocatorChecker Checker(Stack); + if (Checker.Visit(Allocator)) + S.Diag(Allocator->getExprLoc(), + diag::err_omp_allocator_not_in_uses_allocators) + << Allocator->getSourceRange(); + } OMPAllocateDeclAttr::AllocatorTypeTy AllocatorKind = getAllocatorKind(S, Stack, AC->getAllocator()); // OpenMP, 2.11.4 allocate Clause, Restrictions. @@ -5239,6 +5352,7 @@ case OMPC_destroy: case OMPC_inclusive: case OMPC_exclusive: + case OMPC_uses_allocators: continue; case OMPC_allocator: case OMPC_flush: @@ -11365,6 +11479,7 @@ case OMPC_destroy: case OMPC_inclusive: case OMPC_exclusive: + case OMPC_uses_allocators: llvm_unreachable("Clause is not allowed."); } return Res; @@ -12118,6 +12233,7 @@ case OMPC_detach: case OMPC_inclusive: case OMPC_exclusive: + case OMPC_uses_allocators: llvm_unreachable("Unexpected OpenMP clause."); } return CaptureRegion; @@ -12557,6 +12673,7 @@ case OMPC_detach: case OMPC_inclusive: case OMPC_exclusive: + case OMPC_uses_allocators: llvm_unreachable("Clause is not allowed."); } return Res; @@ -12782,6 +12899,7 @@ case OMPC_detach: case OMPC_inclusive: case OMPC_exclusive: + case OMPC_uses_allocators: llvm_unreachable("Clause is not allowed."); } return Res; @@ -13014,6 +13132,7 @@ case OMPC_detach: case OMPC_inclusive: case OMPC_exclusive: + case OMPC_uses_allocators: llvm_unreachable("Clause is not allowed."); } return Res; @@ -13279,6 +13398,7 @@ case OMPC_order: case OMPC_destroy: case OMPC_detach: + case OMPC_uses_allocators: llvm_unreachable("Clause is not allowed."); } return Res; @@ -17719,21 +17839,22 @@ // event-handle is a variable of the omp_event_handle_t type. auto *Ref = dyn_cast(Evt->IgnoreParenImpCasts()); if (!Ref) { - Diag(Evt->getExprLoc(), diag::err_omp_event_var_expected) - << 0 << Evt->getSourceRange(); + Diag(Evt->getExprLoc(), diag::err_omp_var_expected) + << "omp_event_handle_t" << 0 << Evt->getSourceRange(); return nullptr; } auto *VD = dyn_cast_or_null(Ref->getDecl()); if (!VD) { - Diag(Evt->getExprLoc(), diag::err_omp_event_var_expected) - << 0 << Evt->getSourceRange(); + Diag(Evt->getExprLoc(), diag::err_omp_var_expected) + << "omp_event_handle_t" << 0 << Evt->getSourceRange(); return nullptr; } if (!Context.hasSameUnqualifiedType(DSAStack->getOMPEventHandleT(), VD->getType()) || VD->getType().isConstant(Context)) { - Diag(Evt->getExprLoc(), diag::err_omp_event_var_expected) - << 1 << VD->getType() << Evt->getSourceRange(); + Diag(Evt->getExprLoc(), diag::err_omp_var_expected) + << "omp_event_handle_t" << 1 << VD->getType() + << Evt->getSourceRange(); return nullptr; } // OpenMP 5.0, 2.10.1 task Construct @@ -18450,3 +18571,134 @@ return OMPExclusiveClause::Create(Context, StartLoc, LParenLoc, EndLoc, Vars); } + +/// Tries to find omp_alloctrait_t type. +static bool findOMPAlloctraitT(Sema &S, SourceLocation Loc, DSAStackTy *Stack) { + QualType OMPAlloctraitT = Stack->getOMPAlloctraitT(); + if (!OMPAlloctraitT.isNull()) + return true; + IdentifierInfo &II = S.PP.getIdentifierTable().get("omp_alloctrait_t"); + ParsedType PT = S.getTypeName(II, Loc, S.getCurScope()); + if (!PT.getAsOpaquePtr() || PT.get().isNull()) { + S.Diag(Loc, diag::err_omp_implied_type_not_found) << "omp_alloctrait_t"; + return false; + } + Stack->setOMPAlloctraitT(PT.get()); + return true; +} + +OMPClause *Sema::ActOnOpenMPUsesAllocatorClause( + SourceLocation StartLoc, SourceLocation LParenLoc, SourceLocation EndLoc, + ArrayRef Data) { + // OpenMP [2.12.5, target Construct] + // allocator is an identifier of omp_allocator_handle_t type. + if (!findOMPAllocatorHandleT(*this, StartLoc, DSAStack)) + return nullptr; + // OpenMP [2.12.5, target Construct] + // allocator-traits-array is an identifier of const omp_alloctrait_t * type. + if (llvm::any_of( + Data, + [](const UsesAllocatorsData &D) { return D.AllocatorTraits; }) && + !findOMPAlloctraitT(*this, StartLoc, DSAStack)) + return nullptr; + llvm::SmallSet, 4> PredefinedAllocators; + for (int I = OMPAllocateDeclAttr::OMPDefaultMemAlloc; + I < OMPAllocateDeclAttr::OMPUserDefinedMemAlloc; ++I) { + auto AllocatorKind = static_cast(I); + StringRef Allocator = + OMPAllocateDeclAttr::ConvertAllocatorTypeTyToStr(AllocatorKind); + DeclarationName AllocatorName = &Context.Idents.get(Allocator); + PredefinedAllocators.insert(LookupSingleName( + TUScope, AllocatorName, StartLoc, Sema::LookupAnyName)); + } + + SmallVector NewData; + for (const UsesAllocatorsData &D : Data) { + Expr *AllocatorExpr = nullptr; + // Check allocator expression. + if (D.Allocator->isTypeDependent()) { + AllocatorExpr = D.Allocator; + } else { + // Traits were specified - need to assign new allocator to the specified + // allocator, so it must be an lvalue. + AllocatorExpr = D.Allocator->IgnoreParenImpCasts(); + auto *DRE = dyn_cast(AllocatorExpr); + bool IsPredefinedAllocator = false; + if (DRE) + IsPredefinedAllocator = PredefinedAllocators.count(DRE->getDecl()); + if (!DRE || + !(Context.hasSameUnqualifiedType( + AllocatorExpr->getType(), DSAStack->getOMPAllocatorHandleT()) || + Context.typesAreCompatible(AllocatorExpr->getType(), + DSAStack->getOMPAllocatorHandleT(), + /*CompareUnqualified=*/true)) || + (!IsPredefinedAllocator && + (AllocatorExpr->getType().isConstant(Context) || + !AllocatorExpr->isLValue()))) { + Diag(D.Allocator->getExprLoc(), diag::err_omp_var_expected) + << "omp_allocator_handle_t" << (DRE ? 1 : 0) + << AllocatorExpr->getType() << D.Allocator->getSourceRange(); + continue; + } + // OpenMP [2.12.5, target Construct] + // Predefined allocators appearing in a uses_allocators clause cannot have + // traits specified. + if (IsPredefinedAllocator && D.AllocatorTraits) { + Diag(D.AllocatorTraits->getExprLoc(), + diag::err_omp_predefined_allocator_with_traits) + << D.AllocatorTraits->getSourceRange(); + Diag(D.Allocator->getExprLoc(), diag::note_omp_predefined_allocator) + << cast(DRE->getDecl())->getName() + << D.Allocator->getSourceRange(); + continue; + } + // OpenMP [2.12.5, target Construct] + // Non-predefined allocators appearing in a uses_allocators clause must + // have traits specified. + if (!IsPredefinedAllocator && !D.AllocatorTraits) { + Diag(D.Allocator->getExprLoc(), + diag::err_omp_nonpredefined_allocator_without_traits); + continue; + } + // No allocator traits - just convert it to rvalue. + if (!D.AllocatorTraits) + AllocatorExpr = DefaultLvalueConversion(AllocatorExpr).get(); + DSAStack->addUsesAllocatorsDecl(DRE->getDecl()); + } + Expr *AllocatorTraitsExpr = nullptr; + if (D.AllocatorTraits) { + if (D.AllocatorTraits->isTypeDependent()) { + AllocatorTraitsExpr = D.AllocatorTraits; + } else { + // OpenMP [2.12.5, target Construct] + // Arrays that contain allocator traits that appear in a uses_allocators + // clause must be constant arrays, have constant values and be defined + // in the same scope as the construct in which the clause appears. + AllocatorTraitsExpr = D.AllocatorTraits->IgnoreParenImpCasts(); + // Check that traits expr is a constant array. + QualType TraitTy; + if (const ArrayType *Ty = + AllocatorTraitsExpr->getType()->getAsArrayTypeUnsafe()) + if (const auto *ConstArrayTy = dyn_cast(Ty)) + TraitTy = ConstArrayTy->getElementType(); + if (TraitTy.isNull() || + !(Context.hasSameUnqualifiedType(TraitTy, + DSAStack->getOMPAlloctraitT()) || + Context.typesAreCompatible(TraitTy, DSAStack->getOMPAlloctraitT(), + /*CompareUnqualified=*/true))) { + Diag(D.AllocatorTraits->getExprLoc(), + diag::err_omp_expected_array_alloctraits) + << AllocatorTraitsExpr->getType(); + continue; + } + } + } + OMPUsesAllocatorsClause::Data &NewD = NewData.emplace_back(); + NewD.Allocator = AllocatorExpr; + NewD.AllocatorTraits = AllocatorTraitsExpr; + NewD.LParenLoc = D.LParenLoc; + NewD.RParenLoc = D.RParenLoc; + } + return OMPUsesAllocatorsClause::Create(Context, StartLoc, LParenLoc, EndLoc, + NewData); +} diff --git a/clang/lib/Sema/TreeTransform.h b/clang/lib/Sema/TreeTransform.h --- a/clang/lib/Sema/TreeTransform.h +++ b/clang/lib/Sema/TreeTransform.h @@ -2086,6 +2086,17 @@ EndLoc); } + /// Build a new OpenMP 'uses_allocators' clause. + /// + /// By default, performs semantic analysis to build the new OpenMP clause. + /// Subclasses may override this routine to provide different behavior. + OMPClause *RebuildOMPUsesAllocatorsClause( + ArrayRef Data, SourceLocation StartLoc, + SourceLocation LParenLoc, SourceLocation EndLoc) { + return getSema().ActOnOpenMPUsesAllocatorClause(StartLoc, LParenLoc, EndLoc, + Data); + } + /// Build a new OpenMP 'order' clause. /// /// By default, performs semantic analysis to build the new OpenMP clause. @@ -9673,8 +9684,33 @@ } template -OMPClause * -TreeTransform::TransformOMPOrderClause(OMPOrderClause *C) { +OMPClause *TreeTransform::TransformOMPUsesAllocatorsClause( + OMPUsesAllocatorsClause *C) { + SmallVector Data; + Data.reserve(C->getNumberOfAllocators()); + for (unsigned I = 0, E = C->getNumberOfAllocators(); I < E; ++I) { + OMPUsesAllocatorsClause::Data D = C->getAllocatorData(I); + ExprResult Allocator = getDerived().TransformExpr(D.Allocator); + if (Allocator.isInvalid()) + continue; + ExprResult AllocatorTraits; + if (Expr *AT = D.AllocatorTraits) { + AllocatorTraits = getDerived().TransformExpr(AT); + if (AllocatorTraits.isInvalid()) + continue; + } + Sema::UsesAllocatorsData &NewD = Data.emplace_back(); + NewD.Allocator = Allocator.get(); + NewD.AllocatorTraits = AllocatorTraits.get(); + NewD.LParenLoc = D.LParenLoc; + NewD.RParenLoc = D.RParenLoc; + } + return getDerived().RebuildOMPUsesAllocatorsClause( + Data, C->getBeginLoc(), C->getLParenLoc(), C->getEndLoc()); +} + +template +OMPClause *TreeTransform::TransformOMPOrderClause(OMPOrderClause *C) { return getDerived().RebuildOMPOrderClause(C->getKind(), C->getKindKwLoc(), C->getBeginLoc(), C->getLParenLoc(), C->getEndLoc()); diff --git a/clang/lib/Serialization/ASTReader.cpp b/clang/lib/Serialization/ASTReader.cpp --- a/clang/lib/Serialization/ASTReader.cpp +++ b/clang/lib/Serialization/ASTReader.cpp @@ -11882,6 +11882,9 @@ case llvm::omp::OMPC_detach: C = new (Context) OMPDetachClause(); break; + case llvm::omp::OMPC_uses_allocators: + C = OMPUsesAllocatorsClause::CreateEmpty(Context, Record.readInt()); + break; #define OMP_CLAUSE_NO_CLASS(Enum, Str) \ case llvm::omp::Enum: \ break; @@ -12710,6 +12713,21 @@ C->setVarRefs(Vars); } +void OMPClauseReader::VisitOMPUsesAllocatorsClause(OMPUsesAllocatorsClause *C) { + C->setLParenLoc(Record.readSourceLocation()); + unsigned NumOfAllocators = C->getNumberOfAllocators(); + SmallVector Data; + Data.reserve(NumOfAllocators); + for (unsigned I = 0; I != NumOfAllocators; ++I) { + OMPUsesAllocatorsClause::Data &D = Data.emplace_back(); + D.Allocator = Record.readSubExpr(); + D.AllocatorTraits = Record.readSubExpr(); + D.LParenLoc = Record.readSourceLocation(); + D.RParenLoc = Record.readSourceLocation(); + } + C->setAllocatorsData(Data); +} + void OMPClauseReader::VisitOMPOrderClause(OMPOrderClause *C) { C->setKind(Record.readEnum()); C->setLParenLoc(Record.readSourceLocation()); diff --git a/clang/lib/Serialization/ASTWriter.cpp b/clang/lib/Serialization/ASTWriter.cpp --- a/clang/lib/Serialization/ASTWriter.cpp +++ b/clang/lib/Serialization/ASTWriter.cpp @@ -6651,6 +6651,18 @@ Record.AddSourceLocation(C->getKindKwLoc()); } +void OMPClauseWriter::VisitOMPUsesAllocatorsClause(OMPUsesAllocatorsClause *C) { + Record.push_back(C->getNumberOfAllocators()); + Record.AddSourceLocation(C->getLParenLoc()); + for (unsigned I = 0, E = C->getNumberOfAllocators(); I < E; ++I) { + OMPUsesAllocatorsClause::Data Data = C->getAllocatorData(I); + Record.AddStmt(Data.Allocator); + Record.AddStmt(Data.AllocatorTraits); + Record.AddSourceLocation(Data.LParenLoc); + Record.AddSourceLocation(Data.RParenLoc); + } +} + void ASTRecordWriter::writeOMPTraitInfo(const OMPTraitInfo *TI) { writeUInt32(TI->Sets.size()); for (const auto &Set : TI->Sets) { diff --git a/clang/test/OpenMP/target_ast_print.cpp b/clang/test/OpenMP/target_ast_print.cpp --- a/clang/test/OpenMP/target_ast_print.cpp +++ b/clang/test/OpenMP/target_ast_print.cpp @@ -341,6 +341,16 @@ // RUN: %clang_cc1 -DOMP5 -fopenmp-simd -fopenmp-version=50 -x c++ -std=c++11 -emit-pch -o %t %s // RUN: %clang_cc1 -DOMP5 -fopenmp-simd -fopenmp-version=50 -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print | FileCheck %s --check-prefix OMP5 +typedef void **omp_allocator_handle_t; +extern const omp_allocator_handle_t omp_default_mem_alloc; +extern const omp_allocator_handle_t omp_large_cap_mem_alloc; +extern const omp_allocator_handle_t omp_const_mem_alloc; +extern const omp_allocator_handle_t omp_high_bw_mem_alloc; +extern const omp_allocator_handle_t omp_low_lat_mem_alloc; +extern const omp_allocator_handle_t omp_cgroup_mem_alloc; +extern const omp_allocator_handle_t omp_pteam_mem_alloc; +extern const omp_allocator_handle_t omp_thread_mem_alloc; + void foo() {} #pragma omp declare target @@ -1050,8 +1060,12 @@ // OMP5-NEXT: bar(); #pragma omp target defaultmap(none) // OMP5-NEXT: #pragma omp target defaultmap(none) + // OMP5-NEXT: bar(); bar(); +#pragma omp target allocate(omp_default_mem_alloc:argv) uses_allocators(omp_default_mem_alloc,omp_large_cap_mem_alloc) allocate(omp_large_cap_mem_alloc:argc) private(argc, argv) + // OMP5-NEXT: #pragma omp target allocate(omp_default_mem_alloc: argv) uses_allocators(omp_default_mem_alloc,omp_large_cap_mem_alloc) allocate(omp_large_cap_mem_alloc: argc) private(argc,argv) // OMP5-NEXT: bar(); + bar(); return tmain(argc, &argc) + tmain(argv[0][0], argv[0]); } diff --git a/clang/test/OpenMP/target_firstprivate_messages.cpp b/clang/test/OpenMP/target_firstprivate_messages.cpp --- a/clang/test/OpenMP/target_firstprivate_messages.cpp +++ b/clang/test/OpenMP/target_firstprivate_messages.cpp @@ -1,6 +1,6 @@ -// RUN: %clang_cc1 -verify -fopenmp %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=50 %s -Wuninitialized -// RUN: %clang_cc1 -verify -fopenmp-simd %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=50 %s -Wuninitialized void xxx(int argc) { int fp, fp1; // expected-note {{initialize the variable 'fp' to silence this warning}} expected-note {{initialize the variable 'fp1' to silence this warning}} @@ -128,7 +128,7 @@ {} #pragma omp target firstprivate(argv[1]) // expected-error {{expected variable name}} {} -#pragma omp target firstprivate(e, g) allocate(omp_thread_mem_alloc: e) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target' directive}} +#pragma omp target firstprivate(e, g) allocate(omp_thread_mem_alloc: e) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target' directive}} expected-error {{allocator must be specified in the 'uses_allocators' clause}} {} #pragma omp target firstprivate(h) // expected-error {{threadprivate or thread local variable cannot be firstprivate}} {} diff --git a/clang/test/OpenMP/target_parallel_ast_print.cpp b/clang/test/OpenMP/target_parallel_ast_print.cpp --- a/clang/test/OpenMP/target_parallel_ast_print.cpp +++ b/clang/test/OpenMP/target_parallel_ast_print.cpp @@ -10,6 +10,16 @@ #ifndef HEADER #define HEADER +typedef void **omp_allocator_handle_t; +extern const omp_allocator_handle_t omp_default_mem_alloc; +extern const omp_allocator_handle_t omp_large_cap_mem_alloc; +extern const omp_allocator_handle_t omp_const_mem_alloc; +extern const omp_allocator_handle_t omp_high_bw_mem_alloc; +extern const omp_allocator_handle_t omp_low_lat_mem_alloc; +extern const omp_allocator_handle_t omp_cgroup_mem_alloc; +extern const omp_allocator_handle_t omp_pteam_mem_alloc; +extern const omp_allocator_handle_t omp_thread_mem_alloc; + void foo() {} template @@ -41,7 +51,7 @@ T i, j, a[20]; #pragma omp target parallel h=2; -#pragma omp target parallel allocate(argv) default(none), private(argc,b) firstprivate(argv) shared (d) if (parallel:argc > 0) num_threads(C) proc_bind(master) reduction(+:c, arr1[argc]) reduction(max:e, arr[:C][0:10]) +#pragma omp target parallel allocate(omp_large_cap_mem_alloc:argv) default(none), private(argc,b) firstprivate(argv) shared (d) if (parallel:argc > 0) num_threads(C) proc_bind(master) reduction(+:c, arr1[argc]) reduction(max:e, arr[:C][0:10]) uses_allocators(omp_large_cap_mem_alloc) foo(); #pragma omp target parallel if (C) num_threads(s) proc_bind(close) reduction(^:e, f, arr[0:C][:argc]) reduction(&& : g) allocate(g) foo(); @@ -76,7 +86,7 @@ // CHECK-NEXT: T i, j, a[20] // CHECK-NEXT: #pragma omp target parallel{{$}} // CHECK-NEXT: h = 2; -// CHECK-NEXT: #pragma omp target parallel allocate(argv) default(none) private(argc,b) firstprivate(argv) shared(d) if(parallel: argc > 0) num_threads(C) proc_bind(master) reduction(+: c,arr1[argc]) reduction(max: e,arr[:C][0:10]) +// CHECK-NEXT: #pragma omp target parallel allocate(omp_large_cap_mem_alloc: argv) default(none) private(argc,b) firstprivate(argv) shared(d) if(parallel: argc > 0) num_threads(C) proc_bind(master) reduction(+: c,arr1[argc]) reduction(max: e,arr[:C][0:10]) uses_allocators(omp_large_cap_mem_alloc) // CHECK-NEXT: foo() // CHECK-NEXT: #pragma omp target parallel if(C) num_threads(s) proc_bind(close) reduction(^: e,f,arr[0:C][:argc]) reduction(&&: g) allocate(g) // CHECK-NEXT: foo() @@ -108,7 +118,7 @@ // CHECK-NEXT: int i, j, a[20] // CHECK-NEXT: #pragma omp target parallel // CHECK-NEXT: h = 2; -// CHECK-NEXT: #pragma omp target parallel allocate(argv) default(none) private(argc,b) firstprivate(argv) shared(d) if(parallel: argc > 0) num_threads(5) proc_bind(master) reduction(+: c,arr1[argc]) reduction(max: e,arr[:5][0:10]) +// CHECK-NEXT: #pragma omp target parallel allocate(omp_large_cap_mem_alloc: argv) default(none) private(argc,b) firstprivate(argv) shared(d) if(parallel: argc > 0) num_threads(5) proc_bind(master) reduction(+: c,arr1[argc]) reduction(max: e,arr[:5][0:10]) uses_allocators(omp_large_cap_mem_alloc) // CHECK-NEXT: foo() // CHECK-NEXT: #pragma omp target parallel if(5) num_threads(s) proc_bind(close) reduction(^: e,f,arr[0:5][:argc]) reduction(&&: g) allocate(g) // CHECK-NEXT: foo() @@ -140,7 +150,7 @@ // CHECK-NEXT: char i, j, a[20] // CHECK-NEXT: #pragma omp target parallel // CHECK-NEXT: h = 2; -// CHECK-NEXT: #pragma omp target parallel allocate(argv) default(none) private(argc,b) firstprivate(argv) shared(d) if(parallel: argc > 0) num_threads(1) proc_bind(master) reduction(+: c,arr1[argc]) reduction(max: e,arr[:1][0:10]) +// CHECK-NEXT: #pragma omp target parallel allocate(omp_large_cap_mem_alloc: argv) default(none) private(argc,b) firstprivate(argv) shared(d) if(parallel: argc > 0) num_threads(1) proc_bind(master) reduction(+: c,arr1[argc]) reduction(max: e,arr[:1][0:10]) uses_allocators(omp_large_cap_mem_alloc) // CHECK-NEXT: foo() // CHECK-NEXT: #pragma omp target parallel if(1) num_threads(s) proc_bind(close) reduction(^: e,f,arr[0:1][:argc]) reduction(&&: g) allocate(g) // CHECK-NEXT: foo() diff --git a/clang/test/OpenMP/target_parallel_firstprivate_messages.cpp b/clang/test/OpenMP/target_parallel_firstprivate_messages.cpp --- a/clang/test/OpenMP/target_parallel_firstprivate_messages.cpp +++ b/clang/test/OpenMP/target_parallel_firstprivate_messages.cpp @@ -2,6 +2,8 @@ // RUN: %clang_cc1 -verify -fopenmp-simd -ferror-limit 100 %s -Wuninitialized +#pragma omp requires dynamic_allocators + typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; extern const omp_allocator_handle_t omp_large_cap_mem_alloc; diff --git a/clang/test/OpenMP/target_parallel_for_ast_print.cpp b/clang/test/OpenMP/target_parallel_for_ast_print.cpp --- a/clang/test/OpenMP/target_parallel_for_ast_print.cpp +++ b/clang/test/OpenMP/target_parallel_for_ast_print.cpp @@ -10,6 +10,16 @@ #ifndef HEADER #define HEADER +typedef void **omp_allocator_handle_t; +extern const omp_allocator_handle_t omp_default_mem_alloc; +extern const omp_allocator_handle_t omp_large_cap_mem_alloc; +extern const omp_allocator_handle_t omp_const_mem_alloc; +extern const omp_allocator_handle_t omp_high_bw_mem_alloc; +extern const omp_allocator_handle_t omp_low_lat_mem_alloc; +extern const omp_allocator_handle_t omp_cgroup_mem_alloc; +extern const omp_allocator_handle_t omp_pteam_mem_alloc; +extern const omp_allocator_handle_t omp_thread_mem_alloc; + void foo() {} struct S { @@ -195,8 +205,8 @@ // CHECK-NEXT: #pragma omp target parallel for default(none) private(argc,b) firstprivate(argv) shared(d) if(parallel: argc > 0) num_threads(5) proc_bind(master) reduction(+: c,arr1[argc]) reduction(max: e,arr[:5][0:10]) // CHECK-NEXT: for (int i = 0; i < 2; ++i) { // CHECK-NEXT: } -#pragma omp target parallel for if (5) num_threads(s) proc_bind(close) reduction(^:e, f, arr[0:5][:argc]) reduction(&& : h) -// CHECK-NEXT: #pragma omp target parallel for if(5) num_threads(s) proc_bind(close) reduction(^: e,f,arr[0:5][:argc]) reduction(&&: h) +#pragma omp target parallel for if (5) num_threads(s) proc_bind(close) reduction(^:e, f, arr[0:5][:argc]) reduction(&& : h) allocate(omp_const_mem_alloc: h) uses_allocators(omp_const_mem_alloc) +// CHECK-NEXT: #pragma omp target parallel for if(5) num_threads(s) proc_bind(close) reduction(^: e,f,arr[0:5][:argc]) reduction(&&: h) allocate(omp_const_mem_alloc: h) uses_allocators(omp_const_mem_alloc) for (int i = 0; i < 2; ++i) {} // CHECK-NEXT: for (int i = 0; i < 2; ++i) { // CHECK-NEXT: } diff --git a/clang/test/OpenMP/target_parallel_for_firstprivate_messages.cpp b/clang/test/OpenMP/target_parallel_for_firstprivate_messages.cpp --- a/clang/test/OpenMP/target_parallel_for_firstprivate_messages.cpp +++ b/clang/test/OpenMP/target_parallel_for_firstprivate_messages.cpp @@ -2,6 +2,7 @@ // RUN: %clang_cc1 -verify -fopenmp-simd %s -Wuninitialized +#pragma omp requires dynamic_allocators typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; extern const omp_allocator_handle_t omp_large_cap_mem_alloc; diff --git a/clang/test/OpenMP/target_parallel_for_lastprivate_messages.cpp b/clang/test/OpenMP/target_parallel_for_lastprivate_messages.cpp --- a/clang/test/OpenMP/target_parallel_for_lastprivate_messages.cpp +++ b/clang/test/OpenMP/target_parallel_for_lastprivate_messages.cpp @@ -196,7 +196,7 @@ #pragma omp target parallel for lastprivate(2 * 2) // expected-error {{expected variable name}} for (i = 0; i < argc; ++i) foo(); -#pragma omp target parallel for lastprivate(ba) allocate(omp_thread_mem_alloc: ba) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target parallel for' directive}} +#pragma omp target parallel for lastprivate(ba) allocate(omp_thread_mem_alloc: ba) uses_allocators(omp_thread_mem_alloc) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target parallel for' directive}} omp45-error {{unexpected OpenMP clause 'uses_allocators' in directive '#pragma omp target parallel for'}} for (i = 0; i < argc; ++i) foo(); #pragma omp target parallel for lastprivate(ca) // expected-error {{const-qualified variable without mutable fields cannot be lastprivate}} diff --git a/clang/test/OpenMP/target_parallel_for_linear_messages.cpp b/clang/test/OpenMP/target_parallel_for_linear_messages.cpp --- a/clang/test/OpenMP/target_parallel_for_linear_messages.cpp +++ b/clang/test/OpenMP/target_parallel_for_linear_messages.cpp @@ -1,6 +1,6 @@ -// RUN: %clang_cc1 -verify -fopenmp %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=50 %s -Wuninitialized -// RUN: %clang_cc1 -verify -fopenmp-simd %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=50 %s -Wuninitialized typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; @@ -170,7 +170,7 @@ #pragma omp target parallel for linear(argv[1]) // expected-error {{expected variable name}} for (int k = 0; k < argc; ++k) ++k; -#pragma omp target parallel for allocate(omp_thread_mem_alloc: e) linear(e, g) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target parallel for' directive}} +#pragma omp target parallel for allocate(omp_thread_mem_alloc: e) linear(e, g) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target parallel for' directive}} expected-error {{allocator must be specified in the 'uses_allocators' clause}} for (int k = 0; k < argc; ++k) ++k; #pragma omp target parallel for linear(z, h) // expected-error {{threadprivate or thread local variable cannot be linear}} diff --git a/clang/test/OpenMP/target_parallel_for_private_messages.cpp b/clang/test/OpenMP/target_parallel_for_private_messages.cpp --- a/clang/test/OpenMP/target_parallel_for_private_messages.cpp +++ b/clang/test/OpenMP/target_parallel_for_private_messages.cpp @@ -1,6 +1,6 @@ -// RUN: %clang_cc1 -verify -fopenmp %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=50 %s -Wuninitialized -// RUN: %clang_cc1 -verify -fopenmp-simd %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=50 %s -Wuninitialized typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; @@ -138,7 +138,7 @@ #pragma omp target parallel for private(argv[1]) // expected-error {{expected variable name}} for (int k = 0; k < argc; ++k) ++k; -#pragma omp target parallel for private(e, g) allocate(omp_thread_mem_alloc: e) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target parallel for' directive}} +#pragma omp target parallel for private(e, g) allocate(omp_thread_mem_alloc: e) uses_allocators(omp_thread_mem_alloc) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target parallel for' directive}} for (int k = 0; k < argc; ++k) ++k; #pragma omp target parallel for private(h) // expected-error {{threadprivate or thread local variable cannot be private}} diff --git a/clang/test/OpenMP/target_parallel_for_reduction_messages.cpp b/clang/test/OpenMP/target_parallel_for_reduction_messages.cpp --- a/clang/test/OpenMP/target_parallel_for_reduction_messages.cpp +++ b/clang/test/OpenMP/target_parallel_for_reduction_messages.cpp @@ -341,7 +341,7 @@ for (int i = 0; i < 10; ++i) foo(); static int m; -#pragma omp target parallel for allocate(omp_thread_mem_alloc: m) reduction(+ : m) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target parallel for' directive}} +#pragma omp target parallel for allocate(omp_thread_mem_alloc: m) reduction(+ : m) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target parallel for' directive}} omp50-error {{allocator must be specified in the 'uses_allocators' clause}} for (int i = 0; i < 10; ++i) m++; diff --git a/clang/test/OpenMP/target_parallel_for_simd_ast_print.cpp b/clang/test/OpenMP/target_parallel_for_simd_ast_print.cpp --- a/clang/test/OpenMP/target_parallel_for_simd_ast_print.cpp +++ b/clang/test/OpenMP/target_parallel_for_simd_ast_print.cpp @@ -16,6 +16,16 @@ #ifndef HEADER #define HEADER +typedef void **omp_allocator_handle_t; +extern const omp_allocator_handle_t omp_default_mem_alloc; +extern const omp_allocator_handle_t omp_large_cap_mem_alloc; +extern const omp_allocator_handle_t omp_const_mem_alloc; +extern const omp_allocator_handle_t omp_high_bw_mem_alloc; +extern const omp_allocator_handle_t omp_low_lat_mem_alloc; +extern const omp_allocator_handle_t omp_cgroup_mem_alloc; +extern const omp_allocator_handle_t omp_pteam_mem_alloc; +extern const omp_allocator_handle_t omp_thread_mem_alloc; + void foo() {} struct S { @@ -115,13 +125,13 @@ // CHECK-NEXT: } #ifdef OMP5 -#pragma omp target parallel for simd if(target:argc > 0) if (simd: argc) nontemporal(argc, c, d) order(concurrent) +#pragma omp target parallel for simd if(target:argc > 0) if (simd: argc) nontemporal(argc, c, d) order(concurrent) allocate(omp_high_bw_mem_alloc:f) private(f) uses_allocators(omp_high_bw_mem_alloc) #else #pragma omp target parallel for simd if(target:argc > 0) #endif // OMP5 for (T i = 0; i < 2; ++i) {} // OMP45: #pragma omp target parallel for simd if(target: argc > 0) - // OMP50: #pragma omp target parallel for simd if(target: argc > 0) if(simd: argc) nontemporal(argc,c,d) order(concurrent) + // OMP50: #pragma omp target parallel for simd if(target: argc > 0) if(simd: argc) nontemporal(argc,c,d) order(concurrent) allocate(omp_high_bw_mem_alloc: f) private(f) uses_allocators(omp_high_bw_mem_alloc) // CHECK-NEXT: for (T i = 0; i < 2; ++i) { // CHECK-NEXT: } diff --git a/clang/test/OpenMP/target_parallel_for_simd_firstprivate_messages.cpp b/clang/test/OpenMP/target_parallel_for_simd_firstprivate_messages.cpp --- a/clang/test/OpenMP/target_parallel_for_simd_firstprivate_messages.cpp +++ b/clang/test/OpenMP/target_parallel_for_simd_firstprivate_messages.cpp @@ -2,6 +2,8 @@ // RUN: %clang_cc1 -verify -fopenmp-simd %s -Wuninitialized +#pragma omp requires dynamic_allocators + typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; extern const omp_allocator_handle_t omp_large_cap_mem_alloc; diff --git a/clang/test/OpenMP/target_parallel_for_simd_lastprivate_messages.cpp b/clang/test/OpenMP/target_parallel_for_simd_lastprivate_messages.cpp --- a/clang/test/OpenMP/target_parallel_for_simd_lastprivate_messages.cpp +++ b/clang/test/OpenMP/target_parallel_for_simd_lastprivate_messages.cpp @@ -101,7 +101,7 @@ #pragma omp target parallel for simd lastprivate(argc > 0 ? argv[1] : argv[2]) // expected-error {{expected variable name}} for (int k = 0; k < argc; ++k) ++k; -#pragma omp target parallel for simd allocate(omp_thread_mem_alloc: argc) lastprivate(argc) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target parallel for simd' directive}} +#pragma omp target parallel for simd allocate(omp_thread_mem_alloc: argc) lastprivate(argc) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target parallel for simd' directive}} omp50-error {{allocator must be specified in the 'uses_allocators' clause}} for (int k = 0; k < argc; ++k) ++k; #pragma omp target parallel for simd lastprivate(conditional: argc,s) lastprivate(conditional: // omp50-error {{expected expression}} omp45-error 2 {{use of undeclared identifier 'conditional'}} expected-error {{expected ')'}} expected-note {{to match this '('}} omp45-error 2 {{calling a private constructor of class 'S6'}} omp50-error {{expected list item of scalar type in 'lastprivate' clause with 'conditional' modifier}} diff --git a/clang/test/OpenMP/target_parallel_for_simd_linear_messages.cpp b/clang/test/OpenMP/target_parallel_for_simd_linear_messages.cpp --- a/clang/test/OpenMP/target_parallel_for_simd_linear_messages.cpp +++ b/clang/test/OpenMP/target_parallel_for_simd_linear_messages.cpp @@ -2,6 +2,8 @@ // RUN: %clang_cc1 -verify -fopenmp-simd %s -Wuninitialized +#pragma omp requires dynamic_allocators + typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; extern const omp_allocator_handle_t omp_large_cap_mem_alloc; diff --git a/clang/test/OpenMP/target_parallel_for_simd_private_messages.cpp b/clang/test/OpenMP/target_parallel_for_simd_private_messages.cpp --- a/clang/test/OpenMP/target_parallel_for_simd_private_messages.cpp +++ b/clang/test/OpenMP/target_parallel_for_simd_private_messages.cpp @@ -1,6 +1,6 @@ -// RUN: %clang_cc1 -verify -fopenmp %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=50 %s -Wuninitialized -// RUN: %clang_cc1 -verify -fopenmp-simd %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=50 %s -Wuninitialized typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; @@ -68,7 +68,7 @@ S6() : a(0) {} S6(T v) : a(v) { -#pragma omp target parallel for simd allocate(omp_thread_mem_alloc: a) private(a) private(this->a) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target parallel for simd' directive}} +#pragma omp target parallel for simd allocate(omp_thread_mem_alloc: a) private(a) private(this->a) uses_allocators(omp_thread_mem_alloc) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target parallel for simd' directive}} for (int k = 0; k < v; ++k) ++this->a; } diff --git a/clang/test/OpenMP/target_parallel_for_simd_reduction_messages.cpp b/clang/test/OpenMP/target_parallel_for_simd_reduction_messages.cpp --- a/clang/test/OpenMP/target_parallel_for_simd_reduction_messages.cpp +++ b/clang/test/OpenMP/target_parallel_for_simd_reduction_messages.cpp @@ -12,6 +12,8 @@ // RUN: %clang_cc1 -verify=expected,omp50 -fopenmp-simd -fopenmp-version=50 -std=c++98 -ferror-limit 150 -o - %s -Wuninitialized // RUN: %clang_cc1 -verify=expected,omp50 -fopenmp-simd -fopenmp-version=50 -std=c++11 -ferror-limit 150 -o - %s -Wuninitialized +#pragma omp requires dynamic_allocators + typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; extern const omp_allocator_handle_t omp_large_cap_mem_alloc; diff --git a/clang/test/OpenMP/target_parallel_private_messages.cpp b/clang/test/OpenMP/target_parallel_private_messages.cpp --- a/clang/test/OpenMP/target_parallel_private_messages.cpp +++ b/clang/test/OpenMP/target_parallel_private_messages.cpp @@ -1,6 +1,6 @@ -// RUN: %clang_cc1 -verify -fopenmp %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=50 %s -Wuninitialized -// RUN: %clang_cc1 -verify -fopenmp-simd %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=50 %s -Wuninitialized typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; @@ -96,7 +96,7 @@ {} #pragma omp target parallel private(argv[1]) // expected-error {{expected variable name}} {} -#pragma omp target parallel allocate(omp_thread_mem_alloc: ba) private(ba) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target parallel' directive}} +#pragma omp target parallel allocate(omp_thread_mem_alloc: ba) private(ba) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target parallel' directive}} expected-error {{allocator must be specified in the 'uses_allocators' clause}} {} #pragma omp target parallel private(ca) // expected-error {{const-qualified variable without mutable fields cannot be private}} {} diff --git a/clang/test/OpenMP/target_parallel_reduction_messages.cpp b/clang/test/OpenMP/target_parallel_reduction_messages.cpp --- a/clang/test/OpenMP/target_parallel_reduction_messages.cpp +++ b/clang/test/OpenMP/target_parallel_reduction_messages.cpp @@ -179,7 +179,7 @@ #pragma omp for private(fl) for (int i = 0; i < 10; ++i) {} -#pragma omp target parallel reduction(+ : fl) allocate(omp_thread_mem_alloc: fl) // expected-warning 2 {{allocator with the 'thread' trait access has unspecified behavior on 'target parallel' directive}} +#pragma omp target parallel reduction(+ : fl) allocate(omp_thread_mem_alloc: fl) uses_allocators(omp_thread_mem_alloc) // expected-warning 2 {{allocator with the 'thread' trait access has unspecified behavior on 'target parallel' directive}} omp45-error {{unexpected OpenMP clause 'uses_allocators' in directive '#pragma omp target parallel'}} foo(); #pragma omp target parallel #pragma omp for reduction(- : fl) diff --git a/clang/test/OpenMP/target_private_messages.cpp b/clang/test/OpenMP/target_private_messages.cpp --- a/clang/test/OpenMP/target_private_messages.cpp +++ b/clang/test/OpenMP/target_private_messages.cpp @@ -2,6 +2,7 @@ // RUN: %clang_cc1 -verify -fopenmp-simd %s -Wuninitialized +#pragma omp requires dynamic_allocators typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; extern const omp_allocator_handle_t omp_large_cap_mem_alloc; diff --git a/clang/test/OpenMP/target_reduction_messages.cpp b/clang/test/OpenMP/target_reduction_messages.cpp --- a/clang/test/OpenMP/target_reduction_messages.cpp +++ b/clang/test/OpenMP/target_reduction_messages.cpp @@ -177,7 +177,7 @@ #pragma omp parallel #pragma omp for private(fl) for (int i = 0; i < 10; ++i) -#pragma omp target reduction(+ : fl) allocate(omp_thread_mem_alloc: fl) // expected-warning 2 {{allocator with the 'thread' trait access has unspecified behavior on 'target' directive}} +#pragma omp target reduction(+ : fl) allocate(omp_thread_mem_alloc: fl) // expected-warning 2 {{allocator with the 'thread' trait access has unspecified behavior on 'target' directive}} omp50-error 2 {{allocator must be specified in the 'uses_allocators' clause}} foo(); #pragma omp parallel #pragma omp for reduction(- : fl) diff --git a/clang/test/OpenMP/target_simd_ast_print.cpp b/clang/test/OpenMP/target_simd_ast_print.cpp --- a/clang/test/OpenMP/target_simd_ast_print.cpp +++ b/clang/test/OpenMP/target_simd_ast_print.cpp @@ -16,6 +16,16 @@ #ifndef HEADER #define HEADER +typedef void **omp_allocator_handle_t; +extern const omp_allocator_handle_t omp_default_mem_alloc; +extern const omp_allocator_handle_t omp_large_cap_mem_alloc; +extern const omp_allocator_handle_t omp_const_mem_alloc; +extern const omp_allocator_handle_t omp_high_bw_mem_alloc; +extern const omp_allocator_handle_t omp_low_lat_mem_alloc; +extern const omp_allocator_handle_t omp_cgroup_mem_alloc; +extern const omp_allocator_handle_t omp_pteam_mem_alloc; +extern const omp_allocator_handle_t omp_thread_mem_alloc; + void foo() {} struct S { @@ -116,13 +126,13 @@ // CHECK-NEXT: } #ifdef OMP5 -#pragma omp target simd if(target:argc > 0) if (simd:argc) +#pragma omp target simd if(target:argc > 0) if (simd:argc) allocate(omp_default_mem_alloc:f) private(f) uses_allocators(omp_default_mem_alloc) #else #pragma omp target simd if(target:argc > 0) #endif // OMP5 for (T i = 0; i < 2; ++i) {} // OMP45: #pragma omp target simd if(target: argc > 0) - // OMP50: #pragma omp target simd if(target: argc > 0) if(simd: argc) + // OMP50: #pragma omp target simd if(target: argc > 0) if(simd: argc) allocate(omp_default_mem_alloc: f) private(f) uses_allocators(omp_default_mem_alloc) // CHECK-NEXT: for (T i = 0; i < 2; ++i) { // CHECK-NEXT: } diff --git a/clang/test/OpenMP/target_simd_firstprivate_messages.cpp b/clang/test/OpenMP/target_simd_firstprivate_messages.cpp --- a/clang/test/OpenMP/target_simd_firstprivate_messages.cpp +++ b/clang/test/OpenMP/target_simd_firstprivate_messages.cpp @@ -1,6 +1,6 @@ -// RUN: %clang_cc1 -verify -fopenmp %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=50 %s -Wuninitialized -// RUN: %clang_cc1 -verify -fopenmp-simd %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=50 %s -Wuninitialized typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; @@ -125,7 +125,7 @@ { int v = 0; int i; -#pragma omp target simd allocate(omp_thread_mem_alloc: i) firstprivate(i) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target simd' directive}} +#pragma omp target simd allocate(omp_thread_mem_alloc: i) firstprivate(i) uses_allocators(omp_thread_mem_alloc) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target simd' directive}} for (int k = 0; k < argc; ++k) { i = k; v += i; diff --git a/clang/test/OpenMP/target_simd_lastprivate_messages.cpp b/clang/test/OpenMP/target_simd_lastprivate_messages.cpp --- a/clang/test/OpenMP/target_simd_lastprivate_messages.cpp +++ b/clang/test/OpenMP/target_simd_lastprivate_messages.cpp @@ -4,6 +4,7 @@ // RUN: %clang_cc1 -verify=expected,omp45 -fopenmp-version=45 -fopenmp-simd %s -Wuninitialized // RUN: %clang_cc1 -verify=expected,omp50 -fopenmp-version=50 -fopenmp-simd %s -Wuninitialized +#pragma omp requires dynamic_allocators typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; extern const omp_allocator_handle_t omp_large_cap_mem_alloc; diff --git a/clang/test/OpenMP/target_simd_linear_messages.cpp b/clang/test/OpenMP/target_simd_linear_messages.cpp --- a/clang/test/OpenMP/target_simd_linear_messages.cpp +++ b/clang/test/OpenMP/target_simd_linear_messages.cpp @@ -2,6 +2,7 @@ // RUN: %clang_cc1 -verify -fopenmp-simd %s -Wuninitialized +#pragma omp requires dynamic_allocators typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; extern const omp_allocator_handle_t omp_large_cap_mem_alloc; diff --git a/clang/test/OpenMP/target_simd_private_messages.cpp b/clang/test/OpenMP/target_simd_private_messages.cpp --- a/clang/test/OpenMP/target_simd_private_messages.cpp +++ b/clang/test/OpenMP/target_simd_private_messages.cpp @@ -1,6 +1,6 @@ -// RUN: %clang_cc1 -verify -fopenmp %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=50 %s -Wuninitialized -// RUN: %clang_cc1 -verify -fopenmp-simd %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=50 %s -Wuninitialized typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; @@ -68,7 +68,7 @@ S6() : a(0) {} S6(T v) : a(v) { -#pragma omp target simd private(a) private(this->a) allocate(omp_thread_mem_alloc: a) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target simd' directive}} +#pragma omp target simd private(a) private(this->a) allocate(omp_thread_mem_alloc: a) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target simd' directive}} expected-error {{allocator must be specified in the 'uses_allocators' clause}} for (int k = 0; k < v; ++k) ++this->a; } diff --git a/clang/test/OpenMP/target_simd_reduction_messages.cpp b/clang/test/OpenMP/target_simd_reduction_messages.cpp --- a/clang/test/OpenMP/target_simd_reduction_messages.cpp +++ b/clang/test/OpenMP/target_simd_reduction_messages.cpp @@ -37,7 +37,7 @@ } void foobar(int &ref) { -#pragma omp target simd allocate(omp_thread_mem_alloc: ref) reduction(+:ref) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target simd' directive}} +#pragma omp target simd allocate(omp_thread_mem_alloc: ref) reduction(+:ref) uses_allocators(omp_thread_mem_alloc) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target simd' directive}} omp45-error {{unexpected OpenMP clause 'uses_allocators' in directive '#pragma omp target simd'}} for (int i = 0; i < 10; ++i) foo(); } diff --git a/clang/test/OpenMP/target_teams_ast_print.cpp b/clang/test/OpenMP/target_teams_ast_print.cpp --- a/clang/test/OpenMP/target_teams_ast_print.cpp +++ b/clang/test/OpenMP/target_teams_ast_print.cpp @@ -1,15 +1,27 @@ -// RUN: %clang_cc1 -verify -fopenmp -ast-print %s | FileCheck %s -// RUN: %clang_cc1 -fopenmp -x c++ -std=c++11 -emit-pch -o %t %s -// RUN: %clang_cc1 -fopenmp -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print | FileCheck %s +// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=50 -ast-print %s | FileCheck %s +// RUN: %clang_cc1 -fopenmp -fopenmp-version=50 -x c++ -std=c++11 -emit-pch -o %t %s +// RUN: %clang_cc1 -fopenmp -fopenmp-version=50 -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print | FileCheck %s -// RUN: %clang_cc1 -verify -fopenmp-simd -ast-print %s | FileCheck %s -// RUN: %clang_cc1 -fopenmp-simd -x c++ -std=c++11 -emit-pch -o %t %s -// RUN: %clang_cc1 -fopenmp-simd -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print | FileCheck %s +// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=50 -ast-print %s | FileCheck %s +// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -x c++ -std=c++11 -emit-pch -o %t %s +// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print | FileCheck %s // expected-no-diagnostics #ifndef HEADER #define HEADER +struct omp_alloctrait_t {}; + +typedef void **omp_allocator_handle_t; +extern const omp_allocator_handle_t omp_default_mem_alloc; +extern const omp_allocator_handle_t omp_large_cap_mem_alloc; +extern const omp_allocator_handle_t omp_const_mem_alloc; +extern const omp_allocator_handle_t omp_high_bw_mem_alloc; +extern const omp_allocator_handle_t omp_low_lat_mem_alloc; +extern const omp_allocator_handle_t omp_cgroup_mem_alloc; +extern const omp_allocator_handle_t omp_pteam_mem_alloc; +extern const omp_allocator_handle_t omp_thread_mem_alloc; + void foo() {} template @@ -37,11 +49,13 @@ T b = argc, c, d, e, f, g; static T a; S s; + omp_alloctrait_t traits[10]; + omp_allocator_handle_t my_allocator; #pragma omp target teams a=2; #pragma omp target teams default(none), private(argc,b) firstprivate(argv) shared (d) reduction(+:c) reduction(max:e) num_teams(C) thread_limit(d*C) allocate(argv) foo(); -#pragma omp target teams allocate(f) reduction(^:e, f) reduction(&& : g) +#pragma omp target teams allocate(my_allocator:f) reduction(^:e, f) reduction(&& : g) uses_allocators(my_allocator(traits)) foo(); return 0; } @@ -50,31 +64,37 @@ // CHECK-NEXT: T b = argc, c, d, e, f, g; // CHECK-NEXT: static T a; // CHECK-NEXT: S s; +// CHECK-NEXT: omp_alloctrait_t traits[10]; +// CHECK-NEXT: omp_allocator_handle_t my_allocator; // CHECK-NEXT: #pragma omp target teams{{$}} // CHECK-NEXT: a = 2; // CHECK-NEXT: #pragma omp target teams default(none) private(argc,b) firstprivate(argv) shared(d) reduction(+: c) reduction(max: e) num_teams(C) thread_limit(d * C) allocate(argv) // CHECK-NEXT: foo() -// CHECK-NEXT: #pragma omp target teams allocate(f) reduction(^: e,f) reduction(&&: g) +// CHECK-NEXT: #pragma omp target teams allocate(my_allocator: f) reduction(^: e,f) reduction(&&: g) uses_allocators(my_allocator(traits)) // CHECK-NEXT: foo() // CHECK: template<> int tmain(int argc, int *argv) { // CHECK-NEXT: int b = argc, c, d, e, f, g; // CHECK-NEXT: static int a; // CHECK-NEXT: S s; +// CHECK-NEXT: omp_alloctrait_t traits[10]; +// CHECK-NEXT: omp_allocator_handle_t my_allocator; // CHECK-NEXT: #pragma omp target teams // CHECK-NEXT: a = 2; // CHECK-NEXT: #pragma omp target teams default(none) private(argc,b) firstprivate(argv) shared(d) reduction(+: c) reduction(max: e) num_teams(5) thread_limit(d * 5) allocate(argv) // CHECK-NEXT: foo() -// CHECK-NEXT: #pragma omp target teams allocate(f) reduction(^: e,f) reduction(&&: g) +// CHECK-NEXT: #pragma omp target teams allocate(my_allocator: f) reduction(^: e,f) reduction(&&: g) uses_allocators(my_allocator(traits)) // CHECK-NEXT: foo() // CHECK: template<> long tmain(long argc, long *argv) { // CHECK-NEXT: long b = argc, c, d, e, f, g; // CHECK-NEXT: static long a; // CHECK-NEXT: S s; +// CHECK-NEXT: omp_alloctrait_t traits[10]; +// CHECK-NEXT: omp_allocator_handle_t my_allocator; // CHECK-NEXT: #pragma omp target teams // CHECK-NEXT: a = 2; // CHECK-NEXT: #pragma omp target teams default(none) private(argc,b) firstprivate(argv) shared(d) reduction(+: c) reduction(max: e) num_teams(1) thread_limit(d * 1) allocate(argv) // CHECK-NEXT: foo() -// CHECK-NEXT: #pragma omp target teams allocate(f) reduction(^: e,f) reduction(&&: g) +// CHECK-NEXT: #pragma omp target teams allocate(my_allocator: f) reduction(^: e,f) reduction(&&: g) uses_allocators(my_allocator(traits)) // CHECK-NEXT: foo() enum Enum { }; diff --git a/clang/test/OpenMP/target_teams_distribute_ast_print.cpp b/clang/test/OpenMP/target_teams_distribute_ast_print.cpp --- a/clang/test/OpenMP/target_teams_distribute_ast_print.cpp +++ b/clang/test/OpenMP/target_teams_distribute_ast_print.cpp @@ -1,15 +1,25 @@ -// RUN: %clang_cc1 -verify -fopenmp -ast-print %s -Wno-openmp-mapping | FileCheck %s -// RUN: %clang_cc1 -fopenmp -x c++ -std=c++11 -emit-pch -o %t %s -Wno-openmp-mapping -// RUN: %clang_cc1 -fopenmp -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print -Wno-openmp-mapping | FileCheck %s +// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=50 -ast-print %s -Wno-openmp-mapping | FileCheck %s +// RUN: %clang_cc1 -fopenmp -fopenmp-version=50 -x c++ -std=c++11 -emit-pch -o %t %s -Wno-openmp-mapping +// RUN: %clang_cc1 -fopenmp -fopenmp-version=50 -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print -Wno-openmp-mapping | FileCheck %s -// RUN: %clang_cc1 -verify -fopenmp-simd -ast-print %s -Wno-openmp-mapping | FileCheck %s -// RUN: %clang_cc1 -fopenmp-simd -x c++ -std=c++11 -emit-pch -o %t %s -Wno-openmp-mapping -// RUN: %clang_cc1 -fopenmp-simd -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print -Wno-openmp-mapping | FileCheck %s +// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=50 -ast-print %s -Wno-openmp-mapping | FileCheck %s +// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -x c++ -std=c++11 -emit-pch -o %t %s -Wno-openmp-mapping +// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print -Wno-openmp-mapping | FileCheck %s // expected-no-diagnostics #ifndef HEADER #define HEADER +typedef void **omp_allocator_handle_t; +extern const omp_allocator_handle_t omp_default_mem_alloc; +extern const omp_allocator_handle_t omp_large_cap_mem_alloc; +extern const omp_allocator_handle_t omp_const_mem_alloc; +extern const omp_allocator_handle_t omp_high_bw_mem_alloc; +extern const omp_allocator_handle_t omp_low_lat_mem_alloc; +extern const omp_allocator_handle_t omp_cgroup_mem_alloc; +extern const omp_allocator_handle_t omp_pteam_mem_alloc; +extern const omp_allocator_handle_t omp_thread_mem_alloc; + void foo() {} struct S { @@ -40,14 +50,14 @@ void foo() { int b, argv, d, c, e, f; -#pragma omp target teams distribute default(none), private(b) firstprivate(argv) shared(d) reduction(+:c) reduction(max:e) num_teams(f) thread_limit(d) +#pragma omp target teams distribute default(none), private(b) firstprivate(argv) shared(d) reduction(+:c) reduction(max:e) num_teams(f) thread_limit(d) allocate(omp_low_lat_mem_alloc:b) uses_allocators(omp_low_lat_mem_alloc) for (int k = 0; k < a.a; ++k) ++a.a; } }; // CHECK: #pragma omp target teams distribute private(this->a) private(this->a) private(T::a) // CHECK: #pragma omp target teams distribute private(this->a) private(this->a) -// CHECK: #pragma omp target teams distribute default(none) private(b) firstprivate(argv) shared(d) reduction(+: c) reduction(max: e) num_teams(f) thread_limit(d) +// CHECK: #pragma omp target teams distribute default(none) private(b) firstprivate(argv) shared(d) reduction(+: c) reduction(max: e) num_teams(f) thread_limit(d) allocate(omp_low_lat_mem_alloc: b) uses_allocators(omp_low_lat_mem_alloc) // CHECK: #pragma omp target teams distribute private(this->a) private(this->a) private(this->S::a) class S8 : public S7 { diff --git a/clang/test/OpenMP/target_teams_distribute_firstprivate_messages.cpp b/clang/test/OpenMP/target_teams_distribute_firstprivate_messages.cpp --- a/clang/test/OpenMP/target_teams_distribute_firstprivate_messages.cpp +++ b/clang/test/OpenMP/target_teams_distribute_firstprivate_messages.cpp @@ -2,6 +2,7 @@ // RUN: %clang_cc1 -verify -fopenmp-simd %s -Wuninitialized +#pragma omp requires dynamic_allocators typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; extern const omp_allocator_handle_t omp_large_cap_mem_alloc; @@ -30,7 +31,7 @@ extern S1 a; class S2 { mutable int a; - + public: S2() : a(0) {} S2(const S2 &s2) : a(s2.a) {} diff --git a/clang/test/OpenMP/target_teams_distribute_lastprivate_messages.cpp b/clang/test/OpenMP/target_teams_distribute_lastprivate_messages.cpp --- a/clang/test/OpenMP/target_teams_distribute_lastprivate_messages.cpp +++ b/clang/test/OpenMP/target_teams_distribute_lastprivate_messages.cpp @@ -100,7 +100,7 @@ #pragma omp target teams distribute lastprivate(argc > 0 ? argv[1] : argv[2]) // expected-error {{expected variable name}} for (int k = 0; k < argc; ++k) ++k; -#pragma omp target teams distribute allocate(omp_thread_mem_alloc: argc) lastprivate(argc) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target teams distribute' directive}} +#pragma omp target teams distribute allocate(omp_thread_mem_alloc: argc) lastprivate(argc) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target teams distribute' directive}} omp50-error {{allocator must be specified in the 'uses_allocators' clause}} for (int k = 0; k < argc; ++k) ++k; #pragma omp target teams distribute lastprivate(S1) // expected-error {{'S1' does not refer to a value}} diff --git a/clang/test/OpenMP/target_teams_distribute_parallel_for_ast_print.cpp b/clang/test/OpenMP/target_teams_distribute_parallel_for_ast_print.cpp --- a/clang/test/OpenMP/target_teams_distribute_parallel_for_ast_print.cpp +++ b/clang/test/OpenMP/target_teams_distribute_parallel_for_ast_print.cpp @@ -10,6 +10,16 @@ #ifndef HEADER #define HEADER +typedef void **omp_allocator_handle_t; +extern const omp_allocator_handle_t omp_default_mem_alloc; +extern const omp_allocator_handle_t omp_large_cap_mem_alloc; +extern const omp_allocator_handle_t omp_const_mem_alloc; +extern const omp_allocator_handle_t omp_high_bw_mem_alloc; +extern const omp_allocator_handle_t omp_low_lat_mem_alloc; +extern const omp_allocator_handle_t omp_cgroup_mem_alloc; +extern const omp_allocator_handle_t omp_pteam_mem_alloc; +extern const omp_allocator_handle_t omp_thread_mem_alloc; + void foo() {} struct S { @@ -27,7 +37,7 @@ public: S7(typename T::type v) : a(v) { -#pragma omp target teams distribute parallel for private(a) private(this->a) private(T::a) +#pragma omp target teams distribute parallel for private(a) private(this->a) private(T::a) allocate(omp_cgroup_mem_alloc:a) uses_allocators(omp_cgroup_mem_alloc) for (int k = 0; k < a.a; ++k) { ++this->a.a; #pragma omp cancel for @@ -48,7 +58,7 @@ ++a.a; } }; -// CHECK: #pragma omp target teams distribute parallel for private(this->a) private(this->a) private(T::a) +// CHECK: #pragma omp target teams distribute parallel for private(this->a) private(this->a) private(T::a) allocate(omp_cgroup_mem_alloc: this->a) uses_allocators(omp_cgroup_mem_alloc) // CHECK: #pragma omp cancel for // CHECK: #pragma omp target teams distribute parallel for private(this->a) private(this->a) // CHECK: #pragma omp target teams distribute parallel for default(none) private(b) firstprivate(argv) shared(d) reduction(+: c) reduction(max: e) num_teams(f) thread_limit(d) @@ -58,7 +68,7 @@ public: S8(int v) : S7(v){ -#pragma omp target teams distribute parallel for private(a) private(this->a) private(S7::a) +#pragma omp target teams distribute parallel for private(a) private(this->a) private(S7::a) for (int k = 0; k < a.a; ++k) ++this->a.a; } diff --git a/clang/test/OpenMP/target_teams_distribute_parallel_for_firstprivate_messages.cpp b/clang/test/OpenMP/target_teams_distribute_parallel_for_firstprivate_messages.cpp --- a/clang/test/OpenMP/target_teams_distribute_parallel_for_firstprivate_messages.cpp +++ b/clang/test/OpenMP/target_teams_distribute_parallel_for_firstprivate_messages.cpp @@ -1,6 +1,6 @@ -// RUN: %clang_cc1 -verify -fopenmp %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=50 %s -Wuninitialized -// RUN: %clang_cc1 -verify -fopenmp-simd %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=50 %s -Wuninitialized typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; @@ -113,7 +113,7 @@ #pragma omp target teams distribute parallel for firstprivate (argv[1]) // expected-error {{expected variable name}} for (i = 0; i < argc; ++i) foo(); -#pragma omp target teams distribute parallel for firstprivate(ba) allocate(omp_thread_mem_alloc: ba) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target teams distribute parallel for' directive}} +#pragma omp target teams distribute parallel for firstprivate(ba) uses_allocators(omp_thread_mem_alloc) allocate(omp_thread_mem_alloc: ba) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target teams distribute parallel for' directive}} for (i = 0; i < argc; ++i) foo(); #pragma omp target teams distribute parallel for firstprivate(ca) // expected-error {{no matching constructor for initialization of 'S3'}} diff --git a/clang/test/OpenMP/target_teams_distribute_parallel_for_lastprivate_messages.cpp b/clang/test/OpenMP/target_teams_distribute_parallel_for_lastprivate_messages.cpp --- a/clang/test/OpenMP/target_teams_distribute_parallel_for_lastprivate_messages.cpp +++ b/clang/test/OpenMP/target_teams_distribute_parallel_for_lastprivate_messages.cpp @@ -6,6 +6,7 @@ // RUN: %clang_cc1 -verify=expected,le45 -fopenmp-version=45 -fopenmp-simd %s -Wuninitialized // RUN: %clang_cc1 -verify=expected,omp50 -fopenmp-version=50 -fopenmp-simd %s -Wuninitialized +#pragma omp requires dynamic_allocators typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; extern const omp_allocator_handle_t omp_large_cap_mem_alloc; diff --git a/clang/test/OpenMP/target_teams_distribute_parallel_for_private_messages.cpp b/clang/test/OpenMP/target_teams_distribute_parallel_for_private_messages.cpp --- a/clang/test/OpenMP/target_teams_distribute_parallel_for_private_messages.cpp +++ b/clang/test/OpenMP/target_teams_distribute_parallel_for_private_messages.cpp @@ -1,6 +1,6 @@ -// RUN: %clang_cc1 -verify -fopenmp %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=50 %s -Wuninitialized -// RUN: %clang_cc1 -verify -fopenmp-simd %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=50 %s -Wuninitialized typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; @@ -116,7 +116,7 @@ #pragma omp target teams distribute parallel for firstprivate(i), private(i) // expected-error {{firstprivate variable cannot be private}} expected-note {{defined as firstprivate}} for (int k = 0; k < argc; ++k) ++k; -#pragma omp target teams distribute parallel for allocate(omp_thread_mem_alloc: j) private(j) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target teams distribute parallel for' directive}} +#pragma omp target teams distribute parallel for allocate(omp_thread_mem_alloc: j) private(j) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target teams distribute parallel for' directive}} expected-error {{allocator must be specified in the 'uses_allocators' clause}} for (int k = 0; k < argc; ++k) ++k; #pragma omp target teams distribute parallel for reduction(+:i) diff --git a/clang/test/OpenMP/target_teams_distribute_parallel_for_reduction_messages.cpp b/clang/test/OpenMP/target_teams_distribute_parallel_for_reduction_messages.cpp --- a/clang/test/OpenMP/target_teams_distribute_parallel_for_reduction_messages.cpp +++ b/clang/test/OpenMP/target_teams_distribute_parallel_for_reduction_messages.cpp @@ -169,7 +169,7 @@ #pragma omp parallel reduction(min : i) #pragma omp target teams distribute parallel for reduction(max : j) // expected-error 2 {{argument of OpenMP clause 'reduction' must reference the same object in all threads}} for (int j=0; j<100; j++) foo(); -#pragma omp target teams distribute parallel for reduction(+ : fl) allocate(omp_thread_mem_alloc: fl) // expected-warning 2 {{allocator with the 'thread' trait access has unspecified behavior on 'target teams distribute parallel for' directive}} +#pragma omp target teams distribute parallel for reduction(+ : fl) allocate(omp_thread_mem_alloc: fl) uses_allocators(omp_thread_mem_alloc)// expected-warning 2 {{allocator with the 'thread' trait access has unspecified behavior on 'target teams distribute parallel for' directive}} omp45-error {{unexpected OpenMP clause 'uses_allocators' in directive '#pragma omp target teams distribute parallel for'}} for (int j=0; j<100; j++) foo(); return T(); diff --git a/clang/test/OpenMP/target_teams_distribute_parallel_for_simd_ast_print.cpp b/clang/test/OpenMP/target_teams_distribute_parallel_for_simd_ast_print.cpp --- a/clang/test/OpenMP/target_teams_distribute_parallel_for_simd_ast_print.cpp +++ b/clang/test/OpenMP/target_teams_distribute_parallel_for_simd_ast_print.cpp @@ -10,6 +10,16 @@ #ifndef HEADER #define HEADER +typedef void **omp_allocator_handle_t; +extern const omp_allocator_handle_t omp_default_mem_alloc; +extern const omp_allocator_handle_t omp_large_cap_mem_alloc; +extern const omp_allocator_handle_t omp_const_mem_alloc; +extern const omp_allocator_handle_t omp_high_bw_mem_alloc; +extern const omp_allocator_handle_t omp_low_lat_mem_alloc; +extern const omp_allocator_handle_t omp_cgroup_mem_alloc; +extern const omp_allocator_handle_t omp_pteam_mem_alloc; +extern const omp_allocator_handle_t omp_thread_mem_alloc; + void foo() {} struct S { @@ -33,7 +43,7 @@ } S7 &operator=(S7 &s) { int k; -#pragma omp target teams distribute parallel for simd allocate(a) private(a) private(this->a) linear(k) allocate(k) +#pragma omp target teams distribute parallel for simd allocate(omp_pteam_mem_alloc: a) private(a) private(this->a) linear(k) allocate(k) uses_allocators(omp_pteam_mem_alloc) for (k = 0; k < s.a.a; ++k) ++s.a.a; @@ -58,7 +68,7 @@ } }; // CHECK: #pragma omp target teams distribute parallel for simd private(this->a) private(this->a) private(T::a) -// CHECK: #pragma omp target teams distribute parallel for simd allocate(this->a) private(this->a) private(this->a) linear(k) allocate(k) +// CHECK: #pragma omp target teams distribute parallel for simd allocate(omp_pteam_mem_alloc: this->a) private(this->a) private(this->a) linear(k) allocate(k) uses_allocators(omp_pteam_mem_alloc) // CHECK: #pragma omp target teams distribute parallel for simd default(none) private(b) firstprivate(argv) shared(d) reduction(+: c) reduction(max: e) num_teams(f) thread_limit(d) // CHECK: #pragma omp target teams distribute parallel for simd simdlen(slen1) safelen(slen2) aligned(arr: alen) diff --git a/clang/test/OpenMP/target_teams_distribute_parallel_for_simd_firstprivate_messages.cpp b/clang/test/OpenMP/target_teams_distribute_parallel_for_simd_firstprivate_messages.cpp --- a/clang/test/OpenMP/target_teams_distribute_parallel_for_simd_firstprivate_messages.cpp +++ b/clang/test/OpenMP/target_teams_distribute_parallel_for_simd_firstprivate_messages.cpp @@ -2,6 +2,7 @@ // RUN: %clang_cc1 -verify -fopenmp-simd %s -Wuninitialized +#pragma omp requires dynamic_allocators typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; extern const omp_allocator_handle_t omp_large_cap_mem_alloc; diff --git a/clang/test/OpenMP/target_teams_distribute_parallel_for_simd_lastprivate_messages.cpp b/clang/test/OpenMP/target_teams_distribute_parallel_for_simd_lastprivate_messages.cpp --- a/clang/test/OpenMP/target_teams_distribute_parallel_for_simd_lastprivate_messages.cpp +++ b/clang/test/OpenMP/target_teams_distribute_parallel_for_simd_lastprivate_messages.cpp @@ -1,10 +1,10 @@ // RUN: %clang_cc1 -verify=expected,le45 -fopenmp %s -Wuninitialized // RUN: %clang_cc1 -verify=expected,le45 -fopenmp-version=40 -fopenmp %s -Wuninitialized // RUN: %clang_cc1 -verify=expected,le45 -fopenmp-version=45 -fopenmp %s -Wuninitialized -// RUN: %clang_cc1 -verify=expected -fopenmp-version=50 -fopenmp %s -Wuninitialized +// RUN: %clang_cc1 -verify=expected,le50 -fopenmp-version=50 -fopenmp %s -Wuninitialized // RUN: %clang_cc1 -verify=expected,le45 -fopenmp-simd %s -Wuninitialized -// RUN: %clang_cc1 -verify=expected -fopenmp-version=50 -fopenmp-simd %s -Wuninitialized +// RUN: %clang_cc1 -verify=expected,le50 -fopenmp-version=50 -fopenmp-simd %s -Wuninitialized typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; @@ -124,7 +124,7 @@ for (int k = 0; k < argc; ++k) ++k; int v = 0; -#pragma omp target teams distribute parallel for simd lastprivate(i) allocate(omp_thread_mem_alloc: i) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target teams distribute parallel for simd' directive}} +#pragma omp target teams distribute parallel for simd lastprivate(i) allocate(omp_thread_mem_alloc: i) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target teams distribute parallel for simd' directive}} le50-error {{allocator must be specified in the 'uses_allocators' clause}} for (int k = 0; k < argc; ++k) { i = k; v += i; diff --git a/clang/test/OpenMP/target_teams_distribute_parallel_for_simd_private_messages.cpp b/clang/test/OpenMP/target_teams_distribute_parallel_for_simd_private_messages.cpp --- a/clang/test/OpenMP/target_teams_distribute_parallel_for_simd_private_messages.cpp +++ b/clang/test/OpenMP/target_teams_distribute_parallel_for_simd_private_messages.cpp @@ -1,6 +1,6 @@ -// RUN: %clang_cc1 -verify -fopenmp %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=50 %s -Wuninitialized -// RUN: %clang_cc1 -verify -fopenmp-simd %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=50 %s -Wuninitialized typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; @@ -92,7 +92,7 @@ #pragma omp target teams distribute parallel for simd private (argv[1]) // expected-error {{expected variable name}} for (int k = 0; k < argc; ++k) ++k; - #pragma omp target teams distribute parallel for simd private(ba) allocate(omp_thread_mem_alloc: ba) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target teams distribute parallel for simd' directive}} + #pragma omp target teams distribute parallel for simd private(ba) uses_allocators(omp_thread_mem_alloc) allocate(omp_thread_mem_alloc: ba) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target teams distribute parallel for simd' directive}} for (int k = 0; k < argc; ++k) ++k; #pragma omp target teams distribute parallel for simd private(ca) // expected-error {{const-qualified variable without mutable fields cannot be private}} diff --git a/clang/test/OpenMP/target_teams_distribute_parallel_for_simd_reduction_messages.cpp b/clang/test/OpenMP/target_teams_distribute_parallel_for_simd_reduction_messages.cpp --- a/clang/test/OpenMP/target_teams_distribute_parallel_for_simd_reduction_messages.cpp +++ b/clang/test/OpenMP/target_teams_distribute_parallel_for_simd_reduction_messages.cpp @@ -12,6 +12,7 @@ // RUN: %clang_cc1 -verify=expected,omp50 -fopenmp-simd -fopenmp-version=50 -std=c++98 %s -Wuninitialized // RUN: %clang_cc1 -verify=expected,omp50 -fopenmp-simd -fopenmp-version=50 -std=c++11 %s -Wuninitialized +#pragma omp requires dynamic_allocators typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; extern const omp_allocator_handle_t omp_large_cap_mem_alloc; diff --git a/clang/test/OpenMP/target_teams_distribute_private_messages.cpp b/clang/test/OpenMP/target_teams_distribute_private_messages.cpp --- a/clang/test/OpenMP/target_teams_distribute_private_messages.cpp +++ b/clang/test/OpenMP/target_teams_distribute_private_messages.cpp @@ -1,6 +1,6 @@ -// RUN: %clang_cc1 -verify -fopenmp %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=50 %s -Wuninitialized -// RUN: %clang_cc1 -verify -fopenmp-simd %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=50 %s -Wuninitialized typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; @@ -92,7 +92,7 @@ #pragma omp target teams distribute private (argv[1]) // expected-error {{expected variable name}} for (int k = 0; k < argc; ++k) ++k; -#pragma omp target teams distribute private(ba) allocate(omp_thread_mem_alloc: ba) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target teams distribute' directive}} +#pragma omp target teams distribute private(ba) allocate(omp_thread_mem_alloc: ba) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target teams distribute' directive}} expected-error {{allocator must be specified in the 'uses_allocators' clause}} for (int k = 0; k < argc; ++k) ++k; #pragma omp target teams distribute private(ca) // expected-error {{const-qualified variable without mutable fields cannot be private}} diff --git a/clang/test/OpenMP/target_teams_distribute_reduction_messages.cpp b/clang/test/OpenMP/target_teams_distribute_reduction_messages.cpp --- a/clang/test/OpenMP/target_teams_distribute_reduction_messages.cpp +++ b/clang/test/OpenMP/target_teams_distribute_reduction_messages.cpp @@ -174,7 +174,7 @@ #pragma omp parallel reduction(min : i) #pragma omp target teams distribute reduction(max : j) // expected-error 2 {{argument of OpenMP clause 'reduction' must reference the same object in all threads}} for (int j=0; j<100; j++) foo(); -#pragma omp target teams distribute allocate(omp_thread_mem_alloc: fl) reduction(+ : fl) // expected-warning 2 {{allocator with the 'thread' trait access has unspecified behavior on 'target teams distribute' directive}} +#pragma omp target teams distribute uses_allocators(omp_thread_mem_alloc) allocate(omp_thread_mem_alloc: fl) reduction(+ : fl) // expected-warning 2 {{allocator with the 'thread' trait access has unspecified behavior on 'target teams distribute' directive}} omp45-error {{unexpected OpenMP clause 'uses_allocators' in directive '#pragma omp target teams distribute'}} for (int j=0; j<100; j++) foo(); return T(); diff --git a/clang/test/OpenMP/target_teams_distribute_simd_ast_print.cpp b/clang/test/OpenMP/target_teams_distribute_simd_ast_print.cpp --- a/clang/test/OpenMP/target_teams_distribute_simd_ast_print.cpp +++ b/clang/test/OpenMP/target_teams_distribute_simd_ast_print.cpp @@ -16,6 +16,16 @@ #ifndef HEADER #define HEADER +typedef void **omp_allocator_handle_t; +extern const omp_allocator_handle_t omp_default_mem_alloc; +extern const omp_allocator_handle_t omp_large_cap_mem_alloc; +extern const omp_allocator_handle_t omp_const_mem_alloc; +extern const omp_allocator_handle_t omp_high_bw_mem_alloc; +extern const omp_allocator_handle_t omp_low_lat_mem_alloc; +extern const omp_allocator_handle_t omp_cgroup_mem_alloc; +extern const omp_allocator_handle_t omp_pteam_mem_alloc; +extern const omp_allocator_handle_t omp_thread_mem_alloc; + void foo() {} struct S { @@ -133,7 +143,7 @@ foo(); // CHECK: #pragma omp target teams distribute simd // CHECK-NEXT: for (int i = 0; i < 10; ++i) -// CHECK-NEXT: foo(); +// CHECK-NEXT: foo(); #pragma omp target teams distribute simd private(b), firstprivate(argc) shared(d) reduction(+:c) reduction(max:e) num_teams(f) thread_limit(d) for (int k = 0; k < 10; ++k) e += d + argc; @@ -198,14 +208,14 @@ // CHECK-NEXT: for (int k = 0; k < 10; ++k) // CHECK-NEXT: e += d + argc; #ifdef OMP5 -#pragma omp target teams distribute simd safelen(clen-1) aligned(arr:N+6) if(simd:argc) nontemporal(argc, c, d) order(concurrent) +#pragma omp target teams distribute simd safelen(clen-1) aligned(arr:N+6) if(simd:argc) nontemporal(argc, c, d) order(concurrent) allocate(omp_low_lat_mem_alloc:e) firstprivate(e) uses_allocators(omp_low_lat_mem_alloc) #else #pragma omp target teams distribute simd safelen(clen-1) aligned(arr:N+6) #endif // OMP5 for (int k = 0; k < 10; ++k) e += d + argc + arr[k]; // OMP45: #pragma omp target teams distribute simd safelen(clen - 1) aligned(arr: N + 6) -// OMP50: #pragma omp target teams distribute simd safelen(clen - 1) aligned(arr: N + 6) if(simd: argc) nontemporal(argc,c,d) order(concurrent) +// OMP50: #pragma omp target teams distribute simd safelen(clen - 1) aligned(arr: N + 6) if(simd: argc) nontemporal(argc,c,d) order(concurrent) allocate(omp_low_lat_mem_alloc: e) firstprivate(e) uses_allocators(omp_low_lat_mem_alloc) // CHECK-NEXT: for (int k = 0; k < 10; ++k) // CHECK-NEXT: e += d + argc + arr[k]; return (0); diff --git a/clang/test/OpenMP/target_teams_distribute_simd_firstprivate_messages.cpp b/clang/test/OpenMP/target_teams_distribute_simd_firstprivate_messages.cpp --- a/clang/test/OpenMP/target_teams_distribute_simd_firstprivate_messages.cpp +++ b/clang/test/OpenMP/target_teams_distribute_simd_firstprivate_messages.cpp @@ -5,6 +5,7 @@ // RUN: %clang_cc1 -verify=expected,le45 -fopenmp-simd %s -Wuninitialized +#pragma omp requires dynamic_allocators typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; extern const omp_allocator_handle_t omp_large_cap_mem_alloc; diff --git a/clang/test/OpenMP/target_teams_distribute_simd_lastprivate_messages.cpp b/clang/test/OpenMP/target_teams_distribute_simd_lastprivate_messages.cpp --- a/clang/test/OpenMP/target_teams_distribute_simd_lastprivate_messages.cpp +++ b/clang/test/OpenMP/target_teams_distribute_simd_lastprivate_messages.cpp @@ -1,10 +1,10 @@ // RUN: %clang_cc1 -verify=expected,le45 -fopenmp %s -Wuninitialized // RUN: %clang_cc1 -verify=expected,le45 -fopenmp-version=40 -fopenmp %s -Wuninitialized // RUN: %clang_cc1 -verify=expected,le45 -fopenmp-version=45 -fopenmp %s -Wuninitialized -// RUN: %clang_cc1 -verify=expected -fopenmp-version=50 -fopenmp %s -Wuninitialized +// RUN: %clang_cc1 -verify=expected,le50 -fopenmp-version=50 -fopenmp %s -Wuninitialized // RUN: %clang_cc1 -verify=expected,le45 -fopenmp-simd %s -Wuninitialized -// RUN: %clang_cc1 -verify=expected -fopenmp-version=50 -fopenmp-simd %s -Wuninitialized +// RUN: %clang_cc1 -verify=expected,le50 -fopenmp-version=50 -fopenmp-simd %s -Wuninitialized typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; @@ -124,7 +124,7 @@ for (int k = 0; k < argc; ++k) ++k; int v = 0; -#pragma omp target teams distribute simd allocate(omp_thread_mem_alloc: i) lastprivate(i) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target teams distribute simd' directive}} +#pragma omp target teams distribute simd allocate(omp_thread_mem_alloc: i) lastprivate(i) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target teams distribute simd' directive}} le50-error {{allocator must be specified in the 'uses_allocators' clause}} for (int k = 0; k < argc; ++k) { i = k; v += i; diff --git a/clang/test/OpenMP/target_teams_distribute_simd_private_messages.cpp b/clang/test/OpenMP/target_teams_distribute_simd_private_messages.cpp --- a/clang/test/OpenMP/target_teams_distribute_simd_private_messages.cpp +++ b/clang/test/OpenMP/target_teams_distribute_simd_private_messages.cpp @@ -4,6 +4,7 @@ // RUN: %clang_cc1 -verify=expected -fopenmp-version=50 -fopenmp %s -Wuninitialized // RUN: %clang_cc1 -verify=expected,le45 -fopenmp-simd %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=50 %s -Wuninitialized typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; @@ -95,7 +96,7 @@ #pragma omp target teams distribute simd private (k, argv[1]) // expected-error {{expected variable name}} for (int k = 0; k < argc; ++k) ++k; -#pragma omp target teams distribute simd private(ba) allocate(omp_thread_mem_alloc: ba) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target teams distribute simd' directive}} +#pragma omp target teams distribute simd private(ba) allocate(omp_thread_mem_alloc: ba) uses_allocators(omp_thread_mem_alloc) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target teams distribute simd' directive}} le45-error {{unexpected OpenMP clause 'uses_allocators' in directive '#pragma omp target teams distribute simd'}} for (int k = 0; k < argc; ++k) ++k; #pragma omp target teams distribute simd private(ca) // expected-error {{const-qualified variable without mutable fields cannot be private}} diff --git a/clang/test/OpenMP/target_teams_distribute_simd_reduction_messages.cpp b/clang/test/OpenMP/target_teams_distribute_simd_reduction_messages.cpp --- a/clang/test/OpenMP/target_teams_distribute_simd_reduction_messages.cpp +++ b/clang/test/OpenMP/target_teams_distribute_simd_reduction_messages.cpp @@ -12,6 +12,7 @@ // RUN: %clang_cc1 -verify=expected,omp50 -fopenmp-simd -fopenmp-version=50 -std=c++98 %s -Wuninitialized // RUN: %clang_cc1 -verify=expected,omp50 -fopenmp-simd -fopenmp-version=50 -std=c++11 %s -Wuninitialized +#pragma omp requires dynamic_allocators typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; extern const omp_allocator_handle_t omp_large_cap_mem_alloc; diff --git a/clang/test/OpenMP/target_teams_firstprivate_messages.cpp b/clang/test/OpenMP/target_teams_firstprivate_messages.cpp --- a/clang/test/OpenMP/target_teams_firstprivate_messages.cpp +++ b/clang/test/OpenMP/target_teams_firstprivate_messages.cpp @@ -1,6 +1,6 @@ -// RUN: %clang_cc1 -verify -fopenmp %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=50 %s -Wuninitialized -// RUN: %clang_cc1 -verify -fopenmp-simd %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=50 %s -Wuninitialized typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; @@ -103,7 +103,7 @@ foo(); #pragma omp target teams firstprivate(argv[1]) // expected-error {{expected variable name}} foo(); -#pragma omp target teams allocate(omp_thread_mem_alloc: ba) firstprivate(ba) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target teams' directive}} +#pragma omp target teams allocate(omp_thread_mem_alloc: ba) uses_allocators(omp_thread_mem_alloc) firstprivate(ba) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target teams' directive}} foo(); #pragma omp target teams firstprivate(ca, z) foo(); diff --git a/clang/test/OpenMP/target_teams_private_messages.cpp b/clang/test/OpenMP/target_teams_private_messages.cpp --- a/clang/test/OpenMP/target_teams_private_messages.cpp +++ b/clang/test/OpenMP/target_teams_private_messages.cpp @@ -1,6 +1,6 @@ -// RUN: %clang_cc1 -verify -fopenmp %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=50 %s -Wuninitialized -// RUN: %clang_cc1 -verify -fopenmp-simd %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=50 %s -Wuninitialized typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; @@ -108,7 +108,7 @@ foo(); #pragma omp target teams private(j) foo(); -#pragma omp target teams firstprivate(i) allocate(omp_thread_mem_alloc: i) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target teams' directive}} +#pragma omp target teams firstprivate(i) uses_allocators(omp_thread_mem_alloc) allocate(omp_thread_mem_alloc: i) // expected-warning {{allocator with the 'thread' trait access has unspecified behavior on 'target teams' directive}} for (int k = 0; k < 10; ++k) { #pragma omp parallel private(i) foo(); diff --git a/clang/test/OpenMP/target_teams_reduction_messages.cpp b/clang/test/OpenMP/target_teams_reduction_messages.cpp --- a/clang/test/OpenMP/target_teams_reduction_messages.cpp +++ b/clang/test/OpenMP/target_teams_reduction_messages.cpp @@ -12,6 +12,7 @@ // RUN: %clang_cc1 -verify=expected,omp50 -fopenmp-simd -fopenmp-version=50 -std=c++98 -o - %s -Wuninitialized // RUN: %clang_cc1 -verify=expected,omp50 -fopenmp-simd -fopenmp-version=50 -std=c++11 -o - %s -Wuninitialized +#pragma omp requires dynamic_allocators typedef void **omp_allocator_handle_t; extern const omp_allocator_handle_t omp_default_mem_alloc; extern const omp_allocator_handle_t omp_large_cap_mem_alloc; diff --git a/clang/test/OpenMP/target_uses_allocators_messages.cpp b/clang/test/OpenMP/target_uses_allocators_messages.cpp new file mode 100644 --- /dev/null +++ b/clang/test/OpenMP/target_uses_allocators_messages.cpp @@ -0,0 +1,54 @@ +// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=50 %s -Wuninitialized + +// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=50 %s -Wuninitialized + +struct omp_alloctrait_t {}; + +typedef void **omp_allocator_handle_t; +extern const omp_allocator_handle_t omp_default_mem_alloc; +extern const omp_allocator_handle_t omp_large_cap_mem_alloc; +extern const omp_allocator_handle_t omp_const_mem_alloc; +extern const omp_allocator_handle_t omp_high_bw_mem_alloc; +extern const omp_allocator_handle_t omp_low_lat_mem_alloc; +extern const omp_allocator_handle_t omp_cgroup_mem_alloc; +extern const omp_allocator_handle_t omp_pteam_mem_alloc; +extern const omp_allocator_handle_t omp_thread_mem_alloc; + +int main(int argc, char **argv) { + omp_alloctrait_t traits[10]; + omp_alloctrait_t *ptraits; + omp_allocator_handle_t my_alloc = nullptr; + const omp_allocator_handle_t c_my_alloc = my_alloc; +#pragma omp target uses_allocators // expected-error {{expected '(' after 'uses_allocator'}} +{} +#pragma omp target uses_allocators( // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected unqualified-id}} +{} +#pragma omp target uses_allocators() // expected-error {{expected unqualified-id}} +{} +#pragma omp target uses_allocators(omp_default_mem_alloc // expected-error {{expected ',' or ')' in 'uses_allocators' clause}} expected-error {{expected ')'}} expected-note {{to match this '('}} +{} +#pragma omp target uses_allocators(argc, // expected-error {{expected ')'}} expected-error {{expected variable of the 'omp_allocator_handle_t' type, not 'int'}} expected-note {{to match this '('}} +{} +#pragma omp target uses_allocators(argc > 0 ? omp_default_mem_alloc : omp_thread_mem_alloc) // expected-error {{expected ',' or ')' in 'uses_allocators' clause}} expected-error {{expected unqualified-id}} expected-error {{expected variable of the 'omp_allocator_handle_t' type, not 'int'}} +{} +#pragma omp target uses_allocators(omp_default_mem_alloc, omp_large_cap_mem_alloc, omp_const_mem_alloc, omp_high_bw_mem_alloc, omp_low_lat_mem_alloc, omp_cgroup_mem_alloc, omp_pteam_mem_alloc, omp_thread_mem_alloc) +{} +#pragma omp target uses_allocators(omp_default_mem_alloc(traits), omp_large_cap_mem_alloc(traits), omp_const_mem_alloc(traits), omp_high_bw_mem_alloc(traits), omp_low_lat_mem_alloc(traits), omp_cgroup_mem_alloc(traits), omp_pteam_mem_alloc(traits), omp_thread_mem_alloc(traits)) // expected-error 8 {{predefined allocator cannot have traits specified}} expected-note-re 8 {{predefined trait '{{omp_default_mem_alloc|omp_large_cap_mem_alloc|omp_const_mem_alloc|omp_high_bw_mem_alloc|omp_low_lat_mem_alloc|omp_cgroup_mem_alloc|omp_pteam_mem_alloc|omp_thread_mem_alloc}}' used here}} +{} +#pragma omp target uses_allocators(my_alloc, c_my_alloc) // expected-error {{non-predefined allocator must have traits specified}} expected-error {{expected variable of the 'omp_allocator_handle_t' type, not 'const omp_allocator_handle_t' (aka 'void **const')}} +{} +#pragma omp target uses_allocators(my_alloc() // expected-error {{expected unqualified-id}} expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{non-predefined allocator must have traits specified}} +{} +#pragma omp target uses_allocators(my_alloc()) // expected-error {{expected unqualified-id}} expected-error {{non-predefined allocator must have traits specified}} +{} +#pragma omp target uses_allocators(my_alloc(argc > 0 ? argv[0] : argv{1})) // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected constant sized array of 'omp_alloctrait_t' elements, not 'int'}} +{} +#pragma omp target uses_allocators(my_alloc(ptraits)) // expected-error {{expected constant sized array of 'omp_alloctrait_t' elements, not 'omp_alloctrait_t *'}} +{} +#pragma omp target uses_allocators(my_alloc(traits)) private(my_alloc) // expected-error {{allocators used in 'uses_allocators' clause cannot appear in other data-sharing or data-mapping attribute clauses}} expected-note {{defined as private}} +{} +#pragma omp target map(my_alloc, traits) uses_allocators(my_alloc(traits)) // expected-error {{allocators used in 'uses_allocators' clause cannot appear in other data-sharing or data-mapping attribute clauses}} expected-note {{used here}} +{} + return 0; +} + diff --git a/clang/tools/libclang/CIndex.cpp b/clang/tools/libclang/CIndex.cpp --- a/clang/tools/libclang/CIndex.cpp +++ b/clang/tools/libclang/CIndex.cpp @@ -2490,6 +2490,14 @@ Visitor->AddStmt(E); } void OMPClauseEnqueue::VisitOMPOrderClause(const OMPOrderClause *C) {} +void OMPClauseEnqueue::VisitOMPUsesAllocatorsClause( + const OMPUsesAllocatorsClause *C) { + for (unsigned I = 0, E = C->getNumberOfAllocators(); I < E; ++I) { + const OMPUsesAllocatorsClause::Data &D = C->getAllocatorData(I); + Visitor->AddStmt(D.Allocator); + Visitor->AddStmt(D.AllocatorTraits); + } +} } // namespace void EnqueueVisitor::EnqueueChildren(const OMPClause *S) { diff --git a/llvm/include/llvm/Frontend/OpenMP/OMPKinds.def b/llvm/include/llvm/Frontend/OpenMP/OMPKinds.def --- a/llvm/include/llvm/Frontend/OpenMP/OMPKinds.def +++ b/llvm/include/llvm/Frontend/OpenMP/OMPKinds.def @@ -196,6 +196,7 @@ __OMP_CLAUSE(detach, OMPDetachClause) __OMP_CLAUSE(inclusive, OMPInclusiveClause) __OMP_CLAUSE(exclusive, OMPExclusiveClause) +__OMP_CLAUSE(uses_allocators, OMPUsesAllocatorsClause) __OMP_CLAUSE_NO_CLASS(uniform) __OMP_CLAUSE_NO_CLASS(device_type) @@ -887,6 +888,7 @@ __OMP_DIRECTIVE_CLAUSE(target, 1, ~0, is_device_ptr) __OMP_DIRECTIVE_CLAUSE(target, 1, ~0, reduction) __OMP_DIRECTIVE_CLAUSE(target, 1, ~0, allocate) +__OMP_DIRECTIVE_CLAUSE(target, 50, ~0, uses_allocators) __OMP_DIRECTIVE_CLAUSE(requires, 1, ~0, unified_address) __OMP_DIRECTIVE_CLAUSE(requires, 1, ~0, unified_shared_memory) @@ -928,6 +930,7 @@ __OMP_DIRECTIVE_CLAUSE(target_parallel, 1, ~0, reduction) __OMP_DIRECTIVE_CLAUSE(target_parallel, 1, ~0, is_device_ptr) __OMP_DIRECTIVE_CLAUSE(target_parallel, 1, ~0, allocate) +__OMP_DIRECTIVE_CLAUSE(target_parallel, 50, ~0, uses_allocators) __OMP_DIRECTIVE_CLAUSE(target_parallel_for, 1, ~0, if) __OMP_DIRECTIVE_CLAUSE(target_parallel_for, 1, ~0, device) @@ -950,6 +953,7 @@ __OMP_DIRECTIVE_CLAUSE(target_parallel_for, 1, ~0, is_device_ptr) __OMP_DIRECTIVE_CLAUSE(target_parallel_for, 1, ~0, allocate) __OMP_DIRECTIVE_CLAUSE(target_parallel_for, 50, ~0, order) +__OMP_DIRECTIVE_CLAUSE(target_parallel_for, 50, ~0, uses_allocators) __OMP_DIRECTIVE_CLAUSE(target_update, 1, ~0, if) __OMP_DIRECTIVE_CLAUSE(target_update, 1, ~0, device) @@ -1187,6 +1191,7 @@ __OMP_DIRECTIVE_CLAUSE(target_parallel_for_simd, 1, ~0, allocate) __OMP_DIRECTIVE_CLAUSE(target_parallel_for_simd, 50, ~0, nontemporal) __OMP_DIRECTIVE_CLAUSE(target_parallel_for_simd, 50, ~0, order) +__OMP_DIRECTIVE_CLAUSE(target_parallel_for_simd, 50, ~0, uses_allocators) __OMP_DIRECTIVE_CLAUSE(target_simd, 1, ~0, if) __OMP_DIRECTIVE_CLAUSE(target_simd, 1, ~0, device) @@ -1207,6 +1212,7 @@ __OMP_DIRECTIVE_CLAUSE(target_simd, 1, ~0, allocate) __OMP_DIRECTIVE_CLAUSE(target_simd, 50, ~0, nontemporal) __OMP_DIRECTIVE_CLAUSE(target_simd, 50, ~0, order) +__OMP_DIRECTIVE_CLAUSE(target_simd, 50, ~0, uses_allocators) __OMP_DIRECTIVE_CLAUSE(teams_distribute, 1, ~0, default) __OMP_DIRECTIVE_CLAUSE(teams_distribute, 1, ~0, private) @@ -1294,6 +1300,7 @@ __OMP_DIRECTIVE_CLAUSE(target_teams, 1, ~0, num_teams) __OMP_DIRECTIVE_CLAUSE(target_teams, 1, ~0, thread_limit) __OMP_DIRECTIVE_CLAUSE(target_teams, 1, ~0, allocate) +__OMP_DIRECTIVE_CLAUSE(target_teams, 50, ~0, uses_allocators) __OMP_DIRECTIVE_CLAUSE(target_teams_distribute, 1, ~0, if) __OMP_DIRECTIVE_CLAUSE(target_teams_distribute, 1, ~0, device) @@ -1313,6 +1320,7 @@ __OMP_DIRECTIVE_CLAUSE(target_teams_distribute, 1, ~0, collapse) __OMP_DIRECTIVE_CLAUSE(target_teams_distribute, 1, ~0, dist_schedule) __OMP_DIRECTIVE_CLAUSE(target_teams_distribute, 1, ~0, allocate) +__OMP_DIRECTIVE_CLAUSE(target_teams_distribute, 50, ~0, uses_allocators) __OMP_DIRECTIVE_CLAUSE(target_teams_distribute_parallel_for, 1, ~0, if) __OMP_DIRECTIVE_CLAUSE(target_teams_distribute_parallel_for, 1, ~0, device) @@ -1340,6 +1348,8 @@ __OMP_DIRECTIVE_CLAUSE(target_teams_distribute_parallel_for, 1, ~0, schedule) __OMP_DIRECTIVE_CLAUSE(target_teams_distribute_parallel_for, 1, ~0, allocate) __OMP_DIRECTIVE_CLAUSE(target_teams_distribute_parallel_for, 50, ~0, order) +__OMP_DIRECTIVE_CLAUSE(target_teams_distribute_parallel_for, 50, ~0, + uses_allocators) __OMP_DIRECTIVE_CLAUSE(target_teams_distribute_parallel_for_simd, 1, ~0, if) __OMP_DIRECTIVE_CLAUSE(target_teams_distribute_parallel_for_simd, 1, ~0, device) @@ -1387,6 +1397,8 @@ __OMP_DIRECTIVE_CLAUSE(target_teams_distribute_parallel_for_simd, 50, ~0, nontemporal) __OMP_DIRECTIVE_CLAUSE(target_teams_distribute_parallel_for_simd, 50, ~0, order) +__OMP_DIRECTIVE_CLAUSE(target_teams_distribute_parallel_for_simd, 50, ~0, + uses_allocators) __OMP_DIRECTIVE_CLAUSE(target_teams_distribute_simd, 1, ~0, if) __OMP_DIRECTIVE_CLAUSE(target_teams_distribute_simd, 1, ~0, device) @@ -1411,6 +1423,7 @@ __OMP_DIRECTIVE_CLAUSE(target_teams_distribute_simd, 1, ~0, allocate) __OMP_DIRECTIVE_CLAUSE(target_teams_distribute_simd, 50, ~0, nontemporal) __OMP_DIRECTIVE_CLAUSE(target_teams_distribute_simd, 50, ~0, order) +__OMP_DIRECTIVE_CLAUSE(target_teams_distribute_simd, 50, ~0, uses_allocators) __OMP_DIRECTIVE_CLAUSE(taskgroup, 1, ~0, task_reduction) __OMP_DIRECTIVE_CLAUSE(taskgroup, 1, ~0, allocate)