Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

[SYCL] Turning off the deprecated attribute propagation and Sycl2017Compat option specification #14984

Merged
merged 5 commits into from
Aug 14, 2024
Merged
Show file tree
Hide file tree
Changes from all commits
Commits
File filter

Filter by extension

Filter by extension

Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
3 changes: 1 addition & 2 deletions clang/include/clang/Basic/DiagnosticGroups.td
Original file line number Diff line number Diff line change
Expand Up @@ -1353,9 +1353,8 @@ def OpenMP : DiagGroup<"openmp", [
]>;

// SYCL warnings
def Sycl2017Compat : DiagGroup<"sycl-2017-compat">;
def Sycl2020Compat : DiagGroup<"sycl-2020-compat">;
def SyclStrict : DiagGroup<"sycl-strict", [ Sycl2017Compat, Sycl2020Compat]>;
def SyclStrict : DiagGroup<"sycl-strict", [ Sycl2020Compat]>;
def SyclTarget : DiagGroup<"sycl-target">;
def SyclFPGAMismatch : DiagGroup<"sycl-fpga-mismatch">;
def SyclAspectMismatch : DiagGroup<"sycl-aspect-mismatch">;
Expand Down
11 changes: 0 additions & 11 deletions clang/include/clang/Basic/DiagnosticSemaKinds.td
Original file line number Diff line number Diff line change
Expand Up @@ -12444,14 +12444,6 @@ def err_sycl_invalid_aspect_argument : Error<
def warn_sycl_pass_by_value_deprecated
: Warning<"passing kernel functions by value is deprecated in SYCL 2020">,
InGroup<Sycl2020Compat>, ShowInSystemHeader;
def warn_sycl_pass_by_reference_future
: Warning<"passing of kernel functions by reference is a SYCL 2020 extension">,
InGroup<Sycl2017Compat>, ShowInSystemHeader;
def warn_sycl_implicit_decl
: Warning<"SYCL 1.2.1 specification requires an explicit forward "
"declaration for a kernel type name; your program may not "
"be portable">,
InGroup<SyclStrict>, ShowInSystemHeader, DefaultIgnore;
def warn_sycl_potentially_invalid_as_cast : Warning<
"explicit cast from %0 to %1 potentially leads to an invalid address space"
" cast in the resulting code">, InGroup<SyclStrict>,
Expand Down Expand Up @@ -12487,9 +12479,6 @@ def err_sycl_mismatch_group_size
def note_sycl_kernel_declared_here : Note<"kernel declared here">;
def err_sycl_expected_finalize_method : Error<
"expected a 'finalize' method for the 'stream' class">;
def ext_sycl_2020_attr_spelling : ExtWarn<
"use of attribute %0 is a SYCL 2020 extension">,
InGroup<Sycl2017Compat>;
def err_sycl_esimd_not_supported_for_type : Error<
"type %0 is not supported in ESIMD context">;
def err_sycl_taking_address_of_wrong_function : Error<
Expand Down
2 changes: 0 additions & 2 deletions clang/include/clang/Basic/LangOptions.h
Original file line number Diff line number Diff line change
Expand Up @@ -146,15 +146,13 @@ class LangOptionsBase {

enum SYCLMajorVersion {
SYCL_None,
SYCL_2017,
SYCL_2020,
// The "default" SYCL version to be used when none is specified on the
// frontend command line.
SYCL_Default = SYCL_2020
};

enum class SYCLVersionList {
sycl_1_2_1,
undefined
};

Expand Down
22 changes: 4 additions & 18 deletions clang/lib/Sema/SemaDeclAttr.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -156,8 +156,8 @@ void Sema::CheckDeprecatedSYCLAttributeSpelling(const ParsedAttr &A,
return;
}

// Diagnose SYCL 2017 spellings in later SYCL modes.
if (LangOpts.getSYCLVersion() > LangOptions::SYCL_2017) {
// Diagnose SYCL 2020 spellings in later SYCL modes.
if (LangOpts.getSYCLVersion() >= LangOptions::SYCL_2020) {
// All attributes in the cl vendor namespace are deprecated in favor of a
// name in the sycl namespace as of SYCL 2020.
if (A.hasScope() && A.getScopeName()->isStr("cl")) {
Expand All @@ -182,14 +182,6 @@ void Sema::CheckDeprecatedSYCLAttributeSpelling(const ParsedAttr &A,
return;
}
}

// Diagnose SYCL 2020 spellings used in earlier SYCL modes as being an
// extension.
if (LangOpts.getSYCLVersion() == LangOptions::SYCL_2017 && A.hasScope() &&
A.getScopeName()->isStr("sycl")) {
Diag(A.getLoc(), diag::ext_sycl_2020_attr_spelling) << A;
return;
}
}

/// Check if IdxExpr is a valid parameter index for a function or
Expand Down Expand Up @@ -4378,19 +4370,13 @@ static void handleSYCLIntelLoopFuseAttr(Sema &S, Decl *D, const ParsedAttr &A) {
}

static void handleVecTypeHint(Sema &S, Decl *D, const ParsedAttr &AL) {
// This attribute is deprecated without replacement in SYCL 2020 mode.
// Given attribute is deprecated without replacement in SYCL 2020 mode.
// Ignore the attribute in SYCL 2020.
if (S.LangOpts.getSYCLVersion() > LangOptions::SYCL_2017) {
if (S.LangOpts.getSYCLVersion() >= LangOptions::SYCL_2020) {
S.Diag(AL.getLoc(), diag::warn_attribute_deprecated_ignored) << AL;
return;
}

// If the attribute is used with the [[sycl::vec_type_hint]] spelling in SYCL
// 2017 mode, we want to warn about using the newer name in the older
// standard as a compatibility extension.
if (S.LangOpts.getSYCLVersion() == LangOptions::SYCL_2017 && AL.hasScope())
S.Diag(AL.getLoc(), diag::ext_sycl_2020_attr_spelling) << AL;

if (!AL.hasParsedType()) {
S.Diag(AL.getLoc(), diag::err_attribute_wrong_number_arguments) << AL << 1;
return;
Expand Down
30 changes: 7 additions & 23 deletions clang/lib/Sema/SemaSYCL.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -539,11 +539,9 @@ static void collectSYCLAttributes(SemaSYCL &S, FunctionDecl *FD,
if (!FD->hasAttrs())
return;

// In SYCL 1.2.1 mode, the attributes are propagated from the function they
// are applied to onto the kernel which calls the function.
// In SYCL 2020 mode, the attributes are not propagated to the kernel.
if (DirectlyCalled || S.getASTContext().getLangOpts().getSYCLVersion() <
LangOptions::SYCL_2020) {
// In SYCL 2020 mode, the attributes aren't propagated from the function they
// are applied on to the kernel which calls the function.
if (DirectlyCalled) {
llvm::copy_if(FD->getAttrs(), std::back_inserter(Attrs), [](Attr *A) {
// FIXME: Make this list self-adapt as new SYCL attributes are added.
return isa<IntelReqdSubGroupSizeAttr, IntelNamedSubGroupSizeAttr,
Expand All @@ -553,14 +551,8 @@ static void collectSYCLAttributes(SemaSYCL &S, FunctionDecl *FD,
SYCLIntelMaxWorkGroupSizeAttr, SYCLIntelMaxGlobalWorkDimAttr,
SYCLIntelMinWorkGroupsPerComputeUnitAttr,
SYCLIntelMaxWorkGroupsPerMultiprocessorAttr,
SYCLIntelNoGlobalWorkOffsetAttr, SYCLSimdAttr>(A);
});
}

// Attributes that should not be propagated from device functions to a kernel.
if (DirectlyCalled) {
llvm::copy_if(FD->getAttrs(), std::back_inserter(Attrs), [](Attr *A) {
return isa<SYCLIntelLoopFuseAttr, SYCLIntelMaxConcurrencyAttr,
SYCLIntelNoGlobalWorkOffsetAttr, SYCLSimdAttr,
SYCLIntelLoopFuseAttr, SYCLIntelMaxConcurrencyAttr,
SYCLIntelDisableLoopPipeliningAttr,
SYCLIntelInitiationIntervalAttr,
SYCLIntelUseStallEnableClustersAttr, SYCLDeviceHasAttr,
Expand Down Expand Up @@ -4868,10 +4860,6 @@ class SYCLKernelNameTypeVisitor
scope */
0 << KernelNameType;
IsInvalid = true;
} else {
S.Diag(KernelInvocationFuncLoc, diag::warn_sycl_implicit_decl);
S.Diag(DeclNamed->getLocation(), diag::note_previous_decl)
<< DeclNamed->getName();
}
}
}
Expand Down Expand Up @@ -4942,13 +4930,9 @@ void SemaSYCL::CheckSYCLKernelCall(FunctionDecl *KernelFunc,

// check that calling kernel conforms to spec
QualType KernelParamTy = KernelFunc->getParamDecl(0)->getType();
if (KernelParamTy->isReferenceType()) {
// passing by reference, so emit warning if not using SYCL 2020
if (SemaRef.LangOpts.getSYCLVersion() < LangOptions::SYCL_2020)
Diag(KernelFunc->getLocation(), diag::warn_sycl_pass_by_reference_future);
} else {
if (not KernelParamTy->isReferenceType()) {
// passing by value. emit warning if using SYCL 2020 or greater
if (SemaRef.LangOpts.getSYCLVersion() > LangOptions::SYCL_2017)
if (SemaRef.LangOpts.getSYCLVersion() >= LangOptions::SYCL_2020)
Diag(KernelFunc->getLocation(), diag::warn_sycl_pass_by_value_deprecated);
}

Expand Down
26 changes: 0 additions & 26 deletions clang/test/SemaSYCL/implicit_kernel_type.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -29,32 +29,6 @@ int main() {
h.single_task<InvalidKernelName1>([]() {});
});

#if defined(WARN)
// expected-warning@#KernelSingleTask {{SYCL 1.2.1 specification requires an explicit forward declaration for a kernel type name; your program may not be portable}}
#elif defined(ERROR)
// expected-error@#KernelSingleTask {{SYCL 1.2.1 specification requires an explicit forward declaration for a kernel type name; your program may not be portable}}
#endif

q.submit([&](handler &h) {
// expected-note@+2 {{fake_kernel declared here}}
// expected-note@+1 {{in instantiation of function template specialization}}
h.single_task<class fake_kernel>([]() { function(); });
});

#if defined(WARN)
// expected-warning@#KernelSingleTask {{SYCL 1.2.1 specification requires an explicit forward declaration for a kernel type name; your program may not be portable}}
#elif defined(ERROR)
// expected-error@#KernelSingleTask {{SYCL 1.2.1 specification requires an explicit forward declaration for a kernel type name; your program may not be portable}}
#endif

q.submit([&](handler &h) {
// expected-note@+2 {{fake_kernel2 declared here}}
// expected-note@+1 {{in instantiation of function template specialization}}
h.single_task<class fake_kernel2>([]() {
auto l = [](auto f) { f(); };
});
});

q.submit([&](handler &h) {
h.single_task<class myWrapper>([]() { function(); });
});
Expand Down
Loading