[OpenMP] Update map clause SEMA to support OpenMP 4.5 possible list items.

Summary:
Extend support in the map clause SEMA for the expressions supported in the OpenMP 4.5 specification, namely member expressions.  

Fix some bugs in the previous implementation of SEMA related with expressions that do not consist of single variable references.

Fix bug in parsing when the expression in the map clause do not start with an identifier: accept any expression in the map clause and check for validity in SEMA instead of just ignoring it.

Reviewers: hfinkel, kkwli0, arpith-jacob, carlo.bertolli, ABataev

Subscribers: cfe-commits, fraggamuffin, caomhin

Differential Revision: http://reviews.llvm.org/D16385

llvm-svn: 258543

1 files changed, 467 insertions, 97 deletions

diff --git a/clang/lib/Sema/SemaOpenMP.cpp b/clang/lib/Sema/SemaOpenMP.cpp
index 40c8c18a50d..3157da4c208 100644
--- a/clang/lib/Sema/SemaOpenMP.cpp
+++ b/clang/lib/Sema/SemaOpenMP.cpp
@@ -77,12 +77,9 @@ public:
           ImplicitDSALoc() {}
   };
 
-public:
-  struct MapInfo {
-    Expr *RefExpr;
-  };
-
 private:
+  typedef SmallVector<Expr *, 4> MapInfo;
+
   struct DSAInfo {
     OpenMPClauseKind Attributes;
     Expr *RefExpr;
@@ -337,30 +334,39 @@ public:
   Scope *getCurScope() { return Stack.back().CurScope; }
   SourceLocation getConstructLoc() { return Stack.back().ConstructLoc; }
 
-  MapInfo getMapInfoForVar(ValueDecl *VD) {
-    MapInfo VarMI = {0};
-    for (auto Cnt = Stack.size() - 1; Cnt > 0; --Cnt) {
-      if (Stack[Cnt].MappedDecls.count(VD)) {
-        VarMI = Stack[Cnt].MappedDecls[VD];
-        break;
-      }
+  // Do the check specified in MapInfoCheck and return true if any issue is
+  // found.
+  template <class MapInfoCheck>
+  bool checkMapInfoForVar(ValueDecl *VD, bool CurrentRegionOnly,
+                          MapInfoCheck Check) {
+    auto SI = Stack.rbegin();
+    auto SE = Stack.rend();
+
+    if (SI == SE)
+      return false;
+
+    if (CurrentRegionOnly) {
+      SE = std::next(SI);
+    } else {
+      ++SI;
     }
-    return VarMI;
-  }
 
-  void addMapInfoForVar(ValueDecl *VD, MapInfo MI) {
-    if (Stack.size() > 1) {
-      Stack.back().MappedDecls[VD] = MI;
+    for (; SI != SE; ++SI) {
+      auto MI = SI->MappedDecls.find(VD);
+      if (MI != SI->MappedDecls.end()) {
+        for (Expr *E : MI->second) {
+          if (Check(E))
+            return true;
+        }
+      }
     }
+    return false;
   }
 
-  MapInfo IsMappedInCurrentRegion(ValueDecl *VD) {
-    assert(Stack.size() > 1 && "Target level is 0");
-    MapInfo VarMI = {0};
-    if (Stack.size() > 1 && Stack.back().MappedDecls.count(VD)) {
-      VarMI = Stack.back().MappedDecls[VD];
+  void addExprToVarMapInfo(ValueDecl *VD, Expr *E) {
+    if (Stack.size() > 1) {
+      Stack.back().MappedDecls[VD].push_back(E);
     }
-    return VarMI;
   }
 };
 bool isParallelOrTaskRegion(OpenMPDirectiveKind DKind) {
@@ -8563,6 +8569,387 @@ static bool CheckTypeMappable(SourceLocation SL, SourceRange SR, Sema &SemaRef,
   return true;
 }
 
+// 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) {
+  SourceLocation ELoc = E->getExprLoc();
+  SourceRange ERange = E->getSourceRange();
+
+  // The base of elements of list in a map clause have to be either:
+  //  - a reference to variable or field.
+  //  - a member expression.
+  //  - an array expression.
+  //
+  // E.g. if we have the expression 'r.S.Arr[:12]', we want to retrieve the
+  // reference to 'r'.
+  //
+  // If we have:
+  //
+  // struct SS {
+  //   Bla S;
+  //   foo() {
+  //     #pragma omp target map (S.Arr[:12]);
+  //   }
+  // }
+  //
+  // We want to retrieve the member expression 'this->S';
+
+  Expr *RelevantExpr = nullptr;
+
+  // Flags to help capture some memory
+
+  // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, p.2]
+  //  If a list item is an array section, it must specify contiguous storage.
+  //
+  // For this restriction it is sufficient that we make sure only references
+  // to variables or fields and array expressions, and that no array sections
+  // exist except in the rightmost expression. E.g. these would be invalid:
+  //
+  //   r.ArrS[3:5].Arr[6:7]
+  //
+  //   r.ArrS[3:5].x
+  //
+  // but these would be valid:
+  //   r.ArrS[3].Arr[6:7]
+  //
+  //   r.ArrS[3].x
+
+  bool IsRightMostExpression = true;
+
+  while (!RelevantExpr) {
+    auto AllowArraySection = IsRightMostExpression;
+    IsRightMostExpression = false;
+
+    E = E->IgnoreParenImpCasts();
+
+    if (auto *CurE = dyn_cast<DeclRefExpr>(E)) {
+      if (!isa<VarDecl>(CurE->getDecl()))
+        break;
+
+      RelevantExpr = CurE;
+      continue;
+    }
+
+    if (auto *CurE = dyn_cast<MemberExpr>(E)) {
+      auto *BaseE = CurE->getBase()->IgnoreParenImpCasts();
+
+      if (isa<CXXThisExpr>(BaseE))
+        // We found a base expression: this->Val.
+        RelevantExpr = CurE;
+      else
+        E = BaseE;
+
+      if (!isa<FieldDecl>(CurE->getMemberDecl())) {
+        SemaRef.Diag(ELoc, diag::err_omp_expected_access_to_data_field)
+            << CurE->getSourceRange();
+        break;
+      }
+
+      auto *FD = cast<FieldDecl>(CurE->getMemberDecl());
+
+      // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, C/C++, p.3]
+      //  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();
+        break;
+      }
+
+      // 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 CurType = BaseE->getType();
+      if (CurType->isReferenceType())
+        CurType = CurType->getPointeeType();
+
+      // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, C/C++, p.2]
+      //  A list item cannot be a variable that is a member of a structure with
+      //  a union type.
+      //
+      if (auto *RT = CurType->getAs<RecordType>())
+        if (RT->isUnionType()) {
+          SemaRef.Diag(ELoc, diag::err_omp_union_type_not_allowed)
+              << CurE->getSourceRange();
+          break;
+        }
+
+      continue;
+    }
+
+    if (auto *CurE = dyn_cast<ArraySubscriptExpr>(E)) {
+      E = CurE->getBase()->IgnoreParenImpCasts();
+
+      if (!E->getType()->isAnyPointerType() && !E->getType()->isArrayType()) {
+        SemaRef.Diag(ELoc, diag::err_omp_expected_base_var_name)
+            << 0 << CurE->getSourceRange();
+        break;
+      }
+      continue;
+    }
+
+    if (auto *CurE = dyn_cast<OMPArraySectionExpr>(E)) {
+      // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, p.7]
+      //  If a list item is an element of a structure, only the rightmost symbol
+      //  of the variable reference can be an array section.
+      //
+      if (!AllowArraySection) {
+        SemaRef.Diag(ELoc, diag::err_omp_array_section_in_rightmost_expression)
+            << CurE->getSourceRange();
+        break;
+      }
+
+      E = CurE->getBase()->IgnoreParenImpCasts();
+
+      // 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 CurType = E->getType();
+      if (CurType->isReferenceType())
+        CurType = CurType->getPointeeType();
+
+      if (!CurType->isAnyPointerType() && !CurType->isArrayType()) {
+        SemaRef.Diag(ELoc, diag::err_omp_expected_base_var_name)
+            << 0 << CurE->getSourceRange();
+        break;
+      }
+
+      continue;
+    }
+
+    // If nothing else worked, this is not a valid map clause expression.
+    SemaRef.Diag(ELoc,
+                 diag::err_omp_expected_named_var_member_or_array_expression)
+        << ERange;
+    break;
+  }
+
+  return RelevantExpr;
+}
+
+// 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) {
+  assert(VD && E);
+
+  // Types used to organize the components of a valid map clause.
+  typedef std::pair<Expr *, ValueDecl *> MapExpressionComponent;
+  typedef SmallVector<MapExpressionComponent, 4> MapExpressionComponents;
+
+  // Helper to extract the components in the map clause expression E and store
+  // them into MEC. This assumes that E is a valid map clause expression, i.e.
+  // it has already passed the single clause checks.
+  auto ExtractMapExpressionComponents = [](Expr *TE,
+                                           MapExpressionComponents &MEC) {
+    while (true) {
+      TE = TE->IgnoreParenImpCasts();
+
+      if (auto *CurE = dyn_cast<DeclRefExpr>(TE)) {
+        MEC.push_back(
+            MapExpressionComponent(CurE, cast<VarDecl>(CurE->getDecl())));
+        break;
+      }
+
+      if (auto *CurE = dyn_cast<MemberExpr>(TE)) {
+        auto *BaseE = CurE->getBase()->IgnoreParenImpCasts();
+
+        MEC.push_back(MapExpressionComponent(
+            CurE, cast<FieldDecl>(CurE->getMemberDecl())));
+        if (isa<CXXThisExpr>(BaseE))
+          break;
+
+        TE = BaseE;
+        continue;
+      }
+
+      if (auto *CurE = dyn_cast<ArraySubscriptExpr>(TE)) {
+        MEC.push_back(MapExpressionComponent(CurE, nullptr));
+        TE = CurE->getBase()->IgnoreParenImpCasts();
+        continue;
+      }
+
+      if (auto *CurE = dyn_cast<OMPArraySectionExpr>(TE)) {
+        MEC.push_back(MapExpressionComponent(CurE, nullptr));
+        TE = CurE->getBase()->IgnoreParenImpCasts();
+        continue;
+      }
+
+      llvm_unreachable(
+          "Expecting only valid map clause expressions at this point!");
+    }
+  };
+
+  SourceLocation ELoc = E->getExprLoc();
+  SourceRange ERange = E->getSourceRange();
+
+  // In order to easily check the conflicts we need to match each component of
+  // the expression under test with the components of the expressions that are
+  // already in the stack.
+
+  MapExpressionComponents CurComponents;
+  ExtractMapExpressionComponents(E, CurComponents);
+
+  assert(!CurComponents.empty() && "Map clause expression with no components!");
+  assert(CurComponents.back().second == VD &&
+         "Map clause expression with unexpected base!");
+
+  // Variables to help detecting enclosing problems in data environment nests.
+  bool IsEnclosedByDataEnvironmentExpr = false;
+  Expr *EnclosingExpr = nullptr;
+
+  bool FoundError =
+      DSAS->checkMapInfoForVar(VD, CurrentRegionOnly, [&](Expr *RE) -> bool {
+        MapExpressionComponents StackComponents;
+        ExtractMapExpressionComponents(RE, StackComponents);
+        assert(!StackComponents.empty() &&
+               "Map clause expression with no components!");
+        assert(StackComponents.back().second == VD &&
+               "Map clause expression with unexpected base!");
+
+        // Expressions must start from the same base. Here we detect at which
+        // point both expressions diverge from each other and see if we can
+        // detect if the memory referred to both expressions is contiguous and
+        // do not overlap.
+        auto CI = CurComponents.rbegin();
+        auto CE = CurComponents.rend();
+        auto SI = StackComponents.rbegin();
+        auto SE = StackComponents.rend();
+        for (; CI != CE && SI != SE; ++CI, ++SI) {
+
+          // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, p.3]
+          //  At most one list item can be an array item derived from a given
+          //  variable in map clauses of the same construct.
+          if (CurrentRegionOnly && (isa<ArraySubscriptExpr>(CI->first) ||
+                                    isa<OMPArraySectionExpr>(CI->first)) &&
+              (isa<ArraySubscriptExpr>(SI->first) ||
+               isa<OMPArraySectionExpr>(SI->first))) {
+            SemaRef.Diag(CI->first->getExprLoc(),
+                         diag::err_omp_multiple_array_items_in_map_clause)
+                << CI->first->getSourceRange();
+            ;
+            SemaRef.Diag(SI->first->getExprLoc(), diag::note_used_here)
+                << SI->first->getSourceRange();
+            return true;
+          }
+
+          // Do both expressions have the same kind?
+          if (CI->first->getStmtClass() != SI->first->getStmtClass())
+            break;
+
+          // Are we dealing with different variables/fields?
+          if (CI->second != SI->second)
+            break;
+        }
+
+        // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, p.4]
+        //  List items of map clauses in the same construct must not share
+        //  original storage.
+        //
+        // If the expressions are exactly the same or one is a subset of the
+        // other, it means they are sharing storage.
+        if (CI == CE && SI == SE) {
+          if (CurrentRegionOnly) {
+            SemaRef.Diag(ELoc, diag::err_omp_map_shared_storage) << ERange;
+            SemaRef.Diag(RE->getExprLoc(), diag::note_used_here)
+                << RE->getSourceRange();
+            return true;
+          } else {
+            // If we find the same expression in the enclosing data environment,
+            // that is legal.
+            IsEnclosedByDataEnvironmentExpr = true;
+            return false;
+          }
+        }
+
+        QualType DerivedType = std::prev(CI)->first->getType();
+        SourceLocation DerivedLoc = std::prev(CI)->first->getExprLoc();
+
+        // 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.
+        if (DerivedType->isReferenceType())
+          DerivedType = DerivedType->getPointeeType();
+
+        // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, C/C++, p.1]
+        //  A variable for which the type is pointer and an array section
+        //  derived from that variable must not appear as list items of map
+        //  clauses of the same construct.
+        //
+        // Also, cover one of the cases in:
+        // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, p.5]
+        //  If any part of the original storage of a list item has corresponding
+        //  storage in the device data environment, all of the original storage
+        //  must have corresponding storage in the device data environment.
+        //
+        if (DerivedType->isAnyPointerType()) {
+          if (CI == CE || SI == SE) {
+            SemaRef.Diag(
+                DerivedLoc,
+                diag::err_omp_pointer_mapped_along_with_derived_section)
+                << DerivedLoc;
+          } else {
+            assert(CI != CE && SI != SE);
+            SemaRef.Diag(DerivedLoc, diag::err_omp_same_pointer_derreferenced)
+                << DerivedLoc;
+          }
+          SemaRef.Diag(RE->getExprLoc(), diag::note_used_here)
+              << RE->getSourceRange();
+          return true;
+        }
+
+        // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, p.4]
+        //  List items of map clauses in the same construct must not share
+        //  original storage.
+        //
+        // An expression is a subset of the other.
+        if (CurrentRegionOnly && (CI == CE || SI == SE)) {
+          SemaRef.Diag(ELoc, diag::err_omp_map_shared_storage) << ERange;
+          SemaRef.Diag(RE->getExprLoc(), diag::note_used_here)
+              << RE->getSourceRange();
+          return true;
+        }
+
+        // The current expression uses the same base as other expression in the
+        // data environment but does not contain it completelly.
+        if (!CurrentRegionOnly && SI != SE)
+          EnclosingExpr = RE;
+
+        // The current expression is a subset of the expression in the data
+        // environment.
+        IsEnclosedByDataEnvironmentExpr |=
+            (!CurrentRegionOnly && CI != CE && SI == SE);
+
+        return false;
+      });
+
+  if (CurrentRegionOnly)
+    return FoundError;
+
+  // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, p.5]
+  //  If any part of the original storage of a list item has corresponding
+  //  storage in the device data environment, all of the original storage must
+  //  have corresponding storage in the device data environment.
+  // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, p.6]
+  //  If a list item is an element of a structure, and a different element of
+  //  the structure has a corresponding list item in the device data environment
+  //  prior to a task encountering the construct associated with the map clause,
+  //  then the list item must also have a correspnding list item in the device
+  //  data environment prior to the task encountering the construct.
+  //
+  if (EnclosingExpr && !IsEnclosedByDataEnvironmentExpr) {
+    SemaRef.Diag(ELoc,
+                 diag::err_omp_original_storage_is_shared_and_does_not_contain)
+        << ERange;
+    SemaRef.Diag(EnclosingExpr->getExprLoc(), diag::note_used_here)
+        << EnclosingExpr->getSourceRange();
+    return true;
+  }
+
+  return FoundError;
+}
+
 OMPClause *
 Sema::ActOnOpenMPMapClause(OpenMPMapClauseKind MapTypeModifier,
                            OpenMPMapClauseKind MapType, bool IsMapTypeImplicit,
@@ -8580,96 +8967,82 @@ Sema::ActOnOpenMPMapClause(OpenMPMapClauseKind MapTypeModifier,
     }
     SourceLocation ELoc = RE->getExprLoc();
 
-    // OpenMP [2.14.5, Restrictions]
-    //  A variable that is part of another variable (such as field of a
-    //  structure) but is not an array element or an array section cannot appear
-    //  in a map clause.
     auto *VE = RE->IgnoreParenLValueCasts();
 
     if (VE->isValueDependent() || VE->isTypeDependent() ||
         VE->isInstantiationDependent() ||
         VE->containsUnexpandedParameterPack()) {
-      // It will be analyzed later.
+      // We can only analyze this information once the missing information is
+      // resolved.
       Vars.push_back(RE);
       continue;
     }
 
     auto *SimpleExpr = RE->IgnoreParenCasts();
-    auto *DE = dyn_cast<DeclRefExpr>(SimpleExpr);
-    auto *ASE = dyn_cast<ArraySubscriptExpr>(SimpleExpr);
-    auto *OASE = dyn_cast<OMPArraySectionExpr>(SimpleExpr);
-
-    if (!RE->IgnoreParenImpCasts()->isLValue() ||
-        (!OASE && !ASE && !DE) ||
-        (DE && !isa<VarDecl>(DE->getDecl())) ||
-        (ASE && !ASE->getBase()->getType()->isAnyPointerType() &&
-         !ASE->getBase()->getType()->isArrayType())) {
-      Diag(ELoc, diag::err_omp_expected_var_name_member_expr_or_array_item)
-          << 0 << RE->getSourceRange();
+
+    if (!RE->IgnoreParenImpCasts()->isLValue()) {
+      Diag(ELoc, diag::err_omp_expected_named_var_member_or_array_expression)
+          << RE->getSourceRange();
       continue;
     }
 
-    Decl *D = nullptr;
-    if (DE) {
-      D = DE->getDecl();
-    } else if (ASE) {
-      auto *B = ASE->getBase()->IgnoreParenCasts();
-      D = dyn_cast<DeclRefExpr>(B)->getDecl();
-    } else if (OASE) {
-      auto *B = OASE->getBase();
-      D = dyn_cast<DeclRefExpr>(B)->getDecl();
+    // Obtain the array or member expression bases if required.
+    auto *BE = CheckMapClauseExpressionBase(*this, SimpleExpr);
+    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.
+    ValueDecl *D = nullptr;
+    if (auto *DRE = dyn_cast<DeclRefExpr>(BE)) {
+      D = DRE->getDecl();
+    } else {
+      auto *ME = cast<MemberExpr>(BE);
+      assert(isa<CXXThisExpr>(ME->getBase()) && "Unexpected expression!");
+      D = ME->getMemberDecl();
     }
     assert(D && "Null decl on map clause.");
-    auto *VD = cast<VarDecl>(D);
 
-    // OpenMP [2.14.5, Restrictions, p.8]
-    // threadprivate variables cannot appear in a map clause.
-    if (DSAStack->isThreadPrivate(VD)) {
+    auto *VD = dyn_cast<VarDecl>(D);
+    auto *FD = dyn_cast<FieldDecl>(D);
+
+    assert((VD || FD) && "Only variables or fields are expected here!");
+
+    // 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);
       continue;
     }
 
-    // OpenMP [2.14.5, Restrictions, p.2]
-    //  At most one list item can be an array item derived from a given variable
-    //  in map clauses of the same construct.
-    // OpenMP [2.14.5, Restrictions, p.3]
-    //  List items of map clauses in the same construct must not share original
-    //  storage.
-    // OpenMP [2.14.5, Restrictions, C/C++, p.2]
-    //  A variable for which the type is pointer, reference to array, or
-    //  reference to pointer and an array section derived from that variable
-    //  must not appear as list items of map clauses of the same construct.
-    DSAStackTy::MapInfo MI = DSAStack->IsMappedInCurrentRegion(VD);
-    if (MI.RefExpr) {
-      Diag(ELoc, diag::err_omp_map_shared_storage) << ELoc;
-      Diag(MI.RefExpr->getExprLoc(), diag::note_used_here)
-          << MI.RefExpr->getSourceRange();
-      continue;
-    }
+    // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, p.9]
+    //  A list item cannot appear in both a map clause and a data-sharing
+    //  attribute clause on the same construct.
+    //
+    // TODO: Implement this check - it cannot currently be tested because of
+    // 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))
+      break;
+    if (CheckMapConflicts(*this, DSAStack, D, SimpleExpr,
+                          /*CurrentRegionOnly=*/false))
+      break;
 
-    // OpenMP [2.14.5, Restrictions, C/C++, p.3,4]
-    //  A variable for which the type is pointer, reference to array, or
-    //  reference to pointer must not appear as a list item if the enclosing
-    //  device data environment already contains an array section derived from
-    //  that variable.
-    //  An array section derived from a variable for which the type is pointer,
-    //  reference to array, or reference to pointer must not appear as a list
-    //  item if the enclosing device data environment already contains that
-    //  variable.
-    QualType Type = VD->getType();
-    MI = DSAStack->getMapInfoForVar(VD);
-    if (MI.RefExpr && (isa<DeclRefExpr>(MI.RefExpr->IgnoreParenLValueCasts()) !=
-                       isa<DeclRefExpr>(VE)) &&
-        (Type->isPointerType() || Type->isReferenceType())) {
-      Diag(ELoc, diag::err_omp_map_shared_storage) << ELoc;
-      Diag(MI.RefExpr->getExprLoc(), diag::note_used_here)
-          << MI.RefExpr->getSourceRange();
-      continue;
-    }
+    // 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 (Type->isReferenceType())
+      Type = Type->getPointeeType();
 
-    // OpenMP [2.14.5, Restrictions, C/C++, p.7]
+    // 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))
@@ -8686,8 +9059,7 @@ Sema::ActOnOpenMPMapClause(OpenMPMapClauseKind MapTypeModifier,
           << (IsMapTypeImplicit ? 1 : 0)
           << getOpenMPSimpleClauseTypeName(OMPC_map, MapType)
           << getOpenMPDirectiveName(DKind);
-      // Proceed to add the variable in a map clause anyway, to prevent
-      // further spurious messages
+      continue;
     }
 
     // target exit_data
@@ -8702,17 +9074,15 @@ Sema::ActOnOpenMPMapClause(OpenMPMapClauseKind MapTypeModifier,
           << (IsMapTypeImplicit ? 1 : 0)
           << getOpenMPSimpleClauseTypeName(OMPC_map, MapType)
           << getOpenMPDirectiveName(DKind);
-      // Proceed to add the variable in a map clause anyway, to prevent
-      // further spurious messages
+      continue;
     }
 
     Vars.push_back(RE);
-    MI.RefExpr = RE;
-    DSAStack->addMapInfoForVar(VD, MI);
+    DSAStack->addExprToVarMapInfo(D, RE);
   }
-  if (Vars.empty())
-    return nullptr;
 
+  // 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, StartLoc, LParenLoc, EndLoc, Vars,
                               MapTypeModifier, MapType, IsMapTypeImplicit,
                               MapLoc);