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

This patch removes deprecated [[intel::reqd_work_group_size()]] attribute in favor of SYCL 2020 attribute [[sycl::reqd_work_group_size()]].

Signed-off-by: Soumi Manna <[email protected]>

Author: smanna12

clang/include/clang/Basic/Attr.td

@@ -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">,

clang/include/clang/Basic/AttrDocs.td

@@ -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++
 
@@ -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.

clang/lib/Sema/SemaDeclAttr.cpp

@@ -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

clang/test/SemaSYCL/intel-max-global-work-dim-device.cpp

@@ -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 {}
 };

clang/test/SemaSYCL/intel-reqd-work-group-size-device-direct-prop.cpp

@@ -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 {

clang/test/SemaSYCL/intel-reqd-work-group-size-device.cpp

@@ -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'}}

clang/test/SemaSYCL/num_simd_work_items_device.cpp

@@ -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 {
@@ -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
@@ -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}}
@@ -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

clang/test/SemaSYCL/sycl-device-intel-reqd-work-group-size-template.cpp

@@ -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.

sycl/test/kernel_param/attr.cpp

@@ -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)]]{});
     });
   }