[OpenMP] Add support for multidimensional array sections in map claus…

…e SEMA.

Summary: In some cases it can be proved statically that multidimensional array section refer to contiguous storage and can therefore be allowed in a map clause. This patch adds support for those cases in SEMA.

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

Subscribers: cfe-commits, fraggamuffin, caomhin

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

llvm-svn: 263019

clang/include/clang/Basic/DiagnosticSemaKinds.td

@@ -7825,8 +7825,8 @@ def err_omp_expected_named_var_member_or_array_expression: Error<
   "expected expression containing only member accesses and/or array sections based on named variables">;
 def err_omp_bit_fields_forbidden_in_map_clause : Error<
   "bit fields cannot be used to specify storage in a map clause">;
-def err_omp_array_section_in_rightmost_expression : Error<
-  "array section can only be associated with the rightmost variable in a map clause expression">;
+def err_array_section_does_not_specify_contiguous_storage : Error<
+  "array section does not specify contiguous storage">;
 def err_omp_union_type_not_allowed : Error<
   "mapped storage cannot be derived from a union">;
 def err_omp_expected_access_to_data_field : Error<

clang/lib/Sema/SemaOpenMP.cpp

@@ -9104,6 +9104,94 @@ static bool CheckTypeMappable(SourceLocation SL, SourceRange SR, Sema &SemaRef,
   return true;
 }
 
+/// \brief Return true if it can be proven that the provided array expression
+/// (array section or array subscript) does NOT specify the whole size of the
+/// array whose base type is \a BaseQTy.
+static bool CheckArrayExpressionDoesNotReferToWholeSize(Sema &SemaRef,
+                                                        const Expr *E,
+                                                        QualType BaseQTy) {
+  auto *OASE = dyn_cast<OMPArraySectionExpr>(E);
+
+  // If this is an array subscript, it refers to the whole size if the size of
+  // the dimension is constant and equals 1. Also, an array section assumes the
+  // format of an array subscript if no colon is used.
+  if (isa<ArraySubscriptExpr>(E) || (OASE && OASE->getColonLoc().isInvalid())) {
+    if (auto *ATy = dyn_cast<ConstantArrayType>(BaseQTy.getTypePtr()))
+      return ATy->getSize().getSExtValue() != 1;
+    // Size can't be evaluated statically.
+    return false;
+  }
+
+  assert(OASE && "Expecting array section if not an array subscript.");
+  auto *LowerBound = OASE->getLowerBound();
+  auto *Length = OASE->getLength();
+
+  // If there is a lower bound that does not evaluates to zero, we are not
+  // convering the whole dimension.
+  if (LowerBound) {
+    llvm::APSInt ConstLowerBound;
+    if (!LowerBound->EvaluateAsInt(ConstLowerBound, SemaRef.getASTContext()))
+      return false; // Can't get the integer value as a constant.
+    if (ConstLowerBound.getSExtValue())
+      return true;
+  }
+
+  // If we don't have a length we covering the whole dimension.
+  if (!Length)
+    return false;
+
+  // If the base is a pointer, we don't have a way to get the size of the
+  // pointee.
+  if (BaseQTy->isPointerType())
+    return false;
+
+  // We can only check if the length is the same as the size of the dimension
+  // if we have a constant array.
+  auto *CATy = dyn_cast<ConstantArrayType>(BaseQTy.getTypePtr());
+  if (!CATy)
+    return false;
+
+  llvm::APSInt ConstLength;
+  if (!Length->EvaluateAsInt(ConstLength, SemaRef.getASTContext()))
+    return false; // Can't get the integer value as a constant.
+
+  return CATy->getSize().getSExtValue() != ConstLength.getSExtValue();
+}
+
+// Return true if it can be proven that the provided array expression (array
+// section or array subscript) does NOT specify a single element of the array
+// whose base type is \a BaseQTy.
+static bool CheckArrayExpressionDoesNotReferToUnitySize(Sema &SemaRef,
+                                                       const Expr *E,
+                                                       QualType BaseQTy) {
+  auto *OASE = dyn_cast<OMPArraySectionExpr>(E);
+
+  // An array subscript always refer to a single element. Also, an array section
+  // assumes the format of an array subscript if no colon is used.
+  if (isa<ArraySubscriptExpr>(E) || (OASE && OASE->getColonLoc().isInvalid()))
+    return false;
+
+  assert(OASE && "Expecting array section if not an array subscript.");
+  auto *Length = OASE->getLength();
+
+  // If we don't have a length we have to check if the array has unitary size
+  // for this dimension. Also, we should always expect a length if the base type
+  // is pointer.
+  if (!Length) {
+    if (auto *ATy = dyn_cast<ConstantArrayType>(BaseQTy.getTypePtr()))
+      return ATy->getSize().getSExtValue() != 1;
+    // We cannot assume anything.
+    return false;
+  }
+
+  // Check if the length evaluates to 1.
+  llvm::APSInt ConstLength;
+  if (!Length->EvaluateAsInt(ConstLength, SemaRef.getASTContext()))
+    return false; // Can't get the integer value as a constant.
+
+  return ConstLength.getSExtValue() != 1;
+}
+
 // 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.
@@ -9132,14 +9220,13 @@ static Expr *CheckMapClauseExpressionBase(Sema &SemaRef, Expr *E) {
 
   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:
+  // exist except in the rightmost expression (unless they cover the whole
+  // dimension of the array). E.g. these would be invalid:
   //
   //   r.ArrS[3:5].Arr[6:7]
   //
@@ -9150,19 +9237,23 @@ static Expr *CheckMapClauseExpressionBase(Sema &SemaRef, Expr *E) {
   //
   //   r.ArrS[3].x
 
-  bool IsRightMostExpression = true;
-
-  while (!RelevantExpr) {
-    auto AllowArraySection = IsRightMostExpression;
-    IsRightMostExpression = false;
+  bool AllowUnitySizeArraySection = true;
+  bool AllowWholeSizeArraySection = true;
 
+  for (bool IsRightMostExpression = true; !RelevantExpr;
+       IsRightMostExpression = false) {
     E = E->IgnoreParenImpCasts();
 
     if (auto *CurE = dyn_cast<DeclRefExpr>(E)) {
       if (!isa<VarDecl>(CurE->getDecl()))
         break;
 
       RelevantExpr = CurE;
+
+      // If we got a reference to a declaration, we should not expect any array
+      // section before that.
+      AllowUnitySizeArraySection = false;
+      AllowWholeSizeArraySection = false;
       continue;
     }
 
@@ -9195,9 +9286,7 @@ static Expr *CheckMapClauseExpressionBase(Sema &SemaRef, Expr *E) {
       // 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();
+      QualType CurType = BaseE->getType().getNonReferenceType();
 
       // 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
@@ -9210,6 +9299,15 @@ static Expr *CheckMapClauseExpressionBase(Sema &SemaRef, Expr *E) {
           break;
         }
 
+      // If we got a member expression, we should not expect any array section
+      // before that:
+      //
+      // 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.
+      //
+      AllowUnitySizeArraySection = false;
+      AllowWholeSizeArraySection = false;
       continue;
     }
 
@@ -9221,35 +9319,58 @@ static Expr *CheckMapClauseExpressionBase(Sema &SemaRef, Expr *E) {
             << 0 << CurE->getSourceRange();
         break;
       }
+
+      // If we got an array subscript that express the whole dimension we
+      // can have any array expressions before. If it only expressing part of
+      // the dimension, we can only have unitary-size array expressions.
+      if (CheckArrayExpressionDoesNotReferToWholeSize(SemaRef, CurE,
+                                                      E->getType()))
+        AllowWholeSizeArraySection = false;
       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();
 
+      auto CurType =
+          OMPArraySectionExpr::getBaseOriginalType(E).getCanonicalType();
+
       // 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()) {
+      bool IsPointer = CurType->isAnyPointerType();
+
+      if (!IsPointer && !CurType->isArrayType()) {
         SemaRef.Diag(ELoc, diag::err_omp_expected_base_var_name)
             << 0 << CurE->getSourceRange();
         break;
       }
 
+      bool NotWhole =
+          CheckArrayExpressionDoesNotReferToWholeSize(SemaRef, CurE, CurType);
+      bool NotUnity =
+          CheckArrayExpressionDoesNotReferToUnitySize(SemaRef, CurE, CurType);
+
+      if (AllowWholeSizeArraySection && AllowUnitySizeArraySection) {
+        // Any array section is currently allowed.
+        //
+        // If this array section refers to the whole dimension we can still
+        // accept other array sections before this one, except if the base is a
+        // pointer. Otherwise, only unitary sections are accepted.
+        if (NotWhole || IsPointer)
+          AllowWholeSizeArraySection = false;
+      } else if ((AllowUnitySizeArraySection && NotUnity) ||
+                 (AllowWholeSizeArraySection && NotWhole)) {
+        // A unity or whole array section is not allowed and that is not
+        // compatible with the properties of the current array section.
+        SemaRef.Diag(
+            ELoc, diag::err_array_section_does_not_specify_contiguous_storage)
+            << CurE->getSourceRange();
+        break;
+      }
       continue;
     }
 

clang/test/OpenMP/target_map_messages.cpp

@@ -1,4 +1,21 @@
-// RUN: %clang_cc1 -verify -fopenmp -ferror-limit 100 %s
+// RUN: %clang_cc1 -verify -fopenmp -ferror-limit 200 %s
+// RUN: %clang_cc1 -DCCODE -verify -fopenmp -ferror-limit 200 -x c %s
+#ifdef CCODE
+void foo(int arg) {
+  const int n = 0;
+
+  double marr[10][10][10];
+
+  #pragma omp target map(marr[2][0:2][0:2]) // expected-error {{array section does not specify contiguous storage}}
+  {}
+  #pragma omp target map(marr[:][0:][:])
+  {}
+  #pragma omp target map(marr[:][1:][:]) // expected-error {{array section does not specify contiguous storage}}
+  {}
+  #pragma omp target map(marr[:][n:][:])
+  {}
+}
+#else
 template <typename T, int I>
 struct SA {
   static int ss;
@@ -82,17 +99,120 @@ union SD {
 void SAclient(int arg) {
   SA<int,123> s;
   s.func(arg); // expected-note {{in instantiation of member function}}
+  double marr[10][10][10];
+  double marr2[5][10][1];
+  double mvla[5][arg][10];
+  double ***mptr;
+  const int n = 0;
+  const int m = 1;
+  double mvla2[5][arg][m+n+10];
 
   SB *p;
 
   SD u;
   SC r(p),t(p);
   #pragma omp target map(r)
   {}
+  #pragma omp target map(marr[2][0:2][0:2]) // expected-error {{array section does not specify contiguous storage}}
+  {}
+  #pragma omp target map(marr[:][0:2][0:2]) // expected-error {{array section does not specify contiguous storage}}
+  {}
+  #pragma omp target map(marr[2][3][0:2])
+  {}
+  #pragma omp target map(marr[:][:][:])
+  {}
+  #pragma omp target map(marr[:2][:][:])
+  {}
+  #pragma omp target map(marr[arg:][:][:])
+  {}
+  #pragma omp target map(marr[arg:])
+  {}
+  #pragma omp target map(marr[arg:][:arg][:]) // correct if arg is the size of dimension 2
+  {}
+  #pragma omp target map(marr[:arg][:])
+  {}
+  #pragma omp target map(marr[:arg][n:])
+  {}
+  #pragma omp target map(marr[:][:arg][n:]) // correct if arg is the size of  dimension 2
+  {}
+  #pragma omp target map(marr[:][:m][n:]) // expected-error {{array section does not specify contiguous storage}}
+  {}
+  #pragma omp target map(marr[n:m][:arg][n:])
+  {}
+  #pragma omp target map(marr[:2][:1][:]) // expected-error {{array section does not specify contiguous storage}}
+  {}
+  #pragma omp target map(marr[:2][1:][:]) // expected-error {{array section does not specify contiguous storage}}
+  {}
+  #pragma omp target map(marr[:2][:][:1]) // expected-error {{array section does not specify contiguous storage}}
+  {}
+  #pragma omp target map(marr[:2][:][1:]) // expected-error {{array section does not specify contiguous storage}}
+  {}
+  #pragma omp target map(marr[:1][:2][:])
+  {}
+  #pragma omp target map(marr[:1][0][:])
+  {}
+  #pragma omp target map(marr[:arg][:2][:]) // correct if arg is 1
+  {}
+  #pragma omp target map(marr[:1][3:1][:2])
+  {}
+  #pragma omp target map(marr[:1][3:arg][:2]) // correct if arg is 1
+  {}
+  #pragma omp target map(marr[:1][3:2][:2]) // expected-error {{array section does not specify contiguous storage}}
+  {}
+  #pragma omp target map(marr[:2][:10][:])
+  {}
+  #pragma omp target map(marr[:2][:][:5+5])
+  {}
+  #pragma omp target map(marr[:2][2+2-4:][0:5+5])
+  {}
+
+  #pragma omp target map(marr[:1][:2][0]) // expected-error {{array section does not specify contiguous storage}}
+  {}
+  #pragma omp target map(marr2[:1][:2][0])
+  {}
+
+  #pragma omp target map(mvla[:1][:][0]) // correct if the size of dimension 2 is 1.
+  {}
+  #pragma omp target map(mvla[:2][:arg][:]) // correct if arg is the size of dimension 2.
+  {}
+  #pragma omp target map(mvla[:1][:2][0]) // expected-error {{array section does not specify contiguous storage}}
+   {}
+  #pragma omp target map(mvla[1][2:arg][:])
+  {}
+  #pragma omp target map(mvla[:1][:][:])
+  {}
+  #pragma omp target map(mvla2[:1][:2][:11])
+  {}
+  #pragma omp target map(mvla2[:1][:2][:10]) // expected-error {{array section does not specify contiguous storage}}
+  {}
+
+  #pragma omp target map(mptr[:2][2+2-4:1][0:5+5]) // expected-error {{array section does not specify contiguous storage}}
+  {}
+  #pragma omp target map(mptr[:1][:2-1][2:4-3])
+  {}
+  #pragma omp target map(mptr[:1][:arg][2:4-3]) // correct if arg is 1.
+  {}
+  #pragma omp target map(mptr[:1][:2-1][0:2])
+  {}
+  #pragma omp target map(mptr[:1][:2][0:2]) // expected-error {{array section does not specify contiguous storage}}
+  {}
+  #pragma omp target map(mptr[:1][:][0:2]) // expected-error {{section length is unspecified and cannot be inferred because subscripted value is not an array}}
+  {}
+  #pragma omp target map(mptr[:2][:1][0:2]) // expected-error {{array section does not specify contiguous storage}}
+  {}
+
   #pragma omp target map(r.ArrS[0].B)
   {}
+  #pragma omp target map(r.ArrS[:1].B) // expected-error {{OpenMP array section is not allowed here}}
+  {}
+  #pragma omp target map(r.ArrS[:arg].B) // expected-error {{OpenMP array section is not allowed here}}
+  {}
   #pragma omp target map(r.ArrS[0].Arr[1:23])
   {}
+  #pragma omp target map(r.ArrS[0].Arr[1:arg])
+  {}
+  #pragma omp target map(r.ArrS[0].Arr[arg:23])
+  {}
   #pragma omp target map(r.ArrS[0].Error) // expected-error {{no member named 'Error' in 'SB'}}
   {}
   #pragma omp target map(r.ArrS[0].A, r.ArrS[1].A) // expected-error {{multiple array elements associated with the same variable are not allowed in map clauses of the same construct}} expected-note {{used here}}
@@ -382,4 +502,4 @@ int main(int argc, char **argv) {
   foo();
   return tmain<int, 3>(argc)+tmain<from, 4>(argc); // expected-note {{in instantiation of function template specialization 'tmain<int, 3>' requested here}} expected-note {{in instantiation of function template specialization 'tmain<int, 4>' requested here}}
 }
-
+#endif