Index: include/clang/AST/OpenMPClause.h =================================================================== --- include/clang/AST/OpenMPClause.h +++ include/clang/AST/OpenMPClause.h @@ -3471,6 +3471,67 @@ return child_range(child_iterator(), child_iterator()); } }; + +/// \brief This represents clause 'to' in the '#pragma omp ...' +/// directives. +/// +/// \code +/// #pragma omp target update to(a,b) +/// \endcode +/// In this example directive '#pragma omp target update' has clause 'to' +/// with the variables 'a' and 'b'. +/// +class OMPToClause final : public OMPVarListClause, + private llvm::TrailingObjects { + friend TrailingObjects; + friend OMPVarListClause; + friend class OMPClauseReader; + + /// \brief Build clause with number of variables \a N. + /// + /// \param StartLoc Starting location of the clause. + /// \param EndLoc Ending location of the clause. + /// \param N Number of the variables in the clause. + /// + explicit OMPToClause(SourceLocation StartLoc, SourceLocation LParenLoc, + SourceLocation EndLoc, unsigned N) + : OMPVarListClause(OMPC_to, StartLoc, LParenLoc, EndLoc, N) {} + /// \brief Build an empty clause. + /// + /// \param N Number of variables. + /// + explicit OMPToClause(unsigned N) + : OMPVarListClause(OMPC_to, SourceLocation(), SourceLocation(), + SourceLocation(), N) {} + +public: + /// \brief Creates clause with a list of variables \a VL. + /// + /// \param C AST context. + /// \brief StartLoc Starting location of the clause. + /// \brief EndLoc Ending location of the clause. + /// \param VL List of references to the variables. + /// + static OMPToClause *Create(const ASTContext &C, SourceLocation StartLoc, + SourceLocation LParenLoc, SourceLocation EndLoc, + ArrayRef VL); + /// \brief Creates an empty clause with the place for \a N variables. + /// + /// \param C AST context. + /// \param N The number of variables. + /// + static OMPToClause *CreateEmpty(const ASTContext &C, unsigned N); + + static bool classof(const OMPClause *T) { + return T->getClauseKind() == OMPC_to; + } + + child_range children() { + return child_range(reinterpret_cast(varlist_begin()), + reinterpret_cast(varlist_end())); + } +}; + } // end namespace clang #endif // LLVM_CLANG_AST_OPENMPCLAUSE_H Index: include/clang/AST/RecursiveASTVisitor.h =================================================================== --- include/clang/AST/RecursiveASTVisitor.h +++ include/clang/AST/RecursiveASTVisitor.h @@ -2874,6 +2874,12 @@ return true; } +template +bool RecursiveASTVisitor::VisitOMPToClause(OMPToClause *C) { + TRY_TO(VisitOMPClauseList(C)); + return true; +} + // FIXME: look at the following tricky-seeming exprs to see if we // need to recurse on anything. These are ones that have methods // returning decls or qualtypes or nestednamespecifier -- though I'm Index: include/clang/Basic/DiagnosticSemaKinds.td =================================================================== --- include/clang/Basic/DiagnosticSemaKinds.td +++ include/clang/Basic/DiagnosticSemaKinds.td @@ -7835,8 +7835,8 @@ "expected variable name%select{|, data member of current class}0, array element or array section">; def err_omp_expected_named_var_member_or_array_expression: Error< "expected expression containing only member accesses and/or array sections based on named variables">; -def err_omp_bit_fields_forbidden_in_map_clause : Error< - "bit fields cannot be used to specify storage in a map clause">; +def err_omp_bit_fields_forbidden_in_clause : Error< + "bit fields cannot be used to specify storage in a '%0' clause">; def err_array_section_does_not_specify_contiguous_storage : Error< "array section does not specify contiguous storage">; def err_omp_union_type_not_allowed : Error< @@ -7964,6 +7964,8 @@ "arguments of OpenMP clause 'reduction' with bitwise operators cannot be of floating type">; def err_omp_once_referenced : Error< "variable can appear only once in OpenMP '%0' clause">; +def err_omp_once_referenced_in_target_update : Error< + "variable can appear only once in OpenMP 'target update' construct">; def note_omp_referenced : Note< "previously referenced here">; def err_omp_reduction_in_task : Error< @@ -8101,8 +8103,8 @@ "mappable type cannot be polymorphic">; def note_omp_static_member_in_target : Note< "mappable type cannot contain static members">; -def err_omp_threadprivate_in_map : Error< - "threadprivate variables are not allowed in map clause">; +def err_omp_threadprivate_in_clause : Error< + "threadprivate variables are not allowed in '%0' clause">; def err_omp_wrong_ordered_loop_count : Error< "the parameter of the 'ordered' clause must be greater than or equal to the parameter of the 'collapse' clause">; def note_collapse_loop_count : Note< @@ -8149,6 +8151,8 @@ "'ordered' clause with a parameter can not be specified in '#pragma omp %0' directive">; def err_omp_variable_in_map_and_dsa : Error< "%0 variable cannot be in a map clause in '#pragma omp %1' directive">; +def err_omp_at_least_one_motion_clause_required : Error< + "expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'">; } // end of OpenMP category let CategoryName = "Related Result Type Issue" in { Index: include/clang/Basic/OpenMPKinds.def =================================================================== --- include/clang/Basic/OpenMPKinds.def +++ include/clang/Basic/OpenMPKinds.def @@ -204,6 +204,7 @@ OPENMP_CLAUSE(hint, OMPHintClause) OPENMP_CLAUSE(dist_schedule, OMPDistScheduleClause) OPENMP_CLAUSE(defaultmap, OMPDefaultmapClause) +OPENMP_CLAUSE(to, OMPToClause) // Clauses allowed for OpenMP directive 'parallel'. OPENMP_PARALLEL_CLAUSE(if) Index: include/clang/Sema/Sema.h =================================================================== --- include/clang/Sema/Sema.h +++ include/clang/Sema/Sema.h @@ -8317,6 +8317,11 @@ OpenMPDefaultmapClauseModifier M, OpenMPDefaultmapClauseKind Kind, SourceLocation StartLoc, SourceLocation LParenLoc, SourceLocation MLoc, SourceLocation KindLoc, SourceLocation EndLoc); + /// \brief Called on well-formed 'to' clause. + OMPClause *ActOnOpenMPToClause(ArrayRef VarList, + SourceLocation StartLoc, + SourceLocation LParenLoc, + SourceLocation EndLoc); /// \brief The kind of conversion being performed. enum CheckedConversionKind { Index: lib/AST/OpenMPClause.cpp =================================================================== --- lib/AST/OpenMPClause.cpp +++ lib/AST/OpenMPClause.cpp @@ -85,6 +85,7 @@ case OMPC_num_tasks: case OMPC_hint: case OMPC_defaultmap: + case OMPC_to: case OMPC_unknown: break; } @@ -145,6 +146,7 @@ case OMPC_num_tasks: case OMPC_hint: case OMPC_defaultmap: + case OMPC_to: case OMPC_unknown: break; } @@ -550,3 +552,18 @@ void *Mem = C.Allocate(totalSizeToAlloc(N)); return new (Mem) OMPMapClause(N); } + +OMPToClause *OMPToClause::Create(const ASTContext &C, SourceLocation StartLoc, + SourceLocation LParenLoc, SourceLocation EndLoc, + ArrayRef VL) { + void *Mem = C.Allocate(totalSizeToAlloc(VL.size())); + OMPToClause *Clause = new (Mem) OMPToClause(StartLoc, LParenLoc, EndLoc, + VL.size()); + Clause->setVarRefs(VL); + return Clause; +} + +OMPToClause *OMPToClause::CreateEmpty(const ASTContext &C, unsigned N) { + void *Mem = C.Allocate(totalSizeToAlloc(N)); + return new (Mem) OMPToClause(N); +} Index: lib/AST/StmtPrinter.cpp =================================================================== --- lib/AST/StmtPrinter.cpp +++ lib/AST/StmtPrinter.cpp @@ -912,6 +912,14 @@ } } +void OMPClausePrinter::VisitOMPToClause(OMPToClause *Node) { + if (!Node->varlist_empty()) { + OS << "to"; + VisitOMPClauseList(Node, '('); + OS << ")"; + } +} + void OMPClausePrinter::VisitOMPDistScheduleClause(OMPDistScheduleClause *Node) { OS << "dist_schedule(" << getOpenMPSimpleClauseTypeName( OMPC_dist_schedule, Node->getDistScheduleKind()); Index: lib/AST/StmtProfile.cpp =================================================================== --- lib/AST/StmtProfile.cpp +++ lib/AST/StmtProfile.cpp @@ -491,6 +491,9 @@ void OMPClauseProfiler::VisitOMPHintClause(const OMPHintClause *C) { Profiler->VisitStmt(C->getHint()); } +void OMPClauseProfiler::VisitOMPToClause(const OMPToClause *C) { + VisitOMPClauseList(C); +} } void Index: lib/Basic/OpenMPKinds.cpp =================================================================== --- lib/Basic/OpenMPKinds.cpp +++ lib/Basic/OpenMPKinds.cpp @@ -158,6 +158,7 @@ case OMPC_nogroup: case OMPC_num_tasks: case OMPC_hint: + case OMPC_to: break; } llvm_unreachable("Invalid OpenMP simple clause kind"); @@ -292,6 +293,7 @@ case OMPC_nogroup: case OMPC_num_tasks: case OMPC_hint: + case OMPC_to: break; } llvm_unreachable("Invalid OpenMP simple clause kind"); Index: lib/CodeGen/CGStmtOpenMP.cpp =================================================================== --- lib/CodeGen/CGStmtOpenMP.cpp +++ lib/CodeGen/CGStmtOpenMP.cpp @@ -2971,6 +2971,7 @@ case OMPC_hint: case OMPC_dist_schedule: case OMPC_defaultmap: + case OMPC_to: llvm_unreachable("Clause is not allowed in 'omp atomic'."); } } Index: lib/Parse/ParseOpenMP.cpp =================================================================== --- lib/Parse/ParseOpenMP.cpp +++ lib/Parse/ParseOpenMP.cpp @@ -769,7 +769,7 @@ /// update-clause | capture-clause | seq_cst-clause | device-clause | /// simdlen-clause | threads-clause | simd-clause | num_teams-clause | /// thread_limit-clause | priority-clause | grainsize-clause | -/// nogroup-clause | num_tasks-clause | hint-clause +/// nogroup-clause | num_tasks-clause | hint-clause | to-clause /// OMPClause *Parser::ParseOpenMPClause(OpenMPDirectiveKind DKind, OpenMPClauseKind CKind, bool FirstClause) { @@ -893,6 +893,7 @@ case OMPC_flush: case OMPC_depend: case OMPC_map: + case OMPC_to: Clause = ParseOpenMPVarListClause(DKind, CKind); break; case OMPC_unknown: @@ -1210,8 +1211,7 @@ TemplateKWLoc, ReductionId); } -/// \brief Parsing of OpenMP clause 'private', 'firstprivate', 'lastprivate', -/// 'shared', 'copyin', 'copyprivate', 'flush' or 'reduction'. +/// \brief Parsing of OpenMP clause /// /// private-clause: /// 'private' '(' list ')' @@ -1236,6 +1236,8 @@ /// map-clause: /// 'map' '(' [ [ always , ] /// to | from | tofrom | alloc | release | delete ':' ] list ')'; +/// to-clause: +/// 'to' '(' list ')' /// /// For 'linear' clause linear-list may have the following forms: /// list Index: lib/Sema/SemaOpenMP.cpp =================================================================== --- lib/Sema/SemaOpenMP.cpp +++ lib/Sema/SemaOpenMP.cpp @@ -6304,6 +6304,7 @@ case OMPC_nogroup: case OMPC_dist_schedule: case OMPC_defaultmap: + case OMPC_to: case OMPC_unknown: llvm_unreachable("Clause is not allowed."); } @@ -6589,6 +6590,7 @@ case OMPC_hint: case OMPC_dist_schedule: case OMPC_defaultmap: + case OMPC_to: case OMPC_unknown: llvm_unreachable("Clause is not allowed."); } @@ -6739,6 +6741,7 @@ case OMPC_nogroup: case OMPC_num_tasks: case OMPC_hint: + case OMPC_to: case OMPC_unknown: llvm_unreachable("Clause is not allowed."); } @@ -6922,6 +6925,7 @@ case OMPC_hint: case OMPC_dist_schedule: case OMPC_defaultmap: + case OMPC_to: case OMPC_unknown: llvm_unreachable("Clause is not allowed."); } @@ -7036,6 +7040,9 @@ DepLinMapLoc, ColonLoc, VarList, StartLoc, LParenLoc, EndLoc); break; + case OMPC_to: + Res = ActOnOpenMPToClause(VarList, StartLoc, LParenLoc, EndLoc); + break; case OMPC_if: case OMPC_final: case OMPC_num_threads: @@ -7216,8 +7223,9 @@ // 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) { - if(DSAStack->checkMapInfoForVar(VD, /* CurrentRegionOnly = */ true, - [&](Expr *RE) -> bool {return true;})) { + if (DSAStack->checkMapInfoForVar( + VD, /* CurrentRegionOnly = */ true, + [&](Expr *RE) -> bool {return true;})) { Diag(ELoc, diag::err_omp_variable_in_map_and_dsa) << getOpenMPClauseName(OMPC_private) << getOpenMPDirectiveName(DSAStack->getCurrentDirective()); @@ -7461,8 +7469,9 @@ // A list item cannot appear in both a map clause and a data-sharing // attribute clause on the same construct if (CurrDir == OMPD_target) { - if(DSAStack->checkMapInfoForVar(VD, /* CurrentRegionOnly = */ true, - [&](Expr *RE) -> bool {return true;})) { + if (DSAStack->checkMapInfoForVar( + VD, /* CurrentRegionOnly = */ true, + [&](Expr *RE) -> bool {return true;})) { Diag(ELoc, diag::err_omp_variable_in_map_and_dsa) << getOpenMPClauseName(OMPC_firstprivate) << getOpenMPDirectiveName(DSAStack->getCurrentDirective()); @@ -9322,7 +9331,8 @@ // Return the expression of the base of the map clause or null if it cannot // be determined and do all the necessary checks to see if the expression is // valid as a standalone map clause expression. -static Expr *CheckMapClauseExpressionBase(Sema &SemaRef, Expr *E) { +static Expr *CheckMapClauseExpressionBase(Sema &SemaRef, Expr *E, + OpenMPClauseKind CKind) { SourceLocation ELoc = E->getExprLoc(); SourceRange ERange = E->getSourceRange(); @@ -9404,8 +9414,8 @@ // A bit-field cannot appear in a map clause. // if (FD->isBitField()) { - SemaRef.Diag(ELoc, diag::err_omp_bit_fields_forbidden_in_map_clause) - << CurE->getSourceRange(); + SemaRef.Diag(ELoc, diag::err_omp_bit_fields_forbidden_in_clause) + << CurE->getSourceRange() << getOpenMPClauseName(CKind); break; } @@ -9513,7 +9523,8 @@ // Return true if expression E associated with value VD has conflicts with other // map information. static bool CheckMapConflicts(Sema &SemaRef, DSAStackTy *DSAS, ValueDecl *VD, - Expr *E, bool CurrentRegionOnly) { + Expr *E, bool CurrentRegionOnly, + OpenMPClauseKind CKind) { assert(VD && E); // Types used to organize the components of a valid map clause. @@ -9633,7 +9644,13 @@ // other, it means they are sharing storage. if (CI == CE && SI == SE) { if (CurrentRegionOnly) { - SemaRef.Diag(ELoc, diag::err_omp_map_shared_storage) << ERange; + if (CKind == OMPC_map) + SemaRef.Diag(ELoc, diag::err_omp_map_shared_storage) << ERange; + else { + assert(CKind == OMPC_to); + SemaRef.Diag(ELoc, diag::err_omp_once_referenced_in_target_update) + << ERange; + } SemaRef.Diag(RE->getExprLoc(), diag::note_used_here) << RE->getSourceRange(); return true; @@ -9687,7 +9704,13 @@ // // An expression is a subset of the other. if (CurrentRegionOnly && (CI == CE || SI == SE)) { - SemaRef.Diag(ELoc, diag::err_omp_map_shared_storage) << ERange; + if (CKind == OMPC_map) + SemaRef.Diag(ELoc, diag::err_omp_map_shared_storage) << ERange; + else { + assert(CKind == OMPC_to); + SemaRef.Diag(ELoc, diag::err_omp_once_referenced_in_target_update) + << ERange; + } SemaRef.Diag(RE->getExprLoc(), diag::note_used_here) << RE->getSourceRange(); return true; @@ -9732,23 +9755,16 @@ return FoundError; } -OMPClause * -Sema::ActOnOpenMPMapClause(OpenMPMapClauseKind MapTypeModifier, - OpenMPMapClauseKind MapType, bool IsMapTypeImplicit, - SourceLocation MapLoc, SourceLocation ColonLoc, - ArrayRef VarList, SourceLocation StartLoc, - SourceLocation LParenLoc, SourceLocation EndLoc) { - SmallVector Vars; - +static void checkOpenMPToFromMapVars(Sema &SemaRef, DSAStackTy *DSAS, + ArrayRef VarList, + SmallVector &Vars, + SourceLocation StartLoc, + OpenMPClauseKind CKind, + OpenMPMapClauseKind MapType, + bool IsMapTypeImplicit) { for (auto &RE : VarList) { - assert(RE && "Null expr in omp map"); - if (isa(RE)) { - // It will be analyzed later. - Vars.push_back(RE); - continue; - } - SourceLocation ELoc = RE->getExprLoc(); - + assert(RE && "Null expr in omp to/map clause"); + auto ELoc = RE->getExprLoc(); auto *VE = RE->IgnoreParenLValueCasts(); if (VE->isValueDependent() || VE->isTypeDependent() || @@ -9763,18 +9779,19 @@ auto *SimpleExpr = RE->IgnoreParenCasts(); if (!RE->IgnoreParenImpCasts()->isLValue()) { - Diag(ELoc, diag::err_omp_expected_named_var_member_or_array_expression) + SemaRef.Diag(ELoc, + diag::err_omp_expected_named_var_member_or_array_expression) << RE->getSourceRange(); continue; } // Obtain the array or member expression bases if required. - auto *BE = CheckMapClauseExpressionBase(*this, SimpleExpr); + auto *BE = CheckMapClauseExpressionBase(SemaRef, SimpleExpr, CKind); if (!BE) continue; // If the base is a reference to a variable, we rely on that variable for - // the following checks. If it is a 'this' expression we rely on the field. + // the following checks. if it is a 'this' expression we rely on the field. ValueDecl *D = nullptr; if (auto *DRE = dyn_cast(BE)) { D = DRE->getDecl(); @@ -9783,7 +9800,7 @@ assert(isa(ME->getBase()) && "Unexpected expression!"); D = ME->getMemberDecl(); } - assert(D && "Null decl on map clause."); + assert(D && "Null decl on to/map clause."); auto *VD = dyn_cast(D); auto *FD = dyn_cast(D); @@ -9792,11 +9809,14 @@ (void)FD; // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, p.10] - // threadprivate variables cannot appear in a map clause. - if (VD && DSAStack->isThreadPrivate(VD)) { - auto DVar = DSAStack->getTopDSA(VD, false); - Diag(ELoc, diag::err_omp_threadprivate_in_map); - ReportOriginalDSA(*this, DSAStack, VD, DVar); + // threadprivate variables cannot appear in a map clause. + // OpenMP 4.5 [2.10.5, target update Construct] + // threadprivate variables cannot appear in a to clause. + if (VD && DSAS->isThreadPrivate(VD)) { + auto DVar = DSAS->getTopDSA(VD, false); + SemaRef.Diag(ELoc, diag::err_omp_threadprivate_in_clause) + << getOpenMPClauseName(CKind); + ReportOriginalDSA(SemaRef, DSAS, VD, DVar); continue; } @@ -9808,75 +9828,92 @@ // missing implementation of the other data sharing clauses in target // directives. - // Check conflicts with other map clause expressions. We check the conflicts - // with the current construct separately from the enclosing data - // environment, because the restrictions are different. - if (CheckMapConflicts(*this, DSAStack, D, SimpleExpr, - /*CurrentRegionOnly=*/true)) + // Check conflicts with other map or to clause expressions. We check + // the conflicts with the current construct separately from the enclosing + // data environment, because the restrictions are different. + if (CheckMapConflicts(SemaRef, DSAS, D, SimpleExpr, + /*CurrentRegionOnly=*/true, CKind)) break; - if (CheckMapConflicts(*this, DSAStack, D, SimpleExpr, - /*CurrentRegionOnly=*/false)) + if (CKind == OMPC_map && + CheckMapConflicts(SemaRef, DSAS, D, SimpleExpr, + /*CurrentRegionOnly=*/false, CKind)) break; + // OpenMP 4.5 [2.10.5, target update Construct] // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, C++, p.1] - // If the type of a list item is a reference to a type T then the type will - // be considered to be T for all purposes of this clause. - QualType Type = D->getType(); + // If the type of a list item is a reference to a type T then the type will + // be considered to be T for all purpose of this clause. + auto Type = D->getType(); if (Type->isReferenceType()) Type = Type->getPointeeType(); + // OpenMP 4.5 [2.10.5, target update Construct, Restrictions, p.4] + // A list item in a to or from clause must have a mappable type. // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, p.9] - // A list item must have a mappable type. - if (!CheckTypeMappable(VE->getExprLoc(), VE->getSourceRange(), *this, - DSAStack, Type)) - continue; - - // target enter data - // OpenMP [2.10.2, Restrictions, p. 99] - // A map-type must be specified in all map clauses and must be either - // to or alloc. - OpenMPDirectiveKind DKind = DSAStack->getCurrentDirective(); - if (DKind == OMPD_target_enter_data && - !(MapType == OMPC_MAP_to || MapType == OMPC_MAP_alloc)) { - Diag(StartLoc, diag::err_omp_invalid_map_type_for_directive) - << (IsMapTypeImplicit ? 1 : 0) - << getOpenMPSimpleClauseTypeName(OMPC_map, MapType) - << getOpenMPDirectiveName(DKind); + // A list item must have a mappable type. + if (!CheckTypeMappable(VE->getExprLoc(), VE->getSourceRange(), SemaRef, + DSAS, Type)) continue; - } - // 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) - << (IsMapTypeImplicit ? 1 : 0) - << getOpenMPSimpleClauseTypeName(OMPC_map, MapType) - << getOpenMPDirectiveName(DKind); - continue; - } + if (CKind == OMPC_map) { + // target enter data + // OpenMP [2.10.2, Restrictions, p. 99] + // A map-type must be specified in all map clauses and must be either + // to or alloc. + OpenMPDirectiveKind DKind = DSAS->getCurrentDirective(); + if (DKind == OMPD_target_enter_data && + !(MapType == OMPC_MAP_to || MapType == OMPC_MAP_alloc)) { + SemaRef.Diag(StartLoc, diag::err_omp_invalid_map_type_for_directive) + << (IsMapTypeImplicit ? 1 : 0) + << getOpenMPSimpleClauseTypeName(OMPC_map, MapType) + << getOpenMPDirectiveName(DKind); + continue; + } - // 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) { - auto DVar = DSAStack->getTopDSA(VD, false); - if (isOpenMPPrivate(DVar.CKind)) { - Diag(ELoc, diag::err_omp_variable_in_map_and_dsa) - << getOpenMPClauseName(DVar.CKind) - << getOpenMPDirectiveName(DSAStack->getCurrentDirective()); - ReportOriginalDSA(*this, DSAStack, D, DVar); + // 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 = DSAS->getCurrentDirective(); + if (DKind == OMPD_target_exit_data && + !(MapType == OMPC_MAP_from || MapType == OMPC_MAP_release || + MapType == OMPC_MAP_delete)) { + SemaRef.Diag(StartLoc, diag::err_omp_invalid_map_type_for_directive) + << (IsMapTypeImplicit ? 1 : 0) + << getOpenMPSimpleClauseTypeName(OMPC_map, MapType) + << getOpenMPDirectiveName(DKind); continue; } + + // 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) { + auto DVar = DSAS->getTopDSA(VD, false); + if (isOpenMPPrivate(DVar.CKind)) { + SemaRef.Diag(ELoc, diag::err_omp_variable_in_map_and_dsa) + << getOpenMPClauseName(DVar.CKind) + << getOpenMPDirectiveName(DSAS->getCurrentDirective()); + ReportOriginalDSA(SemaRef, DSAS, D, DVar); + continue; + } + } } Vars.push_back(RE); - DSAStack->addExprToVarMapInfo(D, RE); + DSAS->addExprToVarMapInfo(D, RE); } +} + +OMPClause * +Sema::ActOnOpenMPMapClause(OpenMPMapClauseKind MapTypeModifier, + OpenMPMapClauseKind MapType, bool IsMapTypeImplicit, + SourceLocation MapLoc, SourceLocation ColonLoc, + ArrayRef VarList, SourceLocation StartLoc, + SourceLocation LParenLoc, SourceLocation EndLoc) { + SmallVector Vars; + checkOpenMPToFromMapVars(*this, DSAStack, VarList, Vars, StartLoc, OMPC_map, + MapType, IsMapTypeImplicit); // We need to produce a map clause even if we don't have variables so that // other diagnostics related with non-existing map clauses are accurate. @@ -10291,3 +10328,16 @@ return new (Context) OMPDefaultmapClause(StartLoc, LParenLoc, MLoc, KindLoc, EndLoc, Kind, M); } + +OMPClause *Sema::ActOnOpenMPToClause(ArrayRef VarList, + SourceLocation StartLoc, + SourceLocation LParenLoc, + SourceLocation EndLoc) { + SmallVector Vars; + checkOpenMPToFromMapVars(*this, DSAStack, VarList, Vars, StartLoc, OMPC_to, + OMPC_MAP_unknown, /*IsMapTypeImplicit=*/false); + if (Vars.empty()) + return nullptr; + + return OMPToClause::Create(Context, StartLoc, LParenLoc, EndLoc, Vars); +} Index: lib/Sema/TreeTransform.h =================================================================== --- lib/Sema/TreeTransform.h +++ lib/Sema/TreeTransform.h @@ -1751,6 +1751,17 @@ Kind, ChunkSize, StartLoc, LParenLoc, KindLoc, CommaLoc, EndLoc); } + /// \brief Build a new OpenMP 'to' clause. + /// + /// By default, performs semantic analysis to build the new statement. + /// Subclasses may override this routine to provide different behavior. + OMPClause *RebuildOMPToClause(ArrayRef VarList, + SourceLocation StartLoc, + SourceLocation LParenLoc, + SourceLocation EndLoc) { + return getSema().ActOnOpenMPToClause(VarList, StartLoc, LParenLoc, EndLoc); + } + /// \brief Rebuild the operand to an Objective-C \@synchronized statement. /// /// By default, performs semantic analysis to build the new statement. @@ -8025,6 +8036,20 @@ return C; } +template +OMPClause *TreeTransform::TransformOMPToClause(OMPToClause *C) { + llvm::SmallVector Vars; + Vars.reserve(C->varlist_size()); + for (auto *VE : C->varlists()) { + ExprResult EVar = getDerived().TransformExpr(cast(VE)); + if (EVar.isInvalid()) + return 0; + Vars.push_back(EVar.get()); + } + return getDerived().RebuildOMPToClause(Vars, C->getLocStart(), + C->getLParenLoc(), C->getLocEnd()); +} + //===----------------------------------------------------------------------===// // Expression transformation //===----------------------------------------------------------------------===// Index: lib/Serialization/ASTReaderStmt.cpp =================================================================== --- lib/Serialization/ASTReaderStmt.cpp +++ lib/Serialization/ASTReaderStmt.cpp @@ -1888,6 +1888,9 @@ case OMPC_defaultmap: C = new (Context) OMPDefaultmapClause(); break; + case OMPC_to: + C = OMPToClause::CreateEmpty(Context, Record[Idx++]); + break; } Visit(C); C->setLocStart(Reader->ReadSourceLocation(Record, Idx)); @@ -2283,6 +2286,16 @@ C->setDefaultmapKindLoc(Reader->ReadSourceLocation(Record, Idx)); } +void OMPClauseReader::VisitOMPToClause(OMPToClause *C) { + C->setLParenLoc(Reader->ReadSourceLocation(Record, Idx)); + unsigned NumVars = C->varlist_size(); + SmallVector Vars; + Vars.reserve(NumVars); + for (unsigned i = 0; i != NumVars; ++i) + Vars.push_back(Reader->Reader.ReadSubExpr()); + C->setVarRefs(Vars); +} + //===----------------------------------------------------------------------===// // OpenMP Directives. //===----------------------------------------------------------------------===// Index: lib/Serialization/ASTWriterStmt.cpp =================================================================== --- lib/Serialization/ASTWriterStmt.cpp +++ lib/Serialization/ASTWriterStmt.cpp @@ -2071,6 +2071,13 @@ Writer->Writer.AddSourceLocation(C->getDefaultmapKindLoc(), Record); } +void OMPClauseWriter::VisitOMPToClause(OMPToClause *C) { + Record.push_back(C->varlist_size()); + Writer->Writer.AddSourceLocation(C->getLParenLoc(), Record); + for (auto *VE : C->varlists()) + Writer->Writer.AddStmt(VE); +} + //===----------------------------------------------------------------------===// // OpenMP Directives. //===----------------------------------------------------------------------===// Index: test/OpenMP/target_map_messages.cpp =================================================================== --- test/OpenMP/target_map_messages.cpp +++ test/OpenMP/target_map_messages.cpp @@ -40,7 +40,7 @@ #pragma omp target map(arg,a,d[:2]) // expected-error {{subscripted value is not an array or pointer}} {} - #pragma omp target map(to:ss) // expected-error {{threadprivate variables are not allowed in map clause}} + #pragma omp target map(to:ss) // expected-error {{threadprivate variables are not allowed in 'map' clause}} {} #pragma omp target map(to:b,e) @@ -239,7 +239,7 @@ {} #pragma omp target map(r.C, t.C) {} - #pragma omp target map(r.A) // expected-error {{bit fields cannot be used to specify storage in a map clause}} + #pragma omp target map(r.A) // expected-error {{bit fields cannot be used to specify storage in a 'map' clause}} {} #pragma omp target map(r.Arr) {} @@ -407,7 +407,7 @@ #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(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 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(); @@ -476,7 +476,7 @@ #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(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 map(k), map(k[:5]) // expected-error {{pointer cannot be mapped along with a section derived from itself}} expected-note {{used here}} foo(); @@ -502,7 +502,11 @@ foo(); #pragma omp target private(j) map(j) // expected-error {{private variable cannot be in a map clause in '#pragma omp target' directive}} expected-note {{defined as private}} {} -#pragma omp target firstprivate(j) map(j) // expected-error {{firstprivate variable cannot be in a map clause in '#pragma omp target' directive}} expected-note {{defined as firstprivate}} +#pragma omp target map(j) private(j) // expected-error {{private variable cannot be in a map clause in '#pragma omp target' directive}} + {} +#pragma omp target firstprivate(j) map(j) // expected-error {{firstprivate variable cannot be in a map clause in '#pragma omp target' directive}} expected-note {{defined as firstprivate}} + {} +#pragma omp target map(j) firstprivate(j) // expected-error {{firstprivate variable cannot be in a map clause in '#pragma omp target' directive}} {} 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}} } Index: test/OpenMP/target_parallel_for_map_messages.cpp =================================================================== --- test/OpenMP/target_parallel_for_map_messages.cpp +++ test/OpenMP/target_parallel_for_map_messages.cpp @@ -126,7 +126,7 @@ for (i = 0; i < argc; ++i) foo(); #pragma omp target parallel for map(e, g) for (i = 0; i < argc; ++i) foo(); -#pragma omp target parallel for map(h) // expected-error {{threadprivate variables are not allowed in map clause}} +#pragma omp target parallel for map(h) // expected-error {{threadprivate variables are not allowed in 'map' clause}} for (i = 0; i < argc; ++i) foo(); #pragma omp target parallel for map(k), map(k) // expected-error 2 {{variable already marked as mapped in current construct}} expected-note 2 {{used here}} for (i = 0; i < argc; ++i) foo(); @@ -230,7 +230,7 @@ for (i = 0; i < argc; ++i) foo(); #pragma omp target parallel for map(e, g) for (i = 0; i < argc; ++i) foo(); -#pragma omp target parallel for map(h) // expected-error {{threadprivate variables are not allowed in map clause}} +#pragma omp target parallel for map(h) // expected-error {{threadprivate variables are not allowed in 'map' clause}} for (i = 0; i < argc; ++i) foo(); #pragma omp target parallel for map(k), map(k) // expected-error {{variable already marked as mapped in current construct}} expected-note {{used here}} for (i = 0; i < argc; ++i) foo(); Index: test/OpenMP/target_parallel_map_messages.cpp =================================================================== --- test/OpenMP/target_parallel_map_messages.cpp +++ test/OpenMP/target_parallel_map_messages.cpp @@ -126,7 +126,7 @@ foo(); #pragma omp target parallel map(e, g) foo(); -#pragma omp target parallel map(h) // expected-error {{threadprivate variables are not allowed in map clause}} +#pragma omp target parallel map(h) // expected-error {{threadprivate variables are not allowed in 'map' clause}} foo(); #pragma omp target parallel map(k), map(k) // expected-error 2 {{variable already marked as mapped in current construct}} expected-note 2 {{used here}} foo(); @@ -229,7 +229,7 @@ foo(); #pragma omp target parallel map(e, g) foo(); -#pragma omp target parallel map(h) // expected-error {{threadprivate variables are not allowed in map clause}} +#pragma omp target parallel map(h) // expected-error {{threadprivate variables are not allowed in 'map' clause}} foo(); #pragma omp target parallel map(k), map(k) // expected-error {{variable already marked as mapped in current construct}} expected-note {{used here}} foo(); Index: tools/libclang/CIndex.cpp =================================================================== --- tools/libclang/CIndex.cpp +++ tools/libclang/CIndex.cpp @@ -2255,6 +2255,9 @@ } void OMPClauseEnqueue::VisitOMPDefaultmapClause( const OMPDefaultmapClause * /*C*/) {} +void OMPClauseEnqueue::VisitOMPToClause(const OMPToClause *C) { + VisitOMPClauseList(C); +} } void EnqueueVisitor::EnqueueChildren(const OMPClause *S) {