diff --git a/clang/include/clang/AST/OpenMPClause.h b/clang/include/clang/AST/OpenMPClause.h index 79f43fc8ab88..e82a5f09a32d 100644 --- a/clang/include/clang/AST/OpenMPClause.h +++ b/clang/include/clang/AST/OpenMPClause.h @@ -4433,6 +4433,12 @@ class OMPDeviceClause : public OMPClause, public OMPClauseWithPreInit { /// Location of '('. SourceLocation LParenLoc; + /// Device clause modifier. + OpenMPDeviceClauseModifier Modifier = OMPC_DEVICE_unknown; + + /// Location of the modifier. + SourceLocation ModifierLoc; + /// Device number. Stmt *Device = nullptr; @@ -4441,20 +4447,30 @@ class OMPDeviceClause : public OMPClause, public OMPClauseWithPreInit { /// \param E Device number. void setDevice(Expr *E) { Device = E; } + /// Sets modifier. + void setModifier(OpenMPDeviceClauseModifier M) { Modifier = M; } + + /// Setst modifier location. + void setModifierLoc(SourceLocation Loc) { ModifierLoc = Loc; } + public: /// Build 'device' clause. /// + /// \param Modifier Clause modifier. /// \param E Expression associated with this clause. /// \param CaptureRegion Innermost OpenMP region where expressions in this /// clause must be captured. /// \param StartLoc Starting location of the clause. + /// \param ModifierLoc Modifier location. /// \param LParenLoc Location of '('. /// \param EndLoc Ending location of the clause. - OMPDeviceClause(Expr *E, Stmt *HelperE, OpenMPDirectiveKind CaptureRegion, - SourceLocation StartLoc, SourceLocation LParenLoc, + OMPDeviceClause(OpenMPDeviceClauseModifier Modifier, Expr *E, Stmt *HelperE, + OpenMPDirectiveKind CaptureRegion, SourceLocation StartLoc, + SourceLocation LParenLoc, SourceLocation ModifierLoc, SourceLocation EndLoc) : OMPClause(OMPC_device, StartLoc, EndLoc), OMPClauseWithPreInit(this), - LParenLoc(LParenLoc), Device(E) { + LParenLoc(LParenLoc), Modifier(Modifier), ModifierLoc(ModifierLoc), + Device(E) { setPreInitStmt(HelperE, CaptureRegion); } @@ -4475,6 +4491,12 @@ public: /// Return device number. Expr *getDevice() const { return cast(Device); } + /// Gets modifier. + OpenMPDeviceClauseModifier getModifier() const { return Modifier; } + + /// Gets modifier location. + SourceLocation getModifierLoc() const { return ModifierLoc; } + child_range children() { return child_range(&Device, &Device + 1); } const_child_range children() const { diff --git a/clang/include/clang/Basic/OpenMPKinds.def b/clang/include/clang/Basic/OpenMPKinds.def index e6a4aa1d1f58..0488dad6706b 100644 --- a/clang/include/clang/Basic/OpenMPKinds.def +++ b/clang/include/clang/Basic/OpenMPKinds.def @@ -212,6 +212,9 @@ #ifndef OPENMP_DEPOBJ_CLAUSE #define OPENMP_DEPOBJ_CLAUSE(Name) #endif +#ifndef OPENMP_DEVICE_MODIFIER +#define OPENMP_DEVICE_MODIFIER(Name) +#endif // OpenMP clauses. OPENMP_CLAUSE(allocator, OMPAllocatorClause) @@ -366,6 +369,10 @@ OPENMP_SCHEDULE_MODIFIER(monotonic) OPENMP_SCHEDULE_MODIFIER(nonmonotonic) OPENMP_SCHEDULE_MODIFIER(simd) +// Modifiers for 'device' clause. +OPENMP_DEVICE_MODIFIER(ancestor) +OPENMP_DEVICE_MODIFIER(device_num) + // Static attributes for 'defaultmap' clause. OPENMP_DEFAULTMAP_KIND(scalar) OPENMP_DEFAULTMAP_KIND(aggregate) @@ -1091,6 +1098,7 @@ OPENMP_DEPOBJ_CLAUSE(depend) OPENMP_DEPOBJ_CLAUSE(destroy) OPENMP_DEPOBJ_CLAUSE(update) +#undef OPENMP_DEVICE_MODIFIER #undef OPENMP_DEPOBJ_CLAUSE #undef OPENMP_FLUSH_CLAUSE #undef OPENMP_ORDER_KIND diff --git a/clang/include/clang/Basic/OpenMPKinds.h b/clang/include/clang/Basic/OpenMPKinds.h index 43196663c45f..46eeffe999d9 100644 --- a/clang/include/clang/Basic/OpenMPKinds.h +++ b/clang/include/clang/Basic/OpenMPKinds.h @@ -51,6 +51,13 @@ enum OpenMPScheduleClauseModifier { OMPC_SCHEDULE_MODIFIER_last }; +/// OpenMP modifiers for 'device' clause. +enum OpenMPDeviceClauseModifier { +#define OPENMP_DEVICE_MODIFIER(Name) OMPC_DEVICE_##Name, +#include "clang/Basic/OpenMPKinds.def" + OMPC_DEVICE_unknown, +}; + /// OpenMP attributes for 'depend' clause. enum OpenMPDependClauseKind { #define OPENMP_DEPEND_KIND(Name) \ diff --git a/clang/include/clang/Parse/Parser.h b/clang/include/clang/Parse/Parser.h index 85844e2edb07..a9c68a2e231e 100644 --- a/clang/include/clang/Parse/Parser.h +++ b/clang/include/clang/Parse/Parser.h @@ -3029,11 +3029,13 @@ private: /// Parses clause with a single expression and an additional argument /// of a kind \a Kind. /// + /// \param DKind Directive kind. /// \param Kind Kind of current clause. /// \param ParseOnly true to skip the clause's semantic actions and return /// nullptr. /// - OMPClause *ParseOpenMPSingleExprWithArgClause(OpenMPClauseKind Kind, + OMPClause *ParseOpenMPSingleExprWithArgClause(OpenMPDirectiveKind DKind, + OpenMPClauseKind Kind, bool ParseOnly); /// Parses clause without any additional arguments. /// diff --git a/clang/include/clang/Sema/Sema.h b/clang/include/clang/Sema/Sema.h index f29e4f3c227c..28f6705b307f 100644 --- a/clang/include/clang/Sema/Sema.h +++ b/clang/include/clang/Sema/Sema.h @@ -10576,8 +10576,10 @@ public: SourceLocation StartLoc, SourceLocation LParenLoc, SourceLocation EndLoc); /// Called on well-formed 'device' clause. - OMPClause *ActOnOpenMPDeviceClause(Expr *Device, SourceLocation StartLoc, + OMPClause *ActOnOpenMPDeviceClause(OpenMPDeviceClauseModifier Modifier, + Expr *Device, SourceLocation StartLoc, SourceLocation LParenLoc, + SourceLocation ModifierLoc, SourceLocation EndLoc); /// Called on well-formed 'map' clause. OMPClause * diff --git a/clang/lib/AST/OpenMPClause.cpp b/clang/lib/AST/OpenMPClause.cpp index b01aae433763..a95789067056 100644 --- a/clang/lib/AST/OpenMPClause.cpp +++ b/clang/lib/AST/OpenMPClause.cpp @@ -1433,6 +1433,11 @@ void OMPClausePrinter::VisitOMPSIMDClause(OMPSIMDClause *) { OS << "simd"; } void OMPClausePrinter::VisitOMPDeviceClause(OMPDeviceClause *Node) { OS << "device("; + OpenMPDeviceClauseModifier Modifier = Node->getModifier(); + if (Modifier != OMPC_DEVICE_unknown) { + OS << getOpenMPSimpleClauseTypeName(Node->getClauseKind(), Modifier) + << ": "; + } Node->getDevice()->printPretty(OS, nullptr, Policy, 0); OS << ")"; } diff --git a/clang/lib/Basic/OpenMPKinds.cpp b/clang/lib/Basic/OpenMPKinds.cpp index a6e2b9dbf1a1..f2106531d6eb 100644 --- a/clang/lib/Basic/OpenMPKinds.cpp +++ b/clang/lib/Basic/OpenMPKinds.cpp @@ -154,6 +154,11 @@ unsigned clang::getOpenMPSimpleClauseType(OpenMPClauseKind Kind, #define OPENMP_DEPEND_KIND(Name) .Case(#Name, OMPC_DEPEND_##Name) #include "clang/Basic/OpenMPKinds.def" .Default(OMPC_DEPEND_unknown); + case OMPC_device: + return llvm::StringSwitch(Str) +#define OPENMP_DEVICE_MODIFIER(Name) .Case(#Name, OMPC_DEVICE_##Name) +#include "clang/Basic/OpenMPKinds.def" + .Default(OMPC_DEVICE_unknown); case OMPC_unknown: case OMPC_threadprivate: case OMPC_if: @@ -187,7 +192,6 @@ unsigned clang::getOpenMPSimpleClauseType(OpenMPClauseKind Kind, case OMPC_acquire: case OMPC_release: case OMPC_relaxed: - case OMPC_device: case OMPC_threads: case OMPC_simd: case OMPC_num_teams: @@ -380,6 +384,16 @@ const char *clang::getOpenMPSimpleClauseTypeName(OpenMPClauseKind Kind, #include "clang/Basic/OpenMPKinds.def" } llvm_unreachable("Invalid OpenMP 'depend' clause type"); + case OMPC_device: + switch (Type) { + case OMPC_DEVICE_unknown: + return "unknown"; +#define OPENMP_DEVICE_MODIFIER(Name) \ + case OMPC_DEVICE_##Name: \ + return #Name; +#include "clang/Basic/OpenMPKinds.def" + } + llvm_unreachable("Invalid OpenMP 'device' clause modifier"); case OMPC_unknown: case OMPC_threadprivate: case OMPC_if: @@ -413,7 +427,6 @@ const char *clang::getOpenMPSimpleClauseTypeName(OpenMPClauseKind Kind, case OMPC_acquire: case OMPC_release: case OMPC_relaxed: - case OMPC_device: case OMPC_threads: case OMPC_simd: case OMPC_num_teams: diff --git a/clang/lib/CodeGen/CGStmtOpenMP.cpp b/clang/lib/CodeGen/CGStmtOpenMP.cpp index 74fabb1ba04e..2ffe79957ee8 100644 --- a/clang/lib/CodeGen/CGStmtOpenMP.cpp +++ b/clang/lib/CodeGen/CGStmtOpenMP.cpp @@ -4725,8 +4725,11 @@ static void emitCommonOMPTargetDirective(CodeGenFunction &CGF, // Check if we have any device clause associated with the directive. const Expr *Device = nullptr; - if (auto *C = S.getSingleClause()) - Device = C->getDevice(); + if (auto *C = S.getSingleClause()) { + if (C->getModifier() == OMPC_DEVICE_unknown || + C->getModifier() == OMPC_DEVICE_device_num) + Device = C->getDevice(); + } // Check if we have an if clause whose conditional always evaluates to false // or if we do not have any targets specified. If so the target region is not diff --git a/clang/lib/Parse/ParseOpenMP.cpp b/clang/lib/Parse/ParseOpenMP.cpp index 86c5996683bd..84e7b7c9995c 100644 --- a/clang/lib/Parse/ParseOpenMP.cpp +++ b/clang/lib/Parse/ParseOpenMP.cpp @@ -2363,7 +2363,6 @@ OMPClause *Parser::ParseOpenMPClause(OpenMPDirectiveKind DKind, case OMPC_simdlen: case OMPC_collapse: case OMPC_ordered: - case OMPC_device: case OMPC_num_teams: case OMPC_thread_limit: case OMPC_priority: @@ -2379,8 +2378,6 @@ OMPClause *Parser::ParseOpenMPClause(OpenMPDirectiveKind DKind, // Only one safelen clause can appear on a simd directive. // Only one simdlen clause can appear on a simd directive. // Only one collapse clause can appear on a simd directive. - // OpenMP [2.9.1, target data construct, Restrictions] - // At most one device clause can appear on the directive. // OpenMP [2.11.1, task Construct, Restrictions] // At most one if clause can appear on the directive. // At most one final clause can appear on the directive. @@ -2428,6 +2425,7 @@ OMPClause *Parser::ParseOpenMPClause(OpenMPDirectiveKind DKind, Clause = ParseOpenMPSimpleClause(CKind, WrongDirective); break; + case OMPC_device: case OMPC_schedule: case OMPC_dist_schedule: case OMPC_defaultmap: @@ -2435,6 +2433,8 @@ OMPClause *Parser::ParseOpenMPClause(OpenMPDirectiveKind DKind, // Only one schedule clause can appear on a loop directive. // OpenMP 4.5 [2.10.4, Restrictions, p. 106] // At most one defaultmap clause can appear on the directive. + // OpenMP 5.0 [2.12.5, target construct, Restrictions] + // At most one device clause can appear on the directive. if ((getLangOpts().OpenMP < 50 || CKind != OMPC_defaultmap) && !FirstClause) { Diag(Tok, diag::err_omp_more_one_clause) @@ -2443,7 +2443,7 @@ OMPClause *Parser::ParseOpenMPClause(OpenMPDirectiveKind DKind, } LLVM_FALLTHROUGH; case OMPC_if: - Clause = ParseOpenMPSingleExprWithArgClause(CKind, WrongDirective); + Clause = ParseOpenMPSingleExprWithArgClause(DKind, CKind, WrongDirective); break; case OMPC_nowait: case OMPC_untied: @@ -2677,7 +2677,11 @@ OMPClause *Parser::ParseOpenMPClause(OpenMPClauseKind Kind, bool ParseOnly) { /// defaultmap: /// 'defaultmap' '(' modifier ':' kind ')' /// -OMPClause *Parser::ParseOpenMPSingleExprWithArgClause(OpenMPClauseKind Kind, +/// device-clause: +/// 'device' '(' [ device-modifier ':' ] expression ')' +/// +OMPClause *Parser::ParseOpenMPSingleExprWithArgClause(OpenMPDirectiveKind DKind, + OpenMPClauseKind Kind, bool ParseOnly) { SourceLocation Loc = ConsumeToken(); SourceLocation DelimLoc; @@ -2771,6 +2775,21 @@ OMPClause *Parser::ParseOpenMPSingleExprWithArgClause(OpenMPClauseKind Kind, if (Tok.isNot(tok::r_paren) && Tok.isNot(tok::comma) && Tok.isNot(tok::annot_pragma_openmp_end)) ConsumeAnyToken(); + } else if (Kind == OMPC_device) { + // Only target executable directives support extended device construct. + if (isOpenMPTargetExecutionDirective(DKind) && getLangOpts().OpenMP >= 50 && + NextToken().is(tok::colon)) { + // Parse optional ':' + Arg.push_back(getOpenMPSimpleClauseType( + Kind, Tok.isAnnotation() ? "" : PP.getSpelling(Tok))); + KLoc.push_back(Tok.getLocation()); + ConsumeAnyToken(); + // Parse ':' + ConsumeAnyToken(); + } else { + Arg.push_back(OMPC_DEVICE_unknown); + KLoc.emplace_back(); + } } else { assert(Kind == OMPC_if); KLoc.push_back(Tok.getLocation()); @@ -2793,7 +2812,7 @@ OMPClause *Parser::ParseOpenMPSingleExprWithArgClause(OpenMPClauseKind Kind, bool NeedAnExpression = (Kind == OMPC_schedule && DelimLoc.isValid()) || (Kind == OMPC_dist_schedule && DelimLoc.isValid()) || - Kind == OMPC_if; + Kind == OMPC_if || Kind == OMPC_device; if (NeedAnExpression) { SourceLocation ELoc = Tok.getLocation(); ExprResult LHS(ParseCastExpression(AnyCastExpr, false, NotTypeCast)); diff --git a/clang/lib/Sema/SemaOpenMP.cpp b/clang/lib/Sema/SemaOpenMP.cpp index 8c2f8b2a942f..7d0821829daa 100644 --- a/clang/lib/Sema/SemaOpenMP.cpp +++ b/clang/lib/Sema/SemaOpenMP.cpp @@ -11043,9 +11043,6 @@ OMPClause *Sema::ActOnOpenMPSingleExprClause(OpenMPClauseKind Kind, Expr *Expr, case OMPC_ordered: Res = ActOnOpenMPOrderedClause(StartLoc, EndLoc, LParenLoc, Expr); break; - case OMPC_device: - Res = ActOnOpenMPDeviceClause(Expr, StartLoc, LParenLoc, EndLoc); - break; case OMPC_num_teams: Res = ActOnOpenMPNumTeamsClause(Expr, StartLoc, LParenLoc, EndLoc); break; @@ -11070,6 +11067,7 @@ OMPClause *Sema::ActOnOpenMPSingleExprClause(OpenMPClauseKind Kind, Expr *Expr, case OMPC_detach: Res = ActOnOpenMPDetachClause(Expr, StartLoc, LParenLoc, EndLoc); break; + case OMPC_device: case OMPC_if: case OMPC_default: case OMPC_proc_bind: @@ -12440,6 +12438,12 @@ OMPClause *Sema::ActOnOpenMPSingleExprWithArgClause( StartLoc, LParenLoc, ArgumentLoc[Modifier], ArgumentLoc[DefaultmapKind], EndLoc); break; + case OMPC_device: + assert(Argument.size() == 1 && ArgumentLoc.size() == 1); + Res = ActOnOpenMPDeviceClause( + static_cast(Argument.back()), Expr, + StartLoc, LParenLoc, ArgumentLoc.back(), EndLoc); + break; case OMPC_final: case OMPC_num_threads: case OMPC_safelen: @@ -12477,7 +12481,6 @@ OMPClause *Sema::ActOnOpenMPSingleExprWithArgClause( case OMPC_release: case OMPC_relaxed: case OMPC_depend: - case OMPC_device: case OMPC_threads: case OMPC_simd: case OMPC_map: @@ -15641,16 +15644,32 @@ Sema::ActOnOpenMPDependClause(OpenMPDependClauseKind DepKind, return C; } -OMPClause *Sema::ActOnOpenMPDeviceClause(Expr *Device, SourceLocation StartLoc, +OMPClause *Sema::ActOnOpenMPDeviceClause(OpenMPDeviceClauseModifier Modifier, + Expr *Device, SourceLocation StartLoc, SourceLocation LParenLoc, + SourceLocation ModifierLoc, SourceLocation EndLoc) { + assert((ModifierLoc.isInvalid() || LangOpts.OpenMP >= 50) && + "Unexpected device modifier in OpenMP < 50."); + + bool ErrorFound = false; + if (ModifierLoc.isValid() && Modifier == OMPC_DEVICE_unknown) { + std::string Values = + getListOfPossibleValues(OMPC_device, /*First=*/0, OMPC_DEVICE_unknown); + Diag(ModifierLoc, diag::err_omp_unexpected_clause_value) + << Values << getOpenMPClauseName(OMPC_device); + ErrorFound = true; + } + Expr *ValExpr = Device; Stmt *HelperValStmt = nullptr; // OpenMP [2.9.1, Restrictions] // The device expression must evaluate to a non-negative integer value. - if (!isNonNegativeIntegerValue(ValExpr, *this, OMPC_device, - /*StrictlyPositive=*/false)) + ErrorFound = !isNonNegativeIntegerValue(ValExpr, *this, OMPC_device, + /*StrictlyPositive=*/false) || + ErrorFound; + if (ErrorFound) return nullptr; OpenMPDirectiveKind DKind = DSAStack->getCurrentDirective(); @@ -15663,8 +15682,9 @@ OMPClause *Sema::ActOnOpenMPDeviceClause(Expr *Device, SourceLocation StartLoc, HelperValStmt = buildPreInits(Context, Captures); } - return new (Context) OMPDeviceClause(ValExpr, HelperValStmt, CaptureRegion, - StartLoc, LParenLoc, EndLoc); + return new (Context) + OMPDeviceClause(Modifier, ValExpr, HelperValStmt, CaptureRegion, StartLoc, + LParenLoc, ModifierLoc, EndLoc); } static bool checkTypeMappable(SourceLocation SL, SourceRange SR, Sema &SemaRef, diff --git a/clang/lib/Sema/TreeTransform.h b/clang/lib/Sema/TreeTransform.h index c7e90ad982ef..bc1a977dbd5d 100644 --- a/clang/lib/Sema/TreeTransform.h +++ b/clang/lib/Sema/TreeTransform.h @@ -1847,11 +1847,13 @@ public: /// /// By default, performs semantic analysis to build the new statement. /// Subclasses may override this routine to provide different behavior. - OMPClause *RebuildOMPDeviceClause(Expr *Device, SourceLocation StartLoc, + OMPClause *RebuildOMPDeviceClause(OpenMPDeviceClauseModifier Modifier, + Expr *Device, SourceLocation StartLoc, SourceLocation LParenLoc, + SourceLocation ModifierLoc, SourceLocation EndLoc) { - return getSema().ActOnOpenMPDeviceClause(Device, StartLoc, LParenLoc, - EndLoc); + return getSema().ActOnOpenMPDeviceClause(Modifier, Device, StartLoc, + LParenLoc, ModifierLoc, EndLoc); } /// Build a new OpenMP 'map' clause. @@ -9256,8 +9258,9 @@ TreeTransform::TransformOMPDeviceClause(OMPDeviceClause *C) { ExprResult E = getDerived().TransformExpr(C->getDevice()); if (E.isInvalid()) return nullptr; - return getDerived().RebuildOMPDeviceClause(E.get(), C->getBeginLoc(), - C->getLParenLoc(), C->getEndLoc()); + return getDerived().RebuildOMPDeviceClause( + C->getModifier(), E.get(), C->getBeginLoc(), C->getLParenLoc(), + C->getModifierLoc(), C->getEndLoc()); } template diff --git a/clang/lib/Serialization/ASTReader.cpp b/clang/lib/Serialization/ASTReader.cpp index 9f6bfebeabcb..dd0fa9f70daf 100644 --- a/clang/lib/Serialization/ASTReader.cpp +++ b/clang/lib/Serialization/ASTReader.cpp @@ -12294,7 +12294,9 @@ void OMPClauseReader::VisitOMPDependClause(OMPDependClause *C) { void OMPClauseReader::VisitOMPDeviceClause(OMPDeviceClause *C) { VisitOMPClauseWithPreInit(C); + C->setModifier(Record.readEnum()); C->setDevice(Record.readSubExpr()); + C->setModifierLoc(Record.readSourceLocation()); C->setLParenLoc(Record.readSourceLocation()); } diff --git a/clang/lib/Serialization/ASTWriter.cpp b/clang/lib/Serialization/ASTWriter.cpp index b549fe9df016..c96e46543dba 100644 --- a/clang/lib/Serialization/ASTWriter.cpp +++ b/clang/lib/Serialization/ASTWriter.cpp @@ -6376,7 +6376,9 @@ void OMPClauseWriter::VisitOMPDependClause(OMPDependClause *C) { void OMPClauseWriter::VisitOMPDeviceClause(OMPDeviceClause *C) { VisitOMPClauseWithPreInit(C); + Record.writeEnum(C->getModifier()); Record.AddStmt(C->getDevice()); + Record.AddSourceLocation(C->getModifierLoc()); Record.AddSourceLocation(C->getLParenLoc()); } diff --git a/clang/test/OpenMP/target_ast_print.cpp b/clang/test/OpenMP/target_ast_print.cpp index a36b44bd406e..0352ae63ff81 100644 --- a/clang/test/OpenMP/target_ast_print.cpp +++ b/clang/test/OpenMP/target_ast_print.cpp @@ -353,11 +353,11 @@ int a; template T tmain(T argc, T *argv) { T i, j, a[20], always, close; -#pragma omp target +#pragma omp target device(argc) foo(); -#pragma omp target if (target:argc > 0) +#pragma omp target if (target:argc > 0) device(device_num: C) foo(); -#pragma omp target if (C) +#pragma omp target if (C) device(ancestor: argc) foo(); #pragma omp target map(i) foo(); @@ -458,11 +458,11 @@ T tmain(T argc, T *argv) { // OMP5: template T tmain(T argc, T *argv) { // OMP5-NEXT: T i, j, a[20] -// OMP5-NEXT: #pragma omp target{{$}} +// OMP5-NEXT: #pragma omp target device(argc){{$}} // OMP5-NEXT: foo(); -// OMP5-NEXT: #pragma omp target if(target: argc > 0) +// OMP5-NEXT: #pragma omp target if(target: argc > 0) device(device_num: C) // OMP5-NEXT: foo() -// OMP5-NEXT: #pragma omp target if(C) +// OMP5-NEXT: #pragma omp target if(C) device(ancestor: argc) // OMP5-NEXT: foo() // OMP5-NEXT: #pragma omp target map(tofrom: i) // OMP5-NEXT: foo() @@ -650,11 +650,11 @@ T tmain(T argc, T *argv) { // OMP5-NEXT: foo() // OMP5: template<> char tmain(char argc, char *argv) { // OMP5-NEXT: char i, j, a[20] -// OMP5-NEXT: #pragma omp target +// OMP5-NEXT: #pragma omp target device(argc) // OMP5-NEXT: foo(); -// OMP5-NEXT: #pragma omp target if(target: argc > 0) +// OMP5-NEXT: #pragma omp target if(target: argc > 0) device(device_num: 1) // OMP5-NEXT: foo() -// OMP5-NEXT: #pragma omp target if(1) +// OMP5-NEXT: #pragma omp target if(1) device(ancestor: argc) // OMP5-NEXT: foo() // OMP5-NEXT: #pragma omp target map(tofrom: i) // OMP5-NEXT: foo() diff --git a/clang/test/OpenMP/target_data_device_messages.cpp b/clang/test/OpenMP/target_data_device_messages.cpp index 873402d372dd..7610e8413158 100644 --- a/clang/test/OpenMP/target_data_device_messages.cpp +++ b/clang/test/OpenMP/target_data_device_messages.cpp @@ -1,6 +1,6 @@ -// RUN: %clang_cc1 -triple x86_64-apple-macos10.7.0 -verify -fopenmp -ferror-limit 100 -o - %s -Wuninitialized +// RUN: %clang_cc1 -triple x86_64-apple-macos10.7.0 -verify -fopenmp -fopenmp-version=50 -ferror-limit 100 -o - %s -Wuninitialized -// RUN: %clang_cc1 -triple x86_64-apple-macos10.7.0 -verify -fopenmp-simd -ferror-limit 100 -o - %s -Wuninitialized +// RUN: %clang_cc1 -triple x86_64-apple-macos10.7.0 -verify -fopenmp-simd -fopenmp-version=50 -ferror-limit 100 -o - %s -Wuninitialized void foo() { } @@ -24,6 +24,7 @@ int main(int argc, char **argv) { #pragma omp target data map(to: a) device (S1) // expected-error {{'S1' does not refer to a value}} #pragma omp target data map(to: a) device (-2) // expected-error {{argument to 'device' clause must be a non-negative integer value}} #pragma omp target data map(to: a) device (-10u) + #pragma omp target data map(to: a) device (ancestor: -10u) // expected-error {{use of undeclared identifier 'ancestor'}} expected-error {{expected ')'}} expected-note {{to match this '('}} #pragma omp target data map(to: a) device (3.14) // expected-error {{expression must have integral or unscoped enumeration type, not 'double'}} foo(); diff --git a/clang/test/OpenMP/target_device_messages.cpp b/clang/test/OpenMP/target_device_messages.cpp index 8a583e9f3bd7..1ca4a88ba8c5 100644 --- a/clang/test/OpenMP/target_device_messages.cpp +++ b/clang/test/OpenMP/target_device_messages.cpp @@ -1,6 +1,8 @@ -// RUN: %clang_cc1 -triple x86_64-apple-macos10.7.0 -verify -fopenmp -ferror-limit 100 -o - %s -Wuninitialized +// RUN: %clang_cc1 -triple x86_64-apple-macos10.7.0 -verify=expected,omp45 -fopenmp -fopenmp-version=45 -ferror-limit 100 -o - %s -Wuninitialized +// RUN: %clang_cc1 -triple x86_64-apple-macos10.7.0 -verify=expected,omp50 -fopenmp -fopenmp-version=50 -ferror-limit 100 -o - %s -Wuninitialized -// RUN: %clang_cc1 -triple x86_64-apple-macos10.7.0 -verify -fopenmp-simd -ferror-limit 100 -o - %s -Wuninitialized +// RUN: %clang_cc1 -triple x86_64-apple-macos10.7.0 -verify=expected,omp45 -fopenmp-simd -fopenmp-version=45 -ferror-limit 100 -o - %s -Wuninitialized +// RUN: %clang_cc1 -triple x86_64-apple-macos10.7.0 -verify=expected,omp50 -fopenmp-simd -fopenmp-version=50 -ferror-limit 100 -o - %s -Wuninitialized void foo() { } @@ -21,11 +23,13 @@ int main(int argc, char **argv) { foo(); #pragma omp target device (argc // expected-error {{expected ')'}} expected-note {{to match this '('}} foo(); + #pragma omp target device (argc: // expected-error {{expected ')'}} expected-note {{to match this '('}} omp50-error {{expected expression}} + foo(); #pragma omp target device (z-argc)) // expected-warning {{extra tokens at the end of '#pragma omp target' are ignored}} foo(); - #pragma omp target device (argc > 0 ? argv[1] : argv[2]) // expected-error {{expression must have integral or unscoped enumeration type, not 'char *'}} + #pragma omp target device (device_num : argc > 0 ? argv[1] : argv[2]) // omp45-error {{use of undeclared identifier 'device_num'}} omp45-error {{expected ')'}} omp45-note {{to match this '('}} omp50-error {{expression must have integral or unscoped enumeration type, not 'char *'}} foo(); - #pragma omp target device (argc + argc) + #pragma omp target device (argc: argc + argc) // omp45-error {{expected ')'}} omp45-note {{to match this '('}} omp50-error {{expected 'ancestor' or 'device_num' in OpenMP clause 'device'}} foo(); #pragma omp target device (argc), device (argc+1) // expected-error {{directive '#pragma omp target' cannot contain more than one 'device' clause}} foo(); @@ -37,6 +41,8 @@ int main(int argc, char **argv) { foo(); #pragma omp target device (3.14) // expected-error {{expression must have integral or unscoped enumeration type, not 'double'}} foo(); + #pragma omp target device (ancestor) // expected-error {{use of undeclared identifier 'ancestor'}} + foo(); return 0; } diff --git a/clang/test/OpenMP/target_enter_data_device_messages.cpp b/clang/test/OpenMP/target_enter_data_device_messages.cpp index a233a47152dc..7f5356b2fc5f 100644 --- a/clang/test/OpenMP/target_enter_data_device_messages.cpp +++ b/clang/test/OpenMP/target_enter_data_device_messages.cpp @@ -1,6 +1,6 @@ -// RUN: %clang_cc1 -triple x86_64-apple-macos10.7.0 -verify -fopenmp -ferror-limit 100 -o - %s -Wuninitialized +// RUN: %clang_cc1 -triple x86_64-apple-macos10.7.0 -verify -fopenmp -fopenmp-version=50 -ferror-limit 100 -o - %s -Wuninitialized -// RUN: %clang_cc1 -triple x86_64-apple-macos10.7.0 -verify -fopenmp-simd -ferror-limit 100 -o - %s -Wuninitialized +// RUN: %clang_cc1 -triple x86_64-apple-macos10.7.0 -verify -fopenmp-simd -fopenmp-version=50 -ferror-limit 100 -o - %s -Wuninitialized void foo() { } @@ -24,6 +24,7 @@ int main(int argc, char **argv) { #pragma omp target enter data map(to: i) device (S1) // expected-error {{'S1' does not refer to a value}} #pragma omp target enter data map(to: i) device (-2) // expected-error {{argument to 'device' clause must be a non-negative integer value}} #pragma omp target enter data map(to: i) device (-10u) + #pragma omp target enter data map(to: i) device (device_num: -10u) // expected-error {{use of undeclared identifier 'device_num'}} expected-error {{expected ')'}} expected-note {{to match this '('}} #pragma omp target enter data map(to: i) device (3.14) // expected-error {{expression must have integral or unscoped enumeration type, not 'double'}} foo(); diff --git a/clang/test/OpenMP/target_exit_data_device_messages.cpp b/clang/test/OpenMP/target_exit_data_device_messages.cpp index 035bf6f76a61..f896882f6f1b 100644 --- a/clang/test/OpenMP/target_exit_data_device_messages.cpp +++ b/clang/test/OpenMP/target_exit_data_device_messages.cpp @@ -1,6 +1,6 @@ -// RUN: %clang_cc1 -triple x86_64-apple-macos10.7.0 -verify -fopenmp -ferror-limit 100 -o - %s -Wuninitialized +// RUN: %clang_cc1 -triple x86_64-apple-macos10.7.0 -verify -fopenmp -fopenmp-version=50 -ferror-limit 100 -o - %s -Wuninitialized -// RUN: %clang_cc1 -triple x86_64-apple-macos10.7.0 -verify -fopenmp-simd -ferror-limit 100 -o - %s -Wuninitialized +// RUN: %clang_cc1 -triple x86_64-apple-macos10.7.0 -verify -fopenmp-simd -fopenmp-version=50 -ferror-limit 100 -o - %s -Wuninitialized void foo() { } @@ -24,6 +24,7 @@ int main(int argc, char **argv) { #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 + z) + #pragma omp target exit data map(from: i) device (ancestor: -10u + z) // expected-error {{use of undeclared identifier 'ancestor'}} expected-error {{expected ')'}} expected-note {{to match this '('}} #pragma omp target exit data map(from: i) device (3.14) // expected-error {{expression must have integral or unscoped enumeration type, not 'double'}} foo(); diff --git a/clang/test/OpenMP/target_update_device_messages.cpp b/clang/test/OpenMP/target_update_device_messages.cpp index 8b4aefaf580d..fa1c356f9c40 100644 --- a/clang/test/OpenMP/target_update_device_messages.cpp +++ b/clang/test/OpenMP/target_update_device_messages.cpp @@ -1,6 +1,6 @@ -// RUN: %clang_cc1 -verify -fopenmp -ferror-limit 100 %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=50 -ferror-limit 100 %s -Wuninitialized -// RUN: %clang_cc1 -verify -fopenmp-simd -ferror-limit 100 %s -Wuninitialized +// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=50 -ferror-limit 100 %s -Wuninitialized void foo() { } @@ -21,6 +21,7 @@ int tmain(T argc, S **argv) { #pragma omp target update to(i) device (argc)) // expected-warning {{extra tokens at the end of '#pragma omp target update' are ignored}} #pragma omp target update from(i) device (argc > 0 ? argv[1] : argv[2]) // expected-error {{expression must have integral or unscoped enumeration type, not 'char *'}} #pragma omp target update from(i) device (argc + z) +#pragma omp target update from(i) device (device_num: argc + z) // expected-error {{use of undeclared identifier 'device_num'}} expected-error {{expected ')'}} expected-note {{to match this '('}} #pragma omp target update from(i) device (argc), device (argc+1) // expected-error {{directive '#pragma omp target update' cannot contain more than one 'device' clause}} #pragma omp target update from(i) device (S1) // expected-error {{'S1' does not refer to a value}} #pragma omp target update from(i) device (3.14) // expected-error 2 {{expression must have integral or unscoped enumeration type, not 'double'}}