ImplicitDSALoc() {}
};
-public:
- struct MapInfo {
- Expr *RefExpr;
- };
-
private:
+ typedef SmallVector<Expr *, 4> MapInfo;
+
struct DSAInfo {
OpenMPClauseKind Attributes;
Expr *RefExpr;
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) {
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,
}
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))
<< (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
<< (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);
// RUN: %clang_cc1 -verify -fopenmp -ferror-limit 100 %s
+#if 1
+template <typename T, int I>
+struct SA {
+ static int ss;
+ #pragma omp threadprivate(ss) // expected-note {{defined as threadprivate or thread local}}
+ float a;
+ int b[12];
+ float *c;
+ T d;
+ float e[I];
+ T *f;
+ void func(int arg) {
+ #pragma omp target map(arg,a,d)
+ {}
+ #pragma omp target map(arg[2:2],a,d) // expected-error {{subscripted value is not an array or pointer}}
+ {}
+ #pragma omp target map(arg,a*2) // expected-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+ {}
+ #pragma omp target map(arg,(c+1)[2]) // expected-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+ {}
+ #pragma omp target map(arg,a[:2],d) // expected-error {{subscripted value is not an array or pointer}}
+ {}
+ #pragma omp target map(arg,a,d[:2]) // expected-error {{subscripted value is not an array or pointer}}
+ {}
+ #pragma omp target map(to:ss) // expected-error {{threadprivate variables are not allowed in map clause}}
+ {}
+
+ #pragma omp target map(to:b,e)
+ {}
+ #pragma omp target map(to:b,e) map(to:b) // expected-error {{variable already marked as mapped in current construct}} expected-note {{used here}}
+ {}
+ #pragma omp target map(to:b[:2],e)
+ {}
+ #pragma omp target map(to:b,e[:])
+ {}
+
+ #pragma omp target map(always, tofrom: c,f)
+ {}
+ #pragma omp target map(always, tofrom: c[1:2],f)
+ {}
+ #pragma omp target map(always, tofrom: c,f[1:2])
+ {}
+ #pragma omp target map(always, tofrom: c[:],f) // expected-error {{section length is unspecified and cannot be inferred because subscripted value is not an array}}
+ {}
+ #pragma omp target map(always, tofrom: c,f[:]) // expected-error {{section length is unspecified and cannot be inferred because subscripted value is not an array}}
+ {}
+ return;
+ }
+};
+
+struct SB {
+ unsigned A;
+ unsigned B;
+ float Arr[100];
+ float *Ptr;
+ float *foo() {
+ return &Arr[0];
+ }
+};
+
+struct SC {
+ unsigned A : 2;
+ unsigned B : 3;
+ unsigned C;
+ unsigned D;
+ float Arr[100];
+ SB S;
+ SB ArrS[100];
+ SB *PtrS;
+ SB *&RPtrS;
+ float *Ptr;
+
+ SC(SB *&_RPtrS) : RPtrS(_RPtrS) {}
+};
+
+union SD {
+ unsigned A;
+ float B;
+};
+
+void SAclient(int arg) {
+ SA<int,123> s;
+ s.func(arg); // expected-note {{in instantiation of member function}}
+
+ SB *p;
+
+ SD u;
+ SC r(p),t(p);
+ #pragma omp target map(r)
+ {}
+ #pragma omp target map(r.ArrS[0].B)
+ {}
+ #pragma omp target map(r.ArrS[0].Arr[1: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}}
+ {}
+ #pragma omp target map(r.ArrS[0].A, t.ArrS[1].A)
+ {}
+ #pragma omp target map(r.PtrS[0], r.PtrS->B) // expected-error {{same pointer derreferenced in multiple different ways in map clause expressions}} expected-note {{used here}}
+ {}
+ #pragma omp target map(r.RPtrS[0], r.RPtrS->B) // expected-error {{same pointer derreferenced in multiple different ways in map clause expressions}} expected-note {{used here}}
+ {}
+ #pragma omp target map(r.S.Arr[:12])
+ {}
+ #pragma omp target map(r.S.foo()[:12]) // expected-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+ {}
+ #pragma omp target map(r.C, r.D)
+ {}
+ #pragma omp target map(r.C, r.C) // expected-error {{variable already marked as mapped in current construct}} expected-note {{used here}}
+ {}
+ #pragma omp target map(r.C) map(r.C) // expected-error {{variable already marked as mapped in current construct}} expected-note {{used here}}
+ {}
+ #pragma omp target map(r.C, r.S) // this would be an error only caught at runtime - Sema would have to make sure there is not way for the missing data between fields to be mapped somewhere else.
+ {}
+ #pragma omp target map(r, r.S) // expected-error {{variable already marked as mapped in current construct}} expected-note {{used here}}
+ {}
+ #pragma omp target map(r.C, t.C)
+ {}
+ #pragma omp target map(r.A) // expected-error {{bit fields cannot be used to specify storage in a map clause}}
+ {}
+ #pragma omp target map(r.Arr)
+ {}
+ #pragma omp target map(r.Arr[3:5])
+ {}
+ #pragma omp target map(r.Ptr[3:5])
+ {}
+ #pragma omp target map(r.ArrS[3:5].A) // expected-error {{OpenMP array section is not allowed here}}
+ {}
+ #pragma omp target map(r.ArrS[3:5].Arr[6:7]) // expected-error {{OpenMP array section is not allowed here}}
+ {}
+ #pragma omp target map(r.ArrS[3].Arr[6:7])
+ {}
+ #pragma omp target map(r.S.Arr[4:5])
+ {}
+ #pragma omp target map(r.S.Ptr[4:5])
+ {}
+ #pragma omp target map(r.S.Ptr[:]) // expected-error {{section length is unspecified and cannot be inferred because subscripted value is not an array}}
+ {}
+ #pragma omp target map((p+1)->A) // expected-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+ {}
+ #pragma omp target map(u.B) // expected-error {{mapped storage cannot be derived from a union}}
+ {}
+
+ #pragma omp target data map(to: r.C) //expected-note {{used here}}
+ {
+ #pragma omp target map(r.D) // expected-error {{original storage of expression in data environment is shared but data environment do not fully contain mapped expression storage}}
+ {}
+ }
+
+ #pragma omp target data map(to: t.Ptr) //expected-note {{used here}}
+ {
+ #pragma omp target map(t.Ptr[:23]) // expected-error {{pointer cannot be mapped along with a section derived from itself}}
+ {}
+ }
+
+ #pragma omp target data map(to: t.C, t.D)
+ {
+ #pragma omp target data map(to: t.C)
+ {
+ #pragma omp target map(t.D)
+ {}
+ }
+ }
+
+ #pragma omp target data map(to: t)
+ {
+ #pragma omp target data map(to: t.C)
+ {
+ #pragma omp target map(t.D)
+ {}
+ }
+ }
+}
+#endif
void foo() {
}
T y;
T to, tofrom, always;
const T (&l)[5] = da;
-
-
+#if 1
#pragma omp target map // expected-error {{expected '(' after 'map'}}
+ {}
#pragma omp target map( // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected expression}}
+ {}
#pragma omp target map() // expected-error {{expected expression}}
+ {}
#pragma omp target map(alloc) // expected-error {{use of undeclared identifier 'alloc'}}
+ {}
#pragma omp target map(to argc // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected ',' or ')' in 'map' clause}}
+ {}
#pragma omp target map(to:) // expected-error {{expected expression}}
+ {}
#pragma omp target map(from: argc, // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}}
+ {}
#pragma omp target map(x: y) // expected-error {{incorrect map type, expected one of 'to', 'from', 'tofrom', 'alloc', 'release', or 'delete'}}
+ {}
#pragma omp target map(x)
foo();
#pragma omp target map(tofrom: t[:I])
foo();
-#pragma omp target map(T: a) // expected-error {{incorrect map type, expected one of 'to', 'from', 'tofrom', 'alloc', 'release', or 'delete'}}
+#pragma omp target map(T: a) // expected-error {{incorrect map type, expected one of 'to', 'from', 'tofrom', 'alloc', 'release', or 'delete'}} expected-error {{incomplete type 'S1' where a complete type is required}}
foo();
#pragma omp target map(T) // expected-error {{'T' does not refer to a value}}
foo();
-#pragma omp target map(I) // expected-error 2 {{expected variable name, array element or array section}}
+#pragma omp target map(I) // expected-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
foo();
#pragma omp target map(S2::S2s)
foo();
foo();
#pragma omp target map(to, x)
foo();
-#pragma omp target map(to x) // expected-error {{expected ',' or ')' in 'map' clause}}
-#pragma omp target map(tofrom: argc > 0 ? x : y) // expected-error 2 {{expected variable name, array element or array section}}
-#pragma omp target map(argc)
-#pragma omp target map(S1) // expected-error {{'S1' does not refer to a value}}
-#pragma omp target map(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}} expected-error 2 {{type 'S2' is not mappable to target}}
-#pragma omp target map(ba) // expected-error 2 {{type 'S2' is not mappable to target}}
-#pragma omp target map(ca)
-#pragma omp target map(da)
-#pragma omp target map(S2::S2s)
-#pragma omp target map(S2::S2sc)
-#pragma omp target map(e, g)
-#pragma omp target map(h) // expected-error {{threadprivate variables are not allowed in map clause}}
-#pragma omp target map(k), map(k) // expected-error 2 {{variable already marked as mapped in current construct}} expected-note 2 {{used here}}
-#pragma omp target map(k), map(k[:5]) // expected-error 2 {{variable already marked as mapped in current construct}} expected-note 2 {{used here}}
+#endif
+#pragma omp target data map(to x) // expected-error {{expected ',' or ')' in 'map' clause}}
+#pragma omp target data map(tofrom: argc > 0 ? x : y) // expected-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target data map(argc)
+#pragma omp target data map(S1) // expected-error {{'S1' does not refer to a value}}
+#pragma omp target data map(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}} expected-error 2 {{type 'S2' is not mappable to target}}
+#pragma omp target data map(ba) // expected-error 2 {{type 'S2' is not mappable to target}}
+#pragma omp target data map(ca)
+#pragma omp target data map(da)
+#pragma omp target data map(S2::S2s)
+#pragma omp target data map(S2::S2sc)
+#pragma omp target data map(e, g)
+#pragma omp target data map(h) // expected-error {{threadprivate variables are not allowed in map clause}}
+#pragma omp target data map(k) map(k) // expected-error 2 {{variable already marked as mapped in current construct}} expected-note 2 {{used here}}
+#pragma omp target map(k), map(k[:5]) // expected-error 2 {{pointer cannot be mapped along with a section derived from itself}} expected-note 2 {{used here}}
foo();
-#pragma omp target map(da)
+#if 1
+#pragma omp target data map(da)
#pragma omp target map(da[:4])
foo();
-#pragma omp target map(k, j, l) // expected-note 4 {{used here}}
-#pragma omp target map(k[:4]) // expected-error 2 {{variable already marked as mapped in current construct}}
-#pragma omp target map(j)
-#pragma omp target map(l[:5]) // expected-error 2 {{variable already marked as mapped in current construct}}
+#pragma omp target data map(k, j, l) // expected-note 2 {{used here}}
+#pragma omp target data map(k[:4]) // expected-error 2 {{pointer cannot be mapped along with a section derived from itself}}
+#pragma omp target data map(j)
+#pragma omp target map(l) map(l[:5]) // expected-error 2 {{variable already marked as mapped in current construct}} expected-note 2 {{used here}}
foo();
-#pragma omp target map(k[:4], j, l[:5]) // expected-note 4 {{used here}}
-#pragma omp target map(k) // expected-error 2 {{variable already marked as mapped in current construct}}
-#pragma omp target map(j)
-#pragma omp target map(l) // expected-error 2 {{variable already marked as mapped in current construct}}
+#pragma omp target data map(k[:4], j, l[:5]) // expected-note 4 {{used here}}
+#pragma omp target data map(k) // expected-error 2 {{pointer cannot be mapped along with a section derived from itself}}
+#pragma omp target data map(j)
+#pragma omp target map(l) // expected-error 2 {{original storage of expression in data environment is shared but data environment do not fully contain mapped expression storage}}
foo();
-#pragma omp target map(always, tofrom: x)
-#pragma omp target map(always: x) // expected-error {{missing map type}}
-#pragma omp target map(tofrom, always: x) // expected-error {{incorrect map type modifier, expected 'always'}} expected-error {{incorrect map type, expected one of 'to', 'from', 'tofrom', 'alloc', 'release', or 'delete'}}
-#pragma omp target map(always, tofrom: always, tofrom, x)
+#pragma omp target data map(always, tofrom: x)
+#pragma omp target data map(always: x) // expected-error {{missing map type}}
+#pragma omp target data map(tofrom, always: x) // expected-error {{incorrect map type modifier, expected 'always'}} expected-error {{incorrect map type, expected one of 'to', 'from', 'tofrom', 'alloc', 'release', or 'delete'}}
+#pragma omp target data map(always, tofrom: always, tofrom, x)
#pragma omp target map(tofrom j) // expected-error {{expected ',' or ')' in 'map' clause}}
foo();
-
+#endif
return 0;
}
int y;
int to, tofrom, always;
const int (&l)[5] = da;
-#pragma omp target map // expected-error {{expected '(' after 'map'}}
-#pragma omp target map( // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected expression}}
-#pragma omp target map() // expected-error {{expected expression}}
-#pragma omp target map(alloc) // expected-error {{use of undeclared identifier 'alloc'}}
-#pragma omp target map(to argc // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected ',' or ')' in 'map' clause}}
-#pragma omp target map(to:) // expected-error {{expected expression}}
-#pragma omp target map(from: argc, // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}}
-#pragma omp target map(x: y) // expected-error {{incorrect map type, expected one of 'to', 'from', 'tofrom', 'alloc', 'release', or 'delete'}}
+#if 1
+#pragma omp target data map // expected-error {{expected '(' after 'map'}} expected-error {{expected at least one map clause for '#pragma omp target data'}}
+#pragma omp target data map( // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected expression}}
+#pragma omp target data map() // expected-error {{expected expression}}
+#pragma omp target data map(alloc) // expected-error {{use of undeclared identifier 'alloc'}}
+#pragma omp target data map(to argc // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected ',' or ')' in 'map' clause}}
+#pragma omp target data map(to:) // expected-error {{expected expression}}
+#pragma omp target data map(from: argc, // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}}
+#pragma omp target data map(x: y) // expected-error {{incorrect map type, expected one of 'to', 'from', 'tofrom', 'alloc', 'release', or 'delete'}}
#pragma omp target map(x)
foo();
#pragma omp target map(to: x)
foo();
#pragma omp target map(to, x)
foo();
-#pragma omp target map(to x) // expected-error {{expected ',' or ')' in 'map' clause}}
-#pragma omp target map(tofrom: argc > 0 ? argv[1] : argv[2]) // expected-error {{expected variable name, array element or array section}}
-#pragma omp target map(argc)
-#pragma omp target map(S1) // expected-error {{'S1' does not refer to a value}}
-#pragma omp target map(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}} expected-error 2 {{type 'S2' is not mappable to target}}
-#pragma omp target map(argv[1])
-#pragma omp target map(ba) // expected-error 2 {{type 'S2' is not mappable to target}}
-#pragma omp target map(ca)
-#pragma omp target map(da)
-#pragma omp target map(S2::S2s)
-#pragma omp target map(S2::S2sc)
-#pragma omp target map(e, g)
-#pragma omp target map(h) // expected-error {{threadprivate variables are not allowed in map clause}}
-#pragma omp target map(k), map(k) // expected-error {{variable already marked as mapped in current construct}} expected-note {{used here}}
-#pragma omp target map(k), map(k[:5]) // expected-error {{variable already marked as mapped in current construct}} expected-note {{used here}}
+#pragma omp target data map(to x) // expected-error {{expected ',' or ')' in 'map' clause}}
+#pragma omp target data map(tofrom: argc > 0 ? argv[1] : argv[2]) // expected-error {{xpected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target data map(argc)
+#pragma omp target data map(S1) // expected-error {{'S1' does not refer to a value}}
+#pragma omp target data map(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}} expected-error 2 {{type 'S2' is not mappable to target}}
+#pragma omp target data map(argv[1])
+#pragma omp target data map(ba) // expected-error 2 {{type 'S2' is not mappable to target}}
+#pragma omp target data map(ca)
+#pragma omp target data map(da)
+#pragma omp target data map(S2::S2s)
+#pragma omp target data map(S2::S2sc)
+#pragma omp target data map(e, g)
+#pragma omp target data map(h) // expected-error {{threadprivate variables are not allowed in map clause}}
+#pragma omp target data map(k), map(k) // expected-error {{variable already marked as mapped in current construct}} expected-note {{used here}}
+#pragma omp target map(k), map(k[:5]) // expected-error {{pointer cannot be mapped along with a section derived from itself}} expected-note {{used here}}
foo();
-#pragma omp target map(da)
+#pragma omp target data map(da)
#pragma omp target map(da[:4])
foo();
-#pragma omp target map(k, j, l) // expected-note 2 {{used here}}
-#pragma omp target map(k[:4]) // expected-error {{variable already marked as mapped in current construct}}
-#pragma omp target map(j)
-#pragma omp target map(l[:5]) // expected-error {{variable already marked as mapped in current construct}}
+#pragma omp target data map(k, j, l) // expected-note {{used here}}
+#pragma omp target data map(k[:4]) // expected-error {{pointer cannot be mapped along with a section derived from itself}}
+#pragma omp target data map(j)
+#pragma omp target map(l) map(l[:5]) // expected-error {{variable already marked as mapped in current construct}} expected-note {{used here}}
foo();
-#pragma omp target map(k[:4], j, l[:5]) // expected-note 2 {{used here}}
-#pragma omp target map(k) // expected-error {{variable already marked as mapped in current construct}}
-#pragma omp target map(j)
-#pragma omp target map(l) // expected-error {{variable already marked as mapped in current construct}}
+#pragma omp target data map(k[:4], j, l[:5]) // expected-note 2 {{used here}}
+#pragma omp target data map(k) // expected-error {{pointer cannot be mapped along with a section derived from itself}}
+#pragma omp target data map(j)
+#pragma omp target map(l) // expected-error {{original storage of expression in data environment is shared but data environment do not fully contain mapped expression storage}}
foo();
-#pragma omp target map(always, tofrom: x)
-#pragma omp target map(always: x) // expected-error {{missing map type}}
-#pragma omp target map(tofrom, always: x) // expected-error {{incorrect map type modifier, expected 'always'}} expected-error {{incorrect map type, expected one of 'to', 'from', 'tofrom', 'alloc', 'release', or 'delete'}}
-#pragma omp target map(always, tofrom: always, tofrom, x)
+#pragma omp target data map(always, tofrom: x)
+#pragma omp target data map(always: x) // expected-error {{missing map type}}
+#pragma omp target data map(tofrom, always: x) // expected-error {{incorrect map type modifier, expected 'always'}} expected-error {{incorrect map type, expected one of 'to', 'from', 'tofrom', 'alloc', 'release', or 'delete'}}
+#pragma omp target data map(always, tofrom: always, tofrom, x)
#pragma omp target map(tofrom j) // expected-error {{expected ',' or ')' in 'map' clause}}
foo();
-
+#endif
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}}
}
projects / clang / commitdiff