Skip to content

[SYCL] Remove deprecated intel::reqd_work_group_size attribute #4433

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

Merged
merged 6 commits into from
Sep 7, 2021
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
5 changes: 2 additions & 3 deletions clang/include/clang/Basic/Attr.td
Original file line number Diff line number Diff line change
Expand Up @@ -2985,11 +2985,10 @@ def NoDeref : TypeAttr {
let Documentation = [NoDerefDocs];
}

// Default arguments can only be used with the intel::reqd_work_group_size or
// sycl::reqd_work_group_size spellings.
// Default arguments can only be used with the sycl::reqd_work_group_size
// spelling.
def ReqdWorkGroupSize : InheritableAttr {
let Spellings = [GNU<"reqd_work_group_size">,
CXX11<"intel", "reqd_work_group_size">,
CXX11<"cl", "reqd_work_group_size">,
CXX11<"sycl", "reqd_work_group_size">];
let Args = [ExprArgument<"XDim">,
Expand Down
16 changes: 4 additions & 12 deletions clang/include/clang/Basic/AttrDocs.td
Original file line number Diff line number Diff line change
Expand Up @@ -2673,10 +2673,10 @@ allows the Y and Z arguments to be optional. If not provided by the user, the
value of Y and Z defaults to 1. See section 5.8.1 Kernel Attributes for more
details.

In SYCL 1.2.1 mode, the ``intel::reqd_work_group_size``,
``cl::reqd_work_group_size``, and ``sycl::reqd_work_group_size`` 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.
In SYCL 1.2.1 mode, the ``cl::reqd_work_group_size`` and
``sycl::reqd_work_group_size`` 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.

.. code-block:: c++

Expand All @@ -2700,14 +2700,6 @@ The ``[[cl::reqd_work_group_size(X, Y, Z)]]`` and
``__attribute__((reqd_work_group_size(X, Y, Z)))`` spellings are both
deprecated in SYCL 2020.

As an Intel extension, the ``[[intel::reqd_work_group_size(X, Y, Z)]]``
spelling is supported with the same semantics as the
``[[sycl::reqd_work_group_size(X, Y, Z)]]`` spelling.

The ``[[intel::reqd_work_group_size(X, Y, Z)]]`` attribute spelling
is deprecated in favor of the SYCL 2020 attribute spelling
``[[sycl::reqd_work_group_size]]``.

In OpenCL C, this attribute is available with the GNU spelling
(``__attribute__((reqd_work_group_size(X, Y, Z)))``), see section
6.7.2 Optional Attribute Qualifiers of OpenCL 1.2 specification for details.
Expand Down
8 changes: 0 additions & 8 deletions clang/lib/Sema/SemaDeclAttr.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -338,14 +338,6 @@ void Sema::CheckDeprecatedSYCLAttributeSpelling(const ParsedAttr &A,
return;
}

// Deprecate [[intel::reqd_work_group_size]] attribute spelling in favor
// of the SYCL 2020 attribute spelling [[sycl::reqd_work_group_size]].
if (A.getKind() == ParsedAttr::AT_ReqdWorkGroupSize && A.hasScope() &&
A.getScopeName()->isStr("intel")) {
DiagnoseDeprecatedAttribute(A, "sycl", NewName);
return;
}

// All GNU-style spellings are deprecated in favor of a C++-style spelling.
if (A.getSyntax() == ParsedAttr::AS_GNU) {
// Note: we cannot suggest an automatic fix-it because GNU-style
Expand Down
4 changes: 1 addition & 3 deletions clang/test/SemaSYCL/intel-max-global-work-dim-device.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -166,9 +166,7 @@ struct TRIFuncObjBad5 {
};

struct TRIFuncObjBad6 {
[[intel::reqd_work_group_size(4)]] // expected-error{{all 'reqd_work_group_size' attribute arguments must be '1' when the 'max_global_work_dim' attribute argument is '0'}} \
// expected-warning {{attribute 'intel::reqd_work_group_size' is deprecated}} \
// expected-note {{did you mean to use 'sycl::reqd_work_group_size' instead?}}
[[sycl::reqd_work_group_size(4)]] // expected-error{{all 'reqd_work_group_size' attribute arguments must be '1' when the 'max_global_work_dim' attribute argument is '0'}}
[[intel::max_global_work_dim(0)]] void
operator()() const {}
};
Expand Down
Original file line number Diff line number Diff line change
Expand Up @@ -56,8 +56,7 @@ class Functor64 {

class Functor16x16x16 {
public:
[[intel::reqd_work_group_size(16, 16, 16)]] void operator()() const {} // expected-warning {{attribute 'intel::reqd_work_group_size' is deprecated}} \
// expected-note {{did you mean to use 'sycl::reqd_work_group_size' instead?}}
[[sycl::reqd_work_group_size(16, 16, 16)]] void operator()() const {}
};

class FunctorAttr {
Expand Down
2 changes: 2 additions & 0 deletions clang/test/SemaSYCL/intel-reqd-work-group-size-device.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -38,6 +38,8 @@ class Functor33 {
[[sycl::reqd_work_group_size(32, -4)]] void operator()() const {}
};

[[intel::reqd_work_group_size(4, 2, 9)]] void unknown() {} // expected-warning{{unknown attribute 'reqd_work_group_size' ignored}}

class Functor30 {
public:
// expected-warning@+1 2{{implicit conversion changes signedness: 'int' to 'unsigned long long'}}
Expand Down
12 changes: 5 additions & 7 deletions clang/test/SemaSYCL/num_simd_work_items_device.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -42,8 +42,8 @@ struct FuncObj {
};

#ifdef TRIGGER_ERROR
// If the declaration has a [[intel::reqd_work_group_size]]
// [[cl::reqd_work_group_size]] attribute, tests that check
// If the declaration has a [[sycl::reqd_work_group_size]]
// or [[cl::reqd_work_group_size]] attribute, tests that check
// if the work group size attribute argument (the last argument)
// can be evenly divided by the [[intel::num_simd_work_items()]] attribute.
struct TRIFuncObjBad1 {
Expand All @@ -60,7 +60,7 @@ struct TRIFuncObjBad2 {
operator()() const {}
};

// Tests for the default values of [[intel::reqd_work_group_size()]].
// Tests for the default values of [[sycl::reqd_work_group_size()]].

// FIXME: This should be accepted instead of error which turns out to be
// an implementation bug that shouldn't be visible to the user as there
Expand Down Expand Up @@ -131,9 +131,7 @@ struct TRIFuncObjBad8 {
[[intel::num_simd_work_items(2)]] // expected-error{{'num_simd_work_items' attribute must evenly divide the work-group size for the 'reqd_work_group_size' attribute}}
[[sycl::reqd_work_group_size(4, 2, 3)]] void func1(); // expected-note{{conflicting attribute is here}}

[[intel::reqd_work_group_size(4, 2, 3)]] // expected-note{{conflicting attribute is here}} \
// expected-warning {{attribute 'intel::reqd_work_group_size' is deprecated}} \
// expected-note {{did you mean to use 'sycl::reqd_work_group_size' instead?}}
[[sycl::reqd_work_group_size(4, 2, 3)]] // expected-note{{conflicting attribute is here}}
[[intel::num_simd_work_items(2)]] void func2(); // expected-error{{'num_simd_work_items' attribute must evenly divide the work-group size for the 'reqd_work_group_size' attribute}}

[[intel::num_simd_work_items(2)]] // expected-error{{'num_simd_work_items' attribute must evenly divide the work-group size for the 'reqd_work_group_size' attribute}}
Expand Down Expand Up @@ -211,7 +209,7 @@ struct TRIFuncObjBad18 {
};

#endif // TRIGGER_ERROR
// If the declaration has a [[intel::reqd_work_group_size()]]
// If the declaration has a [[sycl::reqd_work_group_size()]]
// or [[cl::reqd_work_group_size()]] or
// __attribute__((reqd_work_group_size)) attribute, check to see
// if the last argument can be evenly divided by the
Expand Down
Original file line number Diff line number Diff line change
@@ -1,6 +1,6 @@
// RUN: %clang_cc1 -fsycl-is-device -fsyntax-only -ast-dump -verify -pedantic %s | FileCheck %s

// Test that checks template parameter support for 'intel::reqd_work_group_size' attribute on sycl device.
// Test that checks template parameter support for 'sycl::reqd_work_group_size' attribute on sycl device.

// Test that checks wrong function template instantiation and ensures that the type
// is checked properly when instantiating from the template definition.
Expand Down
2 changes: 1 addition & 1 deletion sycl/test/kernel_param/attr.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -12,7 +12,7 @@ int main() {
queue myQueue;
myQueue.submit([&](handler &cgh) {
cgh.parallel_for<class C>(Size, [=](item<1> ITEM)
[[intel::reqd_work_group_size(4)]]{});
[[sycl::reqd_work_group_size(4)]]{});
});
}

Expand Down