Index: cfe/trunk/include/clang-c/Index.h =================================================================== --- cfe/trunk/include/clang-c/Index.h +++ cfe/trunk/include/clang-c/Index.h @@ -2354,7 +2354,11 @@ */ CXCursor_OMPTeamsDistributeParallelForDirective = 274, - CXCursor_LastStmt = CXCursor_OMPTeamsDistributeParallelForDirective, + /** \brief OpenMP target teams directive. + */ + CXCursor_OMPTargetTeamsDirective = 275, + + CXCursor_LastStmt = CXCursor_OMPTargetTeamsDirective, /** * \brief Cursor that represents the translation unit itself. Index: cfe/trunk/include/clang/AST/RecursiveASTVisitor.h =================================================================== --- cfe/trunk/include/clang/AST/RecursiveASTVisitor.h +++ cfe/trunk/include/clang/AST/RecursiveASTVisitor.h @@ -2658,6 +2658,9 @@ DEF_TRAVERSE_STMT(OMPTeamsDistributeParallelForDirective, { TRY_TO(TraverseOMPExecutableDirective(S)); }) +DEF_TRAVERSE_STMT(OMPTargetTeamsDirective, + { TRY_TO(TraverseOMPExecutableDirective(S)); }) + // OpenMP clauses. template bool RecursiveASTVisitor::TraverseOMPClause(OMPClause *C) { Index: cfe/trunk/include/clang/AST/StmtOpenMP.h =================================================================== --- cfe/trunk/include/clang/AST/StmtOpenMP.h +++ cfe/trunk/include/clang/AST/StmtOpenMP.h @@ -3439,6 +3439,65 @@ } }; +/// This represents '#pragma omp target teams' directive. +/// +/// \code +/// #pragma omp target teams if(a>0) +/// \endcode +/// In this example directive '#pragma omp target teams' has clause 'if' with +/// condition 'a>0'. +/// +class OMPTargetTeamsDirective final : public OMPExecutableDirective { + friend class ASTStmtReader; + /// Build directive with the given start and end location. + /// + /// \param StartLoc Starting location of the directive kind. + /// \param EndLoc Ending location of the directive. + /// \param NumClauses Number of clauses. + /// + OMPTargetTeamsDirective(SourceLocation StartLoc, SourceLocation EndLoc, + unsigned NumClauses) + : OMPExecutableDirective(this, OMPTargetTeamsDirectiveClass, + OMPD_target_teams, StartLoc, EndLoc, NumClauses, + 1) {} + + /// Build an empty directive. + /// + /// \param NumClauses Number of clauses. + /// + explicit OMPTargetTeamsDirective(unsigned NumClauses) + : OMPExecutableDirective(this, OMPTargetTeamsDirectiveClass, + OMPD_target_teams, SourceLocation(), + SourceLocation(), NumClauses, 1) {} + +public: + /// Creates directive with a list of \a Clauses. + /// + /// \param C AST context. + /// \param StartLoc Starting location of the directive kind. + /// \param EndLoc Ending Location of the directive. + /// \param Clauses List of clauses. + /// \param AssociatedStmt Statement, associated with the directive. + /// + static OMPTargetTeamsDirective *Create(const ASTContext &C, + SourceLocation StartLoc, + SourceLocation EndLoc, + ArrayRef Clauses, + Stmt *AssociatedStmt); + + /// Creates an empty directive with the place for \a NumClauses clauses. + /// + /// \param C AST context. + /// \param NumClauses Number of clauses. + /// + static OMPTargetTeamsDirective *CreateEmpty(const ASTContext &C, + unsigned NumClauses, EmptyShell); + + static bool classof(const Stmt *T) { + return T->getStmtClass() == OMPTargetTeamsDirectiveClass; + } +}; + } // end namespace clang #endif Index: cfe/trunk/include/clang/Basic/OpenMPKinds.h =================================================================== --- cfe/trunk/include/clang/Basic/OpenMPKinds.h +++ cfe/trunk/include/clang/Basic/OpenMPKinds.h @@ -179,10 +179,18 @@ /// otherwise - false. bool isOpenMPTargetDataManagementDirective(OpenMPDirectiveKind DKind); -/// \brief Checks if the specified directive is a teams-kind directive. +/// Checks if the specified composite/combined directive constitutes a teams +/// directive in the outermost nest. For example +/// 'omp teams distribute' or 'omp teams distribute parallel for'. +/// \param DKind Specified directive. +/// \return true - the directive has teams on the outermost nest, otherwise - +/// false. +bool isOpenMPNestingTeamsDirective(OpenMPDirectiveKind DKind); + +/// Checks if the specified directive is a teams-kind directive. For example, +/// 'omp teams distribute' or 'omp target teams'. /// \param DKind Specified directive. -/// \return true - the directive is a teams-like directive like 'omp teams', -/// otherwise - false. +/// \return true - the directive is a teams-like directive, otherwise - false. bool isOpenMPTeamsDirective(OpenMPDirectiveKind DKind); /// \brief Checks if the specified directive is a simd directive. Index: cfe/trunk/include/clang/Basic/OpenMPKinds.def =================================================================== --- cfe/trunk/include/clang/Basic/OpenMPKinds.def +++ cfe/trunk/include/clang/Basic/OpenMPKinds.def @@ -153,6 +153,9 @@ #ifndef OPENMP_TEAMS_DISTRIBUTE_PARALLEL_FOR_CLAUSE #define OPENMP_TEAMS_DISTRIBUTE_PARALLEL_FOR_CLAUSE(Name) #endif +#ifndef OPENMP_TARGET_TEAMS_CLAUSE +#define OPENMP_TARGET_TEAMS_CLAUSE(Name) +#endif // OpenMP directives. OPENMP_DIRECTIVE(threadprivate) @@ -202,6 +205,7 @@ OPENMP_DIRECTIVE_EXT(teams_distribute_simd, "teams distribute simd") OPENMP_DIRECTIVE_EXT(teams_distribute_parallel_for_simd, "teams distribute parallel for simd") OPENMP_DIRECTIVE_EXT(teams_distribute_parallel_for, "teams distribute parallel for") +OPENMP_DIRECTIVE_EXT(target_teams, "target teams") // OpenMP clauses. OPENMP_CLAUSE(if, OMPIfClause) @@ -723,6 +727,22 @@ OPENMP_TEAMS_DISTRIBUTE_PARALLEL_FOR_CLAUSE(num_teams) OPENMP_TEAMS_DISTRIBUTE_PARALLEL_FOR_CLAUSE(thread_limit) +// Clauses allowed for OpenMP directive 'target teams'. +OPENMP_TARGET_TEAMS_CLAUSE(if) +OPENMP_TARGET_TEAMS_CLAUSE(device) +OPENMP_TARGET_TEAMS_CLAUSE(map) +OPENMP_TARGET_TEAMS_CLAUSE(private) +OPENMP_TARGET_TEAMS_CLAUSE(nowait) +OPENMP_TARGET_TEAMS_CLAUSE(depend) +OPENMP_TARGET_TEAMS_CLAUSE(defaultmap) +OPENMP_TARGET_TEAMS_CLAUSE(firstprivate) +OPENMP_TARGET_TEAMS_CLAUSE(is_device_ptr) +OPENMP_TARGET_TEAMS_CLAUSE(default) +OPENMP_TARGET_TEAMS_CLAUSE(shared) +OPENMP_TARGET_TEAMS_CLAUSE(reduction) +OPENMP_TARGET_TEAMS_CLAUSE(num_teams) +OPENMP_TARGET_TEAMS_CLAUSE(thread_limit) + #undef OPENMP_TASKLOOP_SIMD_CLAUSE #undef OPENMP_TASKLOOP_CLAUSE #undef OPENMP_LINEAR_KIND @@ -770,3 +790,4 @@ #undef OPENMP_TEAMS_DISTRIBUTE_SIMD_CLAUSE #undef OPENMP_TEAMS_DISTRIBUTE_PARALLEL_FOR_SIMD_CLAUSE #undef OPENMP_TEAMS_DISTRIBUTE_PARALLEL_FOR_CLAUSE +#undef OPENMP_TARGET_TEAMS_CLAUSE Index: cfe/trunk/include/clang/Basic/StmtNodes.td =================================================================== --- cfe/trunk/include/clang/Basic/StmtNodes.td +++ cfe/trunk/include/clang/Basic/StmtNodes.td @@ -240,3 +240,4 @@ def OMPTeamsDistributeSimdDirective : DStmt; def OMPTeamsDistributeParallelForSimdDirective : DStmt; def OMPTeamsDistributeParallelForDirective : DStmt; +def OMPTargetTeamsDirective : DStmt; Index: cfe/trunk/include/clang/Sema/Sema.h =================================================================== --- cfe/trunk/include/clang/Sema/Sema.h +++ cfe/trunk/include/clang/Sema/Sema.h @@ -8393,6 +8393,12 @@ ArrayRef Clauses, Stmt *AStmt, SourceLocation StartLoc, SourceLocation EndLoc, llvm::DenseMap &VarsWithImplicitDSA); + /// Called on well-formed '\#pragma omp target teams' after parsing of the + /// associated statement. + StmtResult ActOnOpenMPTargetTeamsDirective(ArrayRef Clauses, + Stmt *AStmt, + SourceLocation StartLoc, + SourceLocation EndLoc); /// Checks correctness of linear modifiers. bool CheckOpenMPLinearModifier(OpenMPLinearClauseKind LinKind, Index: cfe/trunk/include/clang/Serialization/ASTBitCodes.h =================================================================== --- cfe/trunk/include/clang/Serialization/ASTBitCodes.h +++ cfe/trunk/include/clang/Serialization/ASTBitCodes.h @@ -1505,6 +1505,7 @@ STMT_OMP_TEAMS_DISTRIBUTE_SIMD_DIRECTIVE, STMT_OMP_TEAMS_DISTRIBUTE_PARALLEL_FOR_SIMD_DIRECTIVE, STMT_OMP_TEAMS_DISTRIBUTE_PARALLEL_FOR_DIRECTIVE, + STMT_OMP_TARGET_TEAMS_DIRECTIVE, EXPR_OMP_ARRAY_SECTION, // ARC Index: cfe/trunk/lib/AST/StmtOpenMP.cpp =================================================================== --- cfe/trunk/lib/AST/StmtOpenMP.cpp +++ cfe/trunk/lib/AST/StmtOpenMP.cpp @@ -1520,3 +1520,26 @@ OMPTeamsDistributeParallelForDirective(CollapsedNum, NumClauses); } +OMPTargetTeamsDirective *OMPTargetTeamsDirective::Create( + const ASTContext &C, SourceLocation StartLoc, SourceLocation EndLoc, + ArrayRef Clauses, Stmt *AssociatedStmt) { + auto Size = + llvm::alignTo(sizeof(OMPTargetTeamsDirective), alignof(OMPClause *)); + void *Mem = + C.Allocate(Size + sizeof(OMPClause *) * Clauses.size() + sizeof(Stmt *)); + OMPTargetTeamsDirective *Dir = + new (Mem) OMPTargetTeamsDirective(StartLoc, EndLoc, Clauses.size()); + Dir->setClauses(Clauses); + Dir->setAssociatedStmt(AssociatedStmt); + return Dir; +} + +OMPTargetTeamsDirective * +OMPTargetTeamsDirective::CreateEmpty(const ASTContext &C, unsigned NumClauses, + EmptyShell) { + auto Size = + llvm::alignTo(sizeof(OMPTargetTeamsDirective), alignof(OMPClause *)); + void *Mem = + C.Allocate(Size + sizeof(OMPClause *) * NumClauses + sizeof(Stmt *)); + return new (Mem) OMPTargetTeamsDirective(NumClauses); +} Index: cfe/trunk/lib/AST/StmtPrinter.cpp =================================================================== --- cfe/trunk/lib/AST/StmtPrinter.cpp +++ cfe/trunk/lib/AST/StmtPrinter.cpp @@ -1227,6 +1227,11 @@ PrintOMPExecutableDirective(Node); } +void StmtPrinter::VisitOMPTargetTeamsDirective(OMPTargetTeamsDirective *Node) { + Indent() << "#pragma omp target teams "; + PrintOMPExecutableDirective(Node); +} + //===----------------------------------------------------------------------===// // Expr printing methods. //===----------------------------------------------------------------------===// Index: cfe/trunk/lib/AST/StmtProfile.cpp =================================================================== --- cfe/trunk/lib/AST/StmtProfile.cpp +++ cfe/trunk/lib/AST/StmtProfile.cpp @@ -748,6 +748,11 @@ VisitOMPLoopDirective(S); } +void StmtProfiler::VisitOMPTargetTeamsDirective( + const OMPTargetTeamsDirective *S) { + VisitOMPExecutableDirective(S); +} + void StmtProfiler::VisitExpr(const Expr *S) { VisitStmt(S); } Index: cfe/trunk/lib/Basic/OpenMPKinds.cpp =================================================================== --- cfe/trunk/lib/Basic/OpenMPKinds.cpp +++ cfe/trunk/lib/Basic/OpenMPKinds.cpp @@ -660,6 +660,16 @@ break; } break; + case OMPD_target_teams: + switch (CKind) { +#define OPENMP_TARGET_TEAMS_CLAUSE(Name) \ + case OMPC_##Name: \ + return true; +#include "clang/Basic/OpenMPKinds.def" + default: + break; + } + break; case OMPD_declare_target: case OMPD_end_declare_target: case OMPD_unknown: @@ -724,10 +734,10 @@ } bool clang::isOpenMPTargetExecutionDirective(OpenMPDirectiveKind DKind) { - // TODO add next directives. return DKind == OMPD_target || DKind == OMPD_target_parallel || DKind == OMPD_target_parallel_for || - DKind == OMPD_target_parallel_for_simd || DKind == OMPD_target_simd; + DKind == OMPD_target_parallel_for_simd || DKind == OMPD_target_simd || + DKind == OMPD_target_teams; } bool clang::isOpenMPTargetDataManagementDirective(OpenMPDirectiveKind DKind) { @@ -735,12 +745,16 @@ DKind == OMPD_target_exit_data || DKind == OMPD_target_update; } -bool clang::isOpenMPTeamsDirective(OpenMPDirectiveKind DKind) { +bool clang::isOpenMPNestingTeamsDirective(OpenMPDirectiveKind DKind) { return DKind == OMPD_teams || DKind == OMPD_teams_distribute || DKind == OMPD_teams_distribute_simd || DKind == OMPD_teams_distribute_parallel_for_simd || DKind == OMPD_teams_distribute_parallel_for; - // TODO add next directives. +} + +bool clang::isOpenMPTeamsDirective(OpenMPDirectiveKind DKind) { + return isOpenMPNestingTeamsDirective(DKind) || + DKind == OMPD_target_teams; } bool clang::isOpenMPSimdDirective(OpenMPDirectiveKind DKind) { Index: cfe/trunk/lib/CodeGen/CGStmt.cpp =================================================================== --- cfe/trunk/lib/CodeGen/CGStmt.cpp +++ cfe/trunk/lib/CodeGen/CGStmt.cpp @@ -315,6 +315,9 @@ EmitOMPTeamsDistributeParallelForDirective( cast(*S)); break; + case Stmt::OMPTargetTeamsDirectiveClass: + EmitOMPTargetTeamsDirective(cast(*S)); + break; } } Index: cfe/trunk/lib/CodeGen/CGStmtOpenMP.cpp =================================================================== --- cfe/trunk/lib/CodeGen/CGStmtOpenMP.cpp +++ cfe/trunk/lib/CodeGen/CGStmtOpenMP.cpp @@ -2003,6 +2003,14 @@ }); } +void CodeGenFunction::EmitOMPTargetTeamsDirective( + const OMPTargetTeamsDirective &S) { + CGM.getOpenMPRuntime().emitInlinedDirective(*this, OMPD_target_teams, + [&S](CodeGenFunction &CGF, PrePostActionTy &) { + CGF.EmitStmt(cast(S.getAssociatedStmt())->getCapturedStmt()); + }); +} + /// \brief Emit a helper variable and return corresponding lvalue. static LValue EmitOMPHelperVar(CodeGenFunction &CGF, const DeclRefExpr *Helper) { Index: cfe/trunk/lib/CodeGen/CodeGenFunction.h =================================================================== --- cfe/trunk/lib/CodeGen/CodeGenFunction.h +++ cfe/trunk/lib/CodeGen/CodeGenFunction.h @@ -2695,6 +2695,7 @@ const OMPTeamsDistributeParallelForSimdDirective &S); void EmitOMPTeamsDistributeParallelForDirective( const OMPTeamsDistributeParallelForDirective &S); + void EmitOMPTargetTeamsDirective(const OMPTargetTeamsDirective &S); /// Emit outlined function for the target directive. static std::pairisParentOrderedRegion()); Recommend = ShouldBeInOrderedRegion; - } else if (isOpenMPTeamsDirective(CurrentRegion)) { + } else if (isOpenMPNestingTeamsDirective(CurrentRegion)) { // OpenMP [2.16, Nesting of Regions] // If specified, a teams construct must be contained within a target // construct. @@ -2032,7 +2034,10 @@ Recommend = ShouldBeInTargetRegion; Stack->setParentTeamsRegionLoc(Stack->getConstructLoc()); } - if (!NestingProhibited && ParentRegion == OMPD_teams) { + if (!NestingProhibited && + !isOpenMPTargetExecutionDirective(CurrentRegion) && + !isOpenMPTargetDataManagementDirective(CurrentRegion) && + (ParentRegion == OMPD_teams || ParentRegion == OMPD_target_teams)) { // OpenMP [2.16, Nesting of Regions] // distribute, parallel, parallel sections, parallel workshare, and the // parallel loop and parallel loop SIMD constructs are the only OpenMP @@ -2046,7 +2051,8 @@ // OpenMP 4.5 [2.17 Nesting of Regions] // The region associated with the distribute construct must be strictly // nested inside a teams region - NestingProhibited = ParentRegion != OMPD_teams; + NestingProhibited = + (ParentRegion != OMPD_teams && ParentRegion != OMPD_target_teams); Recommend = ShouldBeInTeamsRegion; } if (!NestingProhibited && @@ -2414,6 +2420,11 @@ ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA); AllowedNameModifiers.push_back(OMPD_parallel); break; + case OMPD_target_teams: + Res = ActOnOpenMPTargetTeamsDirective(ClausesWithImplicit, AStmt, StartLoc, + EndLoc); + AllowedNameModifiers.push_back(OMPD_target); + break; case OMPD_declare_target: case OMPD_end_declare_target: case OMPD_threadprivate: @@ -6247,6 +6258,27 @@ Context, StartLoc, EndLoc, NestedLoopCount, Clauses, AStmt, B); } +StmtResult Sema::ActOnOpenMPTargetTeamsDirective(ArrayRef Clauses, + Stmt *AStmt, + SourceLocation StartLoc, + SourceLocation EndLoc) { + if (!AStmt) + return StmtError(); + + CapturedStmt *CS = cast(AStmt); + // 1.2.2 OpenMP Language Terminology + // Structured block - An executable statement with a single entry at the + // top and a single exit at the bottom. + // The point of exit cannot be a branch out of the structured block. + // longjmp() and throw() must not violate the entry/exit criteria. + CS->getCapturedDecl()->setNothrow(); + + getCurFunction()->setHasBranchProtectedScope(); + + return OMPTargetTeamsDirective::Create(Context, StartLoc, EndLoc, Clauses, + AStmt); +} + OMPClause *Sema::ActOnOpenMPSingleExprClause(OpenMPClauseKind Kind, Expr *Expr, SourceLocation StartLoc, SourceLocation LParenLoc, @@ -7247,12 +7279,13 @@ continue; } + auto CurrDir = DSAStack->getCurrentDirective(); // Variably modified types are not supported for tasks. if (!Type->isAnyPointerType() && Type->isVariablyModifiedType() && - isOpenMPTaskingDirective(DSAStack->getCurrentDirective())) { + isOpenMPTaskingDirective(CurrDir)) { Diag(ELoc, diag::err_omp_variably_modified_type_not_supported) << getOpenMPClauseName(OMPC_private) << Type - << getOpenMPDirectiveName(DSAStack->getCurrentDirective()); + << getOpenMPDirectiveName(CurrDir); bool IsDecl = !VD || VD->isThisDeclarationADefinition(Context) == VarDecl::DeclarationOnly; @@ -7265,8 +7298,8 @@ // OpenMP 4.5 [2.15.5.1, Restrictions, p.3] // A list item cannot appear in both a map clause and a data-sharing // attribute clause on the same construct - if (DSAStack->getCurrentDirective() == OMPD_target || - DSAStack->getCurrentDirective() == OMPD_target_parallel) { + if (CurrDir == OMPD_target || CurrDir == OMPD_target_parallel || + CurrDir == OMPD_target_teams) { OpenMPClauseKind ConflictKind; if (DSAStack->checkMappableExprComponentListsForDecl( VD, /*CurrentRegionOnly=*/true, @@ -7278,7 +7311,7 @@ Diag(ELoc, diag::err_omp_variable_in_given_clause_and_dsa) << getOpenMPClauseName(OMPC_private) << getOpenMPClauseName(ConflictKind) - << getOpenMPDirectiveName(DSAStack->getCurrentDirective()); + << getOpenMPDirectiveName(CurrDir); ReportOriginalDSA(*this, DSAStack, D, DVar); continue; } @@ -7523,7 +7556,8 @@ // OpenMP 4.5 [2.15.5.1, Restrictions, p.3] // A list item cannot appear in both a map clause and a data-sharing // attribute clause on the same construct - if (CurrDir == OMPD_target || CurrDir == OMPD_target_parallel) { + if (CurrDir == OMPD_target || CurrDir == OMPD_target_parallel || + CurrDir == OMPD_target_teams) { OpenMPClauseKind ConflictKind; if (DSAStack->checkMappableExprComponentListsForDecl( VD, /*CurrentRegionOnly=*/true, @@ -10039,7 +10073,7 @@ // OpenMP 4.5 [2.15.5.1, Restrictions, p.3] // A list item cannot appear in both a map clause and a data-sharing // attribute clause on the same construct - if (DKind == OMPD_target && VD) { + if ((DKind == OMPD_target || DKind == OMPD_target_teams) && VD) { auto DVar = DSAS->getTopDSA(VD, false); if (isOpenMPPrivate(DVar.CKind)) { SemaRef.Diag(ELoc, diag::err_omp_variable_in_given_clause_and_dsa) Index: cfe/trunk/lib/Sema/TreeTransform.h =================================================================== --- cfe/trunk/lib/Sema/TreeTransform.h +++ cfe/trunk/lib/Sema/TreeTransform.h @@ -7727,6 +7727,16 @@ return Res; } +template +StmtResult TreeTransform::TransformOMPTargetTeamsDirective( + OMPTargetTeamsDirective *D) { + DeclarationNameInfo DirName; + getDerived().getSema().StartOpenMPDSABlock(OMPD_target_teams, DirName, + nullptr, D->getLocStart()); + auto Res = getDerived().TransformOMPExecutableDirective(D); + getDerived().getSema().EndOpenMPDSABlock(Res.get()); + return Res; +} //===----------------------------------------------------------------------===// // OpenMP clause transformation Index: cfe/trunk/lib/Serialization/ASTReaderStmt.cpp =================================================================== --- cfe/trunk/lib/Serialization/ASTReaderStmt.cpp +++ cfe/trunk/lib/Serialization/ASTReaderStmt.cpp @@ -2843,6 +2843,13 @@ VisitOMPLoopDirective(D); } +void ASTStmtReader::VisitOMPTargetTeamsDirective(OMPTargetTeamsDirective *D) { + VisitStmt(D); + // The NumClauses field was read in ReadStmtFromStream. + ++Idx; + VisitOMPExecutableDirective(D); +} + //===----------------------------------------------------------------------===// // ASTReader Implementation //===----------------------------------------------------------------------===// @@ -3602,6 +3609,12 @@ break; } + case STMT_OMP_TARGET_TEAMS_DIRECTIVE: { + S = OMPTargetTeamsDirective::CreateEmpty( + Context, Record[ASTStmtReader::NumStmtFields], Empty); + break; + } + case EXPR_CXX_OPERATOR_CALL: S = new (Context) CXXOperatorCallExpr(Context, Empty); break; Index: cfe/trunk/lib/Serialization/ASTWriterStmt.cpp =================================================================== --- cfe/trunk/lib/Serialization/ASTWriterStmt.cpp +++ cfe/trunk/lib/Serialization/ASTWriterStmt.cpp @@ -2537,6 +2537,13 @@ Code = serialization::STMT_OMP_TEAMS_DISTRIBUTE_PARALLEL_FOR_DIRECTIVE; } +void ASTStmtWriter::VisitOMPTargetTeamsDirective(OMPTargetTeamsDirective *D) { + VisitStmt(D); + Record.push_back(D->getNumClauses()); + VisitOMPExecutableDirective(D); + Code = serialization::STMT_OMP_TARGET_TEAMS_DIRECTIVE; +} + //===----------------------------------------------------------------------===// // ASTWriter Implementation //===----------------------------------------------------------------------===// Index: cfe/trunk/lib/StaticAnalyzer/Core/ExprEngine.cpp =================================================================== --- cfe/trunk/lib/StaticAnalyzer/Core/ExprEngine.cpp +++ cfe/trunk/lib/StaticAnalyzer/Core/ExprEngine.cpp @@ -866,6 +866,7 @@ case Stmt::OMPTeamsDistributeSimdDirectiveClass: case Stmt::OMPTeamsDistributeParallelForSimdDirectiveClass: case Stmt::OMPTeamsDistributeParallelForDirectiveClass: + case Stmt::OMPTargetTeamsDirectiveClass: llvm_unreachable("Stmt should not be in analyzer evaluation loop"); case Stmt::ObjCSubscriptRefExprClass: Index: cfe/trunk/test/OpenMP/nesting_of_regions.cpp =================================================================== --- cfe/trunk/test/OpenMP/nesting_of_regions.cpp +++ cfe/trunk/test/OpenMP/nesting_of_regions.cpp @@ -185,6 +185,11 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp parallel + { +#pragma omp target teams // OK + ++a; + } // SIMD DIRECTIVE #pragma omp simd @@ -407,6 +412,11 @@ for (int j = 0; j < 10; ++j) ; } +#pragma omp simd + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // expected-error {{OpenMP constructs may not be nested inside a simd region}} + ++a; + } // FOR DIRECTIVE #pragma omp for @@ -642,6 +652,11 @@ for (int j = 0; j < 10; ++j) ; } +#pragma omp for + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // OK + ++a; + } // FOR SIMD DIRECTIVE #pragma omp for simd @@ -865,6 +880,11 @@ for (int j = 0; j < 10; ++j) ; } +#pragma omp for simd + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // expected-error {{OpenMP constructs may not be nested inside a simd region}} + a++; + } // SECTIONS DIRECTIVE #pragma omp sections @@ -1105,6 +1125,11 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp sections + { +#pragma omp target teams // OK + ++a; + } // SECTION DIRECTIVE #pragma omp section // expected-error {{orphaned 'omp section' directives are prohibited, it must be closely nested to a sections region}} @@ -1407,6 +1432,12 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp sections + { +#pragma omp section +#pragma omp target teams // OK + ++a; + } // SINGLE DIRECTIVE #pragma omp single @@ -1633,6 +1664,11 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp single + { +#pragma omp target teams // OK + ++a; + } // MASTER DIRECTIVE #pragma omp master @@ -1859,6 +1895,11 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp master + { +#pragma omp target teams // OK + a++; + } // CRITICAL DIRECTIVE #pragma omp critical @@ -2099,6 +2140,11 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp critical + { +#pragma omp target teams // OK + a++; + } // PARALLEL FOR DIRECTIVE #pragma omp parallel for @@ -2339,6 +2385,11 @@ for (int j = 0; j < 10; ++j) ; } +#pragma omp parallel for + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // OK + a++; + } // PARALLEL FOR SIMD DIRECTIVE #pragma omp parallel for simd @@ -2580,6 +2631,11 @@ for (int j = 0; j < 10; ++j) ; } +#pragma omp parallel for simd + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // expected-error {{OpenMP constructs may not be nested inside a simd region}} + a++; + } // PARALLEL SECTIONS DIRECTIVE #pragma omp parallel sections @@ -2809,6 +2865,11 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp parallel sections + { +#pragma omp target teams // OK + a++; + } // TASK DIRECTIVE #pragma omp task @@ -2985,6 +3046,11 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp task + { +#pragma omp target teams // OK + a++; + } // ORDERED DIRECTIVE #pragma omp ordered @@ -3232,6 +3298,11 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp ordered + { +#pragma omp target teams // OK + a++; + } // ATOMIC DIRECTIVE #pragma omp atomic @@ -3517,6 +3588,13 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp atomic + // expected-error@+2 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}} + // expected-note@+1 {{expected an expression statement}} + { +#pragma omp target teams // expected-error {{OpenMP constructs may not be nested inside an atomic region}} + a++; + } // TARGET DIRECTIVE #pragma omp target @@ -3709,6 +3787,12 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp target + { +#pragma omp target teams // expected-error {{region cannot be nested inside 'target' region}} + for (int i = 0; i < 10; ++i) + ; + } // TARGET PARALLEL DIRECTIVE #pragma omp target parallel @@ -3895,6 +3979,11 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp target parallel + { +#pragma omp target teams // expected-error {{region cannot be nested inside 'target parallel' region}} + a++; + } // TARGET PARALLEL FOR DIRECTIVE #pragma omp target parallel for @@ -4135,6 +4224,11 @@ for (int j = 0; j < 10; ++j) ; } +#pragma omp target parallel for + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // expected-error {{region cannot be nested inside 'target parallel for' region}} + a++; + } // TEAMS DIRECTIVE #pragma omp teams // expected-error {{orphaned 'omp teams' directives are prohibited; perhaps you forget to enclose the directive into a target region?}} @@ -4248,7 +4342,7 @@ #pragma omp target #pragma omp teams { -#pragma omp target // expected-error {{region cannot be closely nested inside 'teams' region; perhaps you forget to enclose 'omp target' directive into a parallel region?}} +#pragma omp target // expected-error {{region cannot be nested inside 'target' region}} ++a; } #pragma omp target @@ -4265,13 +4359,13 @@ #pragma omp target #pragma omp teams { -#pragma omp target enter data map(to: a) // expected-error {{region cannot be closely nested inside 'teams' region; perhaps you forget to enclose 'omp target enter data' directive into a parallel region?}} +#pragma omp target enter data map(to: a) // expected-error {{region cannot be nested inside 'target' region}} ++a; } #pragma omp target #pragma omp teams { -#pragma omp target exit data map(from: a) // expected-error {{region cannot be closely nested inside 'teams' region; perhaps you forget to enclose 'omp target exit data' directive into a parallel region?}} +#pragma omp target exit data map(from: a) // expected-error {{region cannot be nested inside 'target' region}} ++a; } #pragma omp target @@ -4303,7 +4397,7 @@ #pragma omp target #pragma omp teams { -#pragma omp target update to(a) // expected-error {{region cannot be closely nested inside 'teams' region; perhaps you forget to enclose 'omp target update' directive into a parallel region?}} +#pragma omp target update to(a) // expected-error {{region cannot be nested inside 'target' region}} } #pragma omp target #pragma omp teams @@ -4338,7 +4432,7 @@ #pragma omp target #pragma omp teams { -#pragma omp target simd // expected-error {{region cannot be closely nested inside 'teams' region; perhaps you forget to enclose 'omp target simd' directive into a parallel region?}} +#pragma omp target simd // expected-error {{region cannot be nested inside 'target' region}} for (int i = 0; i < 10; ++i) ; } @@ -4370,6 +4464,12 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp target +#pragma omp teams + { +#pragma omp target teams // expected-error {{region cannot be nested inside 'target' region}} + a++; + } // TASKLOOP DIRECTIVE #pragma omp taskloop @@ -4601,6 +4701,11 @@ for (int j = 0; j < 10; ++j) ++a; } +#pragma omp taskloop + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // OK + ++a; + } // DISTRIBUTE DIRECTIVE #pragma omp target @@ -4866,6 +4971,13 @@ for (int j = 0; j < 10; ++j) ; } +#pragma omp target +#pragma omp teams +#pragma omp distribute + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // expected-error {{region cannot be nested inside 'target' region}} + a++; + } // DISTRIBUTE PARALLEL FOR DIRECTIVE #pragma omp target @@ -5139,6 +5251,13 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp target +#pragma omp teams +#pragma omp distribute parallel for + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // expected-error {{region cannot be nested inside 'target' region}} + a++; + } // DISTRIBUTE PARALLEL FOR SIMD DIRECTIVE #pragma omp target @@ -5420,6 +5539,13 @@ for (int i = 0; i < 10; ++i) ++a; } +#pragma omp target +#pragma omp teams +#pragma omp distribute parallel for simd + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // expected-error {{OpenMP constructs may not be nested inside a simd region}} + ++a; + } // TARGET SIMD DIRECTIVE #pragma omp target simd @@ -6116,6 +6242,12 @@ for (int i = 0; i < 10; ++i) ++a; } +#pragma omp target +#pragma omp teams distribute + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // expected-error {{region cannot be nested inside 'target' region}} + ++a; + } // TEAMS DISTRIBUTE SIMD DIRECTIVE #pragma omp teams distribute simd // expected-error {{orphaned 'omp teams distribute simd' directives are prohibited; perhaps you forget to enclose the directive into a target region?}} @@ -6365,6 +6497,12 @@ for (int i = 0; i < 10; ++i) ++a; } +#pragma omp target +#pragma omp teams distribute simd + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // expected-error {{OpenMP constructs may not be nested inside a simd region}} + ++a; + } // TEAMS DISTRIBUTE PARALLEL FOR SIMD DIRECTIVE #pragma omp teams distribute parallel for simd // expected-error {{orphaned 'omp teams distribute parallel for simd' directives are prohibited; perhaps you forget to enclose the directive into a target region?}} @@ -6614,6 +6752,12 @@ for (int i = 0; i < 10; ++i) ++a; } +#pragma omp target +#pragma omp teams distribute parallel for simd + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // expected-error {{OpenMP constructs may not be nested inside a simd region}} + ++a; + } // TEAMS DISTRIBUTE PARALLEL FOR DIRECTIVE #pragma omp teams distribute parallel for // expected-error {{orphaned 'omp teams distribute parallel for' directives are prohibited; perhaps you forget to enclose the directive into a target region?}} @@ -6863,178 +7007,382 @@ for (int i = 0; i < 10; ++i) ++a; } -} +#pragma omp target +#pragma omp teams distribute parallel for + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // expected-error {{region cannot be nested inside 'target' region}} + ++a; + } -void foo() { - int a = 0; -// PARALLEL DIRECTIVE +// TARGET TEAMS DIRECTIVE +#pragma omp target teams #pragma omp parallel -#pragma omp for + bar(); +#pragma omp target teams +#pragma omp for // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp for' directive into a parallel region?}} for (int i = 0; i < 10; ++i) ; -#pragma omp parallel -#pragma omp simd +#pragma omp target teams +#pragma omp simd // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp simd' directive into a parallel region?}} for (int i = 0; i < 10; ++i) ; -#pragma omp parallel -#pragma omp for simd +#pragma omp target teams +#pragma omp for simd // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp for simd' directive into a parallel region?}} for (int i = 0; i < 10; ++i) ; -#pragma omp parallel -#pragma omp sections +#pragma omp target teams +#pragma omp sections // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp sections' directive into a parallel region?}} { bar(); } -#pragma omp parallel -#pragma omp section // expected-error {{'omp section' directive must be closely nested to a sections region, not a parallel region}} +#pragma omp target teams +#pragma omp section // expected-error {{'omp section' directive must be closely nested to a sections region, not a target teams region}} { bar(); } -#pragma omp parallel -#pragma omp sections +#pragma omp target teams +#pragma omp single // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp single' directive into a parallel region?}} + bar(); + +#pragma omp target teams +#pragma omp master // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp master' directive into a parallel region?}} { bar(); } -#pragma omp parallel -#pragma omp single - bar(); -#pragma omp parallel -#pragma omp master - bar(); -#pragma omp parallel -#pragma omp critical - bar(); -#pragma omp parallel +#pragma omp target teams +#pragma omp critical // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp critical' directive into a parallel region?}} + { + bar(); + } +#pragma omp target teams #pragma omp parallel for for (int i = 0; i < 10; ++i) ; -#pragma omp parallel +#pragma omp target teams #pragma omp parallel for simd for (int i = 0; i < 10; ++i) ; -#pragma omp parallel +#pragma omp target teams #pragma omp parallel sections { bar(); } -#pragma omp parallel -#pragma omp task +#pragma omp target teams +#pragma omp task // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp task' directive into a parallel region?}} { bar(); } -#pragma omp parallel +#pragma omp target teams { -#pragma omp taskyield +#pragma omp taskyield // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp taskyield' directive into a parallel region?}} bar(); } -#pragma omp parallel +#pragma omp target teams { -#pragma omp barrier +#pragma omp barrier // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp barrier' directive into a parallel region?}} bar(); } -#pragma omp parallel +#pragma omp target teams { -#pragma omp taskwait +#pragma omp taskwait // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp taskwait' directive into a parallel region?}} bar(); } -#pragma omp parallel +#pragma omp target teams { -#pragma omp flush +#pragma omp flush // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp flush' directive into a parallel region?}} bar(); } -#pragma omp parallel +#pragma omp target teams { -#pragma omp ordered // expected-error {{region cannot be closely nested inside 'parallel' region; perhaps you forget to enclose 'omp ordered' directive into a for or a parallel for region with 'ordered' clause?}} +#pragma omp ordered // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp ordered' directive into a for or a parallel for region with 'ordered' clause?}} bar(); } -#pragma omp parallel +#pragma omp target teams { -#pragma omp atomic +#pragma omp atomic // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp atomic' directive into a parallel region?}} ++a; } -#pragma omp parallel +#pragma omp target teams { -#pragma omp target +#pragma omp target // expected-error {{region cannot be nested inside 'target teams' region}} ++a; } -#pragma omp parallel +#pragma omp target teams { -#pragma omp target parallel +#pragma omp target parallel // expected-error {{region cannot be nested inside 'target teams' region}} ++a; } -#pragma omp parallel -#pragma omp target parallel for +#pragma omp target teams +#pragma omp target parallel for // expected-error {{region cannot be nested inside 'target teams' region}} for (int i = 0; i < 10; ++i) ; -#pragma omp parallel +#pragma omp target teams { -#pragma omp target enter data map(to: a) - ++a; +#pragma omp target enter data map(to: a) // expected-error {{region cannot be nested inside 'target teams' region}} } -#pragma omp parallel +#pragma omp target teams { -#pragma omp target exit data map(from: a) - ++a; +#pragma omp target exit data map(from: a) // expected-error {{region cannot be nested inside 'target teams' region}} } -#pragma omp parallel +#pragma omp target teams { -#pragma omp teams // expected-error {{region cannot be closely nested inside 'parallel' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} +#pragma omp teams // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} ++a; } -#pragma omp parallel +#pragma omp target teams { -#pragma omp taskloop +#pragma omp taskloop // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp taskloop' directive into a parallel region?}} for (int i = 0; i < 10; ++i) ++a; } -#pragma omp parallel +#pragma omp target teams +#pragma omp distribute + for (int i = 0; i < 10; ++i) + ; +#pragma omp target teams { -#pragma omp distribute // expected-error {{region cannot be closely nested inside 'parallel' region; perhaps you forget to enclose 'omp distribute' directive into a teams region?}} - for (int i = 0; i < 10; ++i) - ; +#pragma omp target update to(a) // expected-error {{region cannot be nested inside 'target teams' region}} + ++a; } -#pragma omp parallel +#pragma omp target teams +#pragma omp distribute parallel for + for (int i = 0; i < 10; ++i) + ; +#pragma omp target teams { -#pragma omp target update to(a) - a++; +#pragma omp distribute parallel for + for (int i = 0; i < 10; ++i) + ; +#pragma omp distribute parallel for + for (int j = 0; j < 10; ++j) + ; } -#pragma omp parallel +#pragma omp target teams { -#pragma omp distribute parallel for // expected-error {{region cannot be closely nested inside 'parallel' region; perhaps you forget to enclose 'omp distribute parallel for' directive into a teams region?}} - for (int i = 0; i < 10; ++i) - ; +#pragma omp distribute parallel for simd + for (int i = 0; i < 10; ++i) + ; +#pragma omp distribute parallel for simd + for (int j = 0; j < 10; ++j) + ; } -#pragma omp parallel +#pragma omp target teams { -#pragma omp distribute parallel for simd // expected-error {{region cannot be closely nested inside 'parallel' region; perhaps you forget to enclose 'omp distribute parallel for simd' directive into a teams region?}} - for (int i = 0; i < 10; ++i) - ; +#pragma omp distribute simd + for (int i = 0; i < 10; ++i) + ; +#pragma omp distribute simd + for (int j = 0; j < 10; ++j) + ; } -#pragma omp parallel +#pragma omp target teams { -#pragma omp distribute simd // expected-error {{region cannot be closely nested inside 'parallel' region; perhaps you forget to enclose 'omp distribute simd' directive into a teams region?}} - for (int i = 0; i < 10; ++i) - ; +#pragma omp target simd // expected-error {{region cannot be nested inside 'target teams' region}} + for (int i = 0; i < 10; ++i) + ; } -#pragma omp parallel +#pragma omp target teams { -#pragma omp target simd // OK - for (int i = 0; i < 10; ++i) - ; +#pragma omp teams distribute // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp teams distribute' directive into a target region?}} + for (int i = 0; i < 10; ++i) + ; } -#pragma omp parallel +#pragma omp target teams { -#pragma omp teams distribute // expected-error {{region cannot be closely nested inside 'parallel' region; perhaps you forget to enclose 'omp teams distribute' directive into a target region?}} - for (int i = 0; i < 10; ++i) - ; +#pragma omp teams distribute simd // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp teams distribute simd' directive into a target region?}} + for (int i = 0; i < 10; ++i) + ; } -#pragma omp parallel +#pragma omp target teams { -#pragma omp teams distribute simd // expected-error {{region cannot be closely nested inside 'parallel' region; perhaps you forget to enclose 'omp teams distribute simd' directive into a target region?}} - for (int i = 0; i < 10; ++i) - ; +#pragma omp teams distribute parallel for simd // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp teams distribute parallel for simd' directive into a target region?}} + for (int i = 0; i < 10; ++i) + ; } -#pragma omp parallel +#pragma omp target teams + { +#pragma omp teams distribute parallel for // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp teams distribute parallel for' directive into a target region?}} + for (int i = 0; i < 10; ++i) + ; + } +#pragma omp target teams + { +#pragma omp target teams // expected-error {{region cannot be nested inside 'target teams' region}} + a++; + } + +} + +void foo() { + int a = 0; +// PARALLEL DIRECTIVE +#pragma omp parallel +#pragma omp for + for (int i = 0; i < 10; ++i) + ; +#pragma omp parallel +#pragma omp simd + for (int i = 0; i < 10; ++i) + ; +#pragma omp parallel +#pragma omp for simd + for (int i = 0; i < 10; ++i) + ; +#pragma omp parallel +#pragma omp sections + { + bar(); + } +#pragma omp parallel +#pragma omp section // expected-error {{'omp section' directive must be closely nested to a sections region, not a parallel region}} + { + bar(); + } +#pragma omp parallel +#pragma omp sections + { + bar(); + } +#pragma omp parallel +#pragma omp single + bar(); +#pragma omp parallel +#pragma omp master + bar(); +#pragma omp parallel +#pragma omp critical + bar(); +#pragma omp parallel +#pragma omp parallel for + for (int i = 0; i < 10; ++i) + ; +#pragma omp parallel +#pragma omp parallel for simd + for (int i = 0; i < 10; ++i) + ; +#pragma omp parallel +#pragma omp parallel sections + { + bar(); + } +#pragma omp parallel +#pragma omp task + { + bar(); + } +#pragma omp parallel + { +#pragma omp taskyield + bar(); + } +#pragma omp parallel + { +#pragma omp barrier + bar(); + } +#pragma omp parallel + { +#pragma omp taskwait + bar(); + } +#pragma omp parallel + { +#pragma omp flush + bar(); + } +#pragma omp parallel + { +#pragma omp ordered // expected-error {{region cannot be closely nested inside 'parallel' region; perhaps you forget to enclose 'omp ordered' directive into a for or a parallel for region with 'ordered' clause?}} + bar(); + } +#pragma omp parallel + { +#pragma omp atomic + ++a; + } +#pragma omp parallel + { +#pragma omp target + ++a; + } +#pragma omp parallel + { +#pragma omp target parallel + ++a; + } +#pragma omp parallel +#pragma omp target parallel for + for (int i = 0; i < 10; ++i) + ; +#pragma omp parallel + { +#pragma omp target enter data map(to: a) + ++a; + } +#pragma omp parallel + { +#pragma omp target exit data map(from: a) + ++a; + } +#pragma omp parallel + { +#pragma omp teams // expected-error {{region cannot be closely nested inside 'parallel' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} + ++a; + } +#pragma omp parallel + { +#pragma omp taskloop + for (int i = 0; i < 10; ++i) + ++a; + } +#pragma omp parallel + { +#pragma omp distribute // expected-error {{region cannot be closely nested inside 'parallel' region; perhaps you forget to enclose 'omp distribute' directive into a teams region?}} + for (int i = 0; i < 10; ++i) + ; + } +#pragma omp parallel + { +#pragma omp target update to(a) + a++; + } +#pragma omp parallel + { +#pragma omp distribute parallel for // expected-error {{region cannot be closely nested inside 'parallel' region; perhaps you forget to enclose 'omp distribute parallel for' directive into a teams region?}} + for (int i = 0; i < 10; ++i) + ; + } +#pragma omp parallel + { +#pragma omp distribute parallel for simd // expected-error {{region cannot be closely nested inside 'parallel' region; perhaps you forget to enclose 'omp distribute parallel for simd' directive into a teams region?}} + for (int i = 0; i < 10; ++i) + ; + } +#pragma omp parallel + { +#pragma omp distribute simd // expected-error {{region cannot be closely nested inside 'parallel' region; perhaps you forget to enclose 'omp distribute simd' directive into a teams region?}} + for (int i = 0; i < 10; ++i) + ; + } +#pragma omp parallel + { +#pragma omp target simd // OK + for (int i = 0; i < 10; ++i) + ; + } +#pragma omp parallel + { +#pragma omp teams distribute // expected-error {{region cannot be closely nested inside 'parallel' region; perhaps you forget to enclose 'omp teams distribute' directive into a target region?}} + for (int i = 0; i < 10; ++i) + ; + } +#pragma omp parallel + { +#pragma omp teams distribute simd // expected-error {{region cannot be closely nested inside 'parallel' region; perhaps you forget to enclose 'omp teams distribute simd' directive into a target region?}} + for (int i = 0; i < 10; ++i) + ; + } +#pragma omp parallel { #pragma omp teams distribute parallel for simd // expected-error {{region cannot be closely nested inside 'parallel' region; perhaps you forget to enclose 'omp teams distribute parallel for simd' directive into a target region?}} for (int i = 0; i < 10; ++i) @@ -7046,6 +7394,11 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp parallel + { +#pragma omp target teams // OK + a++; + } // SIMD DIRECTIVE #pragma omp simd @@ -7252,6 +7605,11 @@ for (int j = 0; j < 10; ++j) ; } +#pragma omp simd + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // expected-error {{OpenMP constructs may not be nested inside a simd region}} + a++; + } // FOR DIRECTIVE #pragma omp for @@ -7478,6 +7836,11 @@ for (int j = 0; j < 10; ++j) ; } +#pragma omp for + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // OK + a++; + } // FOR SIMD DIRECTIVE #pragma omp for simd @@ -7684,6 +8047,11 @@ for (int j = 0; j < 10; ++j) ; } +#pragma omp for simd + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // expected-error {{OpenMP constructs may not be nested inside a simd region}} + a++; + } // SECTIONS DIRECTIVE #pragma omp sections @@ -7899,6 +8267,11 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp sections + { +#pragma omp target teams // OK + a++; + } // SECTION DIRECTIVE #pragma omp section // expected-error {{orphaned 'omp section' directives are prohibited, it must be closely nested to a sections region}} @@ -8211,6 +8584,12 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp sections + { +#pragma omp section +#pragma omp target teams // OK + a++; + } // SINGLE DIRECTIVE #pragma omp single @@ -8427,6 +8806,11 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp single + { +#pragma omp target teams // oK + a++; + } // MASTER DIRECTIVE #pragma omp master @@ -8653,6 +9037,11 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp master + { +#pragma omp target teams // OK + a++; + } // CRITICAL DIRECTIVE #pragma omp critical @@ -8898,6 +9287,11 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp critical + { +#pragma omp target teams // OK + a++; + } // PARALLEL FOR DIRECTIVE #pragma omp parallel for @@ -9139,6 +9533,11 @@ for (int j = 0; j < 10; ++j) ; } +#pragma omp parallel for + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // OK + a++; + } // PARALLEL FOR SIMD DIRECTIVE #pragma omp parallel for simd @@ -9380,6 +9779,11 @@ for (int j = 0; j < 10; ++j) ; } +#pragma omp parallel for simd + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // expected-error {{OpenMP constructs may not be nested inside a simd region}} + a++; + } // PARALLEL SECTIONS DIRECTIVE #pragma omp parallel sections @@ -9605,6 +10009,11 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp parallel sections + { +#pragma omp target teams // OK + a++; + } // TASK DIRECTIVE #pragma omp task @@ -9780,6 +10189,11 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp task + { +#pragma omp target teams // OK + a++; + } // ATOMIC DIRECTIVE #pragma omp atomic @@ -10064,6 +10478,13 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp atomic + // expected-error@+2 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}} + // expected-note@+1 {{expected an expression statement}} + { +#pragma omp target teams // expected-error {{OpenMP constructs may not be nested inside an atomic region}} + a++; + } // TARGET DIRECTIVE #pragma omp target @@ -10253,6 +10674,11 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp target + { +#pragma omp target teams // expected-error {{region cannot be nested inside 'target' region}} + a++; + } // TARGET PARALLEL DIRECTIVE #pragma omp target parallel @@ -10439,6 +10865,11 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp target parallel + { +#pragma omp target teams // expected-error {{region cannot be nested inside 'target parallel' region}} + a++; + } // TARGET PARALLEL FOR DIRECTIVE #pragma omp target parallel for @@ -10680,6 +11111,11 @@ for (int j = 0; j < 10; ++j) ; } +#pragma omp target parallel for + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // expected-error {{region cannot be nested inside 'target parallel for' region}} + a++; + } // TEAMS DIRECTIVE #pragma omp target @@ -10791,7 +11227,7 @@ #pragma omp target #pragma omp teams { -#pragma omp target // expected-error {{region cannot be closely nested inside 'teams' region; perhaps you forget to enclose 'omp target' directive into a parallel region?}} +#pragma omp target // expected-error {{region cannot be nested inside 'target' region}} ++a; } #pragma omp target @@ -10808,12 +11244,12 @@ #pragma omp target #pragma omp teams { -#pragma omp target enter data map(to: a) // expected-error {{region cannot be closely nested inside 'teams' region; perhaps you forget to enclose 'omp target enter data' directive into a parallel region?}} +#pragma omp target enter data map(to: a) // expected-error {{region cannot be nested inside 'target' region}} } #pragma omp target #pragma omp teams { -#pragma omp target exit data map(from: a) // expected-error {{region cannot be closely nested inside 'teams' region; perhaps you forget to enclose 'omp target exit data' directive into a parallel region?}} +#pragma omp target exit data map(from: a) // expected-error {{region cannot be nested inside 'target' region}} } #pragma omp target #pragma omp teams @@ -10844,7 +11280,7 @@ #pragma omp target #pragma omp teams { -#pragma omp target update to(a) // expected-error {{region cannot be closely nested inside 'teams' region; perhaps you forget to enclose 'omp target update' directive into a parallel region?}} +#pragma omp target update to(a) // expected-error {{region cannot be nested inside 'target' region}} ++a; } #pragma omp target @@ -10885,7 +11321,7 @@ #pragma omp target #pragma omp teams { -#pragma omp target simd // expected-error {{region cannot be closely nested inside 'teams' region; perhaps you forget to enclose 'omp target simd' directive into a parallel region?}} +#pragma omp target simd // expected-error {{region cannot be nested inside 'target' region}} for (int i = 0; i < 10; ++i) ; } @@ -10917,6 +11353,12 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp target +#pragma omp teams + { +#pragma omp target teams // expected-error {{region cannot be nested inside 'target' region}} + a++; + } // TASKLOOP DIRECTIVE #pragma omp taskloop @@ -11148,6 +11590,11 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp taskloop + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // OK + a++; + } // DISTRIBUTE DIRECTIVE #pragma omp target @@ -11421,6 +11868,14 @@ for (int i = 0; i < 10; ++i) ++a; } +#pragma omp target +#pragma omp teams +#pragma omp distribute + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // expected-error {{region cannot be nested inside 'target' region}} + for (int i = 0; i < 10; ++i) + ++a; + } // DISTRIBUTE PARALLEL FOR DIRECTIVE #pragma omp target @@ -11703,6 +12158,13 @@ for (int i = 0; i < 10; ++i) ++a; } +#pragma omp target +#pragma omp teams +#pragma omp distribute parallel for + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // expected-error {{region cannot be nested inside 'target' region}} + ++a; + } // DISTRIBUTE PARALLEL FOR SIMD DIRECTIVE #pragma omp target @@ -11976,6 +12438,13 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp target +#pragma omp teams +#pragma omp distribute parallel for simd + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // expected-error {{OpenMP constructs may not be nested inside a simd region}} + a++; + } // DISTRIBUTE SIMD DIRECTIVE #pragma omp target @@ -12249,6 +12718,13 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp target +#pragma omp teams +#pragma omp distribute simd + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // expected-error {{OpenMP constructs may not be nested inside a simd region}} + a++; + } // TARGET SIMD DIRECTIVE #pragma omp target simd @@ -12454,6 +12930,11 @@ for (int i = 0; i < 10; ++i) ; } +#pragma omp target simd + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // expected-error {{OpenMP constructs may not be nested inside a simd region}} + a++; + } // TEAMS DISTRIBUTE DIRECTIVE #pragma omp teams distribute // expected-error {{orphaned 'omp teams distribute' directives are prohibited; perhaps you forget to enclose the directive into a target region?}} @@ -12703,6 +13184,12 @@ for (int i = 0; i < 10; ++i) ++a; } +#pragma omp target +#pragma omp teams distribute + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // expected-error {{region cannot be nested inside 'target' region}} + ++a; + } // TEAMS DISTRIBUTE SIMD DIRECTIVE #pragma omp teams distribute simd // expected-error {{orphaned 'omp teams distribute simd' directives are prohibited; perhaps you forget to enclose the directive into a target region?}} @@ -12952,6 +13439,12 @@ for (int i = 0; i < 10; ++i) ++a; } +#pragma omp target +#pragma omp teams distribute simd + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // expected-error {{OpenMP constructs may not be nested inside a simd region}} + ++a; + } // TEAMS DISTRIBUTE PARALLEL FOR SIMD DIRECTIVE #pragma omp teams distribute parallel for simd // expected-error {{orphaned 'omp teams distribute parallel for simd' directives are prohibited; perhaps you forget to enclose the directive into a target region?}} @@ -13201,8 +13694,12 @@ for (int i = 0; i < 10; ++i) ++a; } - -//-------------------- +#pragma omp target +#pragma omp teams distribute parallel for simd + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // expected-error {{OpenMP constructs may not be nested inside a simd region}} + ++a; + } // TEAMS DISTRIBUTE PARALLEL FOR DIRECTIVE #pragma omp teams distribute parallel for // expected-error {{orphaned 'omp teams distribute parallel for' directives are prohibited; perhaps you forget to enclose the directive into a target region?}} @@ -13452,6 +13949,209 @@ for (int i = 0; i < 10; ++i) ++a; } +#pragma omp target +#pragma omp teams distribute parallel for + for (int i = 0; i < 10; ++i) { +#pragma omp target teams // expected-error {{region cannot be nested inside 'target' region}} + ++a; + } + +// TARGET TEAMS DIRECTIVE +#pragma omp target teams +#pragma omp parallel + bar(); +#pragma omp target teams +#pragma omp for // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp for' directive into a parallel region?}} + for (int i = 0; i < 10; ++i) + ; +#pragma omp target teams +#pragma omp simd // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp simd' directive into a parallel region?}} + for (int i = 0; i < 10; ++i) + ; +#pragma omp target teams +#pragma omp for simd // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp for simd' directive into a parallel region?}} + for (int i = 0; i < 10; ++i) + ; +#pragma omp target teams +#pragma omp sections // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp sections' directive into a parallel region?}} + { + bar(); + } +#pragma omp target teams +#pragma omp section // expected-error {{'omp section' directive must be closely nested to a sections region, not a target teams region}} + { + bar(); + } +#pragma omp target teams +#pragma omp single // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp single' directive into a parallel region?}} + bar(); + +#pragma omp target teams +#pragma omp master // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp master' directive into a parallel region?}} + { + bar(); + } +#pragma omp target teams +#pragma omp critical // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp critical' directive into a parallel region?}} + { + bar(); + } +#pragma omp target teams +#pragma omp parallel for + for (int i = 0; i < 10; ++i) + ; +#pragma omp target teams +#pragma omp parallel for simd + for (int i = 0; i < 10; ++i) + ; +#pragma omp target teams +#pragma omp parallel sections + { + bar(); + } +#pragma omp target teams +#pragma omp task // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp task' directive into a parallel region?}} + { + bar(); + } +#pragma omp target teams + { +#pragma omp taskyield // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp taskyield' directive into a parallel region?}} + bar(); + } +#pragma omp target teams + { +#pragma omp barrier // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp barrier' directive into a parallel region?}} + bar(); + } +#pragma omp target teams + { +#pragma omp taskwait // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp taskwait' directive into a parallel region?}} + bar(); + } +#pragma omp target teams + { +#pragma omp flush // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp flush' directive into a parallel region?}} + bar(); + } +#pragma omp target teams + { +#pragma omp ordered // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp ordered' directive into a for or a parallel for region with 'ordered' clause?}} + bar(); + } +#pragma omp target teams + { +#pragma omp atomic // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp atomic' directive into a parallel region?}} + ++a; + } +#pragma omp target teams + { +#pragma omp target // expected-error {{region cannot be nested inside 'target teams' region}} + ++a; + } +#pragma omp target teams + { +#pragma omp target parallel // expected-error {{region cannot be nested inside 'target teams' region}} + ++a; + } +#pragma omp target teams +#pragma omp target parallel for // expected-error {{region cannot be nested inside 'target teams' region}} + for (int i = 0; i < 10; ++i) + ; +#pragma omp target teams + { +#pragma omp target enter data map(to: a) // expected-error {{region cannot be nested inside 'target teams' region}} + } +#pragma omp target teams + { +#pragma omp target exit data map(from: a) // expected-error {{region cannot be nested inside 'target teams' region}} + } +#pragma omp target teams + { +#pragma omp teams // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} + ++a; + } +#pragma omp target teams + { +#pragma omp taskloop // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp taskloop' directive into a parallel region?}} + for (int i = 0; i < 10; ++i) + ++a; + } +#pragma omp target teams +#pragma omp distribute + for (int i = 0; i < 10; ++i) + ; +#pragma omp target teams + { +#pragma omp target update to(a) // expected-error {{region cannot be nested inside 'target teams' region}} + ++a; + } +#pragma omp target teams +#pragma omp distribute parallel for + for (int i = 0; i < 10; ++i) + ; +#pragma omp target teams + { +#pragma omp distribute parallel for + for (int i = 0; i < 10; ++i) + ; +#pragma omp distribute parallel for + for (int j = 0; j < 10; ++j) + ; + } +#pragma omp target teams + { +#pragma omp distribute parallel for simd + for (int i = 0; i < 10; ++i) + ; +#pragma omp distribute parallel for simd + for (int j = 0; j < 10; ++j) + ; + } +#pragma omp target teams + { +#pragma omp distribute simd + for (int i = 0; i < 10; ++i) + ; +#pragma omp distribute simd + for (int j = 0; j < 10; ++j) + ; + } +#pragma omp target teams + { +#pragma omp target simd // expected-error {{region cannot be nested inside 'target teams' region}} + for (int i = 0; i < 10; ++i) + ; + } +#pragma omp target teams + { +#pragma omp teams distribute // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp teams distribute' directive into a target region?}} + for (int i = 0; i < 10; ++i) + ; + } +#pragma omp target teams + { +#pragma omp teams distribute simd // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp teams distribute simd' directive into a target region?}} + for (int i = 0; i < 10; ++i) + ; + } +#pragma omp target teams + { +#pragma omp teams distribute parallel for simd // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp teams distribute parallel for simd' directive into a target region?}} + for (int i = 0; i < 10; ++i) + ; + } +#pragma omp target teams + { +#pragma omp teams distribute parallel for // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp teams distribute parallel for' directive into a target region?}} + for (int i = 0; i < 10; ++i) + ; + } +#pragma omp target teams + { +#pragma omp target teams // expected-error {{region cannot be nested inside 'target teams' region}} + a++; + } return foo(); } Index: cfe/trunk/test/OpenMP/target_teams_ast_print.cpp =================================================================== --- cfe/trunk/test/OpenMP/target_teams_ast_print.cpp +++ cfe/trunk/test/OpenMP/target_teams_ast_print.cpp @@ -0,0 +1,98 @@ +// 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 +// expected-no-diagnostics + +#ifndef HEADER +#define HEADER + +void foo() {} + +template +struct S { + operator T() {return T();} + static T TS; + #pragma omp threadprivate(TS) +}; + +// CHECK: template struct S { +// CHECK: static T TS; +// CHECK-NEXT: #pragma omp threadprivate(S::TS) +// CHECK: }; +// CHECK: template<> struct S { +// CHECK: static int TS; +// CHECK-NEXT: #pragma omp threadprivate(S::TS) +// CHECK-NEXT: } +// CHECK: template<> struct S { +// CHECK: static long TS; +// CHECK-NEXT: #pragma omp threadprivate(S::TS) +// CHECK-NEXT: } + +template +T tmain(T argc, T *argv) { + T b = argc, c, d, e, f, g; + static T a; + S s; +#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) + foo(); +#pragma omp target teams reduction(^:e, f) reduction(&& : g) + foo(); + return 0; +} + +// CHECK: template T tmain(T argc, T *argv) { +// CHECK-NEXT: T b = argc, c, d, e, f, g; +// CHECK-NEXT: static T a; +// CHECK-NEXT: S s; +// 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) +// CHECK-NEXT: foo() +// CHECK-NEXT: #pragma omp target teams reduction(^: e,f) reduction(&&: g) +// 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: #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) +// CHECK-NEXT: foo() +// CHECK-NEXT: #pragma omp target teams reduction(^: e,f) reduction(&&: g) +// 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: #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) +// CHECK-NEXT: foo() +// CHECK-NEXT: #pragma omp target teams reduction(^: e,f) reduction(&&: g) +// CHECK-NEXT: foo() + +enum Enum { }; + +int main (int argc, char **argv) { + long x; + int b = argc, c, d, e, f, g; + static int a; + #pragma omp threadprivate(a) + Enum ee; +// CHECK: Enum ee; +#pragma omp target teams +// CHECK-NEXT: #pragma omp target teams + a=2; +// CHECK-NEXT: a = 2; +#pragma omp target teams default(none), private(argc,b) num_teams(f) firstprivate(argv) reduction(| : c, d) reduction(* : e) thread_limit(f+g) +// CHECK-NEXT: #pragma omp target teams default(none) private(argc,b) num_teams(f) firstprivate(argv) reduction(|: c,d) reduction(*: e) thread_limit(f + g) + foo(); +// CHECK-NEXT: foo(); + return tmain(b, &b) + tmain(x, &x); +} + +extern template int S::TS; +extern template long S::TS; +#endif Index: cfe/trunk/test/OpenMP/target_teams_default_messages.cpp =================================================================== --- cfe/trunk/test/OpenMP/target_teams_default_messages.cpp +++ cfe/trunk/test/OpenMP/target_teams_default_messages.cpp @@ -0,0 +1,26 @@ +// RUN: %clang_cc1 -verify -fopenmp -o - %s + +void foo(); + +int main(int argc, char **argv) { +#pragma omp target teams default // expected-error {{expected '(' after 'default'}} + foo(); +#pragma omp target teams default ( // expected-error {{expected 'none' or 'shared' in OpenMP clause 'default'}} expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams default () // expected-error {{expected 'none' or 'shared' in OpenMP clause 'default'}} + foo(); +#pragma omp target teams default (none // expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams default (shared), default(shared) // expected-error {{directive '#pragma omp target teams' cannot contain more than one 'default' clause}} + foo(); +#pragma omp target teams default (x) // expected-error {{expected 'none' or 'shared' in OpenMP clause 'default'}} + foo(); + +#pragma omp target teams default(none) + ++argc; // expected-error {{variable 'argc' must have explicitly specified data sharing attributes}} + +#pragma omp target teams default(none) +#pragma omp parallel default(shared) + ++argc; + return 0; +} Index: cfe/trunk/test/OpenMP/target_teams_defaultmap_messages.cpp =================================================================== --- cfe/trunk/test/OpenMP/target_teams_defaultmap_messages.cpp +++ cfe/trunk/test/OpenMP/target_teams_defaultmap_messages.cpp @@ -0,0 +1,56 @@ +// RUN: %clang_cc1 -verify -fopenmp %s + +void foo() { +} + +template +T tmain(T argc, S **argv) { + #pragma omp target teams defaultmap // expected-error {{expected '(' after 'defaultmap'}} + foo(); + #pragma omp target teams defaultmap ( // expected-error {{expected 'tofrom' in OpenMP clause 'defaultmap'}} expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); + #pragma omp target teams defaultmap () // expected-error {{expected 'tofrom' in OpenMP clause 'defaultmap'}} + foo(); + #pragma omp target teams defaultmap (tofrom // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-warning {{missing ':' after defaultmap modifier - ignoring}} expected-error {{expected 'scalar' in OpenMP clause 'defaultmap'}} + foo(); + #pragma omp target teams defaultmap (tofrom: // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected 'scalar' in OpenMP clause 'defaultmap'}} + foo(); + #pragma omp target teams defaultmap (tofrom) // expected-warning {{missing ':' after defaultmap modifier - ignoring}} expected-error {{expected 'scalar' in OpenMP clause 'defaultmap'}} + foo(); + #pragma omp target teams defaultmap (tofrom scalar) // expected-warning {{missing ':' after defaultmap modifier - ignoring}} + foo(); + #pragma omp target teams defaultmap (tofrom, // expected-error {{expected ')'}} expected-error {{expected 'scalar' in OpenMP clause 'defaultmap'}} expected-warning {{missing ':' after defaultmap modifier - ignoring}} expected-note {{to match this '('}} + foo(); + #pragma omp target teams defaultmap (scalar: // expected-error {{expected ')'}} expected-error {{expected 'tofrom' in OpenMP clause 'defaultmap'}} expected-note {{to match this '('}} + foo(); + #pragma omp target teams defaultmap (tofrom, scalar // expected-error {{expected ')'}} expected-warning {{missing ':' after defaultmap modifier - ignoring}} expected-error {{expected 'scalar' in OpenMP clause 'defaultmap'}} expected-note {{to match this '('}} + foo(); + + return argc; +} + +int main(int argc, char **argv) { + #pragma omp target teams defaultmap // expected-error {{expected '(' after 'defaultmap'}} + foo(); + #pragma omp target teams defaultmap ( // expected-error {{expected 'tofrom' in OpenMP clause 'defaultmap'}} expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); + #pragma omp target teams defaultmap () // expected-error {{expected 'tofrom' in OpenMP clause 'defaultmap'}} + foo(); + #pragma omp target teams defaultmap (tofrom // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-warning {{missing ':' after defaultmap modifier - ignoring}} expected-error {{expected 'scalar' in OpenMP clause 'defaultmap'}} + foo(); + #pragma omp target teams defaultmap (tofrom: // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected 'scalar' in OpenMP clause 'defaultmap'}} + foo(); + #pragma omp target teams defaultmap (tofrom) // expected-warning {{missing ':' after defaultmap modifier - ignoring}} expected-error {{expected 'scalar' in OpenMP clause 'defaultmap'}} + foo(); + #pragma omp target teams defaultmap (tofrom scalar) // expected-warning {{missing ':' after defaultmap modifier - ignoring}} + foo(); + #pragma omp target teams defaultmap (tofrom, // expected-error {{expected ')'}} expected-error {{expected 'scalar' in OpenMP clause 'defaultmap'}} expected-warning {{missing ':' after defaultmap modifier - ignoring}} expected-note {{to match this '('}} + foo(); + #pragma omp target teams defaultmap (scalar: // expected-error {{expected ')'}} expected-error {{expected 'tofrom' in OpenMP clause 'defaultmap'}} expected-note {{to match this '('}} + foo(); + #pragma omp target teams defaultmap (tofrom, scalar // expected-error {{expected ')'}} expected-warning {{missing ':' after defaultmap modifier - ignoring}} expected-error {{expected 'scalar' in OpenMP clause 'defaultmap'}} expected-note {{to match this '('}} + foo(); + + return tmain(argc, argv); +} + Index: cfe/trunk/test/OpenMP/target_teams_depend_messages.cpp =================================================================== --- cfe/trunk/test/OpenMP/target_teams_depend_messages.cpp +++ cfe/trunk/test/OpenMP/target_teams_depend_messages.cpp @@ -0,0 +1,89 @@ +// RUN: %clang_cc1 -verify -fopenmp -ferror-limit 100 -o - -std=c++11 %s + +void foo() { +} + +bool foobool(int argc) { + return argc; +} + +struct S1; // expected-note {{declared here}} + +class vector { + public: + int operator[](int index) { return 0; } +}; + +int main(int argc, char **argv, char *env[]) { + vector vec; + typedef float V __attribute__((vector_size(16))); + V a; + auto arr = x; // expected-error {{use of undeclared identifier 'x'}} + +#pragma omp target teams depend // expected-error {{expected '(' after 'depend'}} + foo(); +#pragma omp target teams depend ( // expected-error {{expected 'in', 'out' or 'inout' in OpenMP clause 'depend'}} expected-error {{expected ')'}} expected-note {{to match this '('}} expected-warning {{missing ':' after dependency type - ignoring}} + foo(); +#pragma omp target teams depend () // expected-error {{expected 'in', 'out' or 'inout' in OpenMP clause 'depend'}} expected-warning {{missing ':' after dependency type - ignoring}} + foo(); +#pragma omp target teams depend (argc // expected-error {{expected 'in', 'out' or 'inout' in OpenMP clause 'depend'}} expected-warning {{missing ':' after dependency type - ignoring}} expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams depend (source : argc) // expected-error {{expected 'in', 'out' or 'inout' in OpenMP clause 'depend'}} + foo(); +#pragma omp target teams depend (source) // expected-error {{expected expression}} expected-warning {{missing ':' after dependency type - ignoring}} + foo(); +#pragma omp target teams depend (in : argc)) // expected-warning {{extra tokens at the end of '#pragma omp target teams' are ignored}} + foo(); +#pragma omp target teams depend (out: ) // expected-error {{expected expression}} + foo(); +#pragma omp target teams depend (inout : foobool(argc)), depend (in, argc) // expected-error {{expected variable name, array element or array section}} expected-warning {{missing ':' after dependency type - ignoring}} expected-error {{expected expression}} + foo(); +#pragma omp target teams depend (out :S1) // expected-error {{'S1' does not refer to a value}} + foo(); +#pragma omp target teams depend(in : argv[1][1] = '2') // expected-error {{expected variable name, array element or array section}} + foo(); +#pragma omp target teams depend (in : vec[1]) // expected-error {{expected variable name, array element or array section}} + foo(); +#pragma omp target teams depend (in : argv[0]) + foo(); +#pragma omp target teams depend (in : ) // expected-error {{expected expression}} + foo(); +#pragma omp target teams depend (in : main) // expected-error {{expected variable name, array element or array section}} + foo(); +#pragma omp target teams depend(in : a[0]) // expected-error{{expected variable name, array element or array section}} + foo(); +#pragma omp target teams depend (in : vec[1:2]) // expected-error {{ value is not an array or pointer}} + foo(); +#pragma omp target teams depend (in : argv[ // expected-error {{expected expression}} expected-error {{expected ']'}} expected-error {{expected ')'}} expected-note {{to match this '['}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams depend (in : argv[: // expected-error {{expected expression}} expected-error {{expected ']'}} expected-error {{expected ')'}} expected-note {{to match this '['}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams depend (in : argv[:] // expected-error {{section length is unspecified and cannot be inferred because subscripted value is not an array}} expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams depend (in : argv[argc: // expected-error {{expected expression}} expected-error {{expected ']'}} expected-error {{expected ')'}} expected-note {{to match this '['}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams depend (in : argv[argc:argc] // expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams depend (in : argv[0:-1]) // expected-error {{section length is evaluated to a negative value -1}} + foo(); +#pragma omp target teams depend (in : argv[-1:0]) + foo(); +#pragma omp target teams depend (in : argv[:]) // expected-error {{section length is unspecified and cannot be inferred because subscripted value is not an array}} + foo(); +#pragma omp target teams depend (in : argv[3:4:1]) // expected-error {{expected ']'}} expected-note {{to match this '['}} + foo(); +#pragma omp target teams depend(in:a[0:1]) // expected-error {{subscripted value is not an array or pointer}} + foo(); +#pragma omp target teams depend(in:argv[argv[:2]:1]) // expected-error {{OpenMP array section is not allowed here}} + foo(); +#pragma omp target teams depend(in:argv[0:][:]) // expected-error {{section length is unspecified and cannot be inferred because subscripted value is not an array}} + foo(); +#pragma omp target teams depend(in:env[0:][:]) // expected-error {{section length is unspecified and cannot be inferred because subscripted value is an array of unknown bound}} + foo(); +#pragma omp target teams depend(in : argv[ : argc][1 : argc - 1]) + foo(); +#pragma omp target teams depend(in : arr[0]) + foo(); + + return 0; +} Index: cfe/trunk/test/OpenMP/target_teams_device_messages.cpp =================================================================== --- cfe/trunk/test/OpenMP/target_teams_device_messages.cpp +++ cfe/trunk/test/OpenMP/target_teams_device_messages.cpp @@ -0,0 +1,39 @@ +// RUN: %clang_cc1 -triple x86_64-apple-macos10.7.0 -verify -fopenmp -ferror-limit 100 -o - %s + +void foo() { +} + +bool foobool(int argc) { + return argc; +} + +struct S1; // expected-note {{declared here}} + +int main(int argc, char **argv) { +#pragma omp target teams device // expected-error {{expected '(' after 'device'}} + foo(); +#pragma omp target teams device ( // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams device () // expected-error {{expected expression}} + foo(); +#pragma omp target teams device (argc // expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams device (argc)) // expected-warning {{extra tokens at the end of '#pragma omp target teams' are ignored}} + foo(); +#pragma omp target teams device (argc > 0 ? argv[1] : argv[2]) // expected-error {{expression must have integral or unscoped enumeration type, not 'char *'}} + foo(); +#pragma omp target teams device (argc + argc) + foo(); +#pragma omp target teams device (argc), device (argc+1) // expected-error {{directive '#pragma omp target teams' cannot contain more than one 'device' clause}} + foo(); +#pragma omp target teams device (S1) // expected-error {{'S1' does not refer to a value}} + foo(); +#pragma omp target teams device (-2) // expected-error {{argument to 'device' clause must be a non-negative integer value}} + foo(); +#pragma omp target teams device (-10u) + foo(); +#pragma omp target teams device (3.14) // expected-error {{expression must have integral or unscoped enumeration type, not 'double'}} + foo(); + + return 0; +} Index: cfe/trunk/test/OpenMP/target_teams_firstprivate_messages.cpp =================================================================== --- cfe/trunk/test/OpenMP/target_teams_firstprivate_messages.cpp +++ cfe/trunk/test/OpenMP/target_teams_firstprivate_messages.cpp @@ -0,0 +1,114 @@ +// RUN: %clang_cc1 -verify -fopenmp %s + +void foo() { +} + +bool foobool(int argc) { + return argc; +} + +struct S1; // expected-note {{declared here}} expected-note{{forward declaration of 'S1'}} +extern S1 a; +class S2 { + mutable int a; + +public: + S2() : a(0) {} + S2(const S2 &s2) : a(s2.a) {} + static float S2s; + static const float S2sc; +}; +const float S2::S2sc = 0; +const S2 b; +const S2 ba[5]; +class S3 { + int a; + +public: + S3() : a(0) {} + S3(const S3 &s3) : a(s3.a) {} +}; +const S3 c; +const S3 ca[5]; +extern const int f; +class S4 { + int a; + S4(); + S4(const S4 &s4); // expected-note {{implicitly declared private here}} +public: + S4(int v) : a(v) {} +}; +class S5 { + int a; + S5() : a(0) {} + S5(const S5 &s5) : a(s5.a) {} // expected-note {{implicitly declared private here}} +public: + S5(int v) : a(v) {} +}; + +S3 h; +#pragma omp threadprivate(h) // expected-note {{defined as threadprivate or thread local}} + +namespace A { +double x; +#pragma omp threadprivate(x) // expected-note {{defined as threadprivate or thread local}} +} +namespace B { +using A::x; +} + +int main(int argc, char **argv) { + const int d = 5; + const int da[5] = {0}; + S4 e(4); + S5 g(5); + int i; + int &j = i; +#pragma omp target teams firstprivate // expected-error {{expected '(' after 'firstprivate'}} + foo(); +#pragma omp target teams firstprivate( // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams firstprivate() // expected-error {{expected expression}} + foo(); +#pragma omp target teams firstprivate(argc // expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams firstprivate(argc, // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams firstprivate(argc > 0 ? argv[1] : argv[2]) // expected-error {{expected variable name}} + foo(); +#pragma omp target teams firstprivate(argc) + foo(); +#pragma omp target teams firstprivate(S1) // expected-error {{'S1' does not refer to a value}} + foo(); +#pragma omp target teams firstprivate(a, b, c, d, f) // expected-error {{firstprivate variable with incomplete type 'S1'}} + foo(); +#pragma omp target teams firstprivate(argv[1]) // expected-error {{expected variable name}} + foo(); +#pragma omp target teams firstprivate(ba) + foo(); +#pragma omp target teams firstprivate(ca) + foo(); +#pragma omp target teams firstprivate(da) + foo(); +#pragma omp target teams firstprivate(S2::S2s) + foo(); +#pragma omp target teams firstprivate(S2::S2sc) + foo(); +#pragma omp target teams firstprivate(e, g) // expected-error {{calling a private constructor of class 'S4'}} expected-error {{calling a private constructor of class 'S5'}} + foo(); +#pragma omp target teams firstprivate(h, B::x) // expected-error 2 {{threadprivate or thread local variable cannot be firstprivate}} + foo(); +#pragma omp target teams private(i), firstprivate(i) // expected-error {{private variable cannot be firstprivate}} expected-note{{defined as private}} + foo(); +#pragma omp target teams shared(i) + foo(); +#pragma omp target teams firstprivate(i) + foo(); +#pragma omp target teams firstprivate(j) + foo(); + static int m; +#pragma omp target teams firstprivate(m) // OK + foo(); + + return 0; +} Index: cfe/trunk/test/OpenMP/target_teams_if_messages.cpp =================================================================== --- cfe/trunk/test/OpenMP/target_teams_if_messages.cpp +++ cfe/trunk/test/OpenMP/target_teams_if_messages.cpp @@ -0,0 +1,91 @@ +// RUN: %clang_cc1 -verify -fopenmp %s + +void foo() { +} + +bool foobool(int argc) { + return argc; +} + +struct S1; // expected-note {{declared here}} + +template // expected-note {{declared here}} +int tmain(T argc, S **argv) { +#pragma omp target teams if // expected-error {{expected '(' after 'if'}} + foo(); +#pragma omp target teams if ( // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams if () // expected-error {{expected expression}} + foo(); +#pragma omp target teams if (argc // expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams if (argc)) // expected-warning {{extra tokens at the end of '#pragma omp target teams' are ignored}} + foo(); +#pragma omp target teams if (argc > 0 ? argv[1] : argv[2]) + foo(); +#pragma omp target teams if (foobool(argc)), if (true) // expected-error {{directive '#pragma omp target teams' cannot contain more than one 'if' clause}} + foo(); +#pragma omp target teams if (S) // expected-error {{'S' does not refer to a value}} + foo(); +#pragma omp target teams if (argv[1]=2) // expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams if (argc argc) // expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams if(argc) + foo(); +#pragma omp target teams if(target : // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams if(target : argc // expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams if(target : argc) + foo(); +#pragma omp target teams if(target : argc) if (for:argc) // expected-error {{directive name modifier 'for' is not allowed for '#pragma omp target teams'}} + foo(); +#pragma omp target teams if(target : argc) if (target:argc) // expected-error {{directive '#pragma omp target teams' cannot contain more than one 'if' clause with 'target' name modifier}} + foo(); +#pragma omp target teams if(target : argc) if (argc) // expected-error {{no more 'if' clause is allowed}} expected-note {{previous clause with directive name modifier specified here}} + foo(); + + return 0; +} + +int main(int argc, char **argv) { +#pragma omp target teams if // expected-error {{expected '(' after 'if'}} + foo(); +#pragma omp target teams if ( // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams if () // expected-error {{expected expression}} + foo(); +#pragma omp target teams if (argc // expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams if (argc)) // expected-warning {{extra tokens at the end of '#pragma omp target teams' are ignored}} + foo(); +#pragma omp target teams if (argc > 0 ? argv[1] : argv[2]) + foo(); +#pragma omp target teams if (foobool(argc)), if (true) // expected-error {{directive '#pragma omp target teams' cannot contain more than one 'if' clause}} + foo(); +#pragma omp target teams if (S1) // expected-error {{'S1' does not refer to a value}} + foo(); +#pragma omp target teams if (argv[1]=2) // expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams if (argc argc) // expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams if (1 0) // expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams if(if(tmain(argc, argv) // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams if(target : // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams if(target : argc // expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams if(target : argc) + foo(); +#pragma omp target teams if(target : argc) if (for:argc) // expected-error {{directive name modifier 'for' is not allowed for '#pragma omp target teams'}} + foo(); +#pragma omp target teams if(target : argc) if (target:argc) // expected-error {{directive '#pragma omp target teams' cannot contain more than one 'if' clause with 'target' name modifier}} + foo(); +#pragma omp target teams if(target : argc) if (argc) // expected-error {{no more 'if' clause is allowed}} expected-note {{previous clause with directive name modifier specified here}} + foo(); + + return tmain(argc, argv); +} Index: cfe/trunk/test/OpenMP/target_teams_is_device_ptr_ast_print.cpp =================================================================== --- cfe/trunk/test/OpenMP/target_teams_is_device_ptr_ast_print.cpp +++ cfe/trunk/test/OpenMP/target_teams_is_device_ptr_ast_print.cpp @@ -0,0 +1,291 @@ +// RUN: %clang_cc1 -verify -fopenmp -std=c++11 -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 +// expected-no-diagnostics + +#ifndef HEADER +#define HEADER +struct ST { + int *a; +}; +typedef int arr[10]; +typedef ST STarr[10]; +struct SA { + const int da[5] = { 0 }; + ST g[10]; + STarr &rg = g; + int i; + int &j = i; + int *k = &j; + int *&z = k; + int aa[10]; + arr &raa = aa; + void func(int arg) { +#pragma omp target teams is_device_ptr(k) + {} +#pragma omp target teams is_device_ptr(z) + {} +#pragma omp target teams is_device_ptr(aa) // OK + {} +#pragma omp target teams is_device_ptr(raa) // OK + {} +#pragma omp target teams is_device_ptr(g) // OK + {} +#pragma omp target teams is_device_ptr(rg) // OK + {} +#pragma omp target teams is_device_ptr(da) // OK + {} + return; + } +}; +// CHECK: struct SA +// CHECK-NEXT: const int da[5] = {0}; +// CHECK-NEXT: ST g[10]; +// CHECK-NEXT: STarr &rg = this->g; +// CHECK-NEXT: int i; +// CHECK-NEXT: int &j = this->i; +// CHECK-NEXT: int *k = &this->j; +// CHECK-NEXT: int *&z = this->k; +// CHECK-NEXT: int aa[10]; +// CHECK-NEXT: arr &raa = this->aa; +// CHECK-NEXT: func( +// CHECK-NEXT: #pragma omp target teams is_device_ptr(this->k) +// CHECK-NEXT: { +// CHECK-NEXT: } +// CHECK-NEXT: #pragma omp target teams is_device_ptr(this->z) +// CHECK-NEXT: { +// CHECK-NEXT: } +// CHECK-NEXT: #pragma omp target teams is_device_ptr(this->aa) +// CHECK-NEXT: { +// CHECK-NEXT: } +// CHECK-NEXT: #pragma omp target teams is_device_ptr(this->raa) +// CHECK-NEXT: { +// CHECK-NEXT: } +// CHECK-NEXT: #pragma omp target teams is_device_ptr(this->g) +// CHECK-NEXT: { +// CHECK-NEXT: } +// CHECK-NEXT: #pragma omp target teams is_device_ptr(this->rg) +// CHECK-NEXT: { +// CHECK-NEXT: } +// CHECK-NEXT: #pragma omp target teams is_device_ptr(this->da) + +struct SB { + unsigned A; + unsigned B; + float Arr[100]; + float *Ptr; + float *foo() { + return &Arr[0]; + } +}; + +struct SC { + unsigned A : 2; + unsigned B : 3; + unsigned C; + unsigned D; + float Arr[100]; + SB S; + SB ArrS[100]; + SB *PtrS; + SB *&RPtrS; + float *Ptr; + + SC(SB *&_RPtrS) : RPtrS(_RPtrS) {} +}; + +union SD { + unsigned A; + float B; +}; + +struct S1; +extern S1 a; +class S2 { + mutable int a; +public: + S2():a(0) { } + S2(S2 &s2):a(s2.a) { } + static float S2s; + static const float S2sc; +}; +const float S2::S2sc = 0; +const S2 b; +const S2 ba[5]; +class S3 { + int a; +public: + S3():a(0) { } + S3(S3 &s3):a(s3.a) { } +}; +const S3 c; +const S3 ca[5]; +extern const int f; +class S4 { + int a; + S4(); + S4(const S4 &s4); +public: + S4(int v):a(v) { } +}; +class S5 { + int a; + S5():a(0) {} + S5(const S5 &s5):a(s5.a) { } +public: + S5(int v):a(v) { } +}; + +S3 h; +#pragma omp threadprivate(h) + +typedef struct { + int a; +} S6; + +template +T tmain(T argc) { + const T da[5] = { 0 }; + S6 h[10]; + auto &rh = h; + T i; + T &j = i; + T *k = &j; + T *&z = k; + T aa[10]; + auto &raa = aa; +#pragma omp target teams is_device_ptr(k) + {} +#pragma omp target teams is_device_ptr(z) + {} +#pragma omp target teams is_device_ptr(aa) + {} +#pragma omp target teams is_device_ptr(raa) + {} +#pragma omp target teams is_device_ptr(h) + {} +#pragma omp target teams is_device_ptr(rh) + {} +#pragma omp target teams is_device_ptr(da) + {} + return 0; +} + +// CHECK: template<> int tmain(int argc) { +// CHECK-NEXT: const int da[5] = {0}; +// CHECK-NEXT: S6 h[10]; +// CHECK-NEXT: auto &rh = h; +// CHECK-NEXT: int i; +// CHECK-NEXT: int &j = i; +// CHECK-NEXT: int *k = &j; +// CHECK-NEXT: int *&z = k; +// CHECK-NEXT: int aa[10]; +// CHECK-NEXT: auto &raa = aa; +// CHECK-NEXT: #pragma omp target teams is_device_ptr(k) +// CHECK-NEXT: { +// CHECK-NEXT: } +// CHECK-NEXT: #pragma omp target teams is_device_ptr(z) +// CHECK-NEXT: { +// CHECK-NEXT: } +// CHECK-NEXT: #pragma omp target teams is_device_ptr(aa) +// CHECK-NEXT: { +// CHECK-NEXT: } +// CHECK-NEXT: #pragma omp target teams is_device_ptr(raa) +// CHECK-NEXT: { +// CHECK-NEXT: } +// CHECK-NEXT: #pragma omp target teams is_device_ptr(h) +// CHECK-NEXT: { +// CHECK-NEXT: } +// CHECK-NEXT: #pragma omp target teams is_device_ptr(rh) +// CHECK-NEXT: { +// CHECK-NEXT: } +// CHECK-NEXT: #pragma omp target teams is_device_ptr(da) + +// CHECK: template<> int *tmain(int *argc) { +// CHECK-NEXT: int *const da[5] = {0}; +// CHECK-NEXT: S6 h[10]; +// CHECK-NEXT: auto &rh = h; +// CHECK-NEXT: int *i; +// CHECK-NEXT: int *&j = i; +// CHECK-NEXT: int **k = &j; +// CHECK-NEXT: int **&z = k; +// CHECK-NEXT: int *aa[10]; +// CHECK-NEXT: auto &raa = aa; +// CHECK-NEXT: #pragma omp target teams is_device_ptr(k) +// CHECK-NEXT: { +// CHECK-NEXT: } +// CHECK-NEXT: #pragma omp target teams is_device_ptr(z) +// CHECK-NEXT: { +// CHECK-NEXT: } +// CHECK-NEXT: #pragma omp target teams is_device_ptr(aa) +// CHECK-NEXT: { +// CHECK-NEXT: } +// CHECK-NEXT: #pragma omp target teams is_device_ptr(raa) +// CHECK-NEXT: { +// CHECK-NEXT: } +// CHECK-NEXT: #pragma omp target teams is_device_ptr(h) +// CHECK-NEXT: { +// CHECK-NEXT: } +// CHECK-NEXT: #pragma omp target teams is_device_ptr(rh) +// CHECK-NEXT: { +// CHECK-NEXT: } +// CHECK-NEXT: #pragma omp target teams is_device_ptr(da) + +// CHECK-LABEL: int main(int argc, char **argv) { +int main(int argc, char **argv) { + const int da[5] = { 0 }; + S6 h[10]; + auto &rh = h; + int i; + int &j = i; + int *k = &j; + int *&z = k; + int aa[10]; + auto &raa = aa; +// CHECK-NEXT: const int da[5] = {0}; +// CHECK-NEXT: S6 h[10]; +// CHECK-NEXT: auto &rh = h; +// CHECK-NEXT: int i; +// CHECK-NEXT: int &j = i; +// CHECK-NEXT: int *k = &j; +// CHECK-NEXT: int *&z = k; +// CHECK-NEXT: int aa[10]; +// CHECK-NEXT: auto &raa = aa; +#pragma omp target teams is_device_ptr(k) +// CHECK-NEXT: #pragma omp target teams is_device_ptr(k) + {} +// CHECK-NEXT: { +// CHECK-NEXT: } +#pragma omp target teams is_device_ptr(z) +// CHECK-NEXT: #pragma omp target teams is_device_ptr(z) + {} +// CHECK-NEXT: { +// CHECK-NEXT: } +#pragma omp target teams is_device_ptr(aa) +// CHECK-NEXT: #pragma omp target teams is_device_ptr(aa) + {} +// CHECK-NEXT: { +// CHECK-NEXT: } +#pragma omp target teams is_device_ptr(raa) +// CHECK-NEXT: #pragma omp target teams is_device_ptr(raa) + {} +// CHECK-NEXT: { +// CHECK-NEXT: } +#pragma omp target teams is_device_ptr(h) +// CHECK-NEXT: #pragma omp target teams is_device_ptr(h) + {} +// CHECK-NEXT: { +// CHECK-NEXT: } +#pragma omp target teams is_device_ptr(rh) +// CHECK-NEXT: #pragma omp target teams is_device_ptr(rh) + {} +// CHECK-NEXT: { +// CHECK-NEXT: } +#pragma omp target teams is_device_ptr(da) +// CHECK-NEXT: #pragma omp target teams is_device_ptr(da) + {} +// CHECK-NEXT: { +// CHECK-NEXT: } + return tmain(argc) + *tmain(&argc); +} +#endif Index: cfe/trunk/test/OpenMP/target_teams_is_device_ptr_messages.cpp =================================================================== --- cfe/trunk/test/OpenMP/target_teams_is_device_ptr_messages.cpp +++ cfe/trunk/test/OpenMP/target_teams_is_device_ptr_messages.cpp @@ -0,0 +1,268 @@ +// RUN: %clang_cc1 -std=c++11 -verify -fopenmp -ferror-limit 200 %s +struct ST { + int *a; +}; +typedef int arr[10]; +typedef ST STarr[10]; +struct SA { + const int d = 5; + const int da[5] = { 0 }; + ST e; + ST g[10]; + STarr &rg = g; + int i; + int &j = i; + int *k = &j; + int *&z = k; + int aa[10]; + arr &raa = aa; + void func(int arg) { +#pragma omp target teams is_device_ptr // expected-error {{expected '(' after 'is_device_ptr'}} + {} +#pragma omp target teams is_device_ptr( // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected expression}} + {} +#pragma omp target teams is_device_ptr() // expected-error {{expected expression}} + {} +#pragma omp target teams is_device_ptr(alloc) // expected-error {{use of undeclared identifier 'alloc'}} + {} +#pragma omp target teams is_device_ptr(arg // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}} + {} +#pragma omp target teams is_device_ptr(i) // expected-error {{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}} + {} +#pragma omp target teams is_device_ptr(j) // expected-error {{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}} + {} +#pragma omp target teams is_device_ptr(k) // OK + {} +#pragma omp target teams is_device_ptr(z) // OK + {} +#pragma omp target teams is_device_ptr(aa) // OK + {} +#pragma omp target teams is_device_ptr(raa) // OK + {} +#pragma omp target teams is_device_ptr(e) // expected-error{{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}} + {} +#pragma omp target teams is_device_ptr(g) // OK + {} +#pragma omp target teams is_device_ptr(rg) // OK + {} +#pragma omp target teams is_device_ptr(k,i,j) // expected-error2 {{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}} + {} +#pragma omp target teams is_device_ptr(d) // expected-error{{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}} + {} +#pragma omp target teams is_device_ptr(da) // OK + {} + return; + } +}; +struct SB { + unsigned A; + unsigned B; + float Arr[100]; + float *Ptr; + float *foo() { + return &Arr[0]; + } +}; + +struct SC { + unsigned A : 2; + unsigned B : 3; + unsigned C; + unsigned D; + float Arr[100]; + SB S; + SB ArrS[100]; + SB *PtrS; + SB *&RPtrS; + float *Ptr; + + SC(SB *&_RPtrS) : RPtrS(_RPtrS) {} +}; + +union SD { + unsigned A; + float B; +}; + +struct S1; +extern S1 a; +class S2 { + mutable int a; +public: + S2():a(0) { } + S2(S2 &s2):a(s2.a) { } + static float S2s; + static const float S2sc; +}; +const float S2::S2sc = 0; +const S2 b; +const S2 ba[5]; +class S3 { + int a; +public: + S3():a(0) { } + S3(S3 &s3):a(s3.a) { } +}; +const S3 c; +const S3 ca[5]; +extern const int f; +class S4 { + int a; + S4(); + S4(const S4 &s4); +public: + S4(int v):a(v) { } +}; +class S5 { + int a; + S5():a(0) {} + S5(const S5 &s5):a(s5.a) { } +public: + S5(int v):a(v) { } +}; + +S3 h; +#pragma omp threadprivate(h) + +typedef struct { + int a; +} S6; + +template +T tmain(T argc) { + const T d = 5; + const T da[5] = { 0 }; + S4 e(4); + S5 g(5); + S6 h[10]; + auto &rh = h; + T i; + T &j = i; + T *k = &j; + T *&z = k; + T aa[10]; + auto &raa = aa; + S6 *ps; +#pragma omp target teams is_device_ptr // expected-error {{expected '(' after 'is_device_ptr'}} + {} +#pragma omp target teams is_device_ptr( // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected expression}} + {} +#pragma omp target teams is_device_ptr() // expected-error {{expected expression}} + {} +#pragma omp target teams is_device_ptr(alloc) // expected-error {{use of undeclared identifier 'alloc'}} + {} +#pragma omp target teams is_device_ptr(argc // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error{{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}} + {} +#pragma omp target teams is_device_ptr(i) // expected-error {{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}} + {} +#pragma omp target teams is_device_ptr(j) // expected-error {{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}} + {} +#pragma omp target teams is_device_ptr(k) // OK + {} +#pragma omp target teams is_device_ptr(z) // OK + {} +#pragma omp target teams is_device_ptr(aa) // OK + {} +#pragma omp target teams is_device_ptr(raa) // OK + {} +#pragma omp target teams is_device_ptr(e) // expected-error{{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}} + {} +#pragma omp target teams is_device_ptr(g) // expected-error{{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}} + {} +#pragma omp target teams is_device_ptr(h) // OK + {} +#pragma omp target teams is_device_ptr(rh) // OK + {} +#pragma omp target teams is_device_ptr(k,i,j) // expected-error2 {{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}} + {} +#pragma omp target teams is_device_ptr(d) // expected-error{{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}} + {} +#pragma omp target teams is_device_ptr(da) // OK + {} +#pragma omp target teams map(ps) is_device_ptr(ps) // expected-error{{variable already marked as mapped in current construct}} expected-note{{used here}} + {} +#pragma omp target teams is_device_ptr(ps) map(ps) // expected-error{{variable already marked as mapped in current construct}} expected-note{{used here}} + {} +#pragma omp target teams map(ps->a) is_device_ptr(ps) // expected-error{{variable already marked as mapped in current construct}} expected-note{{used here}} + {} +#pragma omp target teams is_device_ptr(ps) map(ps->a) // expected-error{{pointer cannot be mapped along with a section derived from itself}} expected-note{{used here}} + {} +#pragma omp target teams is_device_ptr(ps) firstprivate(ps) // expected-error{{firstprivate variable cannot be in a is_device_ptr clause in '#pragma omp target teams' directive}} + {} +#pragma omp target teams firstprivate(ps) is_device_ptr(ps) // expected-error{{firstprivate variable cannot be in a is_device_ptr clause in '#pragma omp target teams' directive}} expected-note{{defined as firstprivate}} + {} +#pragma omp target teams is_device_ptr(ps) private(ps) // expected-error{{private variable cannot be in a is_device_ptr clause in '#pragma omp target teams' directive}} + {} +#pragma omp target teams private(ps) is_device_ptr(ps) // expected-error{{private variable cannot be in a is_device_ptr clause in '#pragma omp target teams' directive}} expected-note{{defined as private}} + {} + return 0; +} + +int main(int argc, char **argv) { + const int d = 5; + const int da[5] = { 0 }; + S4 e(4); + S5 g(5); + S6 h[10]; + auto &rh = h; + int i; + int &j = i; + int *k = &j; + int *&z = k; + int aa[10]; + auto &raa = aa; + S6 *ps; +#pragma omp target teams is_device_ptr // expected-error {{expected '(' after 'is_device_ptr'}} + {} +#pragma omp target teams is_device_ptr( // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected expression}} + {} +#pragma omp target teams is_device_ptr() // expected-error {{expected expression}} + {} +#pragma omp target teams is_device_ptr(alloc) // expected-error {{use of undeclared identifier 'alloc'}} + {} +#pragma omp target teams is_device_ptr(argc // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}} + {} +#pragma omp target teams is_device_ptr(i) // expected-error {{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}} + {} +#pragma omp target teams is_device_ptr(j) // expected-error {{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}} + {} +#pragma omp target teams is_device_ptr(k) // OK + {} +#pragma omp target teams is_device_ptr(z) // OK + {} +#pragma omp target teams is_device_ptr(aa) // OK + {} +#pragma omp target teams is_device_ptr(raa) // OK + {} +#pragma omp target teams is_device_ptr(e) // expected-error{{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}} + {} +#pragma omp target teams is_device_ptr(g) // expected-error{{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}} + {} +#pragma omp target teams is_device_ptr(h) // OK + {} +#pragma omp target teams is_device_ptr(rh) // OK + {} +#pragma omp target teams is_device_ptr(k,i,j) // expected-error2 {{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}} + {} +#pragma omp target teams is_device_ptr(d) // expected-error{{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}} + {} +#pragma omp target teams is_device_ptr(da) // OK + {} +#pragma omp target teams map(ps) is_device_ptr(ps) // expected-error{{variable already marked as mapped in current construct}} expected-note{{used here}} + {} +#pragma omp target teams is_device_ptr(ps) map(ps) // expected-error{{variable already marked as mapped in current construct}} expected-note{{used here}} + {} +#pragma omp target teams map(ps->a) is_device_ptr(ps) // expected-error{{variable already marked as mapped in current construct}} expected-note{{used here}} + {} +#pragma omp target teams is_device_ptr(ps) map(ps->a) // expected-error{{pointer cannot be mapped along with a section derived from itself}} expected-note{{used here}} + {} +#pragma omp target teams is_device_ptr(ps) firstprivate(ps) // expected-error{{firstprivate variable cannot be in a is_device_ptr clause in '#pragma omp target teams' directive}} + {} +#pragma omp target teams firstprivate(ps) is_device_ptr(ps) // expected-error{{firstprivate variable cannot be in a is_device_ptr clause in '#pragma omp target teams' directive}} expected-note{{defined as firstprivate}} + {} +#pragma omp target teams is_device_ptr(ps) private(ps) // expected-error{{private variable cannot be in a is_device_ptr clause in '#pragma omp target teams' directive}} + {} +#pragma omp target teams private(ps) is_device_ptr(ps) // expected-error{{private variable cannot be in a is_device_ptr clause in '#pragma omp target teams' directive}} expected-note{{defined as private}} + {} + return tmain(argc); // expected-note {{in instantiation of function template specialization 'tmain' requested here}} +} Index: cfe/trunk/test/OpenMP/target_teams_map_messages.cpp =================================================================== --- cfe/trunk/test/OpenMP/target_teams_map_messages.cpp +++ cfe/trunk/test/OpenMP/target_teams_map_messages.cpp @@ -0,0 +1,538 @@ +// RUN: %clang_cc1 -verify -fopenmp -ferror-limit 200 %s +// rUN: %clang_cc1 -DCCODE -verify -fopenmp -ferror-limit 200 -x c %s +#ifdef CCODE +void foo(int arg) { + const int n = 0; + + double marr[10][10][10]; + + #pragma omp target teams map(marr[2][0:2][0:2]) // expected-error {{array section does not specify contiguous storage}} + {} + #pragma omp target teams map(marr[:][0:][:]) + {} + #pragma omp target teams map(marr[:][1:][:]) // expected-error {{array section does not specify contiguous storage}} + {} + #pragma omp target teams map(marr[:][n:][:]) + {} +} +#else +template +struct SA { + static int ss; + #pragma omp threadprivate(ss) // expected-note {{defined as threadprivate or thread local}} + float a; + int b[12]; + float *c; + T d; + float e[I]; + T *f; + void func(int arg) { + #pragma omp target teams map(arg,a,d) + {} + #pragma omp target teams map(arg[2:2],a,d) // expected-error {{subscripted value is not an array or pointer}} + {} + #pragma omp target teams map(arg,a*2) // expected-error {{expected expression containing only member accesses and/or array sections based on named variables}} + {} + #pragma omp target teams map(arg,(c+1)[2]) // expected-error {{expected expression containing only member accesses and/or array sections based on named variables}} + {} + #pragma omp target teams map(arg,a[:2],d) // expected-error {{subscripted value is not an array or pointer}} + {} + #pragma omp target teams map(arg,a,d[:2]) // expected-error {{subscripted value is not an array or pointer}} + {} + + #pragma omp target teams map(to:ss) // expected-error {{threadprivate variables are not allowed in 'map' clause}} + {} + + #pragma omp target teams map(to:b,e) + {} + #pragma omp target teams map(to:b,e) map(to:b) // expected-error {{variable already marked as mapped in current construct}} expected-note {{used here}} + {} + #pragma omp target teams map(to:b[:2],e) + {} + #pragma omp target teams map(to:b,e[:]) + {} + #pragma omp target teams map(b[-1:]) // expected-error {{array section must be a subset of the original array}} + {} + #pragma omp target teams map(b[:-1]) // expected-error {{section length is evaluated to a negative value -1}} + {} + + #pragma omp target teams map(always, tofrom: c,f) + {} + #pragma omp target teams map(always, tofrom: c[1:2],f) + {} + #pragma omp target teams map(always, tofrom: c,f[1:2]) + {} + #pragma omp target teams map(always, tofrom: c[:],f) // expected-error {{section length is unspecified and cannot be inferred because subscripted value is not an array}} + {} + #pragma omp target teams map(always, tofrom: c,f[:]) // expected-error {{section length is unspecified and cannot be inferred because subscripted value is not an array}} + {} + return; + } +}; + +struct SB { + unsigned A; + unsigned B; + float Arr[100]; + float *Ptr; + float *foo() { + return &Arr[0]; + } +}; + +struct SC { + unsigned A : 2; + unsigned B : 3; + unsigned C; + unsigned D; + float Arr[100]; + SB S; + SB ArrS[100]; + SB *PtrS; + SB *&RPtrS; + float *Ptr; + + SC(SB *&_RPtrS) : RPtrS(_RPtrS) {} +}; + +union SD { + unsigned A; + float B; +}; + +void SAclient(int arg) { + SA s; + s.func(arg); // expected-note {{in instantiation of member function}} + double marr[10][10][10]; + double marr2[5][10][1]; + double mvla[5][arg][10]; + double ***mptr; + const int n = 0; + const int m = 1; + double mvla2[5][arg][m+n+10]; + + SB *p; + + SD u; + SC r(p),t(p); + #pragma omp target teams map(r) + {} + #pragma omp target teams map(marr[2][0:2][0:2]) // expected-error {{array section does not specify contiguous storage}} + {} + #pragma omp target teams map(marr[:][0:2][0:2]) // expected-error {{array section does not specify contiguous storage}} + {} + #pragma omp target teams map(marr[2][3][0:2]) + {} + #pragma omp target teams map(marr[:][:][:]) + {} + #pragma omp target teams map(marr[:2][:][:]) + {} + #pragma omp target teams map(marr[arg:][:][:]) + {} + #pragma omp target teams map(marr[arg:]) + {} + #pragma omp target teams map(marr[arg:][:arg][:]) // correct if arg is the size of dimension 2 + {} + #pragma omp target teams map(marr[:arg][:]) + {} + #pragma omp target teams map(marr[:arg][n:]) + {} + #pragma omp target teams map(marr[:][:arg][n:]) // correct if arg is the size of dimension 2 + {} + #pragma omp target teams map(marr[:][:m][n:]) // expected-error {{array section does not specify contiguous storage}} + {} + #pragma omp target teams map(marr[n:m][:arg][n:]) + {} + #pragma omp target teams map(marr[:2][:1][:]) // expected-error {{array section does not specify contiguous storage}} + {} + #pragma omp target teams map(marr[:2][1:][:]) // expected-error {{array section does not specify contiguous storage}} + {} + #pragma omp target teams map(marr[:2][:][:1]) // expected-error {{array section does not specify contiguous storage}} + {} + #pragma omp target teams map(marr[:2][:][1:]) // expected-error {{array section does not specify contiguous storage}} + {} + #pragma omp target teams map(marr[:1][:2][:]) + {} + #pragma omp target teams map(marr[:1][0][:]) + {} + #pragma omp target teams map(marr[:arg][:2][:]) // correct if arg is 1 + {} + #pragma omp target teams map(marr[:1][3:1][:2]) + {} + #pragma omp target teams map(marr[:1][3:arg][:2]) // correct if arg is 1 + {} + #pragma omp target teams map(marr[:1][3:2][:2]) // expected-error {{array section does not specify contiguous storage}} + {} + #pragma omp target teams map(marr[:2][:10][:]) + {} + #pragma omp target teams map(marr[:2][:][:5+5]) + {} + #pragma omp target teams map(marr[:2][2+2-4:][0:5+5]) + {} + + #pragma omp target teams map(marr[:1][:2][0]) // expected-error {{array section does not specify contiguous storage}} + {} + #pragma omp target teams map(marr2[:1][:2][0]) + {} + + #pragma omp target teams map(mvla[:1][:][0]) // correct if the size of dimension 2 is 1. + {} + #pragma omp target teams map(mvla[:2][:arg][:]) // correct if arg is the size of dimension 2. + {} + #pragma omp target teams map(mvla[:1][:2][0]) // expected-error {{array section does not specify contiguous storage}} + {} + #pragma omp target teams map(mvla[1][2:arg][:]) + {} + #pragma omp target teams map(mvla[:1][:][:]) + {} + #pragma omp target teams map(mvla2[:1][:2][:11]) + {} + #pragma omp target teams map(mvla2[:1][:2][:10]) // expected-error {{array section does not specify contiguous storage}} + {} + + #pragma omp target teams map(mptr[:2][2+2-4:1][0:5+5]) // expected-error {{array section does not specify contiguous storage}} + {} + #pragma omp target teams map(mptr[:1][:2-1][2:4-3]) + {} + #pragma omp target teams map(mptr[:1][:arg][2:4-3]) // correct if arg is 1. + {} + #pragma omp target teams map(mptr[:1][:2-1][0:2]) + {} + #pragma omp target teams map(mptr[:1][:2][0:2]) // expected-error {{array section does not specify contiguous storage}} + {} + #pragma omp target teams map(mptr[:1][:][0:2]) // expected-error {{section length is unspecified and cannot be inferred because subscripted value is not an array}} + {} + #pragma omp target teams map(mptr[:2][:1][0:2]) // expected-error {{array section does not specify contiguous storage}} + {} + + #pragma omp target teams map(r.ArrS[0].B) + {} + #pragma omp target teams map(r.ArrS[:1].B) // expected-error {{OpenMP array section is not allowed here}} + {} + #pragma omp target teams map(r.ArrS[:arg].B) // expected-error {{OpenMP array section is not allowed here}} + {} + #pragma omp target teams map(r.ArrS[0].Arr[1:23]) + {} + #pragma omp target teams map(r.ArrS[0].Arr[1:arg]) + {} + #pragma omp target teams map(r.ArrS[0].Arr[arg:23]) + {} + #pragma omp target teams map(r.ArrS[0].Error) // expected-error {{no member named 'Error' in 'SB'}} + {} + #pragma omp target teams map(r.ArrS[0].A, r.ArrS[1].A) // expected-error {{multiple array elements associated with the same variable are not allowed in map clauses of the same construct}} expected-note {{used here}} + {} + #pragma omp target teams map(r.ArrS[0].A, t.ArrS[1].A) + {} + #pragma omp target teams map(r.PtrS[0], r.PtrS->B) // expected-error {{same pointer derreferenced in multiple different ways in map clause expressions}} expected-note {{used here}} + {} + #pragma omp target teams map(r.RPtrS[0], r.RPtrS->B) // expected-error {{same pointer derreferenced in multiple different ways in map clause expressions}} expected-note {{used here}} + {} + #pragma omp target teams map(r.S.Arr[:12]) + {} + #pragma omp target teams map(r.S.foo()[:12]) // expected-error {{expected expression containing only member accesses and/or array sections based on named variables}} + {} + #pragma omp target teams map(r.C, r.D) + {} + #pragma omp target teams map(r.C, r.C) // expected-error {{variable already marked as mapped in current construct}} expected-note {{used here}} + {} + #pragma omp target teams map(r.C) map(r.C) // expected-error {{variable already marked as mapped in current construct}} expected-note {{used here}} + {} + #pragma omp target teams map(r.C, r.S) // this would be an error only caught at runtime - Sema would have to make sure there is not way for the missing data between fields to be mapped somewhere else. + {} + #pragma omp target teams map(r, r.S) // expected-error {{variable already marked as mapped in current construct}} expected-note {{used here}} + {} + #pragma omp target teams map(r.C, t.C) + {} + #pragma omp target teams map(r.A) // expected-error {{bit fields cannot be used to specify storage in a 'map' clause}} + {} + #pragma omp target teams map(r.Arr) + {} + #pragma omp target teams map(r.Arr[3:5]) + {} + #pragma omp target teams map(r.Ptr[3:5]) + {} + #pragma omp target teams map(r.ArrS[3:5].A) // expected-error {{OpenMP array section is not allowed here}} + {} + #pragma omp target teams map(r.ArrS[3:5].Arr[6:7]) // expected-error {{OpenMP array section is not allowed here}} + {} + #pragma omp target teams map(r.ArrS[3].Arr[6:7]) + {} + #pragma omp target teams map(r.S.Arr[4:5]) + {} + #pragma omp target teams map(r.S.Ptr[4:5]) + {} + #pragma omp target teams map(r.S.Ptr[:]) // expected-error {{section length is unspecified and cannot be inferred because subscripted value is not an array}} + {} + #pragma omp target teams map((p+1)->A) // expected-error {{expected expression containing only member accesses and/or array sections based on named variables}} + {} + #pragma omp target teams map(u.B) // expected-error {{mapped storage cannot be derived from a union}} + {} + + #pragma omp target data map(to: r.C) //expected-note {{used here}} + { + #pragma omp target teams map(r.D) // expected-error {{original storage of expression in data environment is shared but data environment do not fully contain mapped expression storage}} + {} + } + + #pragma omp target data map(to: t.Ptr) //expected-note {{used here}} + { + #pragma omp target teams map(t.Ptr[:23]) // expected-error {{pointer cannot be mapped along with a section derived from itself}} + {} + } + + #pragma omp target data map(to: t.C, t.D) + { + #pragma omp target data map(to: t.C) + { + #pragma omp target teams map(t.D) + {} + } + } + + #pragma omp target data map(to: t) + { + #pragma omp target data map(to: t.C) + { + #pragma omp target teams map(t.D) + {} + } + } +} +void foo() { +} + +bool foobool(int argc) { + return argc; +} + +struct S1; // expected-note 2 {{declared here}} +extern S1 a; +class S2 { + mutable int a; +public: + S2():a(0) { } + S2(S2 &s2):a(s2.a) { } + static float S2s; // expected-note 4 {{mappable type cannot contain static members}} + static const float S2sc; // expected-note 4 {{mappable type cannot contain static members}} +}; +const float S2::S2sc = 0; +const S2 b; +const S2 ba[5]; +class S3 { + int a; +public: + S3():a(0) { } + S3(S3 &s3):a(s3.a) { } +}; +const S3 c; +const S3 ca[5]; +extern const int f; +class S4 { + int a; + S4(); + S4(const S4 &s4); +public: + S4(int v):a(v) { } +}; +class S5 { + int a; + S5():a(0) {} + S5(const S5 &s5):a(s5.a) { } +public: + S5(int v):a(v) { } +}; + +template +struct S6; + +template<> +struct S6 // expected-note {{mappable type cannot be polymorphic}} +{ + virtual void foo(); +}; + +S3 h; +#pragma omp threadprivate(h) // expected-note 2 {{defined as threadprivate or thread local}} + +typedef int from; + +template // expected-note {{declared here}} +T tmain(T argc) { + const T d = 5; + const T da[5] = { 0 }; + S4 e(4); + S5 g(5); + T i, t[20]; + T &j = i; + T *k = &j; + T x; + T y; + T to, tofrom, always; + const T (&l)[5] = da; +#pragma omp target teams map // expected-error {{expected '(' after 'map'}} + {} +#pragma omp target teams map( // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected expression}} + {} +#pragma omp target teams map() // expected-error {{expected expression}} + {} +#pragma omp target teams map(alloc) // expected-error {{use of undeclared identifier 'alloc'}} + {} +#pragma omp target teams map(to argc // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected ',' or ')' in 'map' clause}} + {} +#pragma omp target teams map(to:) // expected-error {{expected expression}} + {} +#pragma omp target teams map(from: argc, // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}} + {} +#pragma omp target teams map(x: y) // expected-error {{incorrect map type, expected one of 'to', 'from', 'tofrom', 'alloc', 'release', or 'delete'}} + {} +#pragma omp target teams map(x) + foo(); +#pragma omp target teams map(tofrom: t[:I]) + foo(); +#pragma omp target teams map(T: a) // expected-error {{incorrect map type, expected one of 'to', 'from', 'tofrom', 'alloc', 'release', or 'delete'}} expected-error {{incomplete type 'S1' where a complete type is required}} + foo(); +#pragma omp target teams map(T) // expected-error {{'T' does not refer to a value}} + foo(); +#pragma omp target teams map(I) // expected-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}} + foo(); +#pragma omp target teams map(S2::S2s) + foo(); +#pragma omp target teams map(S2::S2sc) + foo(); +#pragma omp target teams map(x) + foo(); +#pragma omp target teams map(to: x) + foo(); +#pragma omp target teams map(to: to) + foo(); +#pragma omp target teams map(to) + foo(); +#pragma omp target teams map(to, x) + foo(); + +#pragma omp target data map(to x) // expected-error {{expected ',' or ')' in 'map' clause}} +#pragma omp target data map(tofrom: argc > 0 ? x : y) // expected-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}} +#pragma omp target data map(argc) +#pragma omp target data map(S1) // expected-error {{'S1' does not refer to a value}} +#pragma omp target data map(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}} expected-error 2 {{type 'S2' is not mappable to target}} +#pragma omp target data map(ba) // expected-error 2 {{type 'S2' is not mappable to target}} +#pragma omp target data map(ca) +#pragma omp target data map(da) +#pragma omp target data map(S2::S2s) +#pragma omp target data map(S2::S2sc) +#pragma omp target data map(e, g) +#pragma omp target data map(h) // expected-error {{threadprivate variables are not allowed in 'map' clause}} +#pragma omp target data map(k) map(k) // expected-error 2 {{variable already marked as mapped in current construct}} expected-note 2 {{used here}} +#pragma omp target teams map(k), map(k[:5]) // expected-error 2 {{pointer cannot be mapped along with a section derived from itself}} expected-note 2 {{used here}} + foo(); + +#pragma omp target data map(da) +#pragma omp target teams map(da[:4]) + foo(); + +#pragma omp target data map(k, j, l) // expected-note 2 {{used here}} +#pragma omp target data map(k[:4]) // expected-error 2 {{pointer cannot be mapped along with a section derived from itself}} +#pragma omp target data map(j) +#pragma omp target teams map(l) map(l[:5]) // expected-error 2 {{variable already marked as mapped in current construct}} expected-note 2 {{used here}} + foo(); + +#pragma omp target data map(k[:4], j, l[:5]) // expected-note 2 {{used here}} +#pragma omp target data map(k) // expected-error 2 {{pointer cannot be mapped along with a section derived from itself}} +#pragma omp target data map(j) +#pragma omp target teams map(l) + foo(); + +#pragma omp target data map(always, tofrom: x) +#pragma omp target data map(always: x) // expected-error {{missing map type}} +#pragma omp target data map(tofrom, always: x) // expected-error {{incorrect map type modifier, expected 'always'}} expected-error {{incorrect map type, expected one of 'to', 'from', 'tofrom', 'alloc', 'release', or 'delete'}} +#pragma omp target data map(always, tofrom: always, tofrom, x) +#pragma omp target teams map(tofrom j) // expected-error {{expected ',' or ')' in 'map' clause}} + foo(); + return 0; +} + +int main(int argc, char **argv) { + const int d = 5; + const int da[5] = { 0 }; + S4 e(4); + S5 g(5); + int i; + int &j = i; + int *k = &j; + S6 m; + int x; + int y; + int to, tofrom, always; + const int (&l)[5] = da; +#pragma omp target data map // expected-error {{expected '(' after 'map'}} expected-error {{expected at least one map clause for '#pragma omp target data'}} +#pragma omp target data map( // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected expression}} +#pragma omp target data map() // expected-error {{expected expression}} +#pragma omp target data map(alloc) // expected-error {{use of undeclared identifier 'alloc'}} +#pragma omp target data map(to argc // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected ',' or ')' in 'map' clause}} +#pragma omp target data map(to:) // expected-error {{expected expression}} +#pragma omp target data map(from: argc, // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}} +#pragma omp target data map(x: y) // expected-error {{incorrect map type, expected one of 'to', 'from', 'tofrom', 'alloc', 'release', or 'delete'}} +#pragma omp target teams map(x) + foo(); + +#pragma omp target teams map(to: x) + foo(); +#pragma omp target teams map(to: to) + foo(); +#pragma omp target teams map(to) + foo(); +#pragma omp target teams map(to, x) + foo(); + +#pragma omp target data map(to x) // expected-error {{expected ',' or ')' in 'map' clause}} +#pragma omp target data map(tofrom: argc > 0 ? argv[1] : argv[2]) // expected-error {{xpected expression containing only member accesses and/or array sections based on named variables}} +#pragma omp target data map(argc) +#pragma omp target data map(S1) // expected-error {{'S1' does not refer to a value}} +#pragma omp target data map(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}} expected-error 2 {{type 'S2' is not mappable to target}} +#pragma omp target data map(argv[1]) +#pragma omp target data map(ba) // expected-error 2 {{type 'S2' is not mappable to target}} +#pragma omp target data map(ca) +#pragma omp target data map(da) +#pragma omp target data map(S2::S2s) +#pragma omp target data map(S2::S2sc) +#pragma omp target data map(e, g) +#pragma omp target data map(h) // expected-error {{threadprivate variables are not allowed in 'map' clause}} +#pragma omp target data map(k), map(k) // expected-error {{variable already marked as mapped in current construct}} expected-note {{used here}} +#pragma omp target teams map(k), map(k[:5]) // expected-error {{pointer cannot be mapped along with a section derived from itself}} expected-note {{used here}} + foo(); + +#pragma omp target data map(da) +#pragma omp target teams map(da[:4]) + foo(); + +#pragma omp target data map(k, j, l) // expected-note {{used here}} +#pragma omp target data map(k[:4]) // expected-error {{pointer cannot be mapped along with a section derived from itself}} +#pragma omp target data map(j) +#pragma omp target teams map(l) map(l[:5]) // expected-error {{variable already marked as mapped in current construct}} expected-note {{used here}} + foo(); + +#pragma omp target data map(k[:4], j, l[:5]) // expected-note {{used here}} +#pragma omp target data map(k) // expected-error {{pointer cannot be mapped along with a section derived from itself}} +#pragma omp target data map(j) +#pragma omp target teams map(l) + foo(); + +#pragma omp target data map(always, tofrom: x) +#pragma omp target data map(always: x) // expected-error {{missing map type}} +#pragma omp target data map(tofrom, always: x) // expected-error {{incorrect map type modifier, expected 'always'}} expected-error {{incorrect map type, expected one of 'to', 'from', 'tofrom', 'alloc', 'release', or 'delete'}} +#pragma omp target data map(always, tofrom: always, tofrom, x) +#pragma omp target teams map(tofrom j) // expected-error {{expected ',' or ')' in 'map' clause}} + foo(); + +#pragma omp target teams private(j) map(j) // expected-error {{private variable cannot be in a map clause in '#pragma omp target teams' directive}} expected-note {{defined as private}} + {} + +#pragma omp target teams firstprivate(j) map(j) // expected-error {{firstprivate variable cannot be in a map clause in '#pragma omp target teams' directive}} expected-note {{defined as firstprivate}} + {} + +#pragma omp target teams map(m) // expected-error {{type 'S6' is not mappable to target}} + {} + + return tmain(argc)+tmain(argc); // expected-note {{in instantiation of function template specialization 'tmain' requested here}} expected-note {{in instantiation of function template specialization 'tmain' requested here}} +} +#endif Index: cfe/trunk/test/OpenMP/target_teams_messages.cpp =================================================================== --- cfe/trunk/test/OpenMP/target_teams_messages.cpp +++ cfe/trunk/test/OpenMP/target_teams_messages.cpp @@ -0,0 +1,65 @@ +// RUN: %clang_cc1 -verify -fopenmp -std=c++11 -o - %s + +void foo() { +} + +#pragma omp target teams // expected-error {{unexpected OpenMP directive '#pragma omp target teams'}} + +int main(int argc, char **argv) { +#pragma omp target teams { // expected-warning {{extra tokens at the end of '#pragma omp target teams' are ignored}} + foo(); +#pragma omp target teams ( // expected-warning {{extra tokens at the end of '#pragma omp target teams' are ignored}} + foo(); +#pragma omp target teams [ // expected-warning {{extra tokens at the end of '#pragma omp target teams' are ignored}} + foo(); +#pragma omp target teams ] // expected-warning {{extra tokens at the end of '#pragma omp target teams' are ignored}} + foo(); +#pragma omp target teams ) // expected-warning {{extra tokens at the end of '#pragma omp target teams' are ignored}} + foo(); +#pragma omp target teams } // expected-warning {{extra tokens at the end of '#pragma omp target teams' are ignored}} + foo(); +#pragma omp target teams + foo(); +#pragma omp target teams unknown() // expected-warning {{extra tokens at the end of '#pragma omp target teams' are ignored}} + foo(); + L1: + foo(); +#pragma omp target teams + ; +#pragma omp target teams + { + goto L1; // expected-error {{use of undeclared label 'L1'}} + argc++; + } + + for (int i = 0; i < 10; ++i) { + switch(argc) { + case (0): + #pragma omp target teams + { + foo(); + break; // expected-error {{'break' statement not in loop or switch statement}} + continue; // expected-error {{'continue' statement not in loop statement}} + } + default: + break; + } + } +#pragma omp target teams default(none) + ++argc; // expected-error {{variable 'argc' must have explicitly specified data sharing attributes}} + + goto L2; // expected-error {{use of undeclared label 'L2'}} +#pragma omp target teams + L2: + foo(); +#pragma omp target teams + { + return 1; // expected-error {{cannot return from OpenMP region}} + } + + [[]] // expected-error {{an attribute list cannot appear here}} +#pragma omp target teams + for (int n = 0; n < 100; ++n) {} + + return 0; +} Index: cfe/trunk/test/OpenMP/target_teams_nowait_messages.cpp =================================================================== --- cfe/trunk/test/OpenMP/target_teams_nowait_messages.cpp +++ cfe/trunk/test/OpenMP/target_teams_nowait_messages.cpp @@ -0,0 +1,17 @@ +// RUN: %clang_cc1 -verify -fopenmp -o - %s + +void foo() { +} + +int main(int argc, char **argv) { +#pragma omp target teams nowait( // expected-warning {{extra tokens at the end of '#pragma omp target teams' are ignored}} + foo(); +#pragma omp target teams nowait (argc)) // expected-warning {{extra tokens at the end of '#pragma omp target teams' are ignored}} + foo(); +#pragma omp target teams nowait device (-10u) + foo(); +#pragma omp target teams nowait (3.14) device (-10u) // expected-warning {{extra tokens at the end of '#pragma omp target teams' are ignored}} + foo(); + + return 0; +} Index: cfe/trunk/test/OpenMP/target_teams_num_teams_messages.cpp =================================================================== --- cfe/trunk/test/OpenMP/target_teams_num_teams_messages.cpp +++ cfe/trunk/test/OpenMP/target_teams_num_teams_messages.cpp @@ -0,0 +1,85 @@ +// RUN: %clang_cc1 -verify -fopenmp -std=c++11 -ferror-limit 100 -o - %s + +void foo() { +} + +bool foobool(int argc) { + return argc; +} + +struct S1; // expected-note 2 {{declared here}} + +template // expected-note {{declared here}} +T tmain(T argc) { + char **a; +#pragma omp target teams num_teams(C) + foo(); +#pragma omp target teams num_teams(T) // expected-error {{'T' does not refer to a value}} + foo(); +#pragma omp target teams num_teams // expected-error {{expected '(' after 'num_teams'}} + foo(); +#pragma omp target teams num_teams( // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams num_teams() // expected-error {{expected expression}} + foo(); +#pragma omp target teams num_teams(argc // expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams num_teams(argc)) // expected-warning {{extra tokens at the end of '#pragma omp target teams' are ignored}} + foo(); +#pragma omp target teams num_teams(argc > 0 ? a[1] : a[2]) // expected-error {{expression must have integral or unscoped enumeration type, not 'char *'}} + foo(); +#pragma omp target teams num_teams(argc + argc) + foo(); +#pragma omp target teams num_teams(argc), num_teams (argc+1) // expected-error {{directive '#pragma omp target teams' cannot contain more than one 'num_teams' clause}} + foo(); +#pragma omp target teams num_teams(S1) // expected-error {{'S1' does not refer to a value}} + foo(); +#pragma omp target teams num_teams(-2) // expected-error {{argument to 'num_teams' clause must be a strictly positive integer value}} + foo(); +#pragma omp target teams num_teams(-10u) + foo(); +#pragma omp target teams num_teams(3.14) // expected-error 2 {{expression must have integral or unscoped enumeration type, not 'double'}} + foo(); + + return 0; +} + +int main(int argc, char **argv) { +#pragma omp target teams num_teams // expected-error {{expected '(' after 'num_teams'}} + foo(); + +#pragma omp target teams num_teams ( // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); + +#pragma omp target teams num_teams () // expected-error {{expected expression}} + foo(); + +#pragma omp target teams num_teams (argc // expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); + +#pragma omp target teams num_teams (argc)) // expected-warning {{extra tokens at the end of '#pragma omp target teams' are ignored}} + foo(); + +#pragma omp target teams num_teams (argc > 0 ? argv[1] : argv[2]) // expected-error {{expression must have integral or unscoped enumeration type, not 'char *'}} + foo(); + +#pragma omp target teams num_teams (argc + argc) + foo(); + +#pragma omp target teams num_teams (argc), num_teams (argc+1) // expected-error {{directive '#pragma omp target teams' cannot contain more than one 'num_teams' clause}} + foo(); + +#pragma omp target teams num_teams (S1) // expected-error {{'S1' does not refer to a value}} + foo(); + +#pragma omp target teams num_teams (-2) // expected-error {{argument to 'num_teams' clause must be a strictly positive integer value}} + foo(); + +#pragma omp target teams num_teams (-10u) + foo(); + +#pragma omp target teams num_teams (3.14) // expected-error {{expression must have integral or unscoped enumeration type, not 'double'}} + foo(); + + return tmain(argc); // expected-note {{in instantiation of function template specialization 'tmain' requested here}} +} Index: cfe/trunk/test/OpenMP/target_teams_private_messages.cpp =================================================================== --- cfe/trunk/test/OpenMP/target_teams_private_messages.cpp +++ cfe/trunk/test/OpenMP/target_teams_private_messages.cpp @@ -0,0 +1,109 @@ +// RUN: %clang_cc1 -verify -fopenmp %s + +void foo() { +} + +bool foobool(int argc) { + return argc; +} + +struct S1; // expected-note {{declared here}} expected-note{{forward declaration of 'S1'}} +extern S1 a; +class S2 { + mutable int a; +public: + S2():a(0) { } + static float S2s; // expected-note {{static data member is predetermined as shared}} +}; +const S2 b; +const S2 ba[5]; +class S3 { + int a; +public: + S3():a(0) { } +}; +const S3 c; // expected-note {{global variable is predetermined as shared}} +const S3 ca[5]; // expected-note {{global variable is predetermined as shared}} +extern const int f; // expected-note {{global variable is predetermined as shared}} +class S4 { + int a; + S4(); // expected-note {{implicitly declared private here}} +public: + S4(int v):a(v) { } +}; +class S5 { + int a; + S5():a(0) {} // expected-note {{implicitly declared private here}} +public: + S5(int v):a(v) { } +}; + +int threadvar; +#pragma omp threadprivate(threadvar) // expected-note {{defined as threadprivate or thread local}} + +namespace A { +double x; +#pragma omp threadprivate(x) // expected-note {{defined as threadprivate or thread local}} +} +namespace B { +using A::x; +} + +int main(int argc, char **argv) { + const int d = 5; // expected-note {{constant variable is predetermined as shared}} + const int da[5] = { 0 }; // expected-note {{constant variable is predetermined as shared}} + S4 e(4); + S5 g(5); + int i; + int &j = i; +#pragma omp target teams private // expected-error {{expected '(' after 'private'}} + foo(); +#pragma omp target teams private ( // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams private () // expected-error {{expected expression}} + foo(); +#pragma omp target teams private (argc // expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams private (argc, // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams private (argc > 0 ? argv[1] : argv[2]) // expected-error {{expected variable name}} + foo(); +#pragma omp target teams private (argc argv) // expected-error {{expected ',' or ')' in 'private' clause}} + foo(); +#pragma omp target teams private (S1) // expected-error {{'S1' does not refer to a value}} + foo(); +#pragma omp target teams private (a, b, c, d, f) // expected-error {{a private variable with incomplete type 'S1'}} expected-error 3 {{shared variable cannot be private}} + foo(); +#pragma omp target teams private (argv[1]) // expected-error {{expected variable name}} + foo(); +#pragma omp target teams private(ba) + foo(); +#pragma omp target teams private(ca) // expected-error {{shared variable cannot be private}} + foo(); +#pragma omp target teams private(da) // expected-error {{shared variable cannot be private}} + foo(); +#pragma omp target teams private(S2::S2s) // expected-error {{shared variable cannot be private}} + foo(); +#pragma omp target teams private(e, g) // expected-error {{calling a private constructor of class 'S4'}} expected-error {{calling a private constructor of class 'S5'}} + foo(); +#pragma omp target teams private(threadvar, B::x) // expected-error 2 {{threadprivate or thread local variable cannot be private}} + foo(); +#pragma omp target teams shared(i), private(i) // expected-error {{shared variable cannot be private}} expected-note {{defined as shared}} + foo(); +#pragma omp target teams firstprivate(i) private(i) // expected-error {{firstprivate variable cannot be private}} expected-note {{defined as firstprivate}} + foo(); +#pragma omp target teams private(i) + foo(); +#pragma omp target teams private(j) + foo(); +#pragma omp target teams firstprivate(i) + for (int k = 0; k < 10; ++k) { +#pragma omp parallel private(i) + foo(); + } + static int m; +#pragma omp target teams private(m) // OK + foo(); + + return 0; +} Index: cfe/trunk/test/OpenMP/target_teams_reduction_messages.cpp =================================================================== --- cfe/trunk/test/OpenMP/target_teams_reduction_messages.cpp +++ cfe/trunk/test/OpenMP/target_teams_reduction_messages.cpp @@ -0,0 +1,266 @@ +// RUN: %clang_cc1 -verify -fopenmp -o - %s +// RUN: %clang_cc1 -verify -fopenmp -std=c++98 -o - %s +// RUN: %clang_cc1 -verify -fopenmp -std=c++11 -o - %s + +void foo() { +} + +bool foobool(int argc) { + return argc; +} + +void foobar(int &ref) { +#pragma omp target teams reduction(+:ref) + foo(); +} + +struct S1; // expected-note {{declared here}} expected-note 4 {{forward declaration of 'S1'}} +extern S1 a; +class S2 { + mutable int a; + S2 &operator+(const S2 &arg) { return (*this); } // expected-note 3 {{implicitly declared private here}} + +public: + S2() : a(0) {} + S2(S2 &s2) : a(s2.a) {} + static float S2s; // expected-note 2 {{static data member is predetermined as shared}} + static const float S2sc; +}; +const float S2::S2sc = 0; // expected-note 2 {{'S2sc' defined here}} +S2 b; // expected-note 3 {{'b' defined here}} +const S2 ba[5]; // expected-note 2 {{'ba' defined here}} +class S3 { + int a; + +public: + int b; + S3() : a(0) {} + S3(const S3 &s3) : a(s3.a) {} + S3 operator+(const S3 &arg1) { return arg1; } +}; +int operator+(const S3 &arg1, const S3 &arg2) { return 5; } +S3 c; // expected-note 3 {{'c' defined here}} +const S3 ca[5]; // expected-note 2 {{'ca' defined here}} +extern const int f; // expected-note 4 {{'f' declared here}} +class S4 { + int a; + S4(); // expected-note {{implicitly declared private here}} + S4(const S4 &s4); + S4 &operator+(const S4 &arg) { return (*this); } + +public: + S4(int v) : a(v) {} +}; +S4 &operator&=(S4 &arg1, S4 &arg2) { return arg1; } +class S5 { + int a; + S5() : a(0) {} // expected-note {{implicitly declared private here}} + S5(const S5 &s5) : a(s5.a) {} + S5 &operator+(const S5 &arg); + +public: + S5(int v) : a(v) {} +}; +class S6 { // expected-note 3 {{candidate function (the implicit copy assignment operator) not viable: no known conversion from 'int' to 'const S6' for 1st argument}} +#if __cplusplus >= 201103L // C++11 or later +// expected-note@-2 3 {{candidate function (the implicit move assignment operator) not viable}} +#endif + int a; + +public: + S6() : a(6) {} + operator int() { return 6; } +} o; + +S3 h, k; +#pragma omp threadprivate(h) // expected-note 2 {{defined as threadprivate or thread local}} + +template // expected-note {{declared here}} +T tmain(T argc) { + const T d = T(); // expected-note 4 {{'d' defined here}} + const T da[5] = {T()}; // expected-note 2 {{'da' defined here}} + T qa[5] = {T()}; + T i; + T &j = i; // expected-note 4 {{'j' defined here}} + S3 &p = k; // expected-note 2 {{'p' defined here}} + const T &r = da[(int)i]; // expected-note 2 {{'r' defined here}} + T &q = qa[(int)i]; // expected-note 2 {{'q' defined here}} + T fl; +#pragma omp target teams reduction // expected-error {{expected '(' after 'reduction'}} + foo(); +#pragma omp target teams reduction + // expected-error {{expected '(' after 'reduction'}} expected-warning {{extra tokens at the end of '#pragma omp target teams' are ignored}} + foo(); +#pragma omp target teams reduction( // expected-error {{expected unqualified-id}} expected-warning {{missing ':' after reduction identifier - ignoring}} expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams reduction(- // expected-warning {{missing ':' after reduction identifier - ignoring}} expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams reduction() // expected-error {{expected unqualified-id}} expected-warning {{missing ':' after reduction identifier - ignoring}} + foo(); +#pragma omp target teams reduction(*) // expected-warning {{missing ':' after reduction identifier - ignoring}} expected-error {{expected expression}} + foo(); +#pragma omp target teams reduction(\) // expected-error {{expected unqualified-id}} expected-warning {{missing ':' after reduction identifier - ignoring}} + foo(); +#pragma omp target teams reduction(& : argc // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{invalid operands to binary expression ('float' and 'float')}} + foo(); +#pragma omp target teams reduction(| : argc, // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{invalid operands to binary expression ('float' and 'float')}} + foo(); +#pragma omp target teams reduction(|| : argc ? i : argc) // expected-error 2 {{expected variable name, array element or array section}} + foo(); +#pragma omp target teams reduction(foo : argc) //expected-error {{incorrect reduction identifier, expected one of '+', '-', '*', '&', '|', '^', '&&', '||', 'min' or 'max' or declare reduction for type 'float'}} expected-error {{incorrect reduction identifier, expected one of '+', '-', '*', '&', '|', '^', '&&', '||', 'min' or 'max' or declare reduction for type 'int'}} + foo(); +#pragma omp target teams reduction(&& : argc) + foo(); +#pragma omp target teams reduction(^ : T) // expected-error {{'T' does not refer to a value}} + foo(); +#pragma omp target teams reduction(+ : a, b, c, d, f) // expected-error {{a reduction list item with incomplete type 'S1'}} expected-error 3 {{const-qualified list item cannot be reduction}} expected-error 2 {{'operator+' is a private member of 'S2'}} + foo(); +#pragma omp target teams reduction(min : a, b, c, d, f) // expected-error {{a reduction list item with incomplete type 'S1'}} expected-error 4 {{arguments of OpenMP clause 'reduction' for 'min' or 'max' must be of arithmetic type}} expected-error 3 {{const-qualified list item cannot be reduction}} + foo(); +#pragma omp target teams reduction(max : h.b) // expected-error {{expected variable name, array element or array section}} + foo(); +#pragma omp target teams reduction(+ : ba) // expected-error {{const-qualified list item cannot be reduction}} + foo(); +#pragma omp target teams reduction(* : ca) // expected-error {{const-qualified list item cannot be reduction}} + foo(); +#pragma omp target teams reduction(- : da) // expected-error {{const-qualified list item cannot be reduction}} expected-error {{const-qualified list item cannot be reduction}} + foo(); +#pragma omp target teams reduction(^ : fl) // expected-error {{invalid operands to binary expression ('float' and 'float')}} + foo(); +#pragma omp target teams reduction(&& : S2::S2s) // expected-error {{shared variable cannot be reduction}} + foo(); +#pragma omp target teams reduction(&& : S2::S2sc) // expected-error {{const-qualified list item cannot be reduction}} + foo(); +#pragma omp target teams reduction(+ : h, k) // expected-error {{threadprivate or thread local variable cannot be reduction}} + foo(); +#pragma omp target teams reduction(+ : o) // expected-error 2 {{no viable overloaded '='}} + foo(); +#pragma omp target teams private(i), reduction(+ : j), reduction(+ : q) // expected-error 4 {{argument of OpenMP clause 'reduction' must reference the same object in all threads}} + foo(); +#pragma omp parallel private(k) +#pragma omp target teams reduction(+ : p), reduction(+ : p) // expected-error 2 {{argument of OpenMP clause 'reduction' must reference the same object in all threads}} + foo(); +#pragma omp target teams reduction(+ : p), reduction(+ : p) // expected-error 2 {{variable can appear only once in OpenMP 'reduction' clause}} expected-note 2 {{previously referenced here}} + foo(); +#pragma omp target teams reduction(+ : r) // expected-error 2 {{const-qualified list item cannot be reduction}} + foo(); +#pragma omp parallel shared(i) +#pragma omp parallel reduction(min : i) +#pragma omp target teams reduction(max : j) // expected-error 2 {{argument of OpenMP clause 'reduction' must reference the same object in all threads}} + foo(); +#pragma omp target teams +#pragma omp parallel for private(fl) + for (int i = 0; i < 10; ++i) + {} +#pragma omp target teams reduction(+ : fl) + foo(); +#pragma omp target teams +#pragma omp parallel for reduction(- : fl) + for (int i = 0; i < 10; ++i) + {} +#pragma omp target teams reduction(+ : fl) + foo(); + + return T(); +} + +namespace A { +double x; +#pragma omp threadprivate(x) // expected-note {{defined as threadprivate or thread local}} +} +namespace B { +using A::x; +} + +int main(int argc, char **argv) { + const int d = 5; // expected-note 2 {{'d' defined here}} + const int da[5] = {0}; // expected-note {{'da' defined here}} + int qa[5] = {0}; + S4 e(4); + S5 g(5); + int i; + int &j = i; // expected-note 2 {{'j' defined here}} + S3 &p = k; // expected-note 2 {{'p' defined here}} + const int &r = da[i]; // expected-note {{'r' defined here}} + int &q = qa[i]; // expected-note {{'q' defined here}} + float fl; +#pragma omp target teams reduction // expected-error {{expected '(' after 'reduction'}} + foo(); +#pragma omp target teams reduction + // expected-error {{expected '(' after 'reduction'}} expected-warning {{extra tokens at the end of '#pragma omp target teams' are ignored}} + foo(); +#pragma omp target teams reduction( // expected-error {{expected unqualified-id}} expected-warning {{missing ':' after reduction identifier - ignoring}} expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams reduction(- // expected-warning {{missing ':' after reduction identifier - ignoring}} expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams reduction() // expected-error {{expected unqualified-id}} expected-warning {{missing ':' after reduction identifier - ignoring}} + foo(); +#pragma omp target teams reduction(*) // expected-warning {{missing ':' after reduction identifier - ignoring}} expected-error {{expected expression}} + foo(); +#pragma omp target teams reduction(\) // expected-error {{expected unqualified-id}} expected-warning {{missing ':' after reduction identifier - ignoring}} + foo(); +#pragma omp target teams reduction(foo : argc // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{incorrect reduction identifier, expected one of '+', '-', '*', '&', '|', '^', '&&', '||', 'min' or 'max'}} + foo(); +#pragma omp target teams reduction(| : argc, // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams reduction(|| : argc > 0 ? argv[1] : argv[2]) // expected-error {{expected variable name, array element or array section}} + foo(); +#pragma omp target teams reduction(~ : argc) // expected-error {{expected unqualified-id}} + foo(); +#pragma omp target teams reduction(&& : argc) + foo(); +#pragma omp target teams reduction(^ : S1) // expected-error {{'S1' does not refer to a value}} + foo(); +#pragma omp target teams reduction(+ : a, b, c, d, f) // expected-error {{a reduction list item with incomplete type 'S1'}} expected-error 2 {{const-qualified list item cannot be reduction}} expected-error {{'operator+' is a private member of 'S2'}} + foo(); +#pragma omp target teams reduction(min : a, b, c, d, f) // expected-error {{a reduction list item with incomplete type 'S1'}} expected-error 2 {{arguments of OpenMP clause 'reduction' for 'min' or 'max' must be of arithmetic type}} expected-error 2 {{const-qualified list item cannot be reduction}} + foo(); +#pragma omp target teams reduction(max : h.b) // expected-error {{expected variable name, array element or array section}} + foo(); +#pragma omp target teams reduction(+ : ba) // expected-error {{const-qualified list item cannot be reduction}} + foo(); +#pragma omp target teams reduction(* : ca) // expected-error {{const-qualified list item cannot be reduction}} + foo(); +#pragma omp target teams reduction(- : da) // expected-error {{const-qualified list item cannot be reduction}} + foo(); +#pragma omp target teams reduction(^ : fl) // expected-error {{invalid operands to binary expression ('float' and 'float')}} + foo(); +#pragma omp target teams reduction(&& : S2::S2s) // expected-error {{shared variable cannot be reduction}} + foo(); +#pragma omp target teams reduction(&& : S2::S2sc) // expected-error {{const-qualified list item cannot be reduction}} + foo(); +#pragma omp target teams reduction(& : e, g) // expected-error {{calling a private constructor of class 'S4'}} expected-error {{invalid operands to binary expression ('S4' and 'S4')}} expected-error {{calling a private constructor of class 'S5'}} expected-error {{invalid operands to binary expression ('S5' and 'S5')}} + foo(); +#pragma omp target teams reduction(+ : h, k, B::x) // expected-error 2 {{threadprivate or thread local variable cannot be reduction}} + foo(); +#pragma omp target teams reduction(+ : o) // expected-error {{no viable overloaded '='}} + foo(); +#pragma omp target teams private(i), reduction(+ : j), reduction(+ : q) // expected-error 2 {{argument of OpenMP clause 'reduction' must reference the same object in all threads}} + foo(); +#pragma omp parallel private(k) +#pragma omp target teams reduction(+ : p), reduction(+ : p) // expected-error 2 {{argument of OpenMP clause 'reduction' must reference the same object in all threads}} + foo(); +#pragma omp target teams reduction(+ : p), reduction(+ : p) // expected-error {{variable can appear only once in OpenMP 'reduction' clause}} expected-note {{previously referenced here}} + foo(); +#pragma omp target teams reduction(+ : r) // expected-error {{const-qualified list item cannot be reduction}} + foo(); +#pragma omp parallel shared(i) +#pragma omp parallel reduction(min : i) +#pragma omp target teams reduction(max : j) // expected-error {{argument of OpenMP clause 'reduction' must reference the same object in all threads}} + foo(); +#pragma omp target teams +#pragma omp parallel for private(fl) + for (int i = 0; i < 10; ++i) + {} +#pragma omp target teams reduction(+ : fl) + foo(); +#pragma omp target teams +#pragma omp parallel for reduction(- : fl) + for (int i = 0; i < 10; ++i) + {} +#pragma omp target teams reduction(+ : fl) + foo(); + static int m; +#pragma omp target teams reduction(+ : m) // OK + foo(); + + return tmain(argc) + tmain(fl); // expected-note {{in instantiation of function template specialization 'tmain' requested here}} expected-note {{in instantiation of function template specialization 'tmain' requested here}} +} Index: cfe/trunk/test/OpenMP/target_teams_shared_messages.cpp =================================================================== --- cfe/trunk/test/OpenMP/target_teams_shared_messages.cpp +++ cfe/trunk/test/OpenMP/target_teams_shared_messages.cpp @@ -0,0 +1,110 @@ +// RUN: %clang_cc1 -verify -fopenmp %s + +void foo() { +} + +bool foobool(int argc) { + return argc; +} + +struct S1; // expected-note {{declared here}} +extern S1 a; +class S2 { + mutable int a; +public: + S2():a(0) { } + S2(S2 &s2):a(s2.a) { } +}; +const S2 b; +const S2 ba[5]; +class S3 { + int a; +public: + S3():a(0) { } + S3(S3 &s3):a(s3.a) { } +}; +const S3 c; +const S3 ca[5]; +extern const int f; +class S4 { + int a; + S4(); + S4(const S4 &s4); +public: + S4(int v):a(v) { } +}; +class S5 { + int a; + S5():a(0) {} + S5(const S5 &s5):a(s5.a) { } +public: + S5(int v):a(v) { } +}; + +S3 h; +#pragma omp threadprivate(h) // expected-note {{defined as threadprivate or thread local}} + +namespace A { +double x; +#pragma omp threadprivate(x) // expected-note {{defined as threadprivate or thread local}} +} +namespace B { +using A::x; +} + +int main(int argc, char **argv) { + const int d = 5; + const int da[5] = { 0 }; + S4 e(4); + S5 g(5); + int i; + int &j = i; +#pragma omp target teams shared // expected-error {{expected '(' after 'shared'}} + foo(); +#pragma omp target teams shared ( // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams shared () // expected-error {{expected expression}} + foo(); +#pragma omp target teams shared (argc // expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams shared (argc, // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams shared (argc > 0 ? argv[1] : argv[2]) // expected-error {{expected variable name}} + foo(); +#pragma omp target teams shared (argc) + foo(); +#pragma omp target teams shared (S1) // expected-error {{'S1' does not refer to a value}} + foo(); +#pragma omp target teams shared (a, b, c, d, f) + foo(); +#pragma omp target teams shared (argv[1]) // expected-error {{expected variable name}} + foo(); +#pragma omp target teams shared(ba) + foo(); +#pragma omp target teams shared(ca) + foo(); +#pragma omp target teams shared(da) + foo(); +#pragma omp target teams shared(e, g) + foo(); +#pragma omp target teams shared(h, B::x) // expected-error 2 {{threadprivate or thread local variable cannot be shared}} + foo(); +#pragma omp target teams private(i), shared(i) // expected-error {{private variable cannot be shared}} expected-note {{defined as private}} + foo(); +#pragma omp target teams firstprivate(i), shared(i) // expected-error {{firstprivate variable cannot be shared}} expected-note {{defined as firstprivate}} + foo(); +#pragma omp target teams private(i) + foo(); +#pragma omp target teams shared(i) + foo(); +#pragma omp target teams shared(j) + foo(); +#pragma omp target teams firstprivate(i) + foo(); +#pragma omp target teams shared(i) + foo(); +#pragma omp target teams shared(j) + foo(); + + return 0; +} Index: cfe/trunk/test/OpenMP/target_teams_thread_limit_messages.cpp =================================================================== --- cfe/trunk/test/OpenMP/target_teams_thread_limit_messages.cpp +++ cfe/trunk/test/OpenMP/target_teams_thread_limit_messages.cpp @@ -0,0 +1,85 @@ +// RUN: %clang_cc1 -verify -fopenmp -std=c++11 -ferror-limit 100 -o - %s + +void foo() { +} + +bool foobool(int argc) { + return argc; +} + +struct S1; // expected-note 2 {{declared here}} + +template // expected-note {{declared here}} +T tmain(T argc) { + char **a; +#pragma omp target teams thread_limit(C) + foo(); +#pragma omp target teams thread_limit(T) // expected-error {{'T' does not refer to a value}} + foo(); +#pragma omp target teams thread_limit // expected-error {{expected '(' after 'thread_limit'}} + foo(); +#pragma omp target teams thread_limit( // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams thread_limit() // expected-error {{expected expression}} + foo(); +#pragma omp target teams thread_limit(argc // expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); +#pragma omp target teams thread_limit(argc)) // expected-warning {{extra tokens at the end of '#pragma omp target teams' are ignored}} + foo(); +#pragma omp target teams thread_limit(argc > 0 ? a[1] : a[2]) // expected-error {{expression must have integral or unscoped enumeration type, not 'char *'}} + foo(); +#pragma omp target teams thread_limit(argc + argc) + foo(); +#pragma omp target teams thread_limit(argc), thread_limit (argc+1) // expected-error {{directive '#pragma omp target teams' cannot contain more than one 'thread_limit' clause}} + foo(); +#pragma omp target teams thread_limit(S1) // expected-error {{'S1' does not refer to a value}} + foo(); +#pragma omp target teams thread_limit(-2) // expected-error {{argument to 'thread_limit' clause must be a strictly positive integer value}} + foo(); +#pragma omp target teams thread_limit(-10u) + foo(); +#pragma omp target teams thread_limit(3.14) // expected-error 2 {{expression must have integral or unscoped enumeration type, not 'double'}} + foo(); + + return 0; +} + +int main(int argc, char **argv) { +#pragma omp target teams thread_limit // expected-error {{expected '(' after 'thread_limit'}} + foo(); + +#pragma omp target teams thread_limit ( // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); + +#pragma omp target teams thread_limit () // expected-error {{expected expression}} + foo(); + +#pragma omp target teams thread_limit (argc // expected-error {{expected ')'}} expected-note {{to match this '('}} + foo(); + +#pragma omp target teams thread_limit (argc)) // expected-warning {{extra tokens at the end of '#pragma omp target teams' are ignored}} + foo(); + +#pragma omp target teams thread_limit (argc > 0 ? argv[1] : argv[2]) // expected-error {{expression must have integral or unscoped enumeration type, not 'char *'}} + foo(); + +#pragma omp target teams thread_limit (argc + argc) + foo(); + +#pragma omp target teams thread_limit (argc), thread_limit (argc+1) // expected-error {{directive '#pragma omp target teams' cannot contain more than one 'thread_limit' clause}} + foo(); + +#pragma omp target teams thread_limit (S1) // expected-error {{'S1' does not refer to a value}} + foo(); + +#pragma omp target teams thread_limit (-2) // expected-error {{argument to 'thread_limit' clause must be a strictly positive integer value}} + foo(); + +#pragma omp target teams thread_limit (-10u) + foo(); + +#pragma omp target teams thread_limit (3.14) // expected-error {{expression must have integral or unscoped enumeration type, not 'double'}} + foo(); + + return tmain(argc); // expected-note {{in instantiation of function template specialization 'tmain' requested here}} +} Index: cfe/trunk/tools/libclang/CIndex.cpp =================================================================== --- cfe/trunk/tools/libclang/CIndex.cpp +++ cfe/trunk/tools/libclang/CIndex.cpp @@ -2008,6 +2008,7 @@ const OMPTeamsDistributeParallelForSimdDirective *D); void VisitOMPTeamsDistributeParallelForDirective( const OMPTeamsDistributeParallelForDirective *D); + void VisitOMPTargetTeamsDirective(const OMPTargetTeamsDirective *D); private: void AddDeclarationNameInfo(const Stmt *S); @@ -2812,6 +2813,11 @@ VisitOMPLoopDirective(D); } +void EnqueueVisitor::VisitOMPTargetTeamsDirective( + const OMPTargetTeamsDirective *D) { + VisitOMPExecutableDirective(D); +} + void CursorVisitor::EnqueueWorkList(VisitorWorkList &WL, const Stmt *S) { EnqueueVisitor(WL, MakeCXCursor(S, StmtParent, TU,RegionOfInterest)).Visit(S); } @@ -4971,6 +4977,8 @@ return cxstring::createRef("OMPTeamsDistributeParallelForSimdDirective"); case CXCursor_OMPTeamsDistributeParallelForDirective: return cxstring::createRef("OMPTeamsDistributeParallelForDirective"); + case CXCursor_OMPTargetTeamsDirective: + return cxstring::createRef("OMPTargetTeamsDirective"); case CXCursor_OverloadCandidate: return cxstring::createRef("OverloadCandidate"); case CXCursor_TypeAliasTemplateDecl: Index: cfe/trunk/tools/libclang/CXCursor.cpp =================================================================== --- cfe/trunk/tools/libclang/CXCursor.cpp +++ cfe/trunk/tools/libclang/CXCursor.cpp @@ -667,6 +667,9 @@ case Stmt::OMPTeamsDistributeParallelForDirectiveClass: K = CXCursor_OMPTeamsDistributeParallelForDirective; break; + case Stmt::OMPTargetTeamsDirectiveClass: + K = CXCursor_OMPTargetTeamsDirective; + break; } CXCursor C = { K, 0, { Parent, S, TU } };