Index: include/clang/Basic/DiagnosticSemaKinds.td =================================================================== --- include/clang/Basic/DiagnosticSemaKinds.td +++ include/clang/Basic/DiagnosticSemaKinds.td @@ -7809,8 +7809,8 @@ "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< Index: lib/Sema/SemaOpenMP.cpp =================================================================== --- lib/Sema/SemaOpenMP.cpp +++ lib/Sema/SemaOpenMP.cpp @@ -9067,6 +9067,94 @@ 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(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(E) || (OASE && OASE->getColonLoc().isInvalid())) { + if (auto *ATy = dyn_cast(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(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(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(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(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. @@ -9095,14 +9183,13 @@ 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] // @@ -9113,12 +9200,11 @@ // // 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(E)) { @@ -9126,6 +9212,11 @@ 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; } @@ -9158,9 +9249,7 @@ // 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 @@ -9173,6 +9262,15 @@ 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; } @@ -9184,35 +9282,58 @@ << 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(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; } Index: test/OpenMP/target_map_messages.cpp =================================================================== --- test/OpenMP/target_map_messages.cpp +++ 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 struct SA { static int ss; @@ -82,6 +99,13 @@ void SAclient(int arg) { SA 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; @@ -89,10 +113,106 @@ 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 @@ foo(); return tmain(argc)+tmain(argc); // expected-note {{in instantiation of function template specialization 'tmain' requested here}} expected-note {{in instantiation of function template specialization 'tmain' requested here}} } - +#endif