Skip to content

[SYCL] Tests cleanup for warnings issued due to attributes applied to non-kernel functions #15440

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.

Sign up for GitHub

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

Closed
wants to merge 2 commits into from
Closed

[SYCL] Tests cleanup for warnings issued due to attributes applied to non-kernel functions #15440

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
2 commits
Select commit Hold shift + click to select a range
c4ab882
follow up clean up of tests
dyniols Sep 19, 2024
07f655e
Merge remote-tracking branch 'intel/sycl' into attributes_tests_clean_up
dyniols Sep 19, 2024
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
41 changes: 3 additions & 38 deletions clang/test/SemaSYCL/check-work-group-size-hint-device.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -17,8 +17,7 @@

// Produce a conflicting attribute warning when the args are different.
[[sycl::work_group_size_hint(4, 1, 1)]] void f3(); // expected-note {{previous attribute is here}}
Copy link
Contributor

@elizabethandrews elizabethandrews Sep 20, 2024

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

This can be removed. We already test this in Functor1

Copy link
Contributor

@elizabethandrews elizabethandrews Sep 20, 2024

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Ideally all these tests on device functions should be rewritten on kernel I think since the usage on device function is not correct. I understand that is a lot of work for something of lower priority though. If you can work on that we would appreciate it, but if you cannot, please add a FIXME to all these tests explaining these attributes are incorrectly applied to device functions and the tests should be eventually rewritten.

[[sycl::work_group_size_hint(1, 1, 32)]] void f3() {} // expected-warning {{attribute 'work_group_size_hint' is already applied with different arguments}} \
// expected-warning {{'work_group_size_hint' attribute can only be applied to a SYCL kernel function}}
[[sycl::work_group_size_hint(1, 1, 32)]] void f3(); // expected-warning {{attribute 'work_group_size_hint' is already applied with different arguments}}

// 1 and 2 dim versions
[[sycl::work_group_size_hint(2)]] void f4(); // ok
Expand All @@ -44,13 +43,6 @@ class Functor_1 {
template <int N>
[[sycl::work_group_size_hint(N, 1, 1)]] void f8(); // #f8

// Test that template redeclarations also get diagnosed properly.
template <int X, int Y, int Z>
[[sycl::work_group_size_hint(1, 1, 1)]] void f9(); // #f9prev

template <int X, int Y, int Z>
[[sycl::work_group_size_hint(X, Y, Z)]] void f9() {} // #f9

// Test that a template redeclaration where the difference is known up front is
// diagnosed immediately, even without instantiation.
template <int X, int Y, int Z>
Expand All @@ -70,15 +62,6 @@ void instantiate() {
// expected-error@#f8 {{'work_group_size_hint' attribute requires a positive integral compile time constant expression}}
f8<0>(); // expected-note {{in instantiation}}
#endif

// expected-warning@#f9prev {{'work_group_size_hint' attribute can only be applied to a SYCL kernel function}}
f9<1, 1, 1>(); // OK, args are the same on the redecl.

// expected-warning@#f9 {{attribute 'work_group_size_hint' is already applied with different arguments}}
// expected-note@#f9prev {{previous attribute is here}}
// expected-warning@#f9prev {{'work_group_size_hint' attribute can only be applied to a SYCL kernel function}}

f9<1, 2, 3>(); // expected-note {{in instantiation}}
}

// Show that the attribute works on member functions.
Expand All @@ -101,20 +84,7 @@ class Functor16x2x1 {

class Functor4x4x4 {
public:
[[sycl::work_group_size_hint(4, 4, 4)]] void operator()() const {}; // expected-warning {{'work_group_size_hint' attribute can only be applied to a SYCL kernel function}}
};

// Checking whether propagation of the attribute happens or not, according to the SYCL version.
#if defined(EXPECT_PROP) // if attribute is propagated, then we expect errors here
void f8x8x8(){};
#else // otherwise no error
[[sycl::work_group_size_hint(8, 8, 8)]] void f8x8x8(){}; // expected-warning {{'work_group_size_hint' attribute can only be applied to a SYCL kernel function}}
#endif
class FunctorNoProp {
public:
void operator()() const {
f8x8x8();
};
[[sycl::work_group_size_hint(4, 4, 4)]] void operator()() const {};
};

void invoke() {
Expand All @@ -137,12 +107,6 @@ void invoke() {
// CHECK-NEXT: value: Int 1
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}


FunctorNoProp fNoProp;
h.single_task<class kernel_3>(fNoProp);
// CHECK: FunctionDecl {{.*}} {{.*}}kernel_3
// CHECK-NOT: SYCLWorkGroupSizeHintAttr

h.single_task<class kernel_name4>([]() [[sycl::work_group_size_hint(4,4,4)]] {});
// CHECK: FunctionDecl {{.*}}kernel_name4
// CHECK: SYCLWorkGroupSizeHintAttr {{.*}}
Expand All @@ -156,6 +120,7 @@ void invoke() {
// CHECK-NEXT: value: Int 4
// CHECK-NEXT: IntegerLiteral{{.*}}4{{$}}

h.single_task<class kernel_name5>(f4x4x4);
});

// FIXME: Add tests with the C++23 lambda attribute syntax.
Expand Down
11 changes: 5 additions & 6 deletions clang/test/SemaSYCL/intel-max-work-group-size.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -70,15 +70,14 @@ void instantiate() {
// a declaration along with [[sycl::reqd_work_group_size(X1, Y1, Z1)]]
// attribute, check to see if values of reqd_work_group_size arguments are
// equal or less than values coming from max_work_group_size attribute.
[[sycl::reqd_work_group_size(64, 64, 64)]] // expected-note {{conflicting attribute is here}} // expected-warning {{'reqd_work_group_size' attribute can only be applied to a SYCL kernel function}}
[[sycl::reqd_work_group_size(64, 64, 64)]] // expected-note {{conflicting attribute is here}}
[[intel::max_work_group_size(64, 16, 64)]] // expected-error {{'max_work_group_size' attribute conflicts with 'reqd_work_group_size' attribute}}
void
f9() {}
f9();

[[intel::max_work_group_size(4, 4, 4)]] void f10();
[[sycl::reqd_work_group_size(2, 2, 2)]] void f10(); // OK

[[sycl::reqd_work_group_size(2, 2, 2)]] [[intel::max_work_group_size(4, 4, 4)]] void f11() {} // expected-warning {{'reqd_work_group_size' attribute can only be applied to a SYCL kernel function}}

[[sycl::reqd_work_group_size(64, 64, 64)]] void f12(); // expected-note {{conflicting attribute is here}}
[[intel::max_work_group_size(16, 16, 16)]] void f12(); // expected-error {{'max_work_group_size' attribute conflicts with 'reqd_work_group_size' attribute}}
Expand All @@ -91,14 +90,14 @@ f13() {}
[[sycl::reqd_work_group_size(64, 64, 64)]] void f14(); // expected-error{{'reqd_work_group_size' attribute conflicts with 'max_work_group_size' attribute}}

[[cl::reqd_work_group_size(1, 2, 3)]] // expected-warning {{attribute 'cl::reqd_work_group_size' is deprecated}} \
// expected-note {{did you mean to use 'sycl::reqd_work_group_size' instead?}} // expected-warning {{'reqd_work_group_size' attribute can only be applied to a SYCL kernel function}}
// expected-note {{did you mean to use 'sycl::reqd_work_group_size' instead?}}
[[intel::max_work_group_size(1, 2, 3)]] void
f15() {} // OK
f15(); // OK

[[intel::max_work_group_size(2, 3, 7)]] void f16(); // expected-note {{conflicting attribute is here}}
[[sycl::reqd_work_group_size(7, 3, 2)]] void f16(); // expected-error{{'reqd_work_group_size' attribute conflicts with 'max_work_group_size' attribute}}

[[intel::max_work_group_size(1, 2, 3)]] [[sycl::reqd_work_group_size(1, 2, 3)]] void f17(){}; // expected-warning {{'reqd_work_group_size' attribute can only be applied to a SYCL kernel function}}
[[intel::max_work_group_size(1, 2, 3)]] [[sycl::reqd_work_group_size(1, 2, 3)]] void f17();

[[sycl::reqd_work_group_size(16)]] // expected-note {{conflicting attribute is here}}
[[intel::max_work_group_size(16, 1, 1)]] void // expected-error {{'max_work_group_size' attribute conflicts with 'reqd_work_group_size' attribute}}
Expand Down
25 changes: 4 additions & 21 deletions clang/test/SemaSYCL/reqd-sub-group-size.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -33,22 +33,7 @@ int main() {
return 0;
}
[[intel::reqd_sub_group_size(16)]] SYCL_EXTERNAL void B();
[[intel::reqd_sub_group_size(16)]] void A() // expected-warning {{'reqd_sub_group_size' attribute can only be applied to a SYCL kernel function}}
{
}

[[intel::reqd_sub_group_size(16)]] SYCL_EXTERNAL void B() { // expected-warning {{'reqd_sub_group_size' attribute can only be applied to a SYCL kernel function}}
A();
}
// expected-note@+1 {{conflicting attribute is here}}
[[intel::reqd_sub_group_size(2)]] void sg_size2() {} // expected-warning {{'reqd_sub_group_size' attribute can only be applied to a SYCL kernel function}}

// expected-note@+3 {{conflicting attribute is here}}
// expected-error@+2 {{conflicting attributes applied to a SYCL kernel}}
// expected-warning@+1 {{'reqd_sub_group_size' attribute can only be applied to a SYCL kernel function}}
[[intel::reqd_sub_group_size(4)]] __attribute__((sycl_device)) void sg_size4() {
sg_size2();
}

// Test that checks support and functionality of reqd_sub_group_size attribute support on function.

Expand All @@ -69,7 +54,7 @@ int main() {

// No diagnostic is emitted because the arguments match.
[[intel::reqd_sub_group_size(12)]] void same();
[[intel::reqd_sub_group_size(12)]] void same() {} // expected-warning {{'reqd_sub_group_size' attribute can only be applied to a SYCL kernel function}}
[[intel::reqd_sub_group_size(12)]] void same();

// No diagnostic because the attributes are synonyms with identical behavior.
[[sycl::reqd_sub_group_size(12)]] void same(); // OK
Expand Down Expand Up @@ -119,16 +104,14 @@ int check() {

// Test that checks template parameter support on function.
template <int N>
// expected-error@+2{{'reqd_sub_group_size' attribute requires a positive integral compile time constant expression}}
// expected-warning@+1 {{'reqd_sub_group_size' attribute can only be applied to a SYCL kernel function}}
[[intel::reqd_sub_group_size(N)]] void func3() {}
// expected-error@+1{{'reqd_sub_group_size' attribute requires a positive integral compile time constant expression}}
[[intel::reqd_sub_group_size(N)]] void func3();

template <int N>
// expected-warning@+1 {{'reqd_sub_group_size' attribute can only be applied to a SYCL kernel function}}
[[intel::reqd_sub_group_size(4)]] void func4(); // expected-note {{previous attribute is here}}

template <int N>
[[intel::reqd_sub_group_size(N)]] void func4() {} // expected-warning {{attribute 'reqd_sub_group_size' is already applied with different arguments}}
[[intel::reqd_sub_group_size(N)]] void func4(); // expected-warning {{attribute 'reqd_sub_group_size' is already applied with different arguments}}

int check1() {
// no error expected
Expand Down
64 changes: 18 additions & 46 deletions clang/test/SemaSYCL/reqd_work_group_size.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -22,44 +22,31 @@ class Functor30 {

// Tests for 'reqd_work_group_size' attribute duplication.
// No diagnostic is emitted because the arguments match. Duplicate attribute is silently ignored.
// expected-warning@+1 {{'reqd_work_group_size' attribute can only be applied to a SYCL kernel function}}
[[sycl::reqd_work_group_size(6, 6, 6)]] [[sycl::reqd_work_group_size(6, 6, 6)]] void f2() {}
[[sycl::reqd_work_group_size(6, 6, 6)]] [[sycl::reqd_work_group_size(6, 6, 6)]] void f2();

// No diagnostic is emitted because the arguments match.
[[sycl::reqd_work_group_size(32, 32, 32)]] void f3();
[[sycl::reqd_work_group_size(32, 32, 32)]] void f3(); // OK

// Produce a conflicting attribute warning when the args are different.
[[sycl::reqd_work_group_size(6, 6, 6)]] // expected-note {{previous attribute is here}} // expected-warning {{'reqd_work_group_size' attribute can only be applied to a SYCL kernel function}}
[[sycl::reqd_work_group_size(6, 6, 6)]] // expected-note {{previous attribute is here}}
[[sycl::reqd_work_group_size(16, 16, 16)]] void // expected-error {{attribute 'reqd_work_group_size' is already applied with different arguments}}
f4() {}
f4();

// Catch the easy case where the attributes are all specified at once with
// different arguments.
struct TRIFuncObjGood1 {
// expected-note@+3 {{previous attribute is here}}
// expected-error@+2 {{attribute 'reqd_work_group_size' is already applied with different arguments}}
// expected-warning@+1 {{'reqd_work_group_size' attribute can only be applied to a SYCL kernel function}}
[[sycl::reqd_work_group_size(64)]] [[sycl::reqd_work_group_size(128)]] void operator()() const {}
// expected-note@+2 {{previous attribute is here}}
// expected-error@+1 {{attribute 'reqd_work_group_size' is already applied with different arguments}}
[[sycl::reqd_work_group_size(64)]] [[sycl::reqd_work_group_size(128)]] void operator()() const;
};

struct TRIFuncObjGood2 {
// expected-note@+3 {{previous attribute is here}}
// expected-error@+2 {{attribute 'reqd_work_group_size' is already applied with different arguments}}
// expected-warning@+1 {{'reqd_work_group_size' attribute can only be applied to a SYCL kernel function}}
[[sycl::reqd_work_group_size(64, 64)]] [[sycl::reqd_work_group_size(128, 128)]] void operator()() const {}
// expected-note@+2 {{previous attribute is here}}
// expected-error@+1 {{attribute 'reqd_work_group_size' is already applied with different arguments}}
[[sycl::reqd_work_group_size(64, 64)]] [[sycl::reqd_work_group_size(128, 128)]] void operator()() const;
};

struct TRIFuncObjGood3 {
[[sycl::reqd_work_group_size(8, 8)]] void // expected-note {{previous attribute is here}}
operator()() const;
};

[[sycl::reqd_work_group_size(4, 4)]] // expected-error {{attribute 'reqd_work_group_size' is already applied with different arguments}} \
// expected-warning {{'reqd_work_group_size' attribute can only be applied to a SYCL kernel function}}
void
TRIFuncObjGood3::operator()() const {}

// Show that the attribute works on member functions.
class Functor {
public:
Expand All @@ -77,9 +64,9 @@ class FunctorC {

class Functor32 {
public:
[[sycl::reqd_work_group_size(32, 1, 1)]] // expected-note {{previous attribute is here}} // expected-warning {{'reqd_work_group_size' attribute can only be applied to a SYCL kernel function}}
[[sycl::reqd_work_group_size(32, 1, 1)]] // expected-note {{previous attribute is here}}
[[sycl::reqd_work_group_size(1, 1, 32)]] void // expected-error{{attribute 'reqd_work_group_size' is already applied with different arguments}}
operator()() const {}
operator()() const;
};

// Ensure that template arguments behave appropriately based on instantiations.
Expand All @@ -91,7 +78,7 @@ template <int X, int Y, int Z>
[[sycl::reqd_work_group_size(1, 1, 1)]] void f7(); // #f7prev

template <int X, int Y, int Z>
[[sycl::reqd_work_group_size(X, Y, Z)]] void f7() {} // #f7
[[sycl::reqd_work_group_size(X, Y, Z)]] void f7(); // #f7

// Test that a template redeclaration where the difference is known up front is
// diagnosed immediately, even without instantiation.
Expand All @@ -109,45 +96,31 @@ void instantiate() {
f7<1, 1, 1>(); // OK, args are the same on the redecl.
// expected-error@#f7 {{attribute 'reqd_work_group_size' is already applied with different arguments}}
// expected-note@#f7prev {{previous attribute is here}}
// expected-warning@#f7prev {{'reqd_work_group_size' attribute can only be applied to a SYCL kernel function}}
// expected-warning@#f7prev {{'reqd_work_group_size' attribute can only be applied to a SYCL kernel function}}
f7<2, 2, 2>(); // expected-note {{in instantiation}}
}

// Tests for 'reqd_work_group_size' attribute duplication.

[[sycl::reqd_work_group_size(8)]] // expected-note {{previous attribute is here}} // expected-warning {{'reqd_work_group_size' attribute can only be applied to a SYCL kernel function}}
[[sycl::reqd_work_group_size(8)]] // expected-note {{previous attribute is here}}
[[sycl::reqd_work_group_size(1, 1, 8)]] void // expected-error {{attribute 'reqd_work_group_size' is already applied with different arguments}}
f8(){};
f8();

[[sycl::reqd_work_group_size(32, 32, 1)]] // expected-note {{previous attribute is here}} // expected-warning {{'reqd_work_group_size' attribute can only be applied to a SYCL kernel function}}
[[sycl::reqd_work_group_size(32, 32)]] void f9() {} // expected-error {{attribute 'reqd_work_group_size' is already applied with different arguments}}
[[sycl::reqd_work_group_size(32, 32, 1)]] // expected-note {{previous attribute is here}}
[[sycl::reqd_work_group_size(32, 32)]] void f9(); // expected-error {{attribute 'reqd_work_group_size' is already applied with different arguments}}

// Test that template redeclarations also get diagnosed properly.
template <int X, int Y, int Z>
[[sycl::reqd_work_group_size(64, 1, 1)]] void f10(); // #f10prev
template <int X, int Y, int Z>
[[sycl::reqd_work_group_size(X, Y, Z)]] void f10() {} // #f10err
[[sycl::reqd_work_group_size(X, Y, Z)]] void f10(); // #f10err

void test() {
f10<64, 1, 1>(); // OK, args are the same on the redecl.
// expected-error@#f10err {{attribute 'reqd_work_group_size' is already applied with different arguments}}
// expected-note@#f10prev {{previous attribute is here}}
// expected-warning@#f10prev {{'reqd_work_group_size' attribute can only be applied to a SYCL kernel function}}
// expected-warning@#f10prev {{'reqd_work_group_size' attribute can only be applied to a SYCL kernel function}}
f10<1, 1, 64>(); // expected-note {{in instantiation}}
}

struct TRIFuncObjBad {
[[sycl::reqd_work_group_size(32, 1, 1)]] void // expected-note {{previous attribute is here}}
operator()() const;
};

[[sycl::reqd_work_group_size(1, 1, 32)]] // expected-error {{attribute 'reqd_work_group_size' is already applied with different arguments}} \
// expected-warning {{'reqd_work_group_size' attribute can only be applied to a SYCL kernel function}}
void
TRIFuncObjBad::operator()() const {}

// Test that checks wrong function template instantiation and ensures that the type
// is checked properly when instantiating from the template definition.

Expand Down Expand Up @@ -183,9 +156,8 @@ int main() {
KernelFunctor<16, 1, 1>();
}
// Test that checks template parameter support on function.
// expected-warning@+2 {{'reqd_work_group_size' attribute can only be applied to a SYCL kernel function}}
template <int N, int N1, int N2>
[[sycl::reqd_work_group_size(N, N1, N2)]] void func3() {}
[[sycl::reqd_work_group_size(N, N1, N2)]] void func3();

int check() {
func3<8, 8, 8>();
Expand Down
10 changes: 0 additions & 10 deletions clang/test/SemaSYCL/sub-group-size.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -115,20 +115,10 @@ void calls_kernel_3() {
});
}

[[intel::sub_group_size(10)]] void AttrFunc2() {} // #AttrFunc2
[[intel::sub_group_size(10)]] SYCL_EXTERNAL void AttrExternalDefined2() {} // #AttrExternalDefined2
[[intel::sub_group_size(10)]] SYCL_EXTERNAL void AttrExternalNotDefined2(); // #AttrExternalNotDefined2

void calls_kernel_4() {
sycl::kernel_single_task<class Kernel4>([]() { // #Kernel4
// integer-error@#AttrFunc2{{kernel-called function must have a sub group size that matches the size specified for the kernel}}
// integer-note@#Kernel4{{kernel declared here}}
// expected-warning@#AttrFunc2 {{'sub_group_size' attribute can only be applied to a SYCL kernel function}}
AttrFunc2();
// integer-error@#AttrExternalDefined2{{kernel-called function must have a sub group size that matches the size specified for the kernel}}
// integer-note@#Kernel4{{kernel declared here}}
// expected-warning@#AttrExternalDefined2 {{'sub_group_size' attribute can only be applied to a SYCL kernel function}}
AttrExternalDefined2();
// integer-error@#AttrExternalNotDefined2{{kernel-called function must have a sub group size that matches the size specified for the kernel}}
// integer-note@#Kernel4{{kernel declared here}}
AttrExternalNotDefined2();
Expand Down
6 changes: 2 additions & 4 deletions clang/test/SemaSYCL/sycl-device-intel-reqd-work-group-size-template.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -58,9 +58,8 @@ int main() {
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}

// Test that checks template parameter support on function.
// expected-warning@+2 {{'reqd_work_group_size' attribute can only be applied to a SYCL kernel function}}
template <int N, int N1, int N2>
[[sycl::reqd_work_group_size(N, N1, N2)]] void func3() {}
[[sycl::reqd_work_group_size(N, N1, N2)]] void func3();

int check() {
func3<8, 8, 8>();
Expand All @@ -87,8 +86,7 @@ int check() {
// CHECK-NEXT: IntegerLiteral{{.*}}8{{$}}

// No diagnostic is emitted because the arguments match. Duplicate attribute is silently ignored.
// expected-warning@+1 {{'reqd_work_group_size' attribute can only be applied to a SYCL kernel function}}
[[sycl::reqd_work_group_size(4, 4, 4)]] [[sycl::reqd_work_group_size(4, 4, 4)]] void func4() {}
[[sycl::reqd_work_group_size(4, 4, 4)]] [[sycl::reqd_work_group_size(4, 4, 4)]] void func4();
// CHECK: FunctionDecl {{.*}} {{.*}} func4 'void ()'
// CHECK: SYCLReqdWorkGroupSizeAttr
// CHECK-NEXT: ConstantExpr{{.*}}'int'
Expand Down