diff options
author | Samuel Antao <sfantao@us.ibm.com> | 2016-01-19 20:04:50 +0000 |
---|---|---|
committer | Samuel Antao <sfantao@us.ibm.com> | 2016-01-19 20:04:50 +0000 |
commit | 5243a6feb0e10358b6dd05b0fa1225d6cb51263b (patch) | |
tree | 2179649dab984188c5cd0f78395d67c1c16bce45 | |
parent | d79077e1574d20d7e444ef09b579d040497544bf (diff) | |
download | clang-5243a6feb0e10358b6dd05b0fa1225d6cb51263b.tar.gz |
[OpenMP] Parsing + sema for "target exit data" directive.
Patch by Arpith Jacob. Thanks!
git-svn-id: https://llvm.org/svn/llvm-project/cfe/trunk@258177 91177308-0d34-0410-b5e6-96231b3b80d8
27 files changed, 646 insertions, 5 deletions
diff --git a/include/clang-c/Index.h b/include/clang-c/Index.h index c33ff95215..c884b14b15 100644 --- a/include/clang-c/Index.h +++ b/include/clang-c/Index.h @@ -2278,7 +2278,11 @@ enum CXCursorKind { */ CXCursor_OMPTargetEnterDataDirective = 261, - CXCursor_LastStmt = CXCursor_OMPTargetEnterDataDirective, + /** \brief OpenMP target exit data directive. + */ + CXCursor_OMPTargetExitDataDirective = 262, + + CXCursor_LastStmt = CXCursor_OMPTargetExitDataDirective, /** * \brief Cursor that represents the translation unit itself. diff --git a/include/clang/AST/RecursiveASTVisitor.h b/include/clang/AST/RecursiveASTVisitor.h index 1928d4612d..725cf3bcea 100644 --- a/include/clang/AST/RecursiveASTVisitor.h +++ b/include/clang/AST/RecursiveASTVisitor.h @@ -2436,6 +2436,9 @@ DEF_TRAVERSE_STMT(OMPTargetDataDirective, DEF_TRAVERSE_STMT(OMPTargetEnterDataDirective, { TRY_TO(TraverseOMPExecutableDirective(S)); }) +DEF_TRAVERSE_STMT(OMPTargetExitDataDirective, + { TRY_TO(TraverseOMPExecutableDirective(S)); }) + DEF_TRAVERSE_STMT(OMPTeamsDirective, { TRY_TO(TraverseOMPExecutableDirective(S)); }) diff --git a/include/clang/AST/StmtOpenMP.h b/include/clang/AST/StmtOpenMP.h index d216204b20..6275c5ff58 100644 --- a/include/clang/AST/StmtOpenMP.h +++ b/include/clang/AST/StmtOpenMP.h @@ -2098,6 +2098,66 @@ public: } }; +/// \brief This represents '#pragma omp target exit data' directive. +/// +/// \code +/// #pragma omp target exit data device(0) if(a) map(b[:]) +/// \endcode +/// In this example directive '#pragma omp target exit data' has clauses +/// 'device' with the value '0', 'if' with condition 'a' and 'map' with array +/// section 'b[:]'. +/// +class OMPTargetExitDataDirective : public OMPExecutableDirective { + friend class ASTStmtReader; + /// \brief 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 The number of clauses. + /// + OMPTargetExitDataDirective(SourceLocation StartLoc, SourceLocation EndLoc, + unsigned NumClauses) + : OMPExecutableDirective(this, OMPTargetExitDataDirectiveClass, + OMPD_target_exit_data, StartLoc, EndLoc, + NumClauses, /*NumChildren=*/0) {} + + /// \brief Build an empty directive. + /// + /// \param NumClauses Number of clauses. + /// + explicit OMPTargetExitDataDirective(unsigned NumClauses) + : OMPExecutableDirective(this, OMPTargetExitDataDirectiveClass, + OMPD_target_exit_data, SourceLocation(), + SourceLocation(), NumClauses, + /*NumChildren=*/0) {} + +public: + /// \brief 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 OMPTargetExitDataDirective *Create(const ASTContext &C, + SourceLocation StartLoc, + SourceLocation EndLoc, + ArrayRef<OMPClause *> Clauses); + + /// \brief Creates an empty directive with the place for \a N clauses. + /// + /// \param C AST context. + /// \param N The number of clauses. + /// + static OMPTargetExitDataDirective *CreateEmpty(const ASTContext &C, + unsigned N, EmptyShell); + + static bool classof(const Stmt *T) { + return T->getStmtClass() == OMPTargetExitDataDirectiveClass; + } +}; + /// \brief This represents '#pragma omp teams' directive. /// /// \code diff --git a/include/clang/Basic/OpenMPKinds.def b/include/clang/Basic/OpenMPKinds.def index ffa1872e76..a57d34ca49 100644 --- a/include/clang/Basic/OpenMPKinds.def +++ b/include/clang/Basic/OpenMPKinds.def @@ -63,6 +63,9 @@ #ifndef OPENMP_TARGET_ENTER_DATA_CLAUSE #define OPENMP_TARGET_ENTER_DATA_CLAUSE(Name) #endif +#ifndef OPENMP_TARGET_EXIT_DATA_CLAUSE +#define OPENMP_TARGET_EXIT_DATA_CLAUSE(Name) +#endif #ifndef OPENMP_TEAMS_CLAUSE # define OPENMP_TEAMS_CLAUSE(Name) #endif @@ -132,6 +135,7 @@ OPENMP_DIRECTIVE(teams) OPENMP_DIRECTIVE(cancel) OPENMP_DIRECTIVE_EXT(target_data, "target data") OPENMP_DIRECTIVE_EXT(target_enter_data, "target enter data") +OPENMP_DIRECTIVE_EXT(target_exit_data, "target exit data") OPENMP_DIRECTIVE_EXT(parallel_for, "parallel for") OPENMP_DIRECTIVE_EXT(parallel_for_simd, "parallel for simd") OPENMP_DIRECTIVE_EXT(parallel_sections, "parallel sections") @@ -364,6 +368,12 @@ OPENMP_TARGET_ENTER_DATA_CLAUSE(if) OPENMP_TARGET_ENTER_DATA_CLAUSE(device) OPENMP_TARGET_ENTER_DATA_CLAUSE(map) +// Clauses allowed for OpenMP directive 'target exit data'. +// TODO More clauses for 'target exit data' directive. +OPENMP_TARGET_EXIT_DATA_CLAUSE(if) +OPENMP_TARGET_EXIT_DATA_CLAUSE(device) +OPENMP_TARGET_EXIT_DATA_CLAUSE(map) + // Clauses allowed for OpenMP directive 'teams'. // TODO More clauses for 'teams' directive. OPENMP_TEAMS_CLAUSE(default) @@ -463,6 +473,7 @@ OPENMP_DIST_SCHEDULE_KIND(static) #undef OPENMP_TARGET_CLAUSE #undef OPENMP_TARGET_DATA_CLAUSE #undef OPENMP_TARGET_ENTER_DATA_CLAUSE +#undef OPENMP_TARGET_EXIT_DATA_CLAUSE #undef OPENMP_TEAMS_CLAUSE #undef OPENMP_SIMD_CLAUSE #undef OPENMP_FOR_CLAUSE diff --git a/include/clang/Basic/StmtNodes.td b/include/clang/Basic/StmtNodes.td index 05fc806ce1..7ec97db80a 100644 --- a/include/clang/Basic/StmtNodes.td +++ b/include/clang/Basic/StmtNodes.td @@ -217,6 +217,7 @@ def OMPAtomicDirective : DStmt<OMPExecutableDirective>; def OMPTargetDirective : DStmt<OMPExecutableDirective>; def OMPTargetDataDirective : DStmt<OMPExecutableDirective>; def OMPTargetEnterDataDirective : DStmt<OMPExecutableDirective>; +def OMPTargetExitDataDirective : DStmt<OMPExecutableDirective>; def OMPTeamsDirective : DStmt<OMPExecutableDirective>; def OMPCancellationPointDirective : DStmt<OMPExecutableDirective>; def OMPCancelDirective : DStmt<OMPExecutableDirective>; diff --git a/include/clang/Sema/Sema.h b/include/clang/Sema/Sema.h index e3136b45b3..8eaa98803c 100644 --- a/include/clang/Sema/Sema.h +++ b/include/clang/Sema/Sema.h @@ -7955,6 +7955,11 @@ public: StmtResult ActOnOpenMPTargetEnterDataDirective(ArrayRef<OMPClause *> Clauses, SourceLocation StartLoc, SourceLocation EndLoc); + /// \brief Called on well-formed '\#pragma omp target exit data' after + /// parsing of the associated statement. + StmtResult ActOnOpenMPTargetExitDataDirective(ArrayRef<OMPClause *> Clauses, + SourceLocation StartLoc, + SourceLocation EndLoc); /// \brief Called on well-formed '\#pragma omp teams' after parsing of the /// associated statement. StmtResult ActOnOpenMPTeamsDirective(ArrayRef<OMPClause *> Clauses, diff --git a/include/clang/Serialization/ASTBitCodes.h b/include/clang/Serialization/ASTBitCodes.h index 03bc505895..d7b021eee6 100644 --- a/include/clang/Serialization/ASTBitCodes.h +++ b/include/clang/Serialization/ASTBitCodes.h @@ -1446,6 +1446,7 @@ namespace clang { STMT_OMP_TARGET_DIRECTIVE, STMT_OMP_TARGET_DATA_DIRECTIVE, STMT_OMP_TARGET_ENTER_DATA_DIRECTIVE, + STMT_OMP_TARGET_EXIT_DATA_DIRECTIVE, STMT_OMP_TEAMS_DIRECTIVE, STMT_OMP_TASKGROUP_DIRECTIVE, STMT_OMP_CANCELLATION_POINT_DIRECTIVE, diff --git a/lib/AST/StmtOpenMP.cpp b/lib/AST/StmtOpenMP.cpp index b3dfd25169..4ac2ca2abd 100644 --- a/lib/AST/StmtOpenMP.cpp +++ b/lib/AST/StmtOpenMP.cpp @@ -737,6 +737,28 @@ OMPTargetEnterDataDirective::CreateEmpty(const ASTContext &C, unsigned N, return new (Mem) OMPTargetEnterDataDirective(N); } +OMPTargetExitDataDirective * +OMPTargetExitDataDirective::Create(const ASTContext &C, SourceLocation StartLoc, + SourceLocation EndLoc, + ArrayRef<OMPClause *> Clauses) { + void *Mem = C.Allocate(llvm::alignTo(sizeof(OMPTargetExitDataDirective), + llvm::alignOf<OMPClause *>()) + + sizeof(OMPClause *) * Clauses.size()); + OMPTargetExitDataDirective *Dir = + new (Mem) OMPTargetExitDataDirective(StartLoc, EndLoc, Clauses.size()); + Dir->setClauses(Clauses); + return Dir; +} + +OMPTargetExitDataDirective * +OMPTargetExitDataDirective::CreateEmpty(const ASTContext &C, unsigned N, + EmptyShell) { + void *Mem = C.Allocate(llvm::alignTo(sizeof(OMPTargetExitDataDirective), + llvm::alignOf<OMPClause *>()) + + sizeof(OMPClause *) * N); + return new (Mem) OMPTargetExitDataDirective(N); +} + OMPTeamsDirective *OMPTeamsDirective::Create(const ASTContext &C, SourceLocation StartLoc, SourceLocation EndLoc, diff --git a/lib/AST/StmtPrinter.cpp b/lib/AST/StmtPrinter.cpp index 2e8da360c9..927fdeb8de 100644 --- a/lib/AST/StmtPrinter.cpp +++ b/lib/AST/StmtPrinter.cpp @@ -1067,6 +1067,12 @@ void StmtPrinter::VisitOMPTargetEnterDataDirective( PrintOMPExecutableDirective(Node); } +void StmtPrinter::VisitOMPTargetExitDataDirective( + OMPTargetExitDataDirective *Node) { + Indent() << "#pragma omp target exit data "; + PrintOMPExecutableDirective(Node); +} + void StmtPrinter::VisitOMPTeamsDirective(OMPTeamsDirective *Node) { Indent() << "#pragma omp teams "; PrintOMPExecutableDirective(Node); diff --git a/lib/AST/StmtProfile.cpp b/lib/AST/StmtProfile.cpp index 9ab9c8e121..2ea0243f12 100644 --- a/lib/AST/StmtProfile.cpp +++ b/lib/AST/StmtProfile.cpp @@ -589,6 +589,11 @@ void StmtProfiler::VisitOMPTargetEnterDataDirective( VisitOMPExecutableDirective(S); } +void StmtProfiler::VisitOMPTargetExitDataDirective( + const OMPTargetExitDataDirective *S) { + VisitOMPExecutableDirective(S); +} + void StmtProfiler::VisitOMPTeamsDirective(const OMPTeamsDirective *S) { VisitOMPExecutableDirective(S); } diff --git a/lib/Basic/OpenMPKinds.cpp b/lib/Basic/OpenMPKinds.cpp index 167a700d27..6798387193 100644 --- a/lib/Basic/OpenMPKinds.cpp +++ b/lib/Basic/OpenMPKinds.cpp @@ -423,6 +423,16 @@ bool clang::isAllowedClauseForDirective(OpenMPDirectiveKind DKind, break; } break; + case OMPD_target_exit_data: + switch (CKind) { +#define OPENMP_TARGET_EXIT_DATA_CLAUSE(Name) \ + case OMPC_##Name: \ + return true; +#include "clang/Basic/OpenMPKinds.def" + default: + break; + } + break; case OMPD_teams: switch (CKind) { #define OPENMP_TEAMS_CLAUSE(Name) \ diff --git a/lib/CodeGen/CGStmt.cpp b/lib/CodeGen/CGStmt.cpp index df6766a768..bd4fa8b7a3 100644 --- a/lib/CodeGen/CGStmt.cpp +++ b/lib/CodeGen/CGStmt.cpp @@ -259,6 +259,9 @@ void CodeGenFunction::EmitStmt(const Stmt *S) { case Stmt::OMPTargetEnterDataDirectiveClass: EmitOMPTargetEnterDataDirective(cast<OMPTargetEnterDataDirective>(*S)); break; + case Stmt::OMPTargetExitDataDirectiveClass: + EmitOMPTargetExitDataDirective(cast<OMPTargetExitDataDirective>(*S)); + break; case Stmt::OMPTaskLoopDirectiveClass: EmitOMPTaskLoopDirective(cast<OMPTaskLoopDirective>(*S)); break; diff --git a/lib/CodeGen/CGStmtOpenMP.cpp b/lib/CodeGen/CGStmtOpenMP.cpp index f6b8e5907d..793e1a9731 100644 --- a/lib/CodeGen/CGStmtOpenMP.cpp +++ b/lib/CodeGen/CGStmtOpenMP.cpp @@ -2666,6 +2666,11 @@ void CodeGenFunction::EmitOMPTargetEnterDataDirective( // TODO: codegen for target enter data. } +void CodeGenFunction::EmitOMPTargetExitDataDirective( + const OMPTargetExitDataDirective &S) { + // TODO: codegen for target exit data. +} + void CodeGenFunction::EmitOMPTaskLoopDirective(const OMPTaskLoopDirective &S) { // emit the code inside the construct for now auto CS = cast<CapturedStmt>(S.getAssociatedStmt()); diff --git a/lib/CodeGen/CodeGenFunction.h b/lib/CodeGen/CodeGenFunction.h index d1cf0c9206..129cc9918d 100644 --- a/lib/CodeGen/CodeGenFunction.h +++ b/lib/CodeGen/CodeGenFunction.h @@ -2337,6 +2337,7 @@ public: void EmitOMPTargetDirective(const OMPTargetDirective &S); void EmitOMPTargetDataDirective(const OMPTargetDataDirective &S); void EmitOMPTargetEnterDataDirective(const OMPTargetEnterDataDirective &S); + void EmitOMPTargetExitDataDirective(const OMPTargetExitDataDirective &S); void EmitOMPTeamsDirective(const OMPTeamsDirective &S); void EmitOMPCancellationPointDirective(const OMPCancellationPointDirective &S); diff --git a/lib/Parse/ParseOpenMP.cpp b/lib/Parse/ParseOpenMP.cpp index 4baf5d2661..94d90c0e21 100644 --- a/lib/Parse/ParseOpenMP.cpp +++ b/lib/Parse/ParseOpenMP.cpp @@ -38,6 +38,8 @@ static OpenMPDirectiveKind ParseOpenMPDirectiveKind(Parser &P) { OMPD_unknown /*target enter/exit*/}, {OMPD_unknown /*target enter*/, OMPD_unknown /*data*/, OMPD_target_enter_data}, + {OMPD_unknown /*target exit*/, OMPD_unknown /*data*/, + OMPD_target_exit_data}, {OMPD_for, OMPD_simd, OMPD_for_simd}, {OMPD_parallel, OMPD_for, OMPD_parallel_for}, {OMPD_parallel_for, OMPD_simd, OMPD_parallel_for_simd}, @@ -55,7 +57,9 @@ static OpenMPDirectiveKind ParseOpenMPDirectiveKind(Parser &P) { TokenMatched = ((i == 0) && !P.getPreprocessor().getSpelling(Tok).compare("cancellation")) || - ((i == 3) && !P.getPreprocessor().getSpelling(Tok).compare("enter")); + ((i == 3) && + !P.getPreprocessor().getSpelling(Tok).compare("enter")) || + ((i == 4) && !P.getPreprocessor().getSpelling(Tok).compare("exit")); } else { TokenMatched = DKind == F[i][0] && DKind != OMPD_unknown; } @@ -72,10 +76,11 @@ static OpenMPDirectiveKind ParseOpenMPDirectiveKind(Parser &P) { TokenMatched = ((i == 0) && !P.getPreprocessor().getSpelling(Tok).compare("point")) || - ((i == 1 || i == 3) && + ((i == 1 || i == 3 || i == 4) && !P.getPreprocessor().getSpelling(Tok).compare("data")) || ((i == 2) && - !P.getPreprocessor().getSpelling(Tok).compare("enter")); + (!P.getPreprocessor().getSpelling(Tok).compare("enter") || + !P.getPreprocessor().getSpelling(Tok).compare("exit"))); } else { TokenMatched = SDKind == F[i][1] && SDKind != OMPD_unknown; } @@ -147,6 +152,7 @@ Parser::DeclGroupPtrTy Parser::ParseOpenMPDeclarativeDirective() { case OMPD_cancel: case OMPD_target_data: case OMPD_target_enter_data: + case OMPD_target_exit_data: case OMPD_taskloop: case OMPD_taskloop_simd: case OMPD_distribute: @@ -171,7 +177,8 @@ Parser::DeclGroupPtrTy Parser::ParseOpenMPDeclarativeDirective() { /// 'barrier' | 'taskwait' | 'flush' | 'ordered' | 'atomic' | /// 'for simd' | 'parallel for simd' | 'target' | 'target data' | /// 'taskgroup' | 'teams' | 'taskloop' | 'taskloop simd' {clause} | -/// 'distribute' | 'target enter data' | annot_pragma_openmp_end +/// 'distribute' | 'target enter data' | 'target exit data' +/// annot_pragma_openmp_end /// StmtResult Parser::ParseOpenMPDeclarativeOrExecutableDirective( AllowedContsructsKind Allowed) { @@ -226,6 +233,7 @@ StmtResult Parser::ParseOpenMPDeclarativeOrExecutableDirective( case OMPD_cancellation_point: case OMPD_cancel: case OMPD_target_enter_data: + case OMPD_target_exit_data: if (Allowed == ACK_StatementsOpenMPNonStandalone) { Diag(Tok, diag::err_omp_immediate_directive) << getOpenMPDirectiveName(DKind) << 0; diff --git a/lib/Sema/SemaOpenMP.cpp b/lib/Sema/SemaOpenMP.cpp index a3e33bc658..1b51553638 100644 --- a/lib/Sema/SemaOpenMP.cpp +++ b/lib/Sema/SemaOpenMP.cpp @@ -1611,6 +1611,7 @@ void Sema::ActOnOpenMPRegionStart(OpenMPDirectiveKind DKind, Scope *CurScope) { case OMPD_cancel: case OMPD_flush: case OMPD_target_enter_data: + case OMPD_target_exit_data: llvm_unreachable("OpenMP Directive is not allowed"); case OMPD_unknown: llvm_unreachable("Unknown OpenMP directive"); @@ -1727,6 +1728,8 @@ static bool CheckNestingOfRegions(Sema &SemaRef, DSAStackTy *Stack, // | parallel | target | * | // | parallel | target enter | * | // | | data | | + // | parallel | target exit | * | + // | | data | | // | parallel | teams | + | // | parallel | cancellation | | // | | point | ! | @@ -1758,6 +1761,8 @@ static bool CheckNestingOfRegions(Sema &SemaRef, DSAStackTy *Stack, // | for | target | * | // | for | target enter | * | // | | data | | + // | for | target exit | * | + // | | data | | // | for | teams | + | // | for | cancellation | | // | | point | ! | @@ -1789,6 +1794,8 @@ static bool CheckNestingOfRegions(Sema &SemaRef, DSAStackTy *Stack, // | master | target | * | // | master | target enter | * | // | | data | | + // | master | target exit | * | + // | | data | | // | master | teams | + | // | master | cancellation | | // | | point | | @@ -1819,6 +1826,8 @@ static bool CheckNestingOfRegions(Sema &SemaRef, DSAStackTy *Stack, // | critical | target | * | // | critical | target enter | * | // | | data | | + // | critical | target exit | * | + // | | data | | // | critical | teams | + | // | critical | cancellation | | // | | point | | @@ -1850,6 +1859,8 @@ static bool CheckNestingOfRegions(Sema &SemaRef, DSAStackTy *Stack, // | simd | target | | // | simd | target enter | | // | | data | | + // | simd | target exit | | + // | | data | | // | simd | teams | | // | simd | cancellation | | // | | point | | @@ -1881,6 +1892,8 @@ static bool CheckNestingOfRegions(Sema &SemaRef, DSAStackTy *Stack, // | for simd | target | | // | for simd | target enter | | // | | data | | + // | for simd | target exit | | + // | | data | | // | for simd | teams | | // | for simd | cancellation | | // | | point | | @@ -1912,6 +1925,8 @@ static bool CheckNestingOfRegions(Sema &SemaRef, DSAStackTy *Stack, // | parallel for simd| target | | // | parallel for simd| target enter | | // | | data | | + // | parallel for simd| target exit | | + // | | data | | // | parallel for simd| teams | | // | parallel for simd| cancellation | | // | | point | | @@ -1943,6 +1958,8 @@ static bool CheckNestingOfRegions(Sema &SemaRef, DSAStackTy *Stack, // | sections | target | * | // | sections | target enter | * | // | | data | | + // | sections | target exit | * | + // | | data | | // | sections | teams | + | // | sections | cancellation | | // | | point | ! | @@ -1974,6 +1991,8 @@ static bool CheckNestingOfRegions(Sema &SemaRef, DSAStackTy *Stack, // | section | target | * | // | section | target enter | * | // | | data | | + // | section | target exit | * | + // | | data | | // | section | teams | + | // | section | cancellation | | // | | point | ! | @@ -2005,6 +2024,8 @@ static bool CheckNestingOfRegions(Sema &SemaRef, DSAStackTy *Stack, // | single | target | * | // | single | target enter | * | // | | data | | + // | single | target exit | * | + // | | data | | // | single | teams | + | // | single | cancellation | | // | | point | | @@ -2036,6 +2057,8 @@ static bool CheckNestingOfRegions(Sema &SemaRef, DSAStackTy *Stack, // | parallel for | target | * | // | parallel for | target enter | * | // | | data | | + // | parallel for | target exit | * | + // | | data | | // | parallel for | teams | + | // | parallel for | cancellation | | // | | point | ! | @@ -2067,6 +2090,8 @@ static bool CheckNestingOfRegions(Sema &SemaRef, DSAStackTy *Stack, // | parallel sections| target | * | // | parallel sections| target enter | * | // | | data | | + // | parallel sections| target exit | * | + // | | data | | // | parallel sections| teams | + | // | parallel sections| cancellation | | // | | point | ! | @@ -2098,6 +2123,8 @@ static bool CheckNestingOfRegions(Sema &SemaRef, DSAStackTy *Stack, // | task | target | * | // | task | target enter | * | // | | data | | + // | task | target exit | * | + // | | data | | // | task | teams | + | // | task | cancellation | | // | | point | ! | @@ -2129,6 +2156,8 @@ static bool CheckNestingOfRegions(Sema &SemaRef, DSAStackTy *Stack, // | ordered | target | * | // | ordered | target enter | * | // | | data | | + // | ordered | target exit | * | + // | | data | | // | ordered | teams | + | // | ordered | cancellation | | // | | point | | @@ -2160,6 +2189,8 @@ static bool CheckNestingOfRegions(Sema &SemaRef, DSAStackTy *Stack, // | atomic | target | | // | atomic | target enter | | // | | data | | + // | atomic | target exit | | + // | | data | | // | atomic | teams | | // | atomic | cancellation | | // | | point | | @@ -2191,6 +2222,8 @@ static bool CheckNestingOfRegions(Sema &SemaRef, DSAStackTy *Stack, // | target | target | * | // | target | target enter | * | // | | data | | + // | target | target exit | * | + // | | data | | // | target | teams | * | // | target | cancellation | | // | | point | | @@ -2222,6 +2255,8 @@ static bool CheckNestingOfRegions(Sema &SemaRef, DSAStackTy *Stack, // | teams | target | + | // | teams | target enter | + | // | | data | | + // | teams | target exit | + | + // | | data | | // | teams | teams | + | // | teams | cancellation | | // | | point | | @@ -2253,6 +2288,8 @@ static bool CheckNestingOfRegions(Sema &SemaRef, DSAStackTy *Stack, // | taskloop | target | * | // | taskloop | target enter | * | // | | data | | + // | taskloop | target exit | * | + // | | data | | // | taskloop | teams | + | // | taskloop | cancellation | | // | | point | | @@ -2283,6 +2320,8 @@ static bool CheckNestingOfRegions(Sema &SemaRef, DSAStackTy *Stack, // | taskloop simd | target | | // | taskloop simd | target enter | | // | | data | | + // | taskloop simd | target exit | | + // | | data | | // | taskloop simd | teams | | // | taskloop simd | cancellation | | // | | point | | @@ -2314,6 +2353,8 @@ static bool CheckNestingOfRegions(Sema &SemaRef, DSAStackTy *Stack, // | distribute | target | | // | distribute | target enter | | // | | data | | + // | distribute | target exit | | + // | | data | | // | distribute | teams | | // | distribute | cancellation | + | // | | point | | @@ -2743,6 +2784,11 @@ StmtResult Sema::ActOnOpenMPExecutableDirective( EndLoc); AllowedNameModifiers.push_back(OMPD_target_enter_data); break; + case OMPD_target_exit_data: + Res = ActOnOpenMPTargetExitDataDirective(ClausesWithImplicit, StartLoc, + EndLoc); + AllowedNameModifiers.push_back(OMPD_target_exit_data); + break; case OMPD_taskloop: Res = ActOnOpenMPTaskLoopDirective(ClausesWithImplicit, AStmt, StartLoc, EndLoc, VarsWithInheritedDSA); @@ -5522,6 +5568,21 @@ Sema::ActOnOpenMPTargetEnterDataDirective(ArrayRef<OMPClause *> Clauses, Clauses); } +StmtResult +Sema::ActOnOpenMPTargetExitDataDirective(ArrayRef<OMPClause *> Clauses, + SourceLocation StartLoc, + SourceLocation EndLoc) { + // OpenMP [2.10.3, Restrictions, p. 102] + // At least one map clause must appear on the directive. + if (!HasMapClause(Clauses)) { + Diag(StartLoc, diag::err_omp_no_map_for_directive) + << getOpenMPDirectiveName(OMPD_target_exit_data); + return StmtError(); + } + + return OMPTargetExitDataDirective::Create(Context, StartLoc, EndLoc, Clauses); +} + StmtResult Sema::ActOnOpenMPTeamsDirective(ArrayRef<OMPClause *> Clauses, Stmt *AStmt, SourceLocation StartLoc, SourceLocation EndLoc) { @@ -8539,6 +8600,23 @@ OMPClause *Sema::ActOnOpenMPMapClause( // further spurious messages } + // target exit_data + // OpenMP [2.10.3, Restrictions, p. 102] + // A map-type must be specified in all map clauses and must be either + // from, release, or delete. + DKind = DSAStack->getCurrentDirective(); + if (DKind == OMPD_target_exit_data && + !(MapType == OMPC_MAP_from || MapType == OMPC_MAP_release || + MapType == OMPC_MAP_delete)) { + Diag(StartLoc, diag::err_omp_invalid_map_type_for_directive) + << + // TODO: Need to determine if map type is implicitly determined + 0 << getOpenMPSimpleClauseTypeName(OMPC_map, MapType) + << getOpenMPDirectiveName(DKind); + // Proceed to add the variable in a map clause anyway, to prevent + // further spurious messages + } + Vars.push_back(RE); MI.RefExpr = RE; DSAStack->addMapInfoForVar(VD, MI); diff --git a/lib/Sema/TreeTransform.h b/lib/Sema/TreeTransform.h index 4b614d888c..3b43573195 100644 --- a/lib/Sema/TreeTransform.h +++ b/lib/Sema/TreeTransform.h @@ -7402,6 +7402,17 @@ StmtResult TreeTransform<Derived>::TransformOMPTargetEnterDataDirective( } template <typename Derived> +StmtResult TreeTransform<Derived>::TransformOMPTargetExitDataDirective( + OMPTargetExitDataDirective *D) { + DeclarationNameInfo DirName; + getDerived().getSema().StartOpenMPDSABlock(OMPD_target_exit_data, DirName, + nullptr, D->getLocStart()); + StmtResult Res = getDerived().TransformOMPExecutableDirective(D); + getDerived().getSema().EndOpenMPDSABlock(Res.get()); + return Res; +} + +template <typename Derived> StmtResult TreeTransform<Derived>::TransformOMPTeamsDirective(OMPTeamsDirective *D) { DeclarationNameInfo DirName; diff --git a/lib/Serialization/ASTReaderStmt.cpp b/lib/Serialization/ASTReaderStmt.cpp index 4b312ce757..ae978d5bcb 100644 --- a/lib/Serialization/ASTReaderStmt.cpp +++ b/lib/Serialization/ASTReaderStmt.cpp @@ -2462,6 +2462,13 @@ void ASTStmtReader::VisitOMPTargetEnterDataDirective( VisitOMPExecutableDirective(D); } +void ASTStmtReader::VisitOMPTargetExitDataDirective( + OMPTargetExitDataDirective *D) { + VisitStmt(D); + ++Idx; + VisitOMPExecutableDirective(D); +} + void ASTStmtReader::VisitOMPTeamsDirective(OMPTeamsDirective *D) { VisitStmt(D); // The NumClauses field was read in ReadStmtFromStream. @@ -3110,6 +3117,11 @@ Stmt *ASTReader::ReadStmtFromStream(ModuleFile &F) { Context, Record[ASTStmtReader::NumStmtFields], Empty); break; + case STMT_OMP_TARGET_EXIT_DATA_DIRECTIVE: + S = OMPTargetExitDataDirective::CreateEmpty( + Context, Record[ASTStmtReader::NumStmtFields], Empty); + break; + case STMT_OMP_TEAMS_DIRECTIVE: S = OMPTeamsDirective::CreateEmpty( Context, Record[ASTStmtReader::NumStmtFields], Empty); diff --git a/lib/Serialization/ASTWriterStmt.cpp b/lib/Serialization/ASTWriterStmt.cpp index ed6044d45f..8d1de313d0 100644 --- a/lib/Serialization/ASTWriterStmt.cpp +++ b/lib/Serialization/ASTWriterStmt.cpp @@ -2222,6 +2222,14 @@ void ASTStmtWriter::VisitOMPTargetEnterDataDirective( Code = serialization::STMT_OMP_TARGET_ENTER_DATA_DIRECTIVE; } +void ASTStmtWriter::VisitOMPTargetExitDataDirective( + OMPTargetExitDataDirective *D) { + VisitStmt(D); + Record.push_back(D->getNumClauses()); + VisitOMPExecutableDirective(D); + Code = serialization::STMT_OMP_TARGET_EXIT_DATA_DIRECTIVE; +} + void ASTStmtWriter::VisitOMPTaskyieldDirective(OMPTaskyieldDirective *D) { VisitStmt(D); VisitOMPExecutableDirective(D); diff --git a/lib/StaticAnalyzer/Core/ExprEngine.cpp b/lib/StaticAnalyzer/Core/ExprEngine.cpp index e4e61ab2d6..309caa50ea 100644 --- a/lib/StaticAnalyzer/Core/ExprEngine.cpp +++ b/lib/StaticAnalyzer/Core/ExprEngine.cpp @@ -831,6 +831,7 @@ void ExprEngine::Visit(const Stmt *S, ExplodedNode *Pred, case Stmt::OMPTargetDirectiveClass: case Stmt::OMPTargetDataDirectiveClass: case Stmt::OMPTargetEnterDataDirectiveClass: + case Stmt::OMPTargetExitDataDirectiveClass: case Stmt::OMPTeamsDirectiveClass: case Stmt::OMPCancellationPointDirectiveClass: case Stmt::OMPCancelDirectiveClass: diff --git a/test/OpenMP/nesting_of_regions.cpp b/test/OpenMP/nesting_of_regions.cpp index 0f83a5ee8c..adb59b0d39 100644 --- a/test/OpenMP/nesting_of_regions.cpp +++ b/test/OpenMP/nesting_of_regions.cpp @@ -102,6 +102,11 @@ void foo() { } #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; } @@ -246,6 +251,11 @@ void foo() { } #pragma omp simd for (int i = 0; i < 10; ++i) { +#pragma omp target exit data map(from: a) // expected-error {{OpenMP constructs may not be nested inside a simd region}} + ++a; + } +#pragma omp simd + for (int i = 0; i < 10; ++i) { #pragma omp teams // expected-error {{OpenMP constructs may not be nested inside a simd region}} ++a; } @@ -413,6 +423,11 @@ void foo() { } #pragma omp for for (int i = 0; i < 10; ++i) { +#pragma omp target exit data map(from: a) + ++a; + } +#pragma omp for + for (int i = 0; i < 10; ++i) { #pragma omp teams // expected-error {{region cannot be closely nested inside 'for' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} ++a; } @@ -557,6 +572,11 @@ void foo() { } #pragma omp for simd for (int i = 0; i < 10; ++i) { +#pragma omp target exit data map(from: a) // expected-error {{OpenMP constructs may not be nested inside a simd region}} + ++a; + } +#pragma omp for simd + for (int i = 0; i < 10; ++i) { #pragma omp teams // expected-error {{OpenMP constructs may not be nested inside a simd region}} ++a; } @@ -730,6 +750,10 @@ void foo() { } #pragma omp sections { +#pragma omp target exit data map(from: a) + } +#pragma omp sections + { #pragma omp teams // expected-error {{region cannot be closely nested inside 'sections' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} ++a; } @@ -943,6 +967,14 @@ void foo() { #pragma omp sections { #pragma omp section + { +#pragma omp target exit data map(from: a) + ++a; + } + } +#pragma omp sections + { +#pragma omp section #pragma omp teams // expected-error {{region cannot be closely nested inside 'section' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} ++a; } @@ -1102,6 +1134,11 @@ void foo() { } #pragma omp single { +#pragma omp target exit data map(from: a) + ++a; + } +#pragma omp single + { #pragma omp teams // expected-error {{region cannot be closely nested inside 'single' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} ++a; } @@ -1259,6 +1296,11 @@ void foo() { } #pragma omp master { +#pragma omp target exit data map(from: a) + ++a; + } +#pragma omp master + { #pragma omp teams // expected-error {{region cannot be closely nested inside 'master' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} ++a; } @@ -1430,6 +1472,11 @@ void foo() { } #pragma omp critical { +#pragma omp target exit data map(from: a) + ++a; + } +#pragma omp critical + { #pragma omp teams // expected-error {{region cannot be closely nested inside 'critical' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} ++a; } @@ -1602,6 +1649,11 @@ void foo() { } #pragma omp parallel for for (int i = 0; i < 10; ++i) { +#pragma omp target exit data map(from: a) + ++a; + } +#pragma omp parallel for + for (int i = 0; i < 10; ++i) { #pragma omp teams // expected-error {{region cannot be closely nested inside 'parallel for' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} ++a; } @@ -1774,6 +1826,11 @@ void foo() { } #pragma omp parallel for simd for (int i = 0; i < 10; ++i) { +#pragma omp target exit data map(from: a) // expected-error {{OpenMP constructs may not be nested inside a simd region}} + ++a; + } +#pragma omp parallel for simd + for (int i = 0; i < 10; ++i) { #pragma omp teams // expected-error {{OpenMP constructs may not be nested inside a simd region}} ++a; } @@ -1936,6 +1993,10 @@ void foo() { } #pragma omp parallel sections { +#pragma omp target exit data map(from: a) + } +#pragma omp parallel sections + { #pragma omp teams // expected-error {{region cannot be closely nested inside 'parallel sections' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} ++a; } @@ -2045,6 +2106,11 @@ void foo() { } #pragma omp task { +#pragma omp target exit data map(from: a) + ++a; + } +#pragma omp task + { #pragma omp teams // expected-error {{region cannot be closely nested inside 'task' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} ++a; } @@ -2212,6 +2278,11 @@ void foo() { } #pragma omp ordered { +#pragma omp target exit data map(from: a) + ++a; + } +#pragma omp ordered + { #pragma omp teams // expected-error {{region cannot be closely nested inside 'ordered' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} ++a; } @@ -2400,6 +2471,13 @@ void foo() { // 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 exit data map(from: a) // expected-error {{OpenMP constructs may not be nested inside an atomic region}} + ++a; + } +#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 teams // expected-error {{OpenMP constructs may not be nested inside an atomic region}} ++a; } @@ -2540,6 +2618,10 @@ void foo() { { #pragma omp target enter data map(to: a) } +#pragma omp target + { +#pragma omp target exit data map(from: a) + } // TEAMS DIRECTIVE #pragma omp target @@ -2663,6 +2745,12 @@ void foo() { #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?}} + ++a; + } +#pragma omp target +#pragma omp teams + { #pragma omp teams // expected-error {{region cannot be closely nested inside 'teams' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} ++a; } @@ -2833,6 +2921,11 @@ void foo() { } #pragma omp taskloop for (int i = 0; i < 10; ++i) { +#pragma omp target exit data map(from: a) + ++a; + } +#pragma omp taskloop + for (int i = 0; i < 10; ++i) { #pragma omp teams // expected-error {{region cannot be closely nested inside 'taskloop' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} ++a; } @@ -3034,6 +3127,13 @@ void foo() { #pragma omp teams #pragma omp distribute for (int i = 0; i < 10; ++i) { +#pragma omp target exit data map(from: a) + ++a; + } +#pragma omp target +#pragma omp teams +#pragma omp distribute + for (int i = 0; i < 10; ++i) { #pragma omp teams // expected-error {{region cannot be closely nested inside 'distribute' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} ++a; } @@ -3138,6 +3238,11 @@ void foo() { } #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; } @@ -3275,6 +3380,11 @@ void foo() { } #pragma omp simd for (int i = 0; i < 10; ++i) { +#pragma omp target exit data map(from: a) // expected-error {{OpenMP constructs may not be nested inside a simd region}} + ++a; + } +#pragma omp simd + for (int i = 0; i < 10; ++i) { #pragma omp teams // expected-error {{OpenMP constructs may not be nested inside a simd region}} ++a; } @@ -3432,6 +3542,11 @@ void foo() { } #pragma omp for for (int i = 0; i < 10; ++i) { +#pragma omp target exit data map(from: a) + ++a; + } +#pragma omp for + for (int i = 0; i < 10; ++i) { #pragma omp teams // expected-error {{region cannot be closely nested inside 'for' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} ++a; } @@ -3569,6 +3684,11 @@ void foo() { } #pragma omp for simd for (int i = 0; i < 10; ++i) { +#pragma omp target exit data map(from: a) // expected-error {{OpenMP constructs may not be nested inside a simd region}} + ++a; + } +#pragma omp for simd + for (int i = 0; i < 10; ++i) { #pragma omp teams // expected-error {{OpenMP constructs may not be nested inside a simd region}} ++a; } @@ -3717,6 +3837,10 @@ void foo() { } #pragma omp sections { +#pragma omp target exit data map(from: a) + } +#pragma omp sections + { #pragma omp teams // expected-error {{region cannot be closely nested inside 'sections' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} ++a; } @@ -3935,6 +4059,14 @@ void foo() { { #pragma omp section { +#pragma omp target exit data map(from: a) + ++a; + } + } +#pragma omp sections + { +#pragma omp section + { #pragma omp teams // expected-error {{region cannot be closely nested inside 'section' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} ++a; } @@ -4087,6 +4219,11 @@ void foo() { } #pragma omp single { +#pragma omp target exit data map(from: a) + ++a; + } +#pragma omp single + { #pragma omp teams // expected-error {{region cannot be closely nested inside 'single' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} ++a; } @@ -4244,6 +4381,11 @@ void foo() { } #pragma omp master { +#pragma omp target exit data map(from: a) + ++a; + } +#pragma omp master + { #pragma omp teams // expected-error {{region cannot be closely nested inside 'master' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} ++a; } @@ -4420,6 +4562,11 @@ void foo() { } #pragma omp critical { +#pragma omp target exit data map(from: a) + ++a; + } +#pragma omp critical + { #pragma omp teams // expected-error {{region cannot be closely nested inside 'critical' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} ++a; } @@ -4592,6 +4739,11 @@ void foo() { } #pragma omp parallel for for (int i = 0; i < 10; ++i) { +#pragma omp target exit data map(from: a) + ++a; + } +#pragma omp parallel for + for (int i = 0; i < 10; ++i) { #pragma omp teams // expected-error {{region cannot be closely nested inside 'parallel for' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} ++a; } @@ -4764,6 +4916,11 @@ void foo() { } #pragma omp parallel for simd for (int i = 0; i < 10; ++i) { +#pragma omp target exit data map(from: a) // expected-error {{OpenMP constructs may not be nested inside a simd region}} + ++a; + } +#pragma omp parallel for simd + for (int i = 0; i < 10; ++i) { #pragma omp teams // expected-error {{OpenMP constructs may not be nested inside a simd region}} ++a; } @@ -4922,6 +5079,10 @@ void foo() { } #pragma omp parallel sections { +#pragma omp target exit data map(from: a) + } +#pragma omp parallel sections + { #pragma omp teams // expected-error {{region cannot be closely nested inside 'parallel sections' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} ++a; } @@ -5030,6 +5191,11 @@ void foo() { } #pragma omp task { +#pragma omp target exit data map(from: a) + ++a; + } +#pragma omp task + { #pragma omp teams // expected-error {{region cannot be closely nested inside 'task' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} ++a; } @@ -5218,6 +5384,13 @@ void foo() { // 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 exit data map(from: a) // expected-error {{OpenMP constructs may not be nested inside an atomic region}} + ++a; + } +#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 teams // expected-error {{OpenMP constructs may not be nested inside an atomic region}} ++a; } @@ -5337,6 +5510,10 @@ void foo() { } #pragma omp target { +#pragma omp target exit data map(from: a) + } +#pragma omp target + { #pragma omp teams ++a; } @@ -5480,6 +5657,11 @@ void foo() { #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 +#pragma omp teams + { #pragma omp teams // expected-error {{region cannot be closely nested inside 'teams' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} ++a; } @@ -5650,6 +5832,11 @@ void foo() { } #pragma omp taskloop for (int i = 0; i < 10; ++i) { +#pragma omp target exit data map(from: a) + ++a; + } +#pragma omp taskloop + for (int i = 0; i < 10; ++i) { #pragma omp teams // expected-error {{region cannot be closely nested inside 'taskloop' region; perhaps you forget to enclose 'omp teams' directive into a target region?}} ++a; } @@ -5856,4 +6043,11 @@ void foo() { #pragma omp target enter data map(to: a) ++a; } +#pragma omp target +#pragma omp teams +#pragma omp distribute + for (int i = 0; i < 10; ++i) { +#pragma omp target exit data map(from: a) + ++a; + } } diff --git a/test/OpenMP/target_exit_data_ast_print.cpp b/test/OpenMP/target_exit_data_ast_print.cpp new file mode 100644 index 0000000000..26f9073df1 --- /dev/null +++ b/test/OpenMP/target_exit_data_ast_print.cpp @@ -0,0 +1,100 @@ +// 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 + +template <typename T, int C> +T tmain(T argc, T *argv) { + T i, j, b, c, d, e, x[20]; + + i = argc; +#pragma omp target exit data map(from: i) + +#pragma omp target exit data map(from: i) if (target exit data: j > 0) + +#pragma omp target exit data map(from: i) if (b) + +#pragma omp target exit data map(from: c) + +#pragma omp target exit data map(from: c) if(b>e) + +#pragma omp target exit data map(release: x[0:10], c) + +#pragma omp target exit data map(from: c) map(release: d) + +#pragma omp target exit data map(always,release: e) + + return 0; +} + +// CHECK: template <typename T = int, int C = 5> int tmain(int argc, int *argv) { +// CHECK-NEXT: int i, j, b, c, d, e, x[20]; +// CHECK-NEXT: i = argc; +// CHECK-NEXT: #pragma omp target exit data map(from: i) +// CHECK-NEXT: #pragma omp target exit data map(from: i) if(target exit data: j > 0) +// CHECK-NEXT: #pragma omp target exit data map(from: i) if(b) +// CHECK-NEXT: #pragma omp target exit data map(from: c) +// CHECK-NEXT: #pragma omp target exit data map(from: c) if(b > e) +// CHECK-NEXT: #pragma omp target exit data map(release: x[0:10],c) +// CHECK-NEXT: #pragma omp target exit data map(from: c) map(release: d) +// CHECK-NEXT: #pragma omp target exit data map(always,release: e) +// CHECK: template <typename T = char, int C = 1> char tmain(char argc, char *argv) { +// CHECK-NEXT: char i, j, b, c, d, e, x[20]; +// CHECK-NEXT: i = argc; +// CHECK-NEXT: #pragma omp target exit data map(from: i) +// CHECK-NEXT: #pragma omp target exit data map(from: i) if(target exit data: j > 0) +// CHECK-NEXT: #pragma omp target exit data map(from: i) if(b) +// CHECK-NEXT: #pragma omp target exit data map(from: c) +// CHECK-NEXT: #pragma omp target exit data map(from: c) if(b > e) +// CHECK-NEXT: #pragma omp target exit data map(release: x[0:10],c) +// CHECK-NEXT: #pragma omp target exit data map(from: c) map(release: d) +// CHECK-NEXT: #pragma omp target exit data map(always,release: e) +// CHECK: template <typename T, int C> T tmain(T argc, T *argv) { +// CHECK-NEXT: T i, j, b, c, d, e, x[20]; +// CHECK-NEXT: i = argc; +// CHECK-NEXT: #pragma omp target exit data map(from: i) +// CHECK-NEXT: #pragma omp target exit data map(from: i) if(target exit data: j > 0) +// CHECK-NEXT: #pragma omp target exit data map(from: i) if(b) +// CHECK-NEXT: #pragma omp target exit data map(from: c) +// CHECK-NEXT: #pragma omp target exit data map(from: c) if(b > e) +// CHECK-NEXT: #pragma omp target exit data map(release: x[0:10],c) +// CHECK-NEXT: #pragma omp target exit data map(from: c) map(release: d) +// CHECK-NEXT: #pragma omp target exit data map(always,release: e) + +int main (int argc, char **argv) { + int b = argc, c, d, e, f, g, x[20]; + static int a; +// CHECK: static int a; + +#pragma omp target exit data map(from: a) +// CHECK: #pragma omp target exit data map(from: a) + a=2; +// CHECK-NEXT: a = 2; +#pragma omp target exit data map(from: a) if (target exit data: b) +// CHECK: #pragma omp target exit data map(from: a) if(target exit data: b) + +#pragma omp target exit data map(from: a) if (b > g) +// CHECK: #pragma omp target exit data map(from: a) if(b > g) + +#pragma omp target exit data map(from: c) +// CHECK-NEXT: #pragma omp target exit data map(from: c) + +#pragma omp target exit data map(release: c) if(b>g) +// CHECK-NEXT: #pragma omp target exit data map(release: c) if(b > g) + +#pragma omp target exit data map(from: x[0:10], c) +// CHECK-NEXT: #pragma omp target exit data map(from: x[0:10],c) + +#pragma omp target exit data map(from: c) map(release: d) +// CHECK-NEXT: #pragma omp target exit data map(from: c) map(release: d) + +#pragma omp target exit data map(always,release: e) +// CHECK-NEXT: #pragma omp target exit data map(always,release: e) + + return tmain<int, 5>(argc, &argc) + tmain<char, 1>(argv[0][0], argv[0]); +} + +#endif diff --git a/test/OpenMP/target_exit_data_device_messages.cpp b/test/OpenMP/target_exit_data_device_messages.cpp new file mode 100644 index 0000000000..d9ce0ddd7e --- /dev/null +++ b/test/OpenMP/target_exit_data_device_messages.cpp @@ -0,0 +1,29 @@ +// 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) { + int i; + #pragma omp target exit data map(from: i) device // expected-error {{expected '(' after 'device'}} + #pragma omp target exit data map(from: i) device ( // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}} + #pragma omp target exit data map(from: i) device () // expected-error {{expected expression}} + #pragma omp target exit data map(from: i) device (argc // expected-error {{expected ')'}} expected-note {{to match this '('}} + #pragma omp target exit data map(from: i) device (argc)) // expected-warning {{extra tokens at the end of '#pragma omp target exit data' are ignored}} +#pragma omp target exit data map(from: i) device (argc > 0 ? argv[1] : argv[2]) // expected-error {{expression must have integral or unscoped enumeration type, not 'char *'}} + #pragma omp target exit data map(from: i) device (argc + argc) + #pragma omp target exit data map(from: i) device (argc), device (argc+1) // expected-error {{directive '#pragma omp target exit data' cannot contain more than one 'device' clause}} + #pragma omp target exit data map(from: i) device (S1) // expected-error {{'S1' does not refer to a value}} + #pragma omp target exit data map(from: i) device (-2) // expected-error {{argument to 'device' clause must be a non-negative integer value}} + #pragma omp target exit data map(from: i) device (-10u) + #pragma omp target exit data map(from: i) device (3.14) // expected-error {{expression must have integral or unscoped enumeration type, not 'double'}} + foo(); + + return 0; +} diff --git a/test/OpenMP/target_exit_data_if_messages.cpp b/test/OpenMP/target_exit_data_if_messages.cpp new file mode 100644 index 0000000000..cc674e6933 --- /dev/null +++ b/test/OpenMP/target_exit_data_if_messages.cpp @@ -0,0 +1,35 @@ +// 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) { + int i; + #pragma omp target exit data map(from: i) if // expected-error {{expected '(' after 'if'}} + #pragma omp target exit data map(from: i) if ( // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}} + #pragma omp target exit data map(from: i) if () // expected-error {{expected expression}} + #pragma omp target exit data map(from: i) if (argc // expected-error {{expected ')'}} expected-note {{to match this '('}} + #pragma omp target exit data map(from: i) if (argc)) // expected-warning {{extra tokens at the end of '#pragma omp target exit data' are ignored}} + #pragma omp target exit data map(from: i) if (argc > 0 ? argv[1] : argv[2]) + #pragma omp target exit data map(from: i) if (argc + argc) + #pragma omp target exit data map(from: i) if (foobool(argc)), if (true) // expected-error {{directive '#pragma omp target exit data' cannot contain more than one 'if' clause}} + #pragma omp target exit data map(from: i) if (S1) // expected-error {{'S1' does not refer to a value}} + #pragma omp target exit data map(from: i) if (argv[1]=2) // expected-error {{expected ')'}} expected-note {{to match this '('}} + #pragma omp target exit data map(from: i) if(target data : true) // expected-error {{directive name modifier 'target data' is not allowed for '#pragma omp target exit data'}} + #pragma omp target exit data map(from: i) if(target exit data : // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}} + #pragma omp target exit data map(from: i) if(target exit data : // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}} + #pragma omp target exit data map(from: i) if(target exit data : argc // expected-error {{expected ')'}} expected-note {{to match this '('}} + #pragma omp target exit data map(from: i) if(target exit data : argc) + #pragma omp target exit data map(from: i) if(target exit data : argc) if (for:argc) // expected-error {{directive name modifier 'for' is not allowed for '#pragma omp target exit data'}} + #pragma omp target exit data map(from: i) if(target exit data : argc) if (target exit data:argc) // expected-error {{directive '#pragma omp target exit data' cannot contain more than one 'if' clause with 'target exit data' name modifier}} + #pragma omp target exit data map(from: i) if(target exit data : argc) if (argc) // expected-error {{no more 'if' clause is allowed}} expected-note {{previous clause with directive name modifier specified here}} + foo(); + + return 0; +} diff --git a/test/OpenMP/target_exit_data_map_messages.c b/test/OpenMP/target_exit_data_map_messages.c new file mode 100644 index 0000000000..ec4f576aeb --- /dev/null +++ b/test/OpenMP/target_exit_data_map_messages.c @@ -0,0 +1,17 @@ +// RUN: %clang_cc1 -triple x86_64-apple-macos10.7.0 -verify -fopenmp -ferror-limit 100 -o - %s + +int main(int argc, char **argv) { + + int r; + #pragma omp target exit data // expected-error {{expected at least one map clause for '#pragma omp target exit data'}} + + #pragma omp target exit data map(tofrom: r) // expected-error {{map type 'tofrom' is not allowed for '#pragma omp target exit data'}} + + #pragma omp target exit data map(always, from: r) + #pragma omp target exit data map(delete: r) + #pragma omp target exit data map(release: r) + #pragma omp target exit data map(always, alloc: r) // expected-error {{map type 'alloc' is not allowed for '#pragma omp target exit data'}} + #pragma omp target exit data map(to: r) // expected-error {{map type 'to' is not allowed for '#pragma omp target exit data'}} + + return 0; +} diff --git a/tools/libclang/CIndex.cpp b/tools/libclang/CIndex.cpp index 84ba6b0ab1..5462e94f28 100644 --- a/tools/libclang/CIndex.cpp +++ b/tools/libclang/CIndex.cpp @@ -1954,6 +1954,7 @@ public: void VisitOMPTargetDirective(const OMPTargetDirective *D); void VisitOMPTargetDataDirective(const OMPTargetDataDirective *D); void VisitOMPTargetEnterDataDirective(const OMPTargetEnterDataDirective *D); + void VisitOMPTargetExitDataDirective(const OMPTargetExitDataDirective *D); void VisitOMPTeamsDirective(const OMPTeamsDirective *D); void VisitOMPTaskLoopDirective(const OMPTaskLoopDirective *D); void VisitOMPTaskLoopSimdDirective(const OMPTaskLoopSimdDirective *D); @@ -2639,6 +2640,11 @@ void EnqueueVisitor::VisitOMPTargetEnterDataDirective( VisitOMPExecutableDirective(D); } +void EnqueueVisitor::VisitOMPTargetExitDataDirective( + const OMPTargetExitDataDirective *D) { + VisitOMPExecutableDirective(D); +} + void EnqueueVisitor::VisitOMPTeamsDirective(const OMPTeamsDirective *D) { VisitOMPExecutableDirective(D); } @@ -4846,6 +4852,8 @@ CXString clang_getCursorKindSpelling(enum CXCursorKind Kind) { return cxstring::createRef("OMPTargetDataDirective"); case CXCursor_OMPTargetEnterDataDirective: return cxstring::createRef("OMPTargetEnterDataDirective"); + case CXCursor_OMPTargetExitDataDirective: + return cxstring::createRef("OMPTargetExitDataDirective"); case CXCursor_OMPTeamsDirective: return cxstring::createRef("OMPTeamsDirective"); case CXCursor_OMPCancellationPointDirective: diff --git a/tools/libclang/CXCursor.cpp b/tools/libclang/CXCursor.cpp index cf37c7c32e..317ac17309 100644 --- a/tools/libclang/CXCursor.cpp +++ b/tools/libclang/CXCursor.cpp @@ -603,6 +603,9 @@ CXCursor cxcursor::MakeCXCursor(const Stmt *S, const Decl *Parent, case Stmt::OMPTargetEnterDataDirectiveClass: K = CXCursor_OMPTargetEnterDataDirective; break; + case Stmt::OMPTargetExitDataDirectiveClass: + K = CXCursor_OMPTargetExitDataDirective; + break; case Stmt::OMPTeamsDirectiveClass: K = CXCursor_OMPTeamsDirective; break; |