[OpenMP 5.0] Parsing/sema support for to clause with mapper modifier

[lildmh]

Summary

This patch implements the parsing and sema support for OpenMP to clause with potential user-defined mappers attached. User defined mapper is a new feature in OpenMP 5.0. A to/from clause can have an explicit or implicit associated mapper, which instructs the compiler to generate and use customized mapping functions. An example is shown below:

struct S { int len; int *d; };
#pragma omp declare mapper(id: struct S s) map(s, s.d[0:s.len])
struct S ss;
#pragma omp target update to(mapper(id): ss) // use the mapper with name 'id' to map ss to device

[ABataev]

include/clang/Basic/OpenMPKinds.def

588

Maybe, better to split it. There must be separate modifiers for from clause and for to clause just for possible future changes.

lib/AST/OpenMPClause.cpp

876

auto *

923–924

auto *

1468

Better to split this patch into 2: one for to clause and another one for from clause

lib/Parse/ParseOpenMP.cpp

2154

Remove braces

2173–2174

mapper(...) does not look good here, we never ever used this kind of construct before. Better to use something that was used already.

lib/Sema/SemaOpenMP.cpp

13061

Better to fix this in a separate patch.

13355

Are you sure about this change? This might trigger some asserts.

13413

Again, better to do this in a separate patch

[lildmh]

lib/AST/OpenMPClause.cpp

1468

Hi Alexey,

Thanks for the quick review! The code for to and from clauses are almost identical, so I prefer to have them in the same patch. Besides I don't have commit privilege, so it causes more time to me to have multiple patches. But if you think it's crucial to split it. I can do it. What do you think?

lib/Sema/SemaOpenMP.cpp

13061

I can do that.

13355

Yes. The original code before the mapper patch is like this. By moving it out, to and from clauses will also call buildUserDefinedMapperRef check the mapper as well.

13413

This modification is for to and from clauses, so they also have a default mapper name. I think it is appropriate to have this piece in this patch.

[ABataev]

lib/AST/OpenMPClause.cpp

1468

Hi Lingda, it is clang/LLVM policy to make patches as small as possible (for better "reviewability", reliability, trackability etc.). If it is possible to split the patch into several parts, this must be done.

[lildmh]

get rid of the from clause part.

[ABataev]

lib/Sema/SemaOpenMP.cpp

13195

Is this correct for from clause?

13237

Same here: what about from?

lib/Sema/TreeTransform.h

8825

llvm::SmallVector<Expr *, 16> & -> llvm::SmallVectorImpl<Expr *> &

8827

llvm::SmallVector<Expr *, 16> & -> llvm::SmallVectorImpl<Expr *> &

[lildmh]

lib/Sema/SemaOpenMP.cpp

13195

Yes, it's correct for from, which will never use MVLI.UDMapperList. Regression test is also correct.

[ABataev]

LG

[lildmh]

Thanks a lot!

[lildmh]

Rebase

include/clang/AST/OpenMPClause.h

public:
   Expr *getPostUpdateExpr() { return PostUpdate; }
 
   static OMPClauseWithPostUpdate *get(OMPClause *C);
   static const OMPClauseWithPostUpdate *get(const OMPClause *C);
 };
 
 /// This structure contains most locations needed for by an OMPVarListClause.
 struct OMPVarListLocTy {
+  /// Starting location of the clause (the clause keyword).
   SourceLocation StartLoc;
+  /// Location of '('.
   SourceLocation LParenLoc;
+  /// Ending location of the clause.
   SourceLocation EndLoc;
   OMPVarListLocTy() = default;
   OMPVarListLocTy(SourceLocation StartLoc, SourceLocation LParenLoc,
                   SourceLocation EndLoc)
       : StartLoc(StartLoc), LParenLoc(LParenLoc), EndLoc(EndLoc) {}
 };
 
 /// This represents clauses with the list of variables like 'private',
protected:
   // declarations are expected to be canonical.
   static unsigned
   getUniqueDeclarationsTotalNumber(ArrayRef<const ValueDecl *> Declarations);
 };
 
 /// This structure contains all sizes needed for by an
 /// OMPMappableExprListClause.
 struct OMPMappableExprListSizeTy {
+  /// Number of expressions listed.
   unsigned NumVars;
+  /// Number of unique base declarations.
   unsigned NumUniqueDeclarations;
+  /// Number of component lists.
   unsigned NumComponentLists;
+  /// Total number of expression components.
   unsigned NumComponents;
   OMPMappableExprListSizeTy() = default;
   OMPMappableExprListSizeTy(unsigned NumVars, unsigned NumUniqueDeclarations,
                             unsigned NumComponentLists, unsigned NumComponents)
       : NumVars(NumVars), NumUniqueDeclarations(NumUniqueDeclarations),
         NumComponentLists(NumComponentLists), NumComponents(NumComponents) {}
 };
 
class OMPToClause final : public OMPMappableExprListClause<OMPToClause>,
                               OMPClauseMappableExprCommon::MappableComponent> {
   friend class OMPClauseReader;
   friend OMPMappableExprListClause;
   friend OMPVarListClause;
   friend TrailingObjects;
 
   /// Build clause with number of variables \a NumVars.
   ///
+  /// \param MapperQualifierLoc C++ nested name specifier for the associated
+  /// user-defined mapper.
+  /// \param MapperIdInfo The identifier of associated user-defined mapper.
+  /// \param MapType Map type.
   /// \param Locs Locations needed to build a mappable clause. It includes 1)
   /// StartLoc: starting location of the clause (the clause keyword); 2)
   /// LParenLoc: location of '('; 3) EndLoc: ending location of the clause.
   /// \param Sizes All required sizes to build a mappable clause. It includes 1)
   /// NumVars: number of expressions listed in this clause; 2)
   /// NumUniqueDeclarations: number of unique base declarations in this clause;
   /// 3) NumComponentLists: number of component lists in this clause; and 4)
   /// NumComponents: total number of expression components in the clause.
-  explicit OMPToClause(const OMPVarListLocTy &Locs,
+  explicit OMPToClause(NestedNameSpecifierLoc MapperQualifierLoc,
+                       DeclarationNameInfo MapperIdInfo,
+                       const OMPVarListLocTy &Locs,
                        const OMPMappableExprListSizeTy &Sizes)
-      : OMPMappableExprListClause(OMPC_to, Locs, Sizes) {}
+      : OMPMappableExprListClause(OMPC_to, Locs, Sizes, &MapperQualifierLoc,
+                                  &MapperIdInfo) {}
 
   /// Build an empty clause.
   ///
   /// \param Sizes All required sizes to build a mappable clause. It includes 1)
   /// NumVars: number of expressions listed in this clause; 2)
   /// NumUniqueDeclarations: number of unique base declarations in this clause;
   /// 3) NumComponentLists: number of component lists in this clause; and 4)
   /// NumComponents: total number of expression components in the clause.
   explicit OMPToClause(const OMPMappableExprListSizeTy &Sizes)
       : OMPMappableExprListClause(OMPC_to, OMPVarListLocTy(), Sizes) {}
 
   /// Define the sizes of each trailing object array except the last one. This
   /// is required for TrailingObjects to work properly.
   size_t numTrailingObjects(OverloadToken<Expr *>) const {
-    return varlist_size();
+    // There are varlist_size() of expressions, and varlist_size() of
+    // user-defined mappers.
+    return 2 * varlist_size();
   }
   size_t numTrailingObjects(OverloadToken<ValueDecl *>) const {
     return getUniqueDeclarationsNum();
   }
   size_t numTrailingObjects(OverloadToken<unsigned>) const {
     return getUniqueDeclarationsNum() + getTotalComponentListNum();
   }
 
 public:
   /// Creates clause with a list of variables \a Vars.
   ///
   /// \param C AST context.
   /// \param Locs Locations needed to build a mappable clause. It includes 1)
   /// StartLoc: starting location of the clause (the clause keyword); 2)
   /// LParenLoc: location of '('; 3) EndLoc: ending location of the clause.
   /// \param Vars The original expression used in the clause.
   /// \param Declarations Declarations used in the clause.
   /// \param ComponentLists Component lists used in the clause.
+  /// \param UDMapperRefs References to user-defined mappers associated with
+  /// expressions used in the clause.
+  /// \param UDMQualifierLoc C++ nested name specifier for the associated
+  /// user-defined mapper.
+  /// \param MapperId The identifier of associated user-defined mapper.
   static OMPToClause *Create(const ASTContext &C, const OMPVarListLocTy &Locs,
                              ArrayRef<Expr *> Vars,
                              ArrayRef<ValueDecl *> Declarations,
-                             MappableExprComponentListsRef ComponentLists);
+                             MappableExprComponentListsRef ComponentLists,
+                             ArrayRef<Expr *> UDMapperRefs,
+                             NestedNameSpecifierLoc UDMQualifierLoc,
+                             DeclarationNameInfo MapperId);
 
   /// Creates an empty clause with the place for \a NumVars variables.
   ///
   /// \param C AST context.
   /// \param Sizes All required sizes to build a mappable clause. It includes 1)
   /// NumVars: number of expressions listed in this clause; 2)
   /// NumUniqueDeclarations: number of unique base declarations in this clause;
   /// 3) NumComponentLists: number of component lists in this clause; and 4)
: public OMPMappableExprListClause<OMPFromClause>,
           OMPClauseMappableExprCommon::MappableComponent> {
   friend class OMPClauseReader;
   friend OMPMappableExprListClause;
   friend OMPVarListClause;
   friend TrailingObjects;
 
   /// Build clause with number of variables \a NumVars.
   ///
+  /// \param MapperQualifierLoc C++ nested name specifier for the associated
+  /// user-defined mapper.
+  /// \param MapperIdInfo The identifier of associated user-defined mapper.
+  /// \param MapType Map type.
   /// \param Locs Locations needed to build a mappable clause. It includes 1)
   /// StartLoc: starting location of the clause (the clause keyword); 2)
   /// LParenLoc: location of '('; 3) EndLoc: ending location of the clause.
   /// \param Sizes All required sizes to build a mappable clause. It includes 1)
   /// NumVars: number of expressions listed in this clause; 2)
   /// NumUniqueDeclarations: number of unique base declarations in this clause;
   /// 3) NumComponentLists: number of component lists in this clause; and 4)
   /// NumComponents: total number of expression components in the clause.
-  explicit OMPFromClause(const OMPVarListLocTy &Locs,
+  explicit OMPFromClause(NestedNameSpecifierLoc MapperQualifierLoc,
+                         DeclarationNameInfo MapperIdInfo,
+                         const OMPVarListLocTy &Locs,
                          const OMPMappableExprListSizeTy &Sizes)
-      : OMPMappableExprListClause(OMPC_from, Locs, Sizes) {}
+      : OMPMappableExprListClause(OMPC_from, Locs, Sizes, &MapperQualifierLoc,
+                                  &MapperIdInfo) {}
 
   /// Build an empty clause.
   ///
   /// \param Sizes All required sizes to build a mappable clause. It includes 1)
   /// NumVars: number of expressions listed in this clause; 2)
   /// NumUniqueDeclarations: number of unique base declarations in this clause;
   /// 3) NumComponentLists: number of component lists in this clause; and 4)
   /// NumComponents: total number of expression components in the clause.
   explicit OMPFromClause(const OMPMappableExprListSizeTy &Sizes)
       : OMPMappableExprListClause(OMPC_from, OMPVarListLocTy(), Sizes) {}
 
   /// Define the sizes of each trailing object array except the last one. This
   /// is required for TrailingObjects to work properly.
   size_t numTrailingObjects(OverloadToken<Expr *>) const {
-    return varlist_size();
+    // There are varlist_size() of expressions, and varlist_size() of
+    // user-defined mappers.
+    return 2 * varlist_size();
   }
   size_t numTrailingObjects(OverloadToken<ValueDecl *>) const {
     return getUniqueDeclarationsNum();
   }
   size_t numTrailingObjects(OverloadToken<unsigned>) const {
     return getUniqueDeclarationsNum() + getTotalComponentListNum();
   }
 
 public:
   /// Creates clause with a list of variables \a Vars.
   ///
   /// \param C AST context.
   /// \param Locs Locations needed to build a mappable clause. It includes 1)
   /// StartLoc: starting location of the clause (the clause keyword); 2)
   /// LParenLoc: location of '('; 3) EndLoc: ending location of the clause.
   /// \param Vars The original expression used in the clause.
   /// \param Declarations Declarations used in the clause.
   /// \param ComponentLists Component lists used in the clause.
+  /// \param UDMapperRefs References to user-defined mappers associated with
+  /// expressions used in the clause.
+  /// \param UDMQualifierLoc C++ nested name specifier for the associated
+  /// user-defined mapper.
+  /// \param MapperId The identifier of associated user-defined mapper.
   static OMPFromClause *Create(const ASTContext &C, const OMPVarListLocTy &Locs,
                                ArrayRef<Expr *> Vars,
                                ArrayRef<ValueDecl *> Declarations,
-                               MappableExprComponentListsRef ComponentLists);
+                               MappableExprComponentListsRef ComponentLists,
+                               ArrayRef<Expr *> UDMapperRefs,
+                               NestedNameSpecifierLoc UDMQualifierLoc,
+                               DeclarationNameInfo MapperId);
 
   /// Creates an empty clause with the place for \a NumVars variables.
   ///
   /// \param C AST context.
   /// \param Sizes All required sizes to build a mappable clause. It includes 1)
   /// NumVars: number of expressions listed in this clause; 2)
   /// NumUniqueDeclarations: number of unique base declarations in this clause;
   /// 3) NumComponentLists: number of component lists in this clause; and 4)

include/clang/Basic/DiagnosticParseKinds.td

def err_omp_declare_simd_inbranch_notinbranch : Error<
   "unexpected '%0' clause, '%1' is specified already">;
 def err_expected_end_declare_target : Error<
   "expected '#pragma omp end declare target'">;
 def err_omp_declare_target_unexpected_clause: Error<
   "unexpected '%0' clause, only 'to' or 'link' clauses expected">;
 def err_omp_expected_clause: Error<
   "expected at least one clause on '#pragma omp %0' directive">;
 def err_omp_mapper_illegal_identifier : Error<
-  "illegal identifier on 'omp declare mapper' directive">;
+  "illegal OpenMP user-defined mapper identifier">;
 def err_omp_mapper_expected_declarator : Error<
   "expected declarator on 'omp declare mapper' directive">;
 
 // Pragma loop support.
 def err_pragma_loop_missing_argument : Error<
   "missing argument; expected %select{an integer value|"
   "'enable'%select{|, 'full'}1%select{|, 'assume_safety'}2 or 'disable'}0">;
 def err_pragma_loop_invalid_option : Error<

include/clang/Basic/OpenMPKinds.h

 enum OpenMPMapModifierKind {
   OMPC_MAP_MODIFIER_unknown = OMPC_MAP_unknown,
 #define OPENMP_MAP_MODIFIER_KIND(Name) \
   OMPC_MAP_MODIFIER_##Name,
 #include "clang/Basic/OpenMPKinds.def"
   OMPC_MAP_MODIFIER_last
 };
 
+/// OpenMP modifier kind for 'to' clause.
+enum OpenMPToModifierKind {
+#define OPENMP_TO_MODIFIER_KIND(Name) \
+  OMPC_TO_MODIFIER_##Name,
+#include "clang/Basic/OpenMPKinds.def"
+  OMPC_TO_MODIFIER_unknown
+};
+
+/// OpenMP modifier kind for 'from' clause.
+enum OpenMPFromModifierKind {
+#define OPENMP_FROM_MODIFIER_KIND(Name) \
+  OMPC_FROM_MODIFIER_##Name,
+#include "clang/Basic/OpenMPKinds.def"
+  OMPC_FROM_MODIFIER_unknown
+};
+
 /// OpenMP attributes for 'dist_schedule' clause.
 enum OpenMPDistScheduleClauseKind {
 #define OPENMP_DIST_SCHEDULE_KIND(Name) OMPC_DIST_SCHEDULE_##Name,
 #include "clang/Basic/OpenMPKinds.def"
   OMPC_DIST_SCHEDULE_unknown
 };
 
 /// OpenMP attributes for 'defaultmap' clause.

include/clang/Basic/OpenMPKinds.def

 #define OPENMP_LINEAR_KIND(Name)
 #endif
 #ifndef OPENMP_MAP_KIND
 #define OPENMP_MAP_KIND(Name)
 #endif
 #ifndef OPENMP_MAP_MODIFIER_KIND
 #define OPENMP_MAP_MODIFIER_KIND(Name)
 #endif
+#ifndef OPENMP_TO_MODIFIER_KIND
+#define OPENMP_TO_MODIFIER_KIND(Name)
+#endif
+#ifndef OPENMP_FROM_MODIFIER_KIND
+#define OPENMP_FROM_MODIFIER_KIND(Name)
+#endif
 #ifndef OPENMP_DIST_SCHEDULE_KIND
 #define OPENMP_DIST_SCHEDULE_KIND(Name)
 #endif
 #ifndef OPENMP_DEFAULTMAP_KIND
 #define OPENMP_DEFAULTMAP_KIND(Name)
 #endif
 #ifndef OPENMP_ATOMIC_DEFAULT_MEM_ORDER_KIND
 #define OPENMP_ATOMIC_DEFAULT_MEM_ORDER_KIND(Name)

 OPENMP_MAP_KIND(delete)
 OPENMP_MAP_KIND(release)
 
 // Map-type-modifiers for 'map' clause.
 OPENMP_MAP_MODIFIER_KIND(always)
 OPENMP_MAP_MODIFIER_KIND(close)
 OPENMP_MAP_MODIFIER_KIND(mapper)
 
+// Modifiers for 'to' clause.
+OPENMP_TO_MODIFIER_KIND(mapper)
+
+// Modifiers for 'from' clause.
+OPENMP_FROM_MODIFIER_KIND(mapper)
+
 // Clauses allowed for OpenMP directive 'taskloop'.
 OPENMP_TASKLOOP_CLAUSE(if)
 OPENMP_TASKLOOP_CLAUSE(shared)
 OPENMP_TASKLOOP_CLAUSE(private)
 OPENMP_TASKLOOP_CLAUSE(firstprivate)
 OPENMP_TASKLOOP_CLAUSE(lastprivate)
 OPENMP_TASKLOOP_CLAUSE(default)
 OPENMP_TASKLOOP_CLAUSE(collapse)

 #undef OPENMP_TARGET_PARALLEL_CLAUSE
 #undef OPENMP_TARGET_PARALLEL_FOR_CLAUSE
 #undef OPENMP_TEAMS_CLAUSE
 #undef OPENMP_SIMD_CLAUSE
 #undef OPENMP_FOR_CLAUSE
 #undef OPENMP_FOR_SIMD_CLAUSE
 #undef OPENMP_MAP_KIND
 #undef OPENMP_MAP_MODIFIER_KIND
+#undef OPENMP_TO_MODIFIER_KIND
+#undef OPENMP_FROM_MODIFIER_KIND
 #undef OPENMP_DISTRIBUTE_CLAUSE
 #undef OPENMP_DIST_SCHEDULE_KIND
 #undef OPENMP_DEFAULTMAP_KIND
 #undef OPENMP_DEFAULTMAP_MODIFIER
 #undef OPENMP_TARGET_UPDATE_CLAUSE
 #undef OPENMP_DISTRIBUTE_PARALLEL_FOR_CLAUSE
 #undef OPENMP_DISTRIBUTE_PARALLEL_FOR_SIMD_CLAUSE
 #undef OPENMP_DISTRIBUTE_SIMD_CLAUSE

include/clang/Parse/Parser.h

bool ParseOpenMPVarList(OpenMPDirectiveKind DKind, OpenMPClauseKind Kind,
                           OpenMPVarListDataTy &Data);
   bool ParseUnqualifiedId(CXXScopeSpec &SS, bool EnteringContext,
                           bool AllowDestructorName,
                           bool AllowConstructorName,
                           bool AllowDeductionGuide,
                           ParsedType ObjectType,
                           SourceLocation *TemplateKWLoc,
                           UnqualifiedId &Result);
+  /// Parses the mapper modifier in map, to, and from clauses.
+  bool parseMapperModifier(OpenMPVarListDataTy &Data);
   /// Parses map-type-modifiers in map clause.
   /// map([ [map-type-modifier[,] [map-type-modifier[,] ...] map-type : ] list)
   /// where, map-type-modifier ::= always | close | mapper(mapper-identifier)
   bool parseMapTypeModifiers(OpenMPVarListDataTy &Data);
 
 private:
   //===--------------------------------------------------------------------===//
   // C++ 14: Templates [temp]

include/clang/Sema/Sema.h

OMPClause *ActOnOpenMPDistScheduleClause(
       SourceLocation StartLoc, SourceLocation LParenLoc, SourceLocation KindLoc,
       SourceLocation CommaLoc, SourceLocation EndLoc);
   /// Called on well-formed 'defaultmap' clause.
   OMPClause *ActOnOpenMPDefaultmapClause(
       OpenMPDefaultmapClauseModifier M, OpenMPDefaultmapClauseKind Kind,
       SourceLocation StartLoc, SourceLocation LParenLoc, SourceLocation MLoc,
       SourceLocation KindLoc, SourceLocation EndLoc);
   /// Called on well-formed 'to' clause.
-  OMPClause *ActOnOpenMPToClause(ArrayRef<Expr *> VarList,
-                                 const OMPVarListLocTy &Locs);
+  OMPClause *
+  ActOnOpenMPToClause(ArrayRef<Expr *> VarList, CXXScopeSpec &MapperIdScopeSpec,
+                      DeclarationNameInfo &MapperId,
+                      const OMPVarListLocTy &Locs,
+                      ArrayRef<Expr *> UnresolvedMappers = llvm::None);
   /// Called on well-formed 'from' clause.
-  OMPClause *ActOnOpenMPFromClause(ArrayRef<Expr *> VarList,
-                                   const OMPVarListLocTy &Locs);
+  OMPClause *ActOnOpenMPFromClause(
+      ArrayRef<Expr *> VarList, CXXScopeSpec &MapperIdScopeSpec,
+      DeclarationNameInfo &MapperId, const OMPVarListLocTy &Locs,
+      ArrayRef<Expr *> UnresolvedMappers = llvm::None);
   /// Called on well-formed 'use_device_ptr' clause.
   OMPClause *ActOnOpenMPUseDevicePtrClause(ArrayRef<Expr *> VarList,
                                            const OMPVarListLocTy &Locs);
   /// Called on well-formed 'is_device_ptr' clause.
   OMPClause *ActOnOpenMPIsDevicePtrClause(ArrayRef<Expr *> VarList,
                                           const OMPVarListLocTy &Locs);
 
   /// The kind of conversion being performed.

lib/AST/OpenMPClause.cpp

void *Mem = C.Allocate(
       totalSizeToAlloc<Expr *, ValueDecl *, unsigned,
                        OMPClauseMappableExprCommon::MappableComponent>(
           2 * Sizes.NumVars, Sizes.NumUniqueDeclarations,
           Sizes.NumUniqueDeclarations + Sizes.NumComponentLists,
           Sizes.NumComponents));
   return new (Mem) OMPMapClause(Sizes);
 }
 
-OMPToClause *OMPToClause::Create(const ASTContext &C,
-                                 const OMPVarListLocTy &Locs,
-                                 ArrayRef<Expr *> Vars,
-                                 ArrayRef<ValueDecl *> Declarations,
-                                 MappableExprComponentListsRef ComponentLists) {
+OMPToClause *OMPToClause::Create(
+    const ASTContext &C, const OMPVarListLocTy &Locs, ArrayRef<Expr *> Vars,
+    MappableExprComponentListsRef ComponentLists, ArrayRef<Expr *> UDMapperRefs,
+    NestedNameSpecifierLoc UDMQualifierLoc, DeclarationNameInfo MapperId) {
   OMPMappableExprListSizeTy Sizes;
   Sizes.NumVars = Vars.size();
   Sizes.NumUniqueDeclarations = getUniqueDeclarationsTotalNumber(Declarations);
   Sizes.NumComponentLists = ComponentLists.size();
   Sizes.NumComponents = getComponentsTotalNumber(ComponentLists);
 
   // We need to allocate:
-  // NumVars x Expr* - we have an original list expression for each clause list
-  // entry.
+  // 2 x NumVars x Expr* - we have an original list expression and an associated
+  // user-defined mapper for each clause list entry.
   // NumUniqueDeclarations x ValueDecl* - unique base declarations associated
   // with each component list.
   // (NumUniqueDeclarations + NumComponentLists) x unsigned - we specify the
   // number of lists for each unique declaration and the size of each component
   // list.
   // NumComponents x MappableComponent - the total of all the components in all
   // the lists.
   void *Mem = C.Allocate(
       totalSizeToAlloc<Expr *, ValueDecl *, unsigned,
                        OMPClauseMappableExprCommon::MappableComponent>(
-          Sizes.NumVars, Sizes.NumUniqueDeclarations,
+          2 * Sizes.NumVars, Sizes.NumUniqueDeclarations,
           Sizes.NumUniqueDeclarations + Sizes.NumComponentLists,
           Sizes.NumComponents));
 
-  OMPToClause *Clause = new (Mem) OMPToClause(Locs, Sizes);
+  auto *Clause = new (Mem) OMPToClause(UDMQualifierLoc, MapperId, Locs, Sizes);
 
   Clause->setVarRefs(Vars);
+  Clause->setUDMapperRefs(UDMapperRefs);
   Clause->setClauseInfo(Declarations, ComponentLists);
   return Clause;
 }
 
 OMPToClause *OMPToClause::CreateEmpty(const ASTContext &C,
                                       const OMPMappableExprListSizeTy &Sizes) {
   void *Mem = C.Allocate(
       totalSizeToAlloc<Expr *, ValueDecl *, unsigned,
                        OMPClauseMappableExprCommon::MappableComponent>(
-          Sizes.NumVars, Sizes.NumUniqueDeclarations,
+          2 * Sizes.NumVars, Sizes.NumUniqueDeclarations,
           Sizes.NumUniqueDeclarations + Sizes.NumComponentLists,
           Sizes.NumComponents));
   return new (Mem) OMPToClause(Sizes);
 }
 
-OMPFromClause *
-OMPFromClause::Create(const ASTContext &C, const OMPVarListLocTy &Locs,
-                      ArrayRef<Expr *> Vars, ArrayRef<ValueDecl *> Declarations,
-                      MappableExprComponentListsRef ComponentLists) {
+OMPFromClause *OMPFromClause::Create(
+    const ASTContext &C, const OMPVarListLocTy &Locs, ArrayRef<Expr *> Vars,
+    ArrayRef<ValueDecl *> Declarations,
+    MappableExprComponentListsRef ComponentLists, ArrayRef<Expr *> UDMapperRefs,
+    NestedNameSpecifierLoc UDMQualifierLoc, DeclarationNameInfo MapperId) {
   OMPMappableExprListSizeTy Sizes;
   Sizes.NumVars = Vars.size();
   Sizes.NumUniqueDeclarations = getUniqueDeclarationsTotalNumber(Declarations);
   Sizes.NumComponentLists = ComponentLists.size();
   Sizes.NumComponents = getComponentsTotalNumber(ComponentLists);
 
   // We need to allocate:
-  // NumVars x Expr* - we have an original list expression for each clause list
-  // entry.
+  // 2 x NumVars x Expr* - we have an original list expression and an associated
+  // user-defined mapper for each clause list entry.
   // NumUniqueDeclarations x ValueDecl* - unique base declarations associated
   // with each component list.
   // (NumUniqueDeclarations + NumComponentLists) x unsigned - we specify the
   // number of lists for each unique declaration and the size of each component
   // list.
   // NumComponents x MappableComponent - the total of all the components in all
   // the lists.
   void *Mem = C.Allocate(
       totalSizeToAlloc<Expr *, ValueDecl *, unsigned,
                        OMPClauseMappableExprCommon::MappableComponent>(
-          Sizes.NumVars, Sizes.NumUniqueDeclarations,
+          2 * Sizes.NumVars, Sizes.NumUniqueDeclarations,
           Sizes.NumUniqueDeclarations + Sizes.NumComponentLists,
           Sizes.NumComponents));
 
-  OMPFromClause *Clause = new (Mem) OMPFromClause(Locs, Sizes);
+  auto *Clause =
+      new (Mem) OMPFromClause(UDMQualifierLoc, MapperId, Locs, Sizes);
 
   Clause->setVarRefs(Vars);
+  Clause->setUDMapperRefs(UDMapperRefs);
   Clause->setClauseInfo(Declarations, ComponentLists);
   return Clause;
 }
 
 OMPFromClause *
 OMPFromClause::CreateEmpty(const ASTContext &C,
                            const OMPMappableExprListSizeTy &Sizes) {
   void *Mem = C.Allocate(
       totalSizeToAlloc<Expr *, ValueDecl *, unsigned,
                        OMPClauseMappableExprCommon::MappableComponent>(
-          Sizes.NumVars, Sizes.NumUniqueDeclarations,
+          2 * Sizes.NumVars, Sizes.NumUniqueDeclarations,
           Sizes.NumUniqueDeclarations + Sizes.NumComponentLists,
           Sizes.NumComponents));
   return new (Mem) OMPFromClause(Sizes);
 }
 
 void OMPUseDevicePtrClause::setPrivateCopies(ArrayRef<Expr *> VL) {
   assert(VL.size() == varlist_size() &&
          "Number of private copies is not the same as the preallocated buffer");
if (!Node->varlist_empty()) {
     VisitOMPClauseList(Node, ' ');
     OS << ")";
   }
 }
 
 void OMPClausePrinter::VisitOMPToClause(OMPToClause *Node) {
   if (!Node->varlist_empty()) {
     OS << "to";
+    DeclarationNameInfo MapperId = Node->getMapperIdInfo();
+    if (MapperId.getName() && !MapperId.getName().isEmpty()) {
+      OS << '(';
+      OS << "mapper(";
+      NestedNameSpecifier *MapperNNS =
+          Node->getMapperQualifierLoc().getNestedNameSpecifier();
+      if (MapperNNS)
+        MapperNNS->print(OS, Policy);
+      OS << MapperId << "):";
+      VisitOMPClauseList(Node, ' ');
+    } else {
       VisitOMPClauseList(Node, '(');
+    }
     OS << ")";
   }
 }
 
 void OMPClausePrinter::VisitOMPFromClause(OMPFromClause *Node) {
   if (!Node->varlist_empty()) {
     OS << "from";
+    DeclarationNameInfo MapperId = Node->getMapperIdInfo();
+    if (MapperId.getName() && !MapperId.getName().isEmpty()) {
+      OS << '(';
+      OS << "mapper(";
+      NestedNameSpecifier *MapperNNS =
+          Node->getMapperQualifierLoc().getNestedNameSpecifier();
+      if (MapperNNS)
+        MapperNNS->print(OS, Policy);
+      OS << MapperId << "):";
+      VisitOMPClauseList(Node, ' ');
+    } else {
       VisitOMPClauseList(Node, '(');
+    }
     OS << ")";
   }
 }
 
 void OMPClausePrinter::VisitOMPDistScheduleClause(OMPDistScheduleClause *Node) {
   OS << "dist_schedule(" << getOpenMPSimpleClauseTypeName(
                            OMPC_dist_schedule, Node->getDistScheduleKind());
   if (auto *E = Node->getChunkSize()) {

lib/Basic/OpenMPKinds.cpp

#include "clang/Basic/OpenMPKinds.def"
   case OMPC_map:
     return llvm::StringSwitch<unsigned>(Str)
 #define OPENMP_MAP_KIND(Name)                                                  \
   .Case(#Name, static_cast<unsigned>(OMPC_MAP_##Name))
 #define OPENMP_MAP_MODIFIER_KIND(Name)                                         \
   .Case(#Name, static_cast<unsigned>(OMPC_MAP_MODIFIER_##Name))
 #include "clang/Basic/OpenMPKinds.def"
         .Default(OMPC_MAP_unknown);
+  case OMPC_to:
+    return llvm::StringSwitch<unsigned>(Str)
+#define OPENMP_TO_MODIFIER_KIND(Name)                                          \
+  .Case(#Name, static_cast<unsigned>(OMPC_TO_MODIFIER_##Name))
+#include "clang/Basic/OpenMPKinds.def"
+        .Default(OMPC_TO_MODIFIER_unknown);
+  case OMPC_from:
+    return llvm::StringSwitch<unsigned>(Str)
+#define OPENMP_FROM_MODIFIER_KIND(Name)                                     \
+  .Case(#Name, static_cast<unsigned>(OMPC_FROM_MODIFIER_##Name))
+#include "clang/Basic/OpenMPKinds.def"
+        .Default(OMPC_FROM_MODIFIER_unknown);
   case OMPC_dist_schedule:
     return llvm::StringSwitch<OpenMPDistScheduleClauseKind>(Str)
 #define OPENMP_DIST_SCHEDULE_KIND(Name) .Case(#Name, OMPC_DIST_SCHEDULE_##Name)
 #include "clang/Basic/OpenMPKinds.def"
         .Default(OMPC_DIST_SCHEDULE_unknown);
   case OMPC_defaultmap:
     return llvm::StringSwitch<unsigned>(Str)
 #define OPENMP_DEFAULTMAP_KIND(Name)                                           \
#include "clang/Basic/OpenMPKinds.def"
   case OMPC_num_teams:
   case OMPC_thread_limit:
   case OMPC_priority:
   case OMPC_grainsize:
   case OMPC_nogroup:
   case OMPC_num_tasks:
   case OMPC_hint:
   case OMPC_uniform:
-  case OMPC_to:
-  case OMPC_from:
   case OMPC_use_device_ptr:
   case OMPC_is_device_ptr:
   case OMPC_unified_address:
   case OMPC_unified_shared_memory:
   case OMPC_reverse_offload:
   case OMPC_dynamic_allocators:
     break;
   }

 #define OPENMP_MAP_MODIFIER_KIND(Name)                                       \
   case OMPC_MAP_MODIFIER_##Name:                                             \
     return #Name;
 #include "clang/Basic/OpenMPKinds.def"
     default:
       break;
     }
     llvm_unreachable("Invalid OpenMP 'map' clause type");
+  case OMPC_to:
+    switch (Type) {
+    case OMPC_TO_MODIFIER_unknown:
+      return "unknown";
+#define OPENMP_TO_MODIFIER_KIND(Name)                                          \
+  case OMPC_TO_MODIFIER_##Name:                                                \
+    return #Name;
+#include "clang/Basic/OpenMPKinds.def"
+    default:
+      break;
+    }
+    llvm_unreachable("Invalid OpenMP 'to' clause type");
+  case OMPC_from:
+    switch (Type) {
+    case OMPC_FROM_MODIFIER_unknown:
+      return "unknown";
+#define OPENMP_FROM_MODIFIER_KIND(Name)                                        \
+  case OMPC_FROM_MODIFIER_##Name:                                              \
+    return #Name;
+#include "clang/Basic/OpenMPKinds.def"
+    default:
+      break;
+    }
+    llvm_unreachable("Invalid OpenMP 'from' clause type");
   case OMPC_dist_schedule:
     switch (Type) {
     case OMPC_DIST_SCHEDULE_unknown:
       return "unknown";
 #define OPENMP_DIST_SCHEDULE_KIND(Name)                                      \
   case OMPC_DIST_SCHEDULE_##Name:                                            \
     return #Name;
 #include "clang/Basic/OpenMPKinds.def"
#include "clang/Basic/OpenMPKinds.def"
   case OMPC_num_teams:
   case OMPC_thread_limit:
   case OMPC_priority:
   case OMPC_grainsize:
   case OMPC_nogroup:
   case OMPC_num_tasks:
   case OMPC_hint:
   case OMPC_uniform:
-  case OMPC_to:
-  case OMPC_from:
   case OMPC_use_device_ptr:
   case OMPC_is_device_ptr:
   case OMPC_unified_address:
   case OMPC_unified_shared_memory:
   case OMPC_reverse_offload:
   case OMPC_dynamic_allocators:
     break;
   }

lib/Parse/ParseOpenMP.cpp

if (!Tok.is(tok::identifier))
     return OMPC_MAP_MODIFIER_unknown;
 
   Preprocessor &PP = P.getPreprocessor();
   OpenMPMapModifierKind TypeModifier = static_cast<OpenMPMapModifierKind>(
       getOpenMPSimpleClauseType(OMPC_map, PP.getSpelling(Tok)));
   return TypeModifier;
 }
 
-/// Parse map-type-modifiers in map clause.
-/// map([ [map-type-modifier[,] [map-type-modifier[,] ...] map-type : ] list)
-/// where, map-type-modifier ::= always | close | mapper(mapper-identifier)
-bool Parser::parseMapTypeModifiers(OpenMPVarListDataTy &Data) {
-  while (getCurToken().isNot(tok::colon)) {
-    OpenMPMapModifierKind TypeModifier = isMapModifier(*this);
-    if (TypeModifier == OMPC_MAP_MODIFIER_always ||
-        TypeModifier == OMPC_MAP_MODIFIER_close) {
-      Data.MapTypeModifiers.push_back(TypeModifier);
-      Data.MapTypeModifiersLoc.push_back(Tok.getLocation());
-      ConsumeToken();
-    } else if (TypeModifier == OMPC_MAP_MODIFIER_mapper) {
-      Data.MapTypeModifiers.push_back(TypeModifier);
-      Data.MapTypeModifiersLoc.push_back(Tok.getLocation());
-      ConsumeToken();
-      // Parse '('.
-      BalancedDelimiterTracker T(*this, tok::l_paren, tok::colon);
-      if (T.expectAndConsume(diag::err_expected_lparen_after,
-                             getOpenMPSimpleClauseTypeName(
-                                 OMPC_map, OMPC_MAP_MODIFIER_mapper))) {
-        SkipUntil(tok::colon, tok::annot_pragma_openmp_end, StopBeforeMatch);
-        return true;
-      }
-      // Parse mapper-identifier
-      if (getLangOpts().CPlusPlus)
-        ParseOptionalCXXScopeSpecifier(Data.ReductionOrMapperIdScopeSpec,
-                                       /*ObjectType=*/nullptr,
-                                       /*EnteringContext=*/false);
-      if (Tok.isNot(tok::identifier) && Tok.isNot(tok::kw_default)) {
-        Diag(Tok.getLocation(), diag::err_omp_mapper_illegal_identifier);
-        SkipUntil(tok::colon, tok::annot_pragma_openmp_end, StopBeforeMatch);
-        return true;
-      }
-      auto &DeclNames = Actions.getASTContext().DeclarationNames;
-      Data.ReductionOrMapperId = DeclarationNameInfo(
-          DeclNames.getIdentifier(Tok.getIdentifierInfo()), Tok.getLocation());
-      ConsumeToken();
-      // Parse ')'.
-      T.consumeClose();
+/// Parse the mapper modifier in map, to, and from clauses.
+bool Parser::parseMapperModifier(OpenMPVarListDataTy &Data) {
+  if (T.expectAndConsume(diag::err_expected_lparen_after, "mapper")) {
+    SkipUntil(tok::colon, tok::r_paren, tok::annot_pragma_openmp_end,
+              StopBeforeMatch);
+    SkipUntil(tok::colon, tok::r_paren, tok::annot_pragma_openmp_end,
+              StopBeforeMatch);
+  return T.consumeClose();
+}
+
+/// Parse map-type-modifiers in map clause.
+/// map([ [map-type-modifier[,] [map-type-modifier[,] ...] map-type : ] list)
+/// where, map-type-modifier ::= always | close | mapper(mapper-identifier)
+bool Parser::parseMapTypeModifiers(OpenMPVarListDataTy &Data) {
+  while (getCurToken().isNot(tok::colon)) {
+    OpenMPMapModifierKind TypeModifier = isMapModifier(*this);
+    if (TypeModifier == OMPC_MAP_MODIFIER_always ||
+        TypeModifier == OMPC_MAP_MODIFIER_close) {
+      Data.MapTypeModifiers.push_back(TypeModifier);
+      Data.MapTypeModifiersLoc.push_back(Tok.getLocation());
+      ConsumeToken();
+    } else if (TypeModifier == OMPC_MAP_MODIFIER_mapper) {
+      Data.MapTypeModifiers.push_back(TypeModifier);
+      Data.MapTypeModifiersLoc.push_back(Tok.getLocation());
+      ConsumeToken();
+      if (parseMapperModifier(Data))
+        return true;
     } else {
       // For the case of unknown map-type-modifier or a map-type.
       // Map-type is followed by a colon; the function returns when it
       // encounters a token followed by a colon.
       if (Tok.is(tok::comma)) {
         Diag(Tok, diag::err_omp_map_type_modifier_missing);
         ConsumeToken();
         continue;

 
 /// Parses clauses with list.
 bool Parser::ParseOpenMPVarList(OpenMPDirectiveKind DKind,
                                 OpenMPClauseKind Kind,
                                 SmallVectorImpl<Expr *> &Vars,
                                 OpenMPVarListDataTy &Data) {
   UnqualifiedId UnqualifiedReductionId;
   bool InvalidReductionId = false;
-  bool IsInvalidMapModifier = false;
+  bool IsInvalidMapperModifier = false;
 
   // Parse '('.
   BalancedDelimiterTracker T(*this, tok::l_paren, tok::annot_pragma_openmp_end);
   if (T.expectAndConsume(diag::err_expected_lparen_after,
                          getOpenMPClauseName(Kind)))
     return true;
 
   bool NeedRParenForLinear = false;
if (SkipUntil(tok::colon, tok::r_paren, tok::annot_pragma_openmp_end,
         StopBeforeMatch)) {
       if (Tok.is(tok::colon))
         ColonPresent = true;
     }
     TPA.Revert();
     // Only parse map-type-modifier[s] and map-type if a colon is present in
     // the map clause.
     if (ColonPresent) {
-      IsInvalidMapModifier = parseMapTypeModifiers(Data);
-      if (!IsInvalidMapModifier)
+      IsInvalidMapperModifier = parseMapTypeModifiers(Data);
+      if (!IsInvalidMapperModifier)
         parseMapType(*this, Data);
+      else
+        SkipUntil(tok::colon, tok::annot_pragma_openmp_end, StopBeforeMatch);
     }
     if (Data.MapType == OMPC_MAP_unknown) {
       Data.MapType = OMPC_MAP_tofrom;
       Data.IsMapTypeImplicit = true;
     }
 
     if (Tok.is(tok::colon))
       Data.ColonLoc = ConsumeToken();
+  } else if (Kind == OMPC_to || Kind == OMPC_from) {
+    if (Tok.is(tok::identifier)) {
+      bool IsMapperModifier = false;
+      if (Kind == OMPC_to) {
+        auto Modifier = static_cast<OpenMPToModifierKind>(
+            getOpenMPSimpleClauseType(Kind, PP.getSpelling(Tok)));
+        if (Modifier == OMPC_TO_MODIFIER_mapper)
+          IsMapperModifier = true;
+      } else {
+        auto Modifier = static_cast<OpenMPFromModifierKind>(
+            getOpenMPSimpleClauseType(Kind, PP.getSpelling(Tok)));
+        if (Modifier == OMPC_FROM_MODIFIER_mapper)
+          IsMapperModifier = true;
+      }
+      if (IsMapperModifier) {
+        // Parse the mapper modifier.
+        ConsumeToken();
+        IsInvalidMapperModifier = parseMapperModifier(Data);
+        if (Tok.isNot(tok::colon)) {
+          if (!IsInvalidMapperModifier)
+            Diag(Tok, diag::warn_pragma_expected_colon) << ")";
+          SkipUntil(tok::colon, tok::r_paren, tok::annot_pragma_openmp_end,
+                    StopBeforeMatch);
+        }
+        // Consume ':'.
+        if (Tok.is(tok::colon))
+          ConsumeToken();
+      }
+    }
   }
 
   bool IsComma =
       (Kind != OMPC_reduction && Kind != OMPC_task_reduction &&
        Kind != OMPC_in_reduction && Kind != OMPC_depend && Kind != OMPC_map) ||
       (Kind == OMPC_reduction && !InvalidReductionId) ||
       (Kind == OMPC_map && Data.MapType != OMPC_MAP_unknown) ||
       (Kind == OMPC_depend && Data.DepKind != OMPC_DEPEND_unknown);
bool Parser::ParseOpenMPVarList(OpenMPDirectiveKind DKind,
   // Parse ')'.
   Data.RLoc = Tok.getLocation();
   if (!T.consumeClose())
     Data.RLoc = T.getCloseLocation();
   return (Kind == OMPC_depend && Data.DepKind != OMPC_DEPEND_unknown &&
           Vars.empty()) ||
          (Kind != OMPC_depend && Kind != OMPC_map && Vars.empty()) ||
          (MustHaveTail && !Data.TailExpr) || InvalidReductionId ||
-         IsInvalidMapModifier;
+         IsInvalidMapperModifier;
 }
 
 /// Parsing of OpenMP clause 'private', 'firstprivate', 'lastprivate',
 /// 'shared', 'copyin', 'copyprivate', 'flush', 'reduction', 'task_reduction' or
 /// 'in_reduction'.
 ///
 ///    private-clause:
 ///       'private' '(' list ')'

 ///    copyprivate-clause:
 ///       'copyprivate' '(' list ')'
 ///    flush-clause:
 ///       'flush' '(' list ')'
 ///    depend-clause:
 ///       'depend' '(' in | out | inout : list | source ')'
 ///    map-clause:
 ///       'map' '(' [ [ always [,] ] [ close [,] ]
-///          [ mapper(mapper-identifier) [,] ]
+///          [ mapper '(' mapper-identifier ')' [,] ]
 ///          to | from | tofrom | alloc | release | delete ':' ] list ')';
 ///    to-clause:
-///       'to' '(' list ')'
+///       'to' '(' [ mapper '(' mapper-identifier ')' ':' ] list ')'
 ///    from-clause:
-///       'from' '(' list ')'
+///       'from' '(' [ mapper '(' mapper-identifier ')' ':' ] list ')'
 ///    use_device_ptr-clause:
 ///       'use_device_ptr' '(' list ')'
 ///    is_device_ptr-clause:
 ///       'is_device_ptr' '(' list ')'
 ///
 /// For 'linear' clause linear-list may have the following forms:
 ///  list
 ///  modifier(list)

lib/Sema/SemaOpenMP.cpp

case OMPC_depend:
     break;
   case OMPC_map:
     Res = ActOnOpenMPMapClause(MapTypeModifiers, MapTypeModifiersLoc,
                                ReductionOrMapperIdScopeSpec,
                                ReductionOrMapperId, MapType, IsMapTypeImplicit,
                                DepLinMapLoc, ColonLoc, VarList, Locs);
     break;
   case OMPC_to:
-    Res = ActOnOpenMPToClause(VarList, Locs);
+    Res = ActOnOpenMPToClause(VarList, ReductionOrMapperIdScopeSpec,
+                              ReductionOrMapperId, Locs);
     break;
   case OMPC_from:
-    Res = ActOnOpenMPFromClause(VarList, Locs);
+    Res = ActOnOpenMPFromClause(VarList, ReductionOrMapperIdScopeSpec,
+                                ReductionOrMapperId, Locs);
     break;
   case OMPC_use_device_ptr:
     Res = ActOnOpenMPUseDevicePtrClause(VarList, Locs);
     break;
   case OMPC_is_device_ptr:
     Res = ActOnOpenMPIsDevicePtrClause(VarList, Locs);
     break;
   case OMPC_if:
if (SemaRef.CurContext->isDependentContext() || Type->isDependentType() ||
     for (const UnresolvedSet<8> &Set : Lookups) {
       if (Set.empty())
         continue;
       URS.append(Set.begin(), Set.end());
     }
     return UnresolvedLookupExpr::Create(
         SemaRef.Context, /*NamingClass=*/nullptr,
         MapperIdScopeSpec.getWithLocInContext(SemaRef.Context), MapperId,
-        /*ADL=*/false, /*Overloaded=*/true, URS.begin(), URS.end());
+        /*ADL=*/true, /*Overloaded=*/true, URS.begin(), URS.end());
   }
   // [OpenMP 5.0], 2.19.7.3 declare mapper Directive, Restrictions
   //  The type must be of struct, union or class type in C and C++
   if (!Type->isStructureOrClassType() && !Type->isUnionType())
     return ExprEmpty();
   SourceLocation Loc = MapperId.getLoc();
   // Perform argument dependent lookup.
   if (SemaRef.getLangOpts().CPlusPlus && !MapperIdScopeSpec.isSet())

 // \a CKind. In the check process the valid expressions, mappable expression
 // components, variables, and user-defined mappers are extracted and used to
 // fill \a ProcessedVarList, \a VarComponents, \a VarBaseDeclarations, and \a
 // UDMapperList in MVLI. \a MapType, \a IsMapTypeImplicit, \a MapperIdScopeSpec,
 // and \a MapperId are expected to be valid if the clause kind is 'map'.
 static void checkMappableExpressionList(
     Sema &SemaRef, DSAStackTy *DSAS, OpenMPClauseKind CKind,
     MappableVarListInfo &MVLI, SourceLocation StartLoc,
+    CXXScopeSpec &MapperIdScopeSpec, DeclarationNameInfo MapperId,
+    ArrayRef<Expr *> UnresolvedMappers,
     OpenMPMapClauseKind MapType = OMPC_MAP_unknown,
-    bool IsMapTypeImplicit = false, CXXScopeSpec *MapperIdScopeSpec = nullptr,
-    const DeclarationNameInfo *MapperId = nullptr,
-    ArrayRef<Expr *> UnresolvedMappers = llvm::None) {
+    bool IsMapTypeImplicit = false) {
   // We only expect mappable expressions in 'to', 'from', and 'map' clauses.
   assert((CKind == OMPC_map || CKind == OMPC_to || CKind == OMPC_from) &&
          "Unexpected clause kind with mappable expressions!");
-  assert(
-      ((CKind == OMPC_map && MapperIdScopeSpec && MapperId) ||
-       (CKind != OMPC_map && !MapperIdScopeSpec && !MapperId)) &&
-      "Map clauses and only map clauses have user-defined mapper identifiers.");
+
+  // If the identifier of user-defined mapper is not specified, it is "default".
+  // We do not change the actual name in this clause to distinguish whether a
+  // mapper is specified explicitly, i.e., it is not explicitly specified when
+  // MapperId.getName() is empty.
+  if (!MapperId.getName() || MapperId.getName().isEmpty()) {
+    auto &DeclNames = SemaRef.getASTContext().DeclarationNames;
+    MapperId.setName(DeclNames.getIdentifier(
+        &SemaRef.getASTContext().Idents.get("default")));
+  }
 
   // Iterators to find the current unresolved mapper expression.
   auto UMIt = UnresolvedMappers.begin(), UMEnd = UnresolvedMappers.end();
   bool UpdateUMIt = false;
   Expr *UnresolvedMapper = nullptr;
 
   // Keep track of the mappable components and base declarations in this clause.
   // Each entry in the list is going to have a list of components associated. We
for (Expr *RE : MVLI.VarList) {
     if (UMIt != UMEnd)
       UnresolvedMapper = *UMIt;
 
     const Expr *VE = RE->IgnoreParenLValueCasts();
 
     if (VE->isValueDependent() || VE->isTypeDependent() ||
         VE->isInstantiationDependent() ||
         VE->containsUnexpandedParameterPack()) {
-      if (CKind == OMPC_map) {
-        // Try to find the associated user-defined mapper.
-        ExprResult ER = buildUserDefinedMapperRef(
-            SemaRef, DSAS->getCurScope(), *MapperIdScopeSpec, *MapperId,
-            VE->getType().getCanonicalType(), UnresolvedMapper);
-        if (ER.isInvalid())
-          continue;
-        MVLI.UDMapperList.push_back(ER.get());
-      } else {
-        MVLI.UDMapperList.push_back(nullptr);
-      }
+          SemaRef, DSAS->getCurScope(), MapperIdScopeSpec, MapperId,
       // We can only analyze this information once the missing information is
       // resolved.
       MVLI.ProcessedVarList.push_back(RE);
       continue;
     }
 
     Expr *SimpleExpr = RE->IgnoreParenCasts();
 
if (!BE)
       continue;
 
     assert(!CurComponents.empty() &&
            "Invalid mappable expression information.");
 
     if (const auto *TE = dyn_cast<CXXThisExpr>(BE)) {
       // Add store "this" pointer to class in DSAStackTy for future checking
       DSAS->addMappedClassesQualTypes(TE->getType());
-      if (CKind == OMPC_map) {
-        // Try to find the associated user-defined mapper.
-        ExprResult ER = buildUserDefinedMapperRef(
-            SemaRef, DSAS->getCurScope(), *MapperIdScopeSpec, *MapperId,
-            VE->getType().getCanonicalType(), UnresolvedMapper);
-        if (ER.isInvalid())
-          continue;
-        MVLI.UDMapperList.push_back(ER.get());
-      } else {
-        MVLI.UDMapperList.push_back(nullptr);
-      }
+          SemaRef, DSAS->getCurScope(), MapperIdScopeSpec, MapperId,
       // Skip restriction checking for variable or field declarations
       MVLI.ProcessedVarList.push_back(RE);
       MVLI.VarComponents.resize(MVLI.VarComponents.size() + 1);
       MVLI.VarComponents.back().append(CurComponents.begin(),
                                        CurComponents.end());
       MVLI.VarBaseDeclarations.push_back(nullptr);
       continue;
     }
if (CKind == OMPC_map) {
           SemaRef.Diag(ELoc, diag::err_omp_variable_in_given_clause_and_dsa)
               << getOpenMPClauseName(DVar.CKind)
               << getOpenMPClauseName(OMPC_map)
               << getOpenMPDirectiveName(DSAS->getCurrentDirective());
           reportOriginalDsa(SemaRef, DSAS, CurDeclaration, DVar);
           continue;
         }
       }
+    }
 
-      // Try to find the associated user-defined mapper.
-      ExprResult ER = buildUserDefinedMapperRef(
-          SemaRef, DSAS->getCurScope(), *MapperIdScopeSpec, *MapperId,
-          Type.getCanonicalType(), UnresolvedMapper);
-      if (ER.isInvalid())
-        continue;
-      MVLI.UDMapperList.push_back(ER.get());
-    } else {
-      MVLI.UDMapperList.push_back(nullptr);
-    }
+        SemaRef, DSAS->getCurScope(), MapperIdScopeSpec, MapperId,
 
     // Save the current expression.
     MVLI.ProcessedVarList.push_back(RE);
 
     // Store the components in the stack so that they can be used to check
     // against other clauses later on.
     DSAS->addMappableExpressionComponents(CurDeclaration, CurComponents,
                                           /*WhereFoundClauseKind=*/OMPC_map);
for (unsigned I = 0, E = MapTypeModifiers.size(); I < E; ++I) {
     }
     assert(Count < OMPMapClause::NumberOfModifiers &&
            "Modifiers exceed the allowed number of map type modifiers");
     Modifiers[Count] = MapTypeModifiers[I];
     ModifiersLoc[Count] = MapTypeModifiersLoc[I];
     ++Count;
   }
 
-  // If the identifier of user-defined mapper is not specified, it is "default".
-  if (!MapperId.getName() || MapperId.getName().isEmpty()) {
-    auto &DeclNames = getASTContext().DeclarationNames;
-    MapperId.setName(
-        DeclNames.getIdentifier(&getASTContext().Idents.get("default")));
-  }
-
   MappableVarListInfo MVLI(VarList);
   checkMappableExpressionList(*this, DSAStack, OMPC_map, MVLI, Locs.StartLoc,
-                              MapType, IsMapTypeImplicit, &MapperIdScopeSpec,
-                              &MapperId, UnresolvedMappers);
+                              MapperIdScopeSpec, MapperId, UnresolvedMappers,
+                              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.
   return OMPMapClause::Create(Context, Locs, MVLI.ProcessedVarList,
                               MVLI.VarBaseDeclarations, MVLI.VarComponents,
                               MVLI.UDMapperList, Modifiers, ModifiersLoc,
                               MapperIdScopeSpec.getWithLocInContext(Context),
                               MapperId, MapType, IsMapTypeImplicit, MapLoc);
if (auto *VD = dyn_cast<ValueDecl>(D)) {
     }
   }
   if (!E)
     return;
   checkDeclInTargetContext(E->getExprLoc(), E->getSourceRange(), *this, D);
 }
 
 OMPClause *Sema::ActOnOpenMPToClause(ArrayRef<Expr *> VarList,
-                                     const OMPVarListLocTy &Locs) {
+                                     CXXScopeSpec &MapperIdScopeSpec,
+                                     DeclarationNameInfo &MapperId,
+                                     const OMPVarListLocTy &Locs,
+                                     ArrayRef<Expr *> UnresolvedMappers) {
   MappableVarListInfo MVLI(VarList);
-  checkMappableExpressionList(*this, DSAStack, OMPC_to, MVLI, Locs.StartLoc);
+  checkMappableExpressionList(*this, DSAStack, OMPC_to, MVLI, Locs.StartLoc,
+                              MapperIdScopeSpec, MapperId, UnresolvedMappers);
   if (MVLI.ProcessedVarList.empty())
     return nullptr;
 
-  return OMPToClause::Create(Context, Locs, MVLI.ProcessedVarList,
-                             MVLI.VarBaseDeclarations, MVLI.VarComponents);
+  return OMPToClause::Create(
+      Context, Locs, MVLI.ProcessedVarList, MVLI.VarBaseDeclarations,
+      MVLI.VarComponents, MVLI.UDMapperList,
+      MapperIdScopeSpec.getWithLocInContext(Context), MapperId);
 }
 
 OMPClause *Sema::ActOnOpenMPFromClause(ArrayRef<Expr *> VarList,
-                                       const OMPVarListLocTy &Locs) {
+                                       CXXScopeSpec &MapperIdScopeSpec,
+                                       DeclarationNameInfo &MapperId,
+                                       const OMPVarListLocTy &Locs,
+                                       ArrayRef<Expr *> UnresolvedMappers) {
   MappableVarListInfo MVLI(VarList);
-  checkMappableExpressionList(*this, DSAStack, OMPC_from, MVLI, Locs.StartLoc);
+  checkMappableExpressionList(*this, DSAStack, OMPC_from, MVLI, Locs.StartLoc,
+                              MapperIdScopeSpec, MapperId, UnresolvedMappers);
   if (MVLI.ProcessedVarList.empty())
     return nullptr;
 
-  return OMPFromClause::Create(Context, Locs, MVLI.ProcessedVarList,
-                               MVLI.VarBaseDeclarations, MVLI.VarComponents);
+  return OMPFromClause::Create(
+      Context, Locs, MVLI.ProcessedVarList, MVLI.VarBaseDeclarations,
+      MVLI.VarComponents, MVLI.UDMapperList,
+      MapperIdScopeSpec.getWithLocInContext(Context), MapperId);
 }
 
 OMPClause *Sema::ActOnOpenMPUseDevicePtrClause(ArrayRef<Expr *> VarList,
                                                const OMPVarListLocTy &Locs) {
   MappableVarListInfo MVLI(VarList);
   SmallVector<Expr *, 8> PrivateCopies;
   SmallVector<Expr *, 8> Inits;
 

lib/Sema/TreeTransform.h

return getSema().ActOnOpenMPDistScheduleClause(
         Kind, ChunkSize, StartLoc, LParenLoc, KindLoc, CommaLoc, EndLoc);
   }
 
   /// 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<Expr *> VarList,
-                                const OMPVarListLocTy &Locs) {
-    return getSema().ActOnOpenMPToClause(VarList, Locs);
+                                CXXScopeSpec &MapperIdScopeSpec,
+                                DeclarationNameInfo &MapperId,
+                                const OMPVarListLocTy &Locs,
+                                ArrayRef<Expr *> UnresolvedMappers) {
+    return getSema().ActOnOpenMPToClause(VarList, MapperIdScopeSpec, MapperId,
+                                         Locs, UnresolvedMappers);
   }
 
   /// Build a new OpenMP 'from' clause.
   ///
   /// By default, performs semantic analysis to build the new statement.
   /// Subclasses may override this routine to provide different behavior.
   OMPClause *RebuildOMPFromClause(ArrayRef<Expr *> VarList,
-                                  const OMPVarListLocTy &Locs) {
-    return getSema().ActOnOpenMPFromClause(VarList, Locs);
+                                  CXXScopeSpec &MapperIdScopeSpec,
+                                  DeclarationNameInfo &MapperId,
+                                  const OMPVarListLocTy &Locs,
+                                  ArrayRef<Expr *> UnresolvedMappers) {
+    return getSema().ActOnOpenMPFromClause(VarList, MapperIdScopeSpec, MapperId,
+                                           Locs, UnresolvedMappers);
   }
 
   /// Build a new OpenMP 'use_device_ptr' clause.
   ///
   /// By default, performs semantic analysis to build the new OpenMP clause.
   /// Subclasses may override this routine to provide different behavior.
   OMPClause *RebuildOMPUseDevicePtrClause(ArrayRef<Expr *> VarList,
                                           const OMPVarListLocTy &Locs) {

 TreeTransform<Derived>::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());
 }
 
-template <typename Derived>
-OMPClause *TreeTransform<Derived>::TransformOMPMapClause(OMPMapClause *C) {
-  llvm::SmallVector<Expr *, 16> Vars;
+template <typename Derived, class T>
+bool transformOMPMappableExprListClause(
+    TreeTransform<Derived> &TT, OMPMappableExprListClause<T> *C,
+    llvm::SmallVector<Expr *, 16> &Vars, CXXScopeSpec &MapperIdScopeSpec,
+    DeclarationNameInfo &MapperIdInfo,
+    llvm::SmallVector<Expr *, 16> &UnresolvedMappers) {
+  // Transform expressions in the list.
   Vars.reserve(C->varlist_size());
   for (auto *VE : C->varlists()) {
-    ExprResult EVar = getDerived().TransformExpr(cast<Expr>(VE));
+    ExprResult EVar = TT.getDerived().TransformExpr(cast<Expr>(VE));
     if (EVar.isInvalid())
-      return nullptr;
+      return true;
     Vars.push_back(EVar.get());
   }
+  // Transform mapper scope specifier and identifier.
   NestedNameSpecifierLoc QualifierLoc;
   if (C->getMapperQualifierLoc()) {
-    QualifierLoc = getDerived().TransformNestedNameSpecifierLoc(
+    QualifierLoc = TT.getDerived().TransformNestedNameSpecifierLoc(
         C->getMapperQualifierLoc());
     if (!QualifierLoc)
-      return nullptr;
+      return true;
   }
-  CXXScopeSpec MapperIdScopeSpec;
   MapperIdScopeSpec.Adopt(QualifierLoc);
-  DeclarationNameInfo MapperIdInfo = C->getMapperIdInfo();
+  MapperIdInfo = C->getMapperIdInfo();
   if (MapperIdInfo.getName()) {
-    MapperIdInfo = getDerived().TransformDeclarationNameInfo(MapperIdInfo);
+    MapperIdInfo = TT.getDerived().TransformDeclarationNameInfo(MapperIdInfo);
     if (!MapperIdInfo.getName())
-      return nullptr;
+      return true;
   }
   // Build a list of all candidate OMPDeclareMapperDecls, which is provided by
   // the previous user-defined mapper lookup in dependent environment.
-  llvm::SmallVector<Expr *, 16> UnresolvedMappers;
   for (auto *E : C->mapperlists()) {
     // Transform all the decls.
     if (E) {
       auto *ULE = cast<UnresolvedLookupExpr>(E);
       UnresolvedSet<8> Decls;
       for (auto *D : ULE->decls()) {
         NamedDecl *InstD =
-            cast<NamedDecl>(getDerived().TransformDecl(E->getExprLoc(), D));
+            cast<NamedDecl>(TT.getDerived().TransformDecl(E->getExprLoc(), D));
         Decls.addDecl(InstD, InstD->getAccess());
       }
       UnresolvedMappers.push_back(UnresolvedLookupExpr::Create(
-          SemaRef.Context, /*NamingClass=*/nullptr,
-          MapperIdScopeSpec.getWithLocInContext(SemaRef.Context), MapperIdInfo,
-          /*ADL=*/false, ULE->isOverloaded(), Decls.begin(), Decls.end()));
+          TT.getSema().Context, /*NamingClass=*/nullptr,
+          MapperIdScopeSpec.getWithLocInContext(TT.getSema().Context),
+          MapperIdInfo, /*ADL=*/true, ULE->isOverloaded(), Decls.begin(),
+          Decls.end()));
     } else {
       UnresolvedMappers.push_back(nullptr);
     }
   }
+  return false;
+}
+
+template <typename Derived>
+OMPClause *TreeTransform<Derived>::TransformOMPMapClause(OMPMapClause *C) {
   OMPVarListLocTy Locs(C->getBeginLoc(), C->getLParenLoc(), C->getEndLoc());
+  llvm::SmallVector<Expr *, 16> Vars;
+  CXXScopeSpec MapperIdScopeSpec;
+  DeclarationNameInfo MapperIdInfo;
+  llvm::SmallVector<Expr *, 16> UnresolvedMappers;
+  if (transformOMPMappableExprListClause<Derived, OMPMapClause>(
+          *this, C, Vars, MapperIdScopeSpec, MapperIdInfo, UnresolvedMappers))
+    return nullptr;
   return getDerived().RebuildOMPMapClause(
       C->getMapTypeModifiers(), C->getMapTypeModifiersLoc(), MapperIdScopeSpec,
       MapperIdInfo, C->getMapType(), C->isImplicitMapType(), C->getMapLoc(),
       C->getColonLoc(), Vars, Locs, UnresolvedMappers);
 }
 
 template <typename Derived>
 OMPClause *

 template <typename Derived>
 OMPClause *
 TreeTransform<Derived>::TransformOMPDefaultmapClause(OMPDefaultmapClause *C) {
   return C;
 }
 
 template <typename Derived>
 OMPClause *TreeTransform<Derived>::TransformOMPToClause(OMPToClause *C) {
-  llvm::SmallVector<Expr *, 16> Vars;
-  Vars.reserve(C->varlist_size());
-  for (auto *VE : C->varlists()) {
-    ExprResult EVar = getDerived().TransformExpr(cast<Expr>(VE));
-    if (EVar.isInvalid())
-      return 0;
-    Vars.push_back(EVar.get());
-  }
   OMPVarListLocTy Locs(C->getBeginLoc(), C->getLParenLoc(), C->getEndLoc());
-  return getDerived().RebuildOMPToClause(Vars, Locs);
+  llvm::SmallVector<Expr *, 16> Vars;
+  CXXScopeSpec MapperIdScopeSpec;
+  DeclarationNameInfo MapperIdInfo;
+  llvm::SmallVector<Expr *, 16> UnresolvedMappers;
+  if (transformOMPMappableExprListClause<Derived, OMPToClause>(
+          *this, C, Vars, MapperIdScopeSpec, MapperIdInfo, UnresolvedMappers))
+    return nullptr;
+  return getDerived().RebuildOMPToClause(Vars, MapperIdScopeSpec, MapperIdInfo,
+                                         Locs, UnresolvedMappers);
 }
 
 template <typename Derived>
 OMPClause *TreeTransform<Derived>::TransformOMPFromClause(OMPFromClause *C) {
-  llvm::SmallVector<Expr *, 16> Vars;
-  Vars.reserve(C->varlist_size());
-  for (auto *VE : C->varlists()) {
-    ExprResult EVar = getDerived().TransformExpr(cast<Expr>(VE));
-    if (EVar.isInvalid())
-      return 0;
-    Vars.push_back(EVar.get());
-  }
   OMPVarListLocTy Locs(C->getBeginLoc(), C->getLParenLoc(), C->getEndLoc());
-  return getDerived().RebuildOMPFromClause(Vars, Locs);
+  llvm::SmallVector<Expr *, 16> Vars;
+  CXXScopeSpec MapperIdScopeSpec;
+  DeclarationNameInfo MapperIdInfo;
+  llvm::SmallVector<Expr *, 16> UnresolvedMappers;
+  if (transformOMPMappableExprListClause<Derived, OMPFromClause>(
+          *this, C, Vars, MapperIdScopeSpec, MapperIdInfo, UnresolvedMappers))
+    return nullptr;
+  return getDerived().RebuildOMPFromClause(
+      Vars, MapperIdScopeSpec, MapperIdInfo, Locs, UnresolvedMappers);
 }
 
 template <typename Derived>
 OMPClause *TreeTransform<Derived>::TransformOMPUseDevicePtrClause(
     OMPUseDevicePtrClause *C) {
   llvm::SmallVector<Expr *, 16> Vars;
   Vars.reserve(C->varlist_size());
   for (auto *VE : C->varlists()) {

lib/Serialization/ASTReader.cpp

C->setDefaultmapModifier(
       static_cast<OpenMPDefaultmapClauseModifier>(Record.readInt()));
   C->setLParenLoc(Record.readSourceLocation());
   C->setDefaultmapModifierLoc(Record.readSourceLocation());
   C->setDefaultmapKindLoc(Record.readSourceLocation());
 }
 
 void OMPClauseReader::VisitOMPToClause(OMPToClause *C) {
   C->setLParenLoc(Record.readSourceLocation());
+  C->setMapperQualifierLoc(Record.readNestedNameSpecifierLoc());
+  DeclarationNameInfo DNI;
+  Record.readDeclarationNameInfo(DNI);
+  C->setMapperIdInfo(DNI);
   auto NumVars = C->varlist_size();
   auto UniqueDecls = C->getUniqueDeclarationsNum();
   auto TotalLists = C->getTotalComponentListNum();
   auto TotalComponents = C->getTotalComponentsNum();
 
   SmallVector<Expr *, 16> Vars;
   Vars.reserve(NumVars);
   for (unsigned i = 0; i != NumVars; ++i)
     Vars.push_back(Record.readSubExpr());
   C->setVarRefs(Vars);
 
+  SmallVector<Expr *, 16> UDMappers;
+  UDMappers.reserve(NumVars);
+  for (unsigned I = 0; I < NumVars; ++I)
+    UDMappers.push_back(Record.readSubExpr());
+  C->setUDMapperRefs(UDMappers);
+
   SmallVector<ValueDecl *, 16> Decls;
   Decls.reserve(UniqueDecls);
   for (unsigned i = 0; i < UniqueDecls; ++i)
     Decls.push_back(Record.readDeclAs<ValueDecl>());
   C->setUniqueDecls(Decls);
 
   SmallVector<unsigned, 16> ListsPerDecl;
   ListsPerDecl.reserve(UniqueDecls);
for (unsigned i = 0; i < TotalComponents; ++i) {
     Components.push_back(OMPClauseMappableExprCommon::MappableComponent(
         AssociatedExpr, AssociatedDecl));
   }
   C->setComponents(Components, ListSizes);
 }
 
 void OMPClauseReader::VisitOMPFromClause(OMPFromClause *C) {
   C->setLParenLoc(Record.readSourceLocation());
+  C->setMapperQualifierLoc(Record.readNestedNameSpecifierLoc());
+  DeclarationNameInfo DNI;
+  Record.readDeclarationNameInfo(DNI);
+  C->setMapperIdInfo(DNI);
   auto NumVars = C->varlist_size();
   auto UniqueDecls = C->getUniqueDeclarationsNum();
   auto TotalLists = C->getTotalComponentListNum();
   auto TotalComponents = C->getTotalComponentsNum();
 
   SmallVector<Expr *, 16> Vars;
   Vars.reserve(NumVars);
   for (unsigned i = 0; i != NumVars; ++i)
     Vars.push_back(Record.readSubExpr());
   C->setVarRefs(Vars);
 
+  SmallVector<Expr *, 16> UDMappers;
+  UDMappers.reserve(NumVars);
+  for (unsigned I = 0; I < NumVars; ++I)
+    UDMappers.push_back(Record.readSubExpr());
+  C->setUDMapperRefs(UDMappers);
+
   SmallVector<ValueDecl *, 16> Decls;
   Decls.reserve(UniqueDecls);
   for (unsigned i = 0; i < UniqueDecls; ++i)
     Decls.push_back(Record.readDeclAs<ValueDecl>());
   C->setUniqueDecls(Decls);
 
   SmallVector<unsigned, 16> ListsPerDecl;
   ListsPerDecl.reserve(UniqueDecls);

lib/Serialization/ASTWriter.cpp

 }
 
 void OMPClauseWriter::VisitOMPToClause(OMPToClause *C) {
   Record.push_back(C->varlist_size());
   Record.push_back(C->getUniqueDeclarationsNum());
   Record.push_back(C->getTotalComponentListNum());
   Record.push_back(C->getTotalComponentsNum());
   Record.AddSourceLocation(C->getLParenLoc());
+  Record.AddNestedNameSpecifierLoc(C->getMapperQualifierLoc());
+  Record.AddDeclarationNameInfo(C->getMapperIdInfo());
   for (auto *E : C->varlists())
     Record.AddStmt(E);
+  for (auto *E : C->mapperlists())
+    Record.AddStmt(E);
   for (auto *D : C->all_decls())
     Record.AddDeclRef(D);
   for (auto N : C->all_num_lists())
     Record.push_back(N);
   for (auto N : C->all_lists_sizes())
     Record.push_back(N);
   for (auto &M : C->all_components()) {
     Record.AddStmt(M.getAssociatedExpression());
     Record.AddDeclRef(M.getAssociatedDeclaration());
   }
 }
 
 void OMPClauseWriter::VisitOMPFromClause(OMPFromClause *C) {
   Record.push_back(C->varlist_size());
   Record.push_back(C->getUniqueDeclarationsNum());
   Record.push_back(C->getTotalComponentListNum());
   Record.push_back(C->getTotalComponentsNum());
   Record.AddSourceLocation(C->getLParenLoc());
+  Record.AddNestedNameSpecifierLoc(C->getMapperQualifierLoc());
+  Record.AddDeclarationNameInfo(C->getMapperIdInfo());
   for (auto *E : C->varlists())
     Record.AddStmt(E);
+  for (auto *E : C->mapperlists())
+    Record.AddStmt(E);
   for (auto *D : C->all_decls())
     Record.AddDeclRef(D);
   for (auto N : C->all_num_lists())
     Record.push_back(N);
   for (auto N : C->all_lists_sizes())
     Record.push_back(N);
   for (auto &M : C->all_components()) {
     Record.AddStmt(M.getAssociatedExpression());

test/OpenMP/declare_mapper_ast_print.c

// CHECK: #pragma omp declare mapper (id : struct vec v) map(tofrom: v.len)
     struct vec vv;
     struct dat dd[10];
 #pragma omp target map(mapper(id) alloc: vv)
 // CHECK: #pragma omp target map(mapper(id),alloc: vv)
     { vv.len++; }
 #pragma omp target map(mapper(default), from: dd[0:10])
 // CHECK: #pragma omp target map(mapper(default),from: dd[0:10])
     { dd[0].i++; }
+#pragma omp target update to(mapper(id): vv) from(mapper(default): dd[0:10])
+// CHECK: #pragma omp target update to(mapper(id): vv) from(mapper(default): dd[0:10])
   }
   return 0;
 }
 // CHECK: }
 
 #endif

test/OpenMP/declare_mapper_ast_print.cpp

#pragma omp declare mapper(id: N1::vec v) map(v.len)
   {
 #pragma omp declare mapper(id: N1::vec v) map(v.len)
   }
   struct foodat fd;
 #pragma omp target map(mapper(id) alloc: fd)
   { fd.a++; }
 #pragma omp target map(mapper(idd) alloc: fd.b)
   { fd.b.k++; }
+#pragma omp target update to(mapper(id): fd)
+#pragma omp target update to(mapper(idd): fd.b)
+#pragma omp target update from(mapper(id): fd)
+#pragma omp target update from(mapper(idd): fd.b)
   return 0;
 }
 
 // CHECK: template <typename T> T foo(T a) {
 // CHECK: #pragma omp declare mapper (id : struct foodat v) map(tofrom: v.a)
 // CHECK: #pragma omp declare mapper (idd : struct foodatchild v) map(tofrom: v.k)
 // CHECK: #pragma omp declare mapper (id : N1::vec v) map(tofrom: v.len)
 // CHECK: {
 // CHECK: #pragma omp declare mapper (id : N1::vec v) map(tofrom: v.len)
 // CHECK: }
 // CHECK: #pragma omp target map(mapper(id),alloc: fd)
 // CHECK: #pragma omp target map(mapper(idd),alloc: fd.b)
+// CHECK: #pragma omp target update to(mapper(id): fd)
+// CHECK: #pragma omp target update to(mapper(idd): fd.b)
+// CHECK: #pragma omp target update from(mapper(id): fd)
+// CHECK: #pragma omp target update from(mapper(idd): fd.b)
 // CHECK: }
 // CHECK: template<> int foo<int>(int a) {
 // CHECK: #pragma omp declare mapper (id : struct foodat v) map(tofrom: v.a)
 // CHECK: #pragma omp declare mapper (idd : struct foodatchild v) map(tofrom: v.k)
 // CHECK: #pragma omp declare mapper (id : N1::vec v) map(tofrom: v.len)
 // CHECK: {
 // CHECK: #pragma omp declare mapper (id : N1::vec v) map(tofrom: v.len)
 // CHECK: }
 // CHECK: #pragma omp target map(mapper(id),alloc: fd)
 // CHECK: #pragma omp target map(mapper(idd),alloc: fd.b)
+// CHECK: #pragma omp target update to(mapper(id): fd)
+// CHECK: #pragma omp target update to(mapper(idd): fd.b)
+// CHECK: #pragma omp target update from(mapper(id): fd)
+// CHECK: #pragma omp target update from(mapper(idd): fd.b)
 // CHECK: }
 
 // CHECK: int main() {
 int main() {
   N1::vec vv, vvv;
   N1::vecchild vc;
   dat<double> dd;
 #pragma omp target map(mapper(N1::id) tofrom: vv) map(mapper(dat<double>::id) alloc: vvv)
 // CHECK: #pragma omp target map(mapper(N1::id),tofrom: vv) map(mapper(dat<double>::id),alloc: vvv)
   { vv.len++; }
 #pragma omp target map(mapper(N1::id) tofrom: vc)
 // CHECK: #pragma omp target map(mapper(N1::id),tofrom: vc)
   { vc.len++; }
 #pragma omp target map(mapper(default) tofrom: dd)
 // CHECK: #pragma omp target map(mapper(default),tofrom: dd)
   { dd.d++; }
+#pragma omp target update to(mapper(N1::id) : vc)
+// CHECK: #pragma omp target update to(mapper(N1::id): vc)
+#pragma omp target update from(mapper(N1::id) : vc)
+// CHECK: #pragma omp target update from(mapper(N1::id): vc)
+#pragma omp target update to(mapper(dat<double>::id): vvv)
+// CHECK: #pragma omp target update to(mapper(dat<double>::id): vvv)
+#pragma omp target update from(mapper(dat<double>::id): vvv)
+// CHECK: #pragma omp target update from(mapper(dat<double>::id): vvv)
 #pragma omp declare mapper(id: N1::vec v) map(v.len)
 // CHECK: #pragma omp declare mapper (id : N1::vec v) map(tofrom: v.len)
   {
 #pragma omp declare mapper(id: N1::vec v) map(v.len)
 // CHECK: #pragma omp declare mapper (id : N1::vec v) map(tofrom: v.len)
   }
   return foo<int>(0);
 }
 // CHECK: }
 
 #endif

test/OpenMP/declare_mapper_codegen.cpp

 
 class C {
 public:
   int a;
 };
 
 #pragma omp declare mapper(id: C s) map(s.a)
 
-// CHECK-LABEL: @.__omp_offloading_{{.*}}foo{{.*}}_l49.region_id = weak constant i8 0
+// CHECK-LABEL: @.__omp_offloading_{{.*}}foo{{.*}}_l54.region_id = weak constant i8 0
 
-// CHECK-DAG: [[SIZES:@.+]] = {{.+}}constant [1 x i{{64|32}}] [i{{64|32}} 4]
-// CHECK-DAG: [[TYPES:@.+]] = {{.+}}constant [1 x i64] [i64 35]
+// CHECK: [[SIZES:@.+]] = {{.+}}constant [1 x i[[sz:64|32]]] [i{{64|32}} 4]
+// CHECK: [[TYPES:@.+]] = {{.+}}constant [1 x i64] [i64 35]
+// CHECK: [[FSIZES:@.+]] = {{.+}}constant [1 x i[[sz]]] [i[[sz]] 4]
+// CHECK: [[FTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 34]
+// CHECK: [[TSIZES:@.+]] = {{.+}}constant [1 x i[[sz]]] [i[[sz]] 4]
+// CHECK: [[TTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 33]
 
 // CHECK-LABEL: foo{{.*}}(
 void foo(int a){
   int i = a;
   C c;
   c.a = a;
 
   // CHECK-DAG: call i32 @__tgt_target(i64 {{.+}}, i8* {{.+}}, i32 1, i8** [[BPGEP:%[0-9]+]], i8** [[PGEP:%[0-9]+]], {{.+}}[[SIZES]]{{.+}}, {{.+}}[[TYPES]]{{.+}})
   // CHECK-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BPS:%[^,]+]], i32 0, i32 0
   // CHECK-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[PS:%[^,]+]], i32 0, i32 0
   // CHECK-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BPS]], i32 0, i32 0
   // CHECK-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[PS]], i32 0, i32 0
   // CHECK-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to %class.C**
   // CHECK-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to %class.C**
   // CHECK-DAG: store %class.C* [[VAL:%[^,]+]], %class.C** [[CBP1]]
   // CHECK-DAG: store %class.C* [[VAL]], %class.C** [[CP1]]
+  // CHECK: call void [[KERNEL:@.+]](%class.C* [[VAL]])
   #pragma omp target map(mapper(id),tofrom: c)
   {
    ++c.a;
   }
-  // CHECK: call void [[KERNEL:@.+]](%class.C* [[VAL]])
+
+  // CHECK-DAG: call void @__tgt_target_data_update(i64 -1, i32 1, i8** [[FGEPBP:%.+]], i8** [[FGEPP:%.+]], i[[sz]]* getelementptr {{.+}}[1 x i[[sz]]]* [[FSIZES]], i32 0, i32 0), {{.+}}getelementptr {{.+}}[1 x i64]* [[FTYPES]]{{.+}})
+  // CHECK-DAG: [[FGEPBP]] = getelementptr inbounds {{.+}}[[FBP:%[^,]+]], i{{.+}} 0, i{{.+}} 0
+  // CHECK-DAG: [[FGEPP]] = getelementptr inbounds {{.+}}[[FP:%[^,]+]], i{{.+}} 0, i{{.+}} 0
+  // CHECK-DAG: [[FBP0:%.+]] = getelementptr inbounds {{.+}}[[FBP]], i{{.+}} 0, i{{.+}} 0
+  // CHECK-DAG: [[FP0:%.+]] = getelementptr inbounds {{.+}}[[FP]], i{{.+}} 0, i{{.+}} 0
+  // CHECK-DAG: [[FCBP0:%.+]] = bitcast i8** [[FBP0]] to %class.C**
+  // CHECK-DAG: [[FCP0:%.+]] = bitcast i8** [[FP0]] to %class.C**
+  // CHECK-DAG: store %class.C* [[VAL]], %class.C** [[FCBP0]]
+  // CHECK-DAG: store %class.C* [[VAL]], %class.C** [[FCP0]]
+  #pragma omp target update from(mapper(id): c)
+
+  // CHECK-DAG: call void @__tgt_target_data_update(i64 -1, i32 1, i8** [[TGEPBP:%.+]], i8** [[TGEPP:%.+]], i[[sz]]* getelementptr {{.+}}[1 x i[[sz]]]* [[TSIZES]], i32 0, i32 0), {{.+}}getelementptr {{.+}}[1 x i64]* [[TTYPES]]{{.+}})
+  // CHECK-DAG: [[TGEPBP]] = getelementptr inbounds {{.+}}[[TBP:%[^,]+]], i{{.+}} 0, i{{.+}} 0
+  // CHECK-DAG: [[TGEPP]] = getelementptr inbounds {{.+}}[[TP:%[^,]+]], i{{.+}} 0, i{{.+}} 0
+  // CHECK-DAG: [[TBP0:%.+]] = getelementptr inbounds {{.+}}[[TBP]], i{{.+}} 0, i{{.+}} 0
+  // CHECK-DAG: [[TP0:%.+]] = getelementptr inbounds {{.+}}[[TP]], i{{.+}} 0, i{{.+}} 0
+  // CHECK-DAG: [[TCBP0:%.+]] = bitcast i8** [[TBP0]] to %class.C**
+  // CHECK-DAG: [[TCP0:%.+]] = bitcast i8** [[TP0]] to %class.C**
+  // CHECK-DAG: store %class.C* [[VAL]], %class.C** [[TCBP0]]
+  // CHECK-DAG: store %class.C* [[VAL]], %class.C** [[TCP0]]
+  #pragma omp target update to(mapper(id): c)
 }
 
 
 // CHECK: define internal void [[KERNEL]](%class.C* {{.+}}[[ARG:%.+]])
 // CHECK: [[ADDR:%.+]] = alloca %class.C*,
 // CHECK: store %class.C* [[ARG]], %class.C** [[ADDR]]
 // CHECK: [[CADDR:%.+]] = load %class.C*, %class.C** [[ADDR]]
 // CHECK: [[CAADDR:%.+]] = getelementptr inbounds %class.C, %class.C* [[CADDR]], i32 0, i32 0
 // CHECK: [[VAL:%[^,]+]] = load i32, i32* [[CAADDR]]
 // CHECK: {{.+}} = add nsw i32 [[VAL]], 1
 // CHECK: }
 
 #endif

test/OpenMP/declare_mapper_messages.c

 #pragma omp declare mapper(struct v                                     // expected-error {{expected declarator on 'omp declare mapper' directive}}
 #pragma omp declare mapper(struct vec                                   // expected-error {{expected declarator on 'omp declare mapper' directive}}
 #pragma omp declare mapper(S v                                          // expected-error {{unknown type name 'S'}}
 #pragma omp declare mapper(struct vec v                                 // expected-error {{expected ')'}} expected-note {{to match this '('}}
 #pragma omp declare mapper(aa:struct vec v)                             // expected-error {{expected at least one clause on '#pragma omp declare mapper' directive}}
 #pragma omp declare mapper(bb:struct vec v) private(v)                  // expected-error {{expected at least one clause on '#pragma omp declare mapper' directive}} // expected-error {{unexpected OpenMP clause 'private' in directive '#pragma omp declare mapper'}}
 #pragma omp declare mapper(cc:struct vec v) map(v) (                    // expected-warning {{extra tokens at the end of '#pragma omp declare mapper' are ignored}}
 
-#pragma omp declare mapper(++: struct vec v) map(v.len)                 // expected-error {{illegal identifier on 'omp declare mapper' directive}}
+#pragma omp declare mapper(++: struct vec v) map(v.len)                 // expected-error {{illegal OpenMP user-defined mapper identifier}}
 #pragma omp declare mapper(id1: struct vec v) map(v.len, temp)          // expected-error {{only variable v is allowed in map clauses of this 'omp declare mapper' directive}}
 #pragma omp declare mapper(default : struct vec kk) map(kk.data[0:2])   // expected-note {{previous definition is here}}
 #pragma omp declare mapper(struct vec v) map(v.len)                     // expected-error {{redefinition of user-defined mapper for type 'struct vec' with name 'default'}}
 #pragma omp declare mapper(int v) map(v)                                // expected-error {{mapper type must be of struct, union or class type}}
 
 int fun(int arg) {
 #pragma omp declare mapper(id: struct vec v) map(v.len)
   {

 #pragma omp target map(mapper(aa :vv)                                   // expected-error {{use of undeclared identifier 'aa'}} expected-error {{expected ')'}} expected-warning {{implicit declaration of function 'mapper' is invalid in C99}} expected-note {{to match this '('}}
       {}
 #pragma omp target map(mapper(ab) :vv)                                  // expected-error {{missing map type}} expected-error {{cannot find a valid user-defined mapper for type 'struct vec' with name 'ab'}}
       {}
 #pragma omp target map(mapper(aa) :vv)                                  // expected-error {{missing map type}}
       {}
 #pragma omp target map(mapper(aa) to:vv) map(close mapper(aa) from:v1)
       {}
+
+#pragma omp target update to(mapper)                                    // expected-error {{expected '(' after 'mapper'}} expected-error {{expected expression}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(mapper()                                   // expected-error {{illegal OpenMP user-defined mapper identifier}} expected-error {{expected expression}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(mapper:vv)                                 // expected-error {{expected '(' after 'mapper'}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(mapper(:vv)                                // expected-error {{illegal OpenMP user-defined mapper identifier}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(mapper(aa :vv)                             // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(mapper(ab):vv)                             // expected-error {{cannot find a valid user-defined mapper for type 'struct vec' with name 'ab'}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(mapper(aa):vv)
+
+#pragma omp target update from(mapper)                                  // expected-error {{expected '(' after 'mapper'}} expected-error {{expected expression}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(mapper()                                 // expected-error {{illegal OpenMP user-defined mapper identifier}} expected-error {{expected expression}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(mapper:vv)                               // expected-error {{expected '(' after 'mapper'}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(mapper(:vv)                              // expected-error {{illegal OpenMP user-defined mapper identifier}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(mapper(aa :vv)                           // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(mapper(ab):vv)                           // expected-error {{cannot find a valid user-defined mapper for type 'struct vec' with name 'ab'}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(mapper(aa) a:vv)                         // expected-warning {{missing ':' after ) - ignoring}}
+#pragma omp target update from(mapper(aa):vv)
     }
   }
   return arg;
 }

test/OpenMP/declare_mapper_messages.cpp

 #pragma omp declare mapper(v                                            // expected-error {{unknown type name 'v'}} expected-error {{expected declarator on 'omp declare mapper' directive}}
 #pragma omp declare mapper(vec                                          // expected-error {{expected declarator on 'omp declare mapper' directive}}
 #pragma omp declare mapper(S v                                          // expected-error {{unknown type name 'S'}}
 #pragma omp declare mapper(vec v                                        // expected-error {{expected ')'}} expected-note {{to match this '('}}
 #pragma omp declare mapper(aa: vec v)                                   // expected-error {{expected at least one clause on '#pragma omp declare mapper' directive}}
 #pragma omp declare mapper(bb: vec v) private(v)                        // expected-error {{expected at least one clause on '#pragma omp declare mapper' directive}} // expected-error {{unexpected OpenMP clause 'private' in directive '#pragma omp declare mapper'}}
 #pragma omp declare mapper(cc: vec v) map(v) (                          // expected-warning {{extra tokens at the end of '#pragma omp declare mapper' are ignored}}
 
-#pragma omp declare mapper(++: vec v) map(v.len)                        // expected-error {{illegal identifier on 'omp declare mapper' directive}}
+#pragma omp declare mapper(++: vec v) map(v.len)                        // expected-error {{illegal OpenMP user-defined mapper identifier}}
 #pragma omp declare mapper(id1: vec v) map(v.len, temp)                 // expected-error {{only variable v is allowed in map clauses of this 'omp declare mapper' directive}}
 #pragma omp declare mapper(default : vec kk) map(kk.data[0:2])          // expected-note {{previous definition is here}}
 #pragma omp declare mapper(vec v) map(v.len)                            // expected-error {{redefinition of user-defined mapper for type 'vec' with name 'default'}}
 #pragma omp declare mapper(int v) map(v)                                // expected-error {{mapper type must be of struct, union or class type}}
 #pragma omp declare mapper(id2: vec v) map(v.len, v.p)                  // expected-error {{'p' is a private member of 'vec'}}
 
 namespace N1 {
 template <class T>

 #pragma omp target map(mapper:vv)                                       // expected-error {{expected '(' after 'mapper'}}
       {}
 #pragma omp target map(mapper( :vv)                                     // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}}
       {}
 #pragma omp target map(mapper(aa :vv)                                   // expected-error {{use of undeclared identifier 'aa'}} expected-error {{expected ')'}} expected-note {{to match this '('}}
       {}
 #pragma omp target map(mapper(ab) :vv)                                  // expected-error {{missing map type}} expected-error {{cannot find a valid user-defined mapper for type 'vec' with name 'ab'}}
       {}
-#pragma omp target map(mapper(N2::) :vv)                                // expected-error {{use of undeclared identifier 'N2'}} expected-error {{illegal identifier on 'omp declare mapper' directive}}
+#pragma omp target map(mapper(N2::) :vv)                                // expected-error {{use of undeclared identifier 'N2'}} expected-error {{illegal OpenMP user-defined mapper identifier}}
       {}
-#pragma omp target map(mapper(N1::) :vv)                                // expected-error {{illegal identifier on 'omp declare mapper' directive}}
+#pragma omp target map(mapper(N1::) :vv)                                // expected-error {{illegal OpenMP user-defined mapper identifier}}
       {}
 #pragma omp target map(mapper(aa) :vv)                                  // expected-error {{missing map type}}
       {}
 #pragma omp target map(mapper(N1::aa) alloc:vv)                         // expected-error {{cannot find a valid user-defined mapper for type 'vec' with name 'aa'}}
       {}
 #pragma omp target map(mapper(aa) to:vv) map(close mapper(aa) from:v1)
       {}
 #pragma omp target map(mapper(N1::stack<int>::id) to:vv)
       {}
+
+#pragma omp target update to(mapper)                                    // expected-error {{expected '(' after 'mapper'}} expected-error {{expected expression}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(mapper()                                   // expected-error {{illegal OpenMP user-defined mapper identifier}} expected-error {{expected expression}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(mapper:vv)                                 // expected-error {{expected '(' after 'mapper'}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(mapper(:vv)                                // expected-error {{illegal OpenMP user-defined mapper identifier}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(mapper(aa :vv)                             // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(mapper(N2:: :vv)                           // expected-error {{use of undeclared identifier 'N2'}} expected-error {{illegal OpenMP user-defined mapper identifier}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(mapper(N1:: :vv)                           // expected-error {{illegal OpenMP user-defined mapper identifier}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(mapper(N1::aa) :vv)                        // expected-error {{cannot find a valid user-defined mapper for type 'vec' with name 'aa'}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(mapper(ab):vv)                             // expected-error {{cannot find a valid user-defined mapper for type 'vec' with name 'ab'}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(mapper(aa) a:vv)                           // expected-warning {{missing ':' after ) - ignoring}}
+#pragma omp target update to(mapper(aa):vv)
+#pragma omp target update to(mapper(N1::stack<int>::id) :vv)
+
+#pragma omp target update from(mapper)                                  // expected-error {{expected '(' after 'mapper'}} expected-error {{expected expression}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(mapper()                                 // expected-error {{illegal OpenMP user-defined mapper identifier}} expected-error {{expected expression}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(mapper:vv)                               // expected-error {{expected '(' after 'mapper'}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(mapper(:vv)                              // expected-error {{illegal OpenMP user-defined mapper identifier}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(mapper(aa :vv)                           // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(mapper(N2:: :vv)                         // expected-error {{use of undeclared identifier 'N2'}} expected-error {{illegal OpenMP user-defined mapper identifier}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(mapper(N1:: :vv)                         // expected-error {{illegal OpenMP user-defined mapper identifier}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(mapper(N1::aa) :vv)                      // expected-error {{cannot find a valid user-defined mapper for type 'vec' with name 'aa'}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(mapper(ab):vv)                           // expected-error {{cannot find a valid user-defined mapper for type 'vec' with name 'ab'}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(mapper(aa) a:vv)                         // expected-warning {{missing ':' after ) - ignoring}}
+#pragma omp target update from(mapper(aa):vv)
+#pragma omp target update from(mapper(N1::stack<int>::id) :vv)
     }
 #pragma omp declare mapper(id: vec v) map(v.len)                        // expected-error {{redefinition of user-defined mapper for type 'vec' with name 'id'}}
   }
   return arg;
 }