aboutsummaryrefslogtreecommitdiff
path: root/clang/lib/Sema
diff options
context:
space:
mode:
Diffstat (limited to 'clang/lib/Sema')
-rw-r--r--clang/lib/Sema/DeclSpec.cpp2
-rw-r--r--clang/lib/Sema/ParsedAttr.cpp2
-rw-r--r--clang/lib/Sema/SemaAMDGPU.cpp103
-rw-r--r--clang/lib/Sema/SemaDecl.cpp2
-rw-r--r--clang/lib/Sema/SemaExpr.cpp22
-rw-r--r--clang/lib/Sema/SemaOpenACC.cpp6
-rw-r--r--clang/lib/Sema/SemaOpenACCClause.cpp114
7 files changed, 179 insertions, 72 deletions
diff --git a/clang/lib/Sema/DeclSpec.cpp b/clang/lib/Sema/DeclSpec.cpp
index 8756ce5..184d31e 100644
--- a/clang/lib/Sema/DeclSpec.cpp
+++ b/clang/lib/Sema/DeclSpec.cpp
@@ -197,7 +197,7 @@ DeclaratorChunk DeclaratorChunk::getFunction(bool hasProto,
[&](DeclSpec::TQ TypeQual, StringRef PrintName, SourceLocation SL) {
I.Fun.MethodQualifiers->SetTypeQual(TypeQual, SL);
});
- I.Fun.MethodQualifiers->getAttributes().takeAllFrom(attrs);
+ I.Fun.MethodQualifiers->getAttributes().takeAllPrependingFrom(attrs);
I.Fun.MethodQualifiers->getAttributePool().takeAllFrom(attrs.getPool());
}
diff --git a/clang/lib/Sema/ParsedAttr.cpp b/clang/lib/Sema/ParsedAttr.cpp
index 294f88e..2b5ad33 100644
--- a/clang/lib/Sema/ParsedAttr.cpp
+++ b/clang/lib/Sema/ParsedAttr.cpp
@@ -304,7 +304,7 @@ bool ParsedAttr::checkAtMostNumArgs(Sema &S, unsigned Num) const {
void clang::takeAndConcatenateAttrs(ParsedAttributes &First,
ParsedAttributes &&Second) {
- First.takeAllAtEndFrom(Second);
+ First.takeAllAppendingFrom(Second);
if (!First.Range.getBegin().isValid())
First.Range.setBegin(Second.Range.getBegin());
diff --git a/clang/lib/Sema/SemaAMDGPU.cpp b/clang/lib/Sema/SemaAMDGPU.cpp
index 3a0c231..e32f437 100644
--- a/clang/lib/Sema/SemaAMDGPU.cpp
+++ b/clang/lib/Sema/SemaAMDGPU.cpp
@@ -11,6 +11,7 @@
//===----------------------------------------------------------------------===//
#include "clang/Sema/SemaAMDGPU.h"
+#include "clang/Basic/DiagnosticFrontend.h"
#include "clang/Basic/DiagnosticSema.h"
#include "clang/Basic/TargetBuiltins.h"
#include "clang/Sema/Ownership.h"
@@ -111,6 +112,108 @@ bool SemaAMDGPU::CheckAMDGCNBuiltinFunctionCall(unsigned BuiltinID,
case AMDGPU::BI__builtin_amdgcn_cooperative_atomic_store_16x8B:
case AMDGPU::BI__builtin_amdgcn_cooperative_atomic_store_8x16B:
return checkCoopAtomicFunctionCall(TheCall, /*IsStore=*/true);
+ case AMDGPU::BI__builtin_amdgcn_image_load_1d_v4f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_1darray_v4f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_1d_v4f16_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_1darray_v4f16_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_2d_f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_2d_v4f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_2d_v4f16_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_2darray_f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_2darray_v4f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_2darray_v4f16_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_3d_v4f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_3d_v4f16_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_cube_v4f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_cube_v4f16_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_mip_1d_v4f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_mip_1d_v4f16_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_mip_1darray_v4f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_mip_1darray_v4f16_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_mip_2d_f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_mip_2d_v4f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_mip_2d_v4f16_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_mip_2darray_f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_mip_2darray_v4f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_mip_2darray_v4f16_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_mip_3d_v4f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_mip_3d_v4f16_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_mip_cube_v4f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_load_mip_cube_v4f16_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_sample_1d_v4f32_f32:
+ case AMDGPU::BI__builtin_amdgcn_image_sample_1darray_v4f32_f32:
+ case AMDGPU::BI__builtin_amdgcn_image_sample_1d_v4f16_f32:
+ case AMDGPU::BI__builtin_amdgcn_image_sample_1darray_v4f16_f32:
+ case AMDGPU::BI__builtin_amdgcn_image_sample_2d_f32_f32:
+ case AMDGPU::BI__builtin_amdgcn_image_sample_2d_v4f32_f32:
+ case AMDGPU::BI__builtin_amdgcn_image_sample_2d_v4f16_f32:
+ case AMDGPU::BI__builtin_amdgcn_image_sample_2darray_f32_f32:
+ case AMDGPU::BI__builtin_amdgcn_image_sample_2darray_v4f32_f32:
+ case AMDGPU::BI__builtin_amdgcn_image_sample_2darray_v4f16_f32:
+ case AMDGPU::BI__builtin_amdgcn_image_sample_3d_v4f32_f32:
+ case AMDGPU::BI__builtin_amdgcn_image_sample_3d_v4f16_f32:
+ case AMDGPU::BI__builtin_amdgcn_image_sample_cube_v4f32_f32:
+ case AMDGPU::BI__builtin_amdgcn_image_sample_cube_v4f16_f32: {
+ StringRef FeatureList(
+ getASTContext().BuiltinInfo.getRequiredFeatures(BuiltinID));
+ if (!Builtin::evaluateRequiredTargetFeatures(FeatureList,
+ CallerFeatureMap)) {
+ Diag(TheCall->getBeginLoc(), diag::err_builtin_needs_feature)
+ << FD->getDeclName() << FeatureList;
+ return false;
+ }
+
+ unsigned ArgCount = TheCall->getNumArgs() - 1;
+ llvm::APSInt Result;
+
+ return (SemaRef.BuiltinConstantArg(TheCall, 0, Result)) ||
+ (SemaRef.BuiltinConstantArg(TheCall, ArgCount, Result)) ||
+ (SemaRef.BuiltinConstantArg(TheCall, (ArgCount - 1), Result));
+ }
+ case AMDGPU::BI__builtin_amdgcn_image_store_1d_v4f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_1darray_v4f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_1d_v4f16_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_1darray_v4f16_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_2d_f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_2d_v4f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_2d_v4f16_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_2darray_f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_2darray_v4f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_2darray_v4f16_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_3d_v4f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_3d_v4f16_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_cube_v4f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_cube_v4f16_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_mip_1d_v4f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_mip_1d_v4f16_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_mip_1darray_v4f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_mip_1darray_v4f16_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_mip_2d_f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_mip_2d_v4f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_mip_2d_v4f16_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_mip_2darray_f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_mip_2darray_v4f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_mip_2darray_v4f16_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_mip_3d_v4f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_mip_3d_v4f16_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_mip_cube_v4f32_i32:
+ case AMDGPU::BI__builtin_amdgcn_image_store_mip_cube_v4f16_i32: {
+ StringRef FeatureList(
+ getASTContext().BuiltinInfo.getRequiredFeatures(BuiltinID));
+ if (!Builtin::evaluateRequiredTargetFeatures(FeatureList,
+ CallerFeatureMap)) {
+ Diag(TheCall->getBeginLoc(), diag::err_builtin_needs_feature)
+ << FD->getDeclName() << FeatureList;
+ return false;
+ }
+
+ unsigned ArgCount = TheCall->getNumArgs() - 1;
+ llvm::APSInt Result;
+
+ return (SemaRef.BuiltinConstantArg(TheCall, 1, Result)) ||
+ (SemaRef.BuiltinConstantArg(TheCall, ArgCount, Result)) ||
+ (SemaRef.BuiltinConstantArg(TheCall, (ArgCount - 1), Result));
+ }
default:
return false;
}
diff --git a/clang/lib/Sema/SemaDecl.cpp b/clang/lib/Sema/SemaDecl.cpp
index 6eaf7b9..0e83c20 100644
--- a/clang/lib/Sema/SemaDecl.cpp
+++ b/clang/lib/Sema/SemaDecl.cpp
@@ -14637,7 +14637,7 @@ StmtResult Sema::ActOnCXXForRangeIdentifier(Scope *S, SourceLocation IdentLoc,
Declarator D(DS, ParsedAttributesView::none(), DeclaratorContext::ForInit);
D.SetIdentifier(Ident, IdentLoc);
- D.takeAttributes(Attrs);
+ D.takeAttributesAppending(Attrs);
D.AddTypeInfo(DeclaratorChunk::getReference(0, IdentLoc, /*lvalue*/ false),
IdentLoc);
diff --git a/clang/lib/Sema/SemaExpr.cpp b/clang/lib/Sema/SemaExpr.cpp
index 4230ea7..01abc1f 100644
--- a/clang/lib/Sema/SemaExpr.cpp
+++ b/clang/lib/Sema/SemaExpr.cpp
@@ -166,6 +166,11 @@ static void diagnoseUseOfInternalDeclInInlineFunction(Sema &S,
// This is disabled under C++; there are too many ways for this to fire in
// contexts where the warning is a false positive, or where it is technically
// correct but benign.
+ //
+ // WG14 N3622 which removed the constraint entirely in C2y. It is left
+ // enabled in earlier language modes because this is a constraint in those
+ // language modes. But in C2y mode, we still want to issue the "incompatible
+ // with previous standards" diagnostic, too.
if (S.getLangOpts().CPlusPlus)
return;
@@ -190,16 +195,17 @@ static void diagnoseUseOfInternalDeclInInlineFunction(Sema &S,
// This last can give us false negatives, but it's better than warning on
// wrappers for simple C library functions.
const FunctionDecl *UsedFn = dyn_cast<FunctionDecl>(D);
- bool DowngradeWarning = S.getSourceManager().isInMainFile(Loc);
- if (!DowngradeWarning && UsedFn)
- DowngradeWarning = UsedFn->isInlined() || UsedFn->hasAttr<ConstAttr>();
-
- S.Diag(Loc, DowngradeWarning ? diag::ext_internal_in_extern_inline_quiet
- : diag::ext_internal_in_extern_inline)
- << /*IsVar=*/!UsedFn << D;
+ unsigned DiagID;
+ if (S.getLangOpts().C2y)
+ DiagID = diag::warn_c2y_compat_internal_in_extern_inline;
+ else if ((UsedFn && (UsedFn->isInlined() || UsedFn->hasAttr<ConstAttr>())) ||
+ S.getSourceManager().isInMainFile(Loc))
+ DiagID = diag::ext_internal_in_extern_inline_quiet;
+ else
+ DiagID = diag::ext_internal_in_extern_inline;
+ S.Diag(Loc, DiagID) << /*IsVar=*/!UsedFn << D;
S.MaybeSuggestAddingStaticToDecl(Current);
-
S.Diag(D->getCanonicalDecl()->getLocation(), diag::note_entity_declared_at)
<< D;
}
diff --git a/clang/lib/Sema/SemaOpenACC.cpp b/clang/lib/Sema/SemaOpenACC.cpp
index 4824b5a..f3969a9 100644
--- a/clang/lib/Sema/SemaOpenACC.cpp
+++ b/clang/lib/Sema/SemaOpenACC.cpp
@@ -2759,7 +2759,7 @@ OpenACCPrivateRecipe SemaOpenACC::CreatePrivateInitRecipe(const Expr *VarExpr) {
// Array sections are special, and we have to treat them that way.
if (const auto *ASE =
dyn_cast<ArraySectionExpr>(VarExpr->IgnoreParenImpCasts()))
- VarTy = ArraySectionExpr::getBaseOriginalType(ASE);
+ VarTy = ASE->getElementType();
VarDecl *AllocaDecl = CreateAllocaDecl(
getASTContext(), SemaRef.getCurContext(), VarExpr->getBeginLoc(),
@@ -2795,7 +2795,7 @@ SemaOpenACC::CreateFirstPrivateInitRecipe(const Expr *VarExpr) {
// Array sections are special, and we have to treat them that way.
if (const auto *ASE =
dyn_cast<ArraySectionExpr>(VarExpr->IgnoreParenImpCasts()))
- VarTy = ArraySectionExpr::getBaseOriginalType(ASE);
+ VarTy = ASE->getElementType();
VarDecl *AllocaDecl = CreateAllocaDecl(
getASTContext(), SemaRef.getCurContext(), VarExpr->getBeginLoc(),
@@ -2896,7 +2896,7 @@ OpenACCReductionRecipe SemaOpenACC::CreateReductionInitRecipe(
// Array sections are special, and we have to treat them that way.
if (const auto *ASE =
dyn_cast<ArraySectionExpr>(VarExpr->IgnoreParenImpCasts()))
- VarTy = ArraySectionExpr::getBaseOriginalType(ASE);
+ VarTy = ASE->getElementType();
VarDecl *AllocaDecl = CreateAllocaDecl(
getASTContext(), SemaRef.getCurContext(), VarExpr->getBeginLoc(),
diff --git a/clang/lib/Sema/SemaOpenACCClause.cpp b/clang/lib/Sema/SemaOpenACCClause.cpp
index b086929..881e960 100644
--- a/clang/lib/Sema/SemaOpenACCClause.cpp
+++ b/clang/lib/Sema/SemaOpenACCClause.cpp
@@ -1915,51 +1915,34 @@ SemaOpenACC::ActOnClause(ArrayRef<const OpenACCClause *> ExistingClauses,
return Result;
}
-/// OpenACC 3.3 section 2.5.15:
-/// At a mininmum, the supported data types include ... the numerical data types
-/// in C, C++, and Fortran.
-///
-/// If the reduction var is a composite variable, each
-/// member of the composite variable must be a supported datatype for the
-/// reduction operation.
-ExprResult SemaOpenACC::CheckReductionVar(OpenACCDirectiveKind DirectiveKind,
- OpenACCReductionOperator ReductionOp,
- Expr *VarExpr) {
- // For now, we only support 'scalar' types, or composites/arrays of scalar
- // types.
- VarExpr = VarExpr->IgnoreParenCasts();
+bool SemaOpenACC::CheckReductionVarType(Expr *VarExpr) {
SourceLocation VarLoc = VarExpr->getBeginLoc();
SmallVector<PartialDiagnosticAt> Notes;
- QualType CurType = VarExpr->getType();
-
- // For array like things, the expression can either be an array element
- // (subscript expr), array section, or array type. Peel those off, and add
- // notes in case we find an illegal kind. We'll allow scalar or composite of
- // scalars inside of this.
- if (auto *ASE = dyn_cast<ArraySectionExpr>(VarExpr)) {
- QualType BaseType = ArraySectionExpr::getBaseOriginalType(ASE);
+ // The standard isn't clear how many levels of 'array element' or 'subarray'
+ // are permitted, but we can handle as many as we need, so we'll strip them
+ // off here. This will result in CurType being the actual 'type' of the
+ // expression, which is what we are looking to check.
+ QualType CurType = isa<ArraySectionExpr>(VarExpr)
+ ? ArraySectionExpr::getBaseOriginalType(VarExpr)
+ : VarExpr->getType();
+
+ // This can happen when we have a dependent type in an array element that the
+ // above function has tried to 'unwrap'. Since this can only happen with
+ // dependence, just let it go.
+ if (CurType.isNull())
+ return false;
- PartialDiagnostic PD = PDiag(diag::note_acc_reduction_array)
- << diag::OACCReductionArray::Section << BaseType;
- Notes.push_back({ASE->getBeginLoc(), PD});
-
- CurType = getASTContext().getBaseElementType(BaseType);
- } else if (auto *SubExpr = dyn_cast<ArraySubscriptExpr>(VarExpr)) {
- // Array subscript already results in the type of the thing as its type, so
- // there is no type to change here.
- PartialDiagnostic PD =
- PDiag(diag::note_acc_reduction_array)
- << diag::OACCReductionArray::Subscript
- << SubExpr->getBase()->IgnoreParenImpCasts()->getType();
- Notes.push_back({SubExpr->getBeginLoc(), PD});
- } else if (auto *AT = getASTContext().getAsArrayType(CurType)) {
+ // If we are still an array type, we allow 1 level of 'unpeeling' of the
+ // array. The standard isn't clear here whether this is allowed, but
+ // array-of-valid-things makes sense.
+ if (auto *AT = getASTContext().getAsArrayType(CurType)) {
// If we're already the array type, peel off the array and leave the element
// type.
- CurType = getASTContext().getBaseElementType(AT);
PartialDiagnostic PD = PDiag(diag::note_acc_reduction_array)
<< diag::OACCReductionArray::ArrayTy << CurType;
Notes.push_back({VarLoc, PD});
+ CurType = AT->getElementType();
}
auto IsValidMemberOfComposite = [](QualType Ty) {
@@ -1974,31 +1957,26 @@ ExprResult SemaOpenACC::CheckReductionVar(OpenACCDirectiveKind DirectiveKind,
for (auto [Loc, PD] : Notes)
Diag(Loc, PD);
- Diag(VarLoc, diag::note_acc_reduction_type_summary);
+ return Diag(VarLoc, diag::note_acc_reduction_type_summary);
};
// If the type is already scalar, or is dependent, just give up.
if (IsValidMemberOfComposite(CurType)) {
// Nothing to do here, is valid.
} else if (auto *RD = CurType->getAsRecordDecl()) {
- if (!RD->isStruct() && !RD->isClass()) {
- EmitDiags(VarLoc, PDiag(diag::err_acc_reduction_type)
- << RD << diag::OACCReductionTy::NotClassStruct);
- return ExprError();
- }
+ if (!RD->isStruct() && !RD->isClass())
+ return EmitDiags(VarLoc, PDiag(diag::err_acc_reduction_type)
+ << RD
+ << diag::OACCReductionTy::NotClassStruct);
- if (!RD->isCompleteDefinition()) {
- EmitDiags(VarLoc, PDiag(diag::err_acc_reduction_type)
- << RD << diag::OACCReductionTy::NotComplete);
- return ExprError();
- }
+ if (!RD->isCompleteDefinition())
+ return EmitDiags(VarLoc, PDiag(diag::err_acc_reduction_type)
+ << RD << diag::OACCReductionTy::NotComplete);
if (const auto *CXXRD = dyn_cast<CXXRecordDecl>(RD);
- CXXRD && !CXXRD->isAggregate()) {
- EmitDiags(VarLoc, PDiag(diag::err_acc_reduction_type)
- << CXXRD << diag::OACCReductionTy::NotAgg);
- return ExprError();
- }
+ CXXRD && !CXXRD->isAggregate())
+ return EmitDiags(VarLoc, PDiag(diag::err_acc_reduction_type)
+ << CXXRD << diag::OACCReductionTy::NotAgg);
for (FieldDecl *FD : RD->fields()) {
if (!IsValidMemberOfComposite(FD->getType())) {
@@ -2007,17 +1985,37 @@ ExprResult SemaOpenACC::CheckReductionVar(OpenACCDirectiveKind DirectiveKind,
<< FD->getName() << RD->getName();
Notes.push_back({FD->getBeginLoc(), PD});
// TODO: member here.note_acc_reduction_member_of_composite
- EmitDiags(VarLoc, PDiag(diag::err_acc_reduction_type)
- << FD->getType()
- << diag::OACCReductionTy::MemberNotScalar);
- return ExprError();
+ return EmitDiags(VarLoc, PDiag(diag::err_acc_reduction_type)
+ << FD->getType()
+ << diag::OACCReductionTy::MemberNotScalar);
}
}
} else {
- EmitDiags(VarLoc, PDiag(diag::err_acc_reduction_type)
- << CurType << diag::OACCReductionTy::NotScalar);
+ return EmitDiags(VarLoc, PDiag(diag::err_acc_reduction_type)
+ << CurType
+ << diag::OACCReductionTy::NotScalar);
}
+ return false;
+}
+
+/// OpenACC 3.3 section 2.5.15:
+/// At a mininmum, the supported data types include ... the numerical data types
+/// in C, C++, and Fortran.
+///
+/// If the reduction var is a composite variable, each
+/// member of the composite variable must be a supported datatype for the
+/// reduction operation.
+ExprResult SemaOpenACC::CheckReductionVar(OpenACCDirectiveKind DirectiveKind,
+ OpenACCReductionOperator ReductionOp,
+ Expr *VarExpr) {
+ // For now, we only support 'scalar' types, or composites/arrays of scalar
+ // types.
+ VarExpr = VarExpr->IgnoreParenCasts();
+
+ if (CheckReductionVarType(VarExpr))
+ return ExprError();
+
// OpenACC3.3: 2.9.11: Reduction clauses on nested constructs for the same
// reduction 'var' must have the same reduction operator.
if (!VarExpr->isInstantiationDependent()) {
generated by cgit v1.1 at 2025-10-11 10:30:47 +0000