Skip to content

Commit 0db7d55

Browse files
dklochkov-embdklochkov-emb
committed
[SYCL] do not modify tests of clang subgroup
1 parent 916fcba commit 0db7d55

File tree

2 files changed

+93
-51
lines changed

2 files changed

+93
-51
lines changed

‎clang/test/SemaSYCL/reqd-sub-group-size.cpp‎

Lines changed: 74 additions & 34 deletions
Original file line numberDiff line numberDiff line change
@@ -1,6 +1,6 @@
11
// RUN: %clang_cc1 -fsycl-is-device -internal-isystem %S/Inputs -fsyntax-only -verify -pedantic %s
22

3-
// The test checks functionality of [[sycl::reqd_sub_group_size()]] attribute on SYCL kernel and [[intel::reqd_sub_group_size()]] is deprecated.
3+
// The test checks functionality of [[intel::reqd_sub_group_size()]] attribute on SYCL kernel and [[intel::reqd_sub_group_size()]] is deprecated.
44

55

66
#include "sycl.hpp" //clang/test/SemaSYCL/Inputs/sycl.hpp
@@ -32,59 +32,87 @@ int main() {
3232
});
3333
return 0;
3434
}
35-
[[sycl::reqd_sub_group_size(16)]] SYCL_EXTERNAL void B();
36-
[[sycl::reqd_sub_group_size(16)]] void A() // expected-warning {{'reqd_sub_group_size' attribute can only be applied to a SYCL kernel function}}
35+
// expected-note@+2 {{did you mean to use 'intel::reqd_sub_group_size' instead?}}
36+
// expected-warning@+1{{attribute 'intel::reqd_sub_group_size' is deprecated}}
37+
[[intel::reqd_sub_group_size(16)]] SYCL_EXTERNAL void B();
38+
// expected-note@+2 {{did you mean to use 'intel::reqd_sub_group_size' instead?}}
39+
// expected-warning@+1{{attribute 'intel::reqd_sub_group_size' is deprecated}}
40+
[[intel::reqd_sub_group_size(16)]] void A() // expected-warning {{'reqd_sub_group_size' attribute can only be applied to a SYCL kernel function}}
3741
{
3842
}
3943

40-
// expected-note@+2 {{did you mean to use 'sycl::reqd_sub_group_size' instead?}}
44+
// expected-note@+2 {{did you mean to use 'intel::reqd_sub_group_size' instead?}}
4145
// expected-warning@+1{{attribute 'intel::reqd_sub_group_size' is deprecated}}
4246
[[intel::reqd_sub_group_size(16)]] SYCL_EXTERNAL void B();
4347

44-
[[sycl::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}}
48+
// expected-note@+2 {{did you mean to use 'intel::reqd_sub_group_size' instead?}}
49+
// expected-warning@+1{{attribute 'intel::reqd_sub_group_size' is deprecated}}
50+
[[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}}
4551
A();
4652
}
53+
// expected-note@+3 {{did you mean to use 'intel::reqd_sub_group_size' instead?}}
54+
// expected-warning@+2{{attribute 'intel::reqd_sub_group_size' is deprecated}}
4755
// expected-note@+1 {{conflicting attribute is here}}
48-
[[sycl::reqd_sub_group_size(2)]] void sg_size2() {} // expected-warning {{'reqd_sub_group_size' attribute can only be applied to a SYCL kernel function}}
56+
[[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}}
4957

50-
// expected-note@+3 {{conflicting attribute is here}}
51-
// expected-error@+2 {{conflicting attributes applied to a SYCL kernel}}
52-
// expected-warning@+1 {{'reqd_sub_group_size' attribute can only be applied to a SYCL kernel function}}
53-
[[sycl::reqd_sub_group_size(4)]] __attribute__((sycl_device)) void sg_size4() {
58+
// expected-note@+5 {{conflicting attribute is here}}
59+
// expected-error@+4 {{conflicting attributes applied to a SYCL kernel}}
60+
// expected-warning@+3 {{'reqd_sub_group_size' attribute can only be applied to a SYCL kernel function}}
61+
// expected-note@+2 {{did you mean to use 'intel::reqd_sub_group_size' instead?}}
62+
// expected-warning@+1{{attribute 'intel::reqd_sub_group_size' is deprecated}}
63+
[[intel::reqd_sub_group_size(4)]] __attribute__((sycl_device)) void sg_size4() {
5464
sg_size2();
5565
}
5666

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

5969
// Tests for incorrect argument values for Intel reqd_sub_group_size attribute.
60-
[[sycl::reqd_sub_group_size]] void one() {} // expected-error {{'reqd_sub_group_size' attribute takes one argument}}
61-
[[sycl::reqd_sub_group_size(5)]] int a; // expected-error{{'reqd_sub_group_size' attribute only applies to functions}}
62-
[[sycl::reqd_sub_group_size("foo")]] void func() {} // expected-error{{integral constant expression must have integral or unscoped enumeration type, not 'const char[4]'}}
63-
[[sycl::reqd_sub_group_size(-1)]] void func1() {} // expected-error{{'reqd_sub_group_size' attribute requires a positive integral compile time constant expression}}
64-
[[sycl::reqd_sub_group_size(0, 1)]] void arg() {} // expected-error{{'reqd_sub_group_size' attribute takes one argument}}
70+
[[intel::reqd_sub_group_size]] void one() {} // expected-error {{'reqd_sub_group_size' attribute takes one argument}}
71+
[[intel::reqd_sub_group_size(5)]] int a; // expected-error{{'reqd_sub_group_size' attribute only applies to functions}}
72+
// expected-note@+2 {{did you mean to use 'intel::reqd_sub_group_size' instead?}}
73+
// expected-warning@+1{{attribute 'intel::reqd_sub_group_size' is deprecated}}
74+
[[intel::reqd_sub_group_size("foo")]] void func() {} // expected-error{{integral constant expression must have integral or unscoped enumeration type, not 'const char[4]'}}
75+
// expected-note@+2 {{did you mean to use 'intel::reqd_sub_group_size' instead?}}
76+
// expected-warning@+1{{attribute 'intel::reqd_sub_group_size' is deprecated}}
77+
[[intel::reqd_sub_group_size(-1)]] void func1() {} // expected-error{{'reqd_sub_group_size' attribute requires a positive integral compile time constant expression}}
78+
[[intel::reqd_sub_group_size(0, 1)]] void arg() {} // expected-error{{'reqd_sub_group_size' attribute takes one argument}}
6579

6680
// Diagnostic is emitted because the arguments mismatch.
67-
[[sycl::reqd_sub_group_size(12)]] void quux(); // expected-note {{previous attribute is here}}
68-
[[sycl::reqd_sub_group_size(100)]] void quux(); // expected-warning {{attribute 'reqd_sub_group_size' is already applied with different arguments}} expected-note {{previous attribute is here}}
69-
[[sycl::reqd_sub_group_size(200)]] void quux(); // expected-warning {{attribute 'reqd_sub_group_size' is already applied with different arguments}}
81+
// expected-note@+2 {{did you mean to use 'intel::reqd_sub_group_size' instead?}}
82+
// expected-warning@+1{{attribute 'intel::reqd_sub_group_size' is deprecated}}
83+
[[intel::reqd_sub_group_size(12)]] void quux(); // expected-note {{previous attribute is here}}
84+
// expected-note@+2 {{did you mean to use 'intel::reqd_sub_group_size' instead?}}
85+
// expected-warning@+1{{attribute 'intel::reqd_sub_group_size' is deprecated}}
86+
[[intel::reqd_sub_group_size(100)]] void quux(); // expected-warning {{attribute 'reqd_sub_group_size' is already applied with different arguments}} expected-note {{previous attribute is here}}
87+
// expected-note@+2 {{did you mean to use 'intel::reqd_sub_group_size' instead?}}
88+
// expected-warning@+1{{attribute 'intel::reqd_sub_group_size' is deprecated}}
89+
[[intel::reqd_sub_group_size(200)]] void quux(); // expected-warning {{attribute 'reqd_sub_group_size' is already applied with different arguments}}
7090

7191
// Make sure there's at least one argument passed.
72-
[[sycl::reqd_sub_group_size]] void quibble(); // expected-error {{'reqd_sub_group_size' attribute takes one argument}}
92+
[[intel::reqd_sub_group_size]] void quibble(); // expected-error {{'reqd_sub_group_size' attribute takes one argument}}
7393

7494
// No diagnostic is emitted because the arguments match.
75-
[[sycl::reqd_sub_group_size(12)]] void same();
76-
[[sycl::reqd_sub_group_size(12)]] void same() {} // expected-warning {{'reqd_sub_group_size' attribute can only be applied to a SYCL kernel function}}
95+
// expected-note@+2 {{did you mean to use 'intel::reqd_sub_group_size' instead?}}
96+
// expected-warning@+1{{attribute 'intel::reqd_sub_group_size' is deprecated}}
97+
[[intel::reqd_sub_group_size(12)]] void same();
98+
// expected-note@+2 {{did you mean to use 'intel::reqd_sub_group_size' instead?}}
99+
// expected-warning@+1{{attribute 'intel::reqd_sub_group_size' is deprecated}}
100+
[[intel::reqd_sub_group_size(12)]] void same() {} // expected-warning {{'reqd_sub_group_size' attribute can only be applied to a SYCL kernel function}}
77101

78102
// No diagnostic because the attributes are synonyms with identical behavior.
79-
[[sycl::reqd_sub_group_size(12)]] void same(); // OK
103+
// expected-note@+2 {{did you mean to use 'intel::reqd_sub_group_size' instead?}}
104+
// expected-warning@+1{{attribute 'intel::reqd_sub_group_size' is deprecated}}
105+
[[intel::reqd_sub_group_size(12)]] void same(); // OK
80106

81107
// Test that checks wrong function template instantiation and ensures that the type
82108
// is checked properly when instantiating from the template definition.
83109
template <typename Ty>
84-
// expected-error@+3{{'reqd_sub_group_size' attribute requires a positive integral compile time constant expression}}
85-
// expected-error@+2 {{integral constant expression must have integral or unscoped enumeration type, not 'S'}}
86-
// expected-error@+1 {{integral constant expression must have integral or unscoped enumeration type, not 'float'}}
87-
[[sycl::reqd_sub_group_size(Ty{})]] void func() {}
110+
// expected-error@+5{{'reqd_sub_group_size' attribute requires a positive integral compile time constant expression}}
111+
// expected-error@+4 {{integral constant expression must have integral or unscoped enumeration type, not 'S'}}
112+
// expected-error@+3 {{integral constant expression must have integral or unscoped enumeration type, not 'float'}}
113+
// expected-note@+2 {{did you mean to use 'intel::reqd_sub_group_size' instead?}}
114+
// expected-warning@+1{{attribute 'intel::reqd_sub_group_size' is deprecated}}
115+
[[intel::reqd_sub_group_size(Ty{})]] void func() {}
88116

89117
struct S {};
90118
void test() {
@@ -99,20 +127,26 @@ void test() {
99127
// Test that checks expression is not a constant expression.
100128
// expected-note@+1{{declared here}}
101129
int foo1();
102-
// expected-error@+2{{expression is not an integral constant expression}}
103-
// expected-note@+1{{non-constexpr function 'foo1' cannot be used in a constant expression}}
104-
[[sycl::reqd_sub_group_size(foo1() + 12)]] void func1();
130+
// expected-error@+4{{expression is not an integral constant expression}}
131+
// expected-note@+3{{non-constexpr function 'foo1' cannot be used in a constant expression}}
132+
// expected-note@+2 {{did you mean to use 'intel::reqd_sub_group_size' instead?}}
133+
// expected-warning@+1{{attribute 'intel::reqd_sub_group_size' is deprecated}}
134+
[[intel::reqd_sub_group_size(foo1() + 12)]] void func1();
105135

106136
// Test that checks expression is a constant expression.
107137
constexpr int bar1() { return 0; }
108-
[[sycl::reqd_sub_group_size(bar1() + 12)]] void func2(); // OK
138+
// expected-note@+2 {{did you mean to use 'intel::reqd_sub_group_size' instead?}}
139+
// expected-warning@+1{{attribute 'intel::reqd_sub_group_size' is deprecated}}
140+
[[intel::reqd_sub_group_size(bar1() + 12)]] void func2(); // OK
109141

110142
// Test that checks template parameter support on member function of class template.
111143
template <int SIZE>
112144
class KernelFunctor {
113145
public:
146+
// expected-note@+3 {{did you mean to use 'intel::reqd_sub_group_size' instead?}}
147+
// expected-warning@+2{{attribute 'intel::reqd_sub_group_size' is deprecated}}
114148
// expected-error@+1{{'reqd_sub_group_size' attribute requires a positive integral compile time constant expression}}
115-
[[sycl::reqd_sub_group_size(SIZE)]] void operator()() {}
149+
[[intel::reqd_sub_group_size(SIZE)]] void operator()() {}
116150
};
117151

118152
int check() {
@@ -123,16 +157,22 @@ int check() {
123157

124158
// Test that checks template parameter support on function.
125159
template <int N>
160+
// expected-note@+4 {{did you mean to use 'intel::reqd_sub_group_size' instead?}}
161+
// expected-warning@+3{{attribute 'intel::reqd_sub_group_size' is deprecated}}
126162
// expected-error@+2{{'reqd_sub_group_size' attribute requires a positive integral compile time constant expression}}
127163
// expected-warning@+1 {{'reqd_sub_group_size' attribute can only be applied to a SYCL kernel function}}
128-
[[sycl::reqd_sub_group_size(N)]] void func3() {}
164+
[[intel::reqd_sub_group_size(N)]] void func3() {}
129165

130166
template <int N>
167+
// expected-note@+3 {{did you mean to use 'intel::reqd_sub_group_size' instead?}}
168+
// expected-warning@+2{{attribute 'intel::reqd_sub_group_size' is deprecated}}
131169
// expected-warning@+1 {{'reqd_sub_group_size' attribute can only be applied to a SYCL kernel function}}
132-
[[sycl::reqd_sub_group_size(4)]] void func4(); // expected-note {{previous attribute is here}}
170+
[[intel::reqd_sub_group_size(4)]] void func4(); // expected-note {{previous attribute is here}}
133171

134172
template <int N>
135-
[[sycl::reqd_sub_group_size(N)]] void func4() {} // expected-warning {{attribute 'reqd_sub_group_size' is already applied with different arguments}}
173+
// expected-note@+2{{did you mean to use 'intel::reqd_sub_group_size' instead?}}
174+
// expected-warning@+1{{attribute 'intel::reqd_sub_group_size' is deprecated}}
175+
[[intel::reqd_sub_group_size(N)]] void func4() {} // expected-warning {{attribute 'reqd_sub_group_size' is already applied with different arguments}}
136176

137177
int check1() {
138178
// no error expected

‎clang/test/SemaSYCL/sub-group-size.cpp‎

Lines changed: 19 additions & 17 deletions
Original file line numberDiff line numberDiff line change
@@ -3,29 +3,29 @@
33
// RUN: %clang_cc1 -internal-isystem %S/Inputs -fsycl-is-device -fsycl-default-sub-group-size=10 -sycl-std=2020 -internal-isystem %S/Inputs -fsyntax-only -verify=expected,primary %s
44

55
// Validate the semantic analysis checks for the interaction betwen the
6-
// named_sub_group_size and reqd_sub_group_size attributes. These are not able to be
6+
// named_sub_group_size and sub_group_size attributes. These are not able to be
77
// combined, and require that they only be applied to non-sycl-kernel/
88
// non-sycl-device functions if they match the kernel they are being called
99
// from.
1010

1111
#include "Inputs/sycl.hpp"
1212

13-
// expected-error@+2 {{'named_sub_group_size' and 'reqd_sub_group_size' attributes are not compatible}}
13+
// expected-error@+2 {{'named_sub_group_size' and 'sub_group_size' attributes are not compatible}}
1414
// expected-note@+1 {{conflicting attribute is here}}
15-
[[sycl::reqd_sub_group_size(1)]] [[intel::named_sub_group_size(automatic)]] void f1();
16-
// expected-error@+2 {{'reqd_sub_group_size' and 'named_sub_group_size' attributes are not compatible}}
15+
[[intel::sub_group_size(1)]] [[intel::named_sub_group_size(automatic)]] void f1();
16+
// expected-error@+2 {{'sub_group_size' and 'named_sub_group_size' attributes are not compatible}}
1717
// expected-note@+1 {{conflicting attribute is here}}
18-
[[intel::named_sub_group_size(primary)]] [[sycl::reqd_sub_group_size(1)]] void f2();
18+
[[intel::named_sub_group_size(primary)]] [[intel::sub_group_size(1)]] void f2();
1919

2020
// expected-note@+1 {{conflicting attribute is here}}
21-
[[sycl::reqd_sub_group_size(1)]] void f3();
22-
// expected-error@+1 {{'named_sub_group_size' and 'reqd_sub_group_size' attributes are not compatible}}
21+
[[intel::sub_group_size(1)]] void f3();
22+
// expected-error@+1 {{'named_sub_group_size' and 'sub_group_size' attributes are not compatible}}
2323
[[intel::named_sub_group_size(primary)]] void f3();
2424

2525
// expected-note@+1 {{conflicting attribute is here}}
2626
[[intel::named_sub_group_size(primary)]] void f4();
27-
// expected-error@+1 {{'reqd_sub_group_size' and 'named_sub_group_size' attributes are not compatible}}
28-
[[sycl::reqd_sub_group_size(1)]] void f4();
27+
// expected-error@+1 {{'sub_group_size' and 'named_sub_group_size' attributes are not compatible}}
28+
[[intel::sub_group_size(1)]] void f4();
2929

3030
// expected-note@+1 {{previous attribute is here}}
3131
[[intel::named_sub_group_size(automatic)]] void f5();
@@ -43,9 +43,9 @@
4343
// expected-error@+2 {{'named_sub_group_size' and 'sycl_explicit_simd' attributes are not compatible}}
4444
// expected-note@+1 {{conflicting attribute is here}}
4545
[[intel::sycl_explicit_simd]] [[intel::named_sub_group_size(automatic)]] void f8();
46-
// expected-error@+2 {{'reqd_sub_group_size' and 'sycl_explicit_simd' attributes are not compatible}}
46+
// expected-error@+2 {{'sub_group_size' and 'sycl_explicit_simd' attributes are not compatible}}
4747
// expected-note@+1 {{conflicting attribute is here}}
48-
[[intel::sycl_explicit_simd]] [[sycl::reqd_sub_group_size(1)]] void f9();
48+
[[intel::sycl_explicit_simd]] [[intel::sub_group_size(1)]] void f9();
4949

5050
// expected-note@+1 {{conflicting attribute is here}}
5151
[[intel::named_sub_group_size(primary)]] void f10();
@@ -115,19 +115,19 @@ void calls_kernel_3() {
115115
});
116116
}
117117

118-
[[sycl::reqd_sub_group_size(10)]] void AttrFunc2() {} // #AttrFunc2
119-
[[sycl::reqd_sub_group_size(10)]] SYCL_EXTERNAL void AttrExternalDefined2() {} // #AttrExternalDefined2
120-
[[sycl::reqd_sub_group_size(10)]] SYCL_EXTERNAL void AttrExternalNotDefined2(); // #AttrExternalNotDefined2
118+
[[intel::sub_group_size(10)]] void AttrFunc2() {} // #AttrFunc2
119+
[[intel::sub_group_size(10)]] SYCL_EXTERNAL void AttrExternalDefined2() {} // #AttrExternalDefined2
120+
[[intel::sub_group_size(10)]] SYCL_EXTERNAL void AttrExternalNotDefined2(); // #AttrExternalNotDefined2
121121

122122
void calls_kernel_4() {
123123
sycl::kernel_single_task<class Kernel4>([]() { // #Kernel4
124124
// integer-error@#AttrFunc2{{kernel-called function must have a sub group size that matches the size specified for the kernel}}
125125
// integer-note@#Kernel4{{kernel declared here}}
126-
// expected-warning@#AttrFunc2 {{'reqd_sub_group_size' attribute can only be applied to a SYCL kernel function}}
126+
// expected-warning@#AttrFunc2 {{'sub_group_size' attribute can only be applied to a SYCL kernel function}}
127127
AttrFunc2();
128128
// integer-error@#AttrExternalDefined2{{kernel-called function must have a sub group size that matches the size specified for the kernel}}
129129
// integer-note@#Kernel4{{kernel declared here}}
130-
// expected-warning@#AttrExternalDefined2 {{'reqd_sub_group_size' attribute can only be applied to a SYCL kernel function}}
130+
// expected-warning@#AttrExternalDefined2 {{'sub_group_size' attribute can only be applied to a SYCL kernel function}}
131131
AttrExternalDefined2();
132132
// integer-error@#AttrExternalNotDefined2{{kernel-called function must have a sub group size that matches the size specified for the kernel}}
133133
// integer-note@#Kernel4{{kernel declared here}}
@@ -153,7 +153,9 @@ void calls_kernel_5() {
153153

154154
// Don't diag with the old sub-group-size.
155155
void calls_kernel_6() {
156-
sycl::kernel_single_task<class Kernel6>([]() [[sycl::reqd_sub_group_size(10)]] { // #Kernel6
156+
// expected-note@+2 {{did you mean to use 'intel::reqd_sub_group_size' instead?}}
157+
// expected-warning@+1{{attribute 'intel::reqd_sub_group_size' is deprecated}}
158+
sycl::kernel_single_task<class Kernel6>([]() [[intel::reqd_sub_group_size(10)]] { // #Kernel6
157159
NoAttrExternalNotDefined();
158160
});
159161
}

0 commit comments

Comments
 (0)