Skip to content

Commit 331e2e4

Browse files
smanna12smanna12
authored
[SYCL][FPGA][NFC] Refactor [[intel::max_global_work_dim()]] attribute tests (#6371)
This patch updates tests for [[intel::max_global_work_dim()]] attribute: 1. separates AST and diagnostics to make it easier to read/follow-up. 2. removes duplicate test cases. 3. adds missing tests. 4. no compiler changes. Signed-off-by: Soumi Manna [email protected]
1 parent f51357f commit 331e2e4

File tree

3 files changed

+472
-470
lines changed

3 files changed

+472
-470
lines changed

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

Lines changed: 329 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,329 @@
1+
// RUN: %clang_cc1 %s -fsyntax-only -ast-dump -fsycl-is-device -internal-isystem %S/Inputs -sycl-std=2017 -Wno-sycl-2017-compat -triple spir64 | FileCheck %s
2+
3+
// The test checks AST of [[intel::max_global_work_dim()]] attribute.
4+
5+
#include "sycl.hpp"
6+
7+
using namespace cl::sycl;
8+
queue q;
9+
10+
// No diagnostic is emitted because the arguments match. Duplicate attribute is silently ignored.
11+
// CHECK: FunctionDecl {{.*}} {{.*}} funcc 'void ()'
12+
// CHECK: SYCLIntelMaxGlobalWorkDimAttr
13+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
14+
// CHECK-NEXT: value: Int 2
15+
// CHECK-NEXT: IntegerLiteral{{.*}}2{{$}}
16+
// CHECK-NOT: SYCLIntelMaxGlobalWorkDimAttr
17+
[[intel::max_global_work_dim(2)]] [[intel::max_global_work_dim(2)]] void funcc() {}
18+
19+
// Test that checks template parameter support on member function of class template.
20+
// CHECK: ClassTemplateDecl {{.*}} {{.*}} KernelFunctor
21+
// CHECK: ClassTemplateSpecializationDecl {{.*}} {{.*}} class KernelFunctor definition
22+
// CHECK: TemplateArgument integral 2
23+
// CHECK: CXXRecordDecl {{.*}} {{.*}} implicit class KernelFunctor
24+
// CHECK: SYCLIntelMaxGlobalWorkDimAttr
25+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
26+
// CHECK-NEXT: value: Int 2
27+
// CHECK-NEXT: SubstNonTypeTemplateParmExpr {{.*}}
28+
// CHECK-NEXT: NonTypeTemplateParmDecl {{.*}}
29+
// CHECK-NEXT: IntegerLiteral{{.*}}2{{$}}
30+
template <int SIZE>
31+
class KernelFunctor {
32+
public:
33+
[[intel::max_global_work_dim(SIZE)]] void operator()() {}
34+
};
35+
36+
int kernel() {
37+
// no error expected
38+
KernelFunctor<2>();
39+
}
40+
41+
// Test that checks template parameter support on function.
42+
// CHECK: FunctionDecl {{.*}} {{.*}} func1 'void ()'
43+
// CHECK: TemplateArgument integral 3
44+
// CHECK: SYCLIntelMaxGlobalWorkDimAttr
45+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
46+
// CHECK-NEXT: value: Int 3
47+
// CHECK-NEXT: SubstNonTypeTemplateParmExpr {{.*}}
48+
// CHECK-NEXT: NonTypeTemplateParmDecl {{.*}}
49+
// CHECK-NEXT: IntegerLiteral{{.*}}3{{$}}
50+
template <int N>
51+
[[intel::max_global_work_dim(N)]] void func1() {}
52+
53+
int ver() {
54+
func1<3>(); // OK
55+
return 0;
56+
}
57+
58+
[[intel::max_global_work_dim(2)]] void func_do_not_ignore() {}
59+
60+
struct FuncObj {
61+
[[intel::max_global_work_dim(1)]] void operator()() const {}
62+
};
63+
64+
// Checks correctness of mutual usage of different work_group_size attributes:
65+
// reqd_work_group_size, max_work_group_size, and max_global_work_dim.
66+
// In case the value of 'max_global_work_dim' attribute equals to 0 we shall
67+
// ensure that if max_work_group_size and reqd_work_group_size attributes exist,
68+
// they hold equal values (1, 1, 1).
69+
70+
struct TRIFuncObjGood1 {
71+
[[intel::max_global_work_dim(0)]] [[intel::max_work_group_size(1, 1, 1)]] [[sycl::reqd_work_group_size(1, 1, 1)]] void
72+
operator()() const {}
73+
};
74+
75+
struct TRIFuncObjGood2 {
76+
[[intel::max_global_work_dim(3)]] [[intel::max_work_group_size(8, 1, 1)]] [[sycl::reqd_work_group_size(4, 1, 1)]] void
77+
operator()() const {}
78+
};
79+
80+
struct TRIFuncObjGood3 {
81+
[[sycl::reqd_work_group_size(1)]] [[intel::max_global_work_dim(0)]] void
82+
operator()() const {}
83+
};
84+
85+
struct TRIFuncObjGood4 {
86+
[[sycl::reqd_work_group_size(1, 1, 1)]] [[intel::max_global_work_dim(0)]] void
87+
operator()() const {}
88+
};
89+
90+
struct TRIFuncObjGood5 {
91+
[[intel::max_work_group_size(1, 1, 1)]] [[intel::max_global_work_dim(0)]] void
92+
operator()() const {}
93+
};
94+
95+
struct TRIFuncObjGood6 {
96+
[[sycl::reqd_work_group_size(4, 1, 1)]] [[intel::max_global_work_dim(3)]] void
97+
operator()() const {}
98+
};
99+
100+
struct TRIFuncObjGood7 {
101+
[[sycl::reqd_work_group_size(4, 4, 4)]] void // OK
102+
operator()() const;
103+
};
104+
105+
[[intel::max_global_work_dim(1)]] void TRIFuncObjGood7::operator()() const {}
106+
107+
struct TRIFuncObjGood8 {
108+
[[intel::max_work_group_size(8, 1, 1)]] [[intel::max_global_work_dim(3)]] void
109+
operator()() const {}
110+
};
111+
112+
struct TRIFuncObjGood9 {
113+
[[intel::max_work_group_size(4, 4, 4)]] void // OK
114+
operator()() const;
115+
};
116+
117+
[[intel::max_global_work_dim(1)]] void TRIFuncObjGood9::operator()() const {}
118+
119+
int main() {
120+
q.submit([&](handler &h) {
121+
// CHECK-LABEL: FunctionDecl {{.*}}test_kernel1
122+
// CHECK: SYCLIntelMaxGlobalWorkDimAttr
123+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
124+
// CHECK-NEXT: value: Int 1
125+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
126+
h.single_task<class test_kernel1>(FuncObj());
127+
128+
// CHECK-LABEL: FunctionDecl {{.*}}test_kernel2
129+
// CHECK: SYCLIntelMaxGlobalWorkDimAttr
130+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
131+
// CHECK-NEXT: value: Int 2
132+
// CHECK-NEXT: IntegerLiteral{{.*}}2{{$}}
133+
h.single_task<class test_kernel2>(
134+
[]() [[intel::max_global_work_dim(2)]] {});
135+
136+
// CHECK-LABEL: FunctionDecl {{.*}}test_kernel3
137+
// CHECK: SYCLIntelMaxGlobalWorkDimAttr
138+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
139+
// CHECK-NEXT: value: Int 2
140+
// CHECK-NEXT: IntegerLiteral{{.*}}2{{$}}
141+
h.single_task<class test_kernel3>(
142+
[]() { func_do_not_ignore(); });
143+
144+
h.single_task<class test_kernel4>(TRIFuncObjGood1());
145+
// CHECK-LABEL: FunctionDecl {{.*}}test_kernel4
146+
// CHECK: SYCLIntelMaxGlobalWorkDimAttr
147+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
148+
// CHECK-NEXT: value: Int 0
149+
// CHECK-NEXT: IntegerLiteral{{.*}}0{{$}}
150+
// CHECK: SYCLIntelMaxWorkGroupSizeAttr
151+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
152+
// CHECK-NEXT: value: Int 1
153+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
154+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
155+
// CHECK-NEXT: value: Int 1
156+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
157+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
158+
// CHECK-NEXT: value: Int 1
159+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
160+
// CHECK: ReqdWorkGroupSizeAttr {{.*}}
161+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
162+
// CHECK-NEXT: value: Int 1
163+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
164+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
165+
// CHECK-NEXT: value: Int 1
166+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
167+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
168+
// CHECK-NEXT: value: Int 1
169+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
170+
171+
h.single_task<class test_kernel5>(TRIFuncObjGood2());
172+
// CHECK-LABEL: FunctionDecl {{.*}}test_kernel5
173+
// CHECK: SYCLIntelMaxGlobalWorkDimAttr
174+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
175+
// CHECK-NEXT: value: Int 3
176+
// CHECK-NEXT: IntegerLiteral{{.*}}3{{$}}
177+
// CHECK: SYCLIntelMaxWorkGroupSizeAttr
178+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
179+
// CHECK-NEXT: value: Int 8
180+
// CHECK-NEXT: IntegerLiteral{{.*}}8{{$}}
181+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
182+
// CHECK-NEXT: value: Int 1
183+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
184+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
185+
// CHECK-NEXT: value: Int 1
186+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
187+
// CHECK: ReqdWorkGroupSizeAttr
188+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
189+
// CHECK-NEXT: value: Int 4
190+
// CHECK-NEXT: IntegerLiteral{{.*}}4{{$}}
191+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
192+
// CHECK-NEXT: value: Int 1
193+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
194+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
195+
// CHECK-NEXT: value: Int 1
196+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
197+
198+
h.single_task<class test_kernel5>(TRIFuncObjGood3());
199+
// CHECK-LABEL: FunctionDecl {{.*}}test_kernel5
200+
// CHECK: ReqdWorkGroupSizeAttr
201+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
202+
// CHECK-NEXT: value: Int 1
203+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
204+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
205+
// CHECK-NEXT: value: Int 1
206+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
207+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
208+
// CHECK-NEXT: value: Int 1
209+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
210+
// CHECK: SYCLIntelMaxGlobalWorkDimAttr
211+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
212+
// CHECK-NEXT: value: Int 0
213+
// CHECK-NEXT: IntegerLiteral{{.*}}0{{$}}
214+
215+
h.single_task<class test_kernel6>(TRIFuncObjGood4());
216+
// CHECK-LABEL: FunctionDecl {{.*}}test_kernel6
217+
// CHECK: ReqdWorkGroupSizeAttr
218+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
219+
// CHECK-NEXT: value: Int 1
220+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
221+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
222+
// CHECK-NEXT: value: Int 1
223+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
224+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
225+
// CHECK-NEXT: value: Int 1
226+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
227+
// CHECK: SYCLIntelMaxGlobalWorkDimAttr
228+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
229+
// CHECK-NEXT: value: Int 0
230+
// CHECK-NEXT: IntegerLiteral{{.*}}0{{$}}
231+
232+
h.single_task<class test_kernel7>(TRIFuncObjGood5());
233+
// CHECK-LABEL: FunctionDecl {{.*}}test_kernel7
234+
// CHECK: SYCLIntelMaxWorkGroupSizeAttr
235+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
236+
// CHECK-NEXT: value: Int 1
237+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
238+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
239+
// CHECK-NEXT: value: Int 1
240+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
241+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
242+
// CHECK-NEXT: value: Int 1
243+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
244+
// CHECK: SYCLIntelMaxGlobalWorkDimAttr
245+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
246+
// CHECK-NEXT: value: Int 0
247+
// CHECK-NEXT: IntegerLiteral{{.*}}0{{$}}
248+
249+
h.single_task<class test_kernel8>(TRIFuncObjGood6());
250+
// CHECK-LABEL: FunctionDecl {{.*}}test_kernel8
251+
// CHECK: ReqdWorkGroupSizeAttr
252+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
253+
// CHECK-NEXT: value: Int 4
254+
// CHECK-NEXT: IntegerLiteral{{.*}}4{{$}}
255+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
256+
// CHECK-NEXT: value: Int 1
257+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
258+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
259+
// CHECK-NEXT: value: Int 1
260+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
261+
// CHECK: SYCLIntelMaxGlobalWorkDimAttr
262+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
263+
// CHECK-NEXT: value: Int 3
264+
// CHECK-NEXT: IntegerLiteral{{.*}}3{{$}}
265+
266+
h.single_task<class test_kernel9>(TRIFuncObjGood7());
267+
// CHECK-LABEL: FunctionDecl {{.*}}test_kernel9
268+
// CHECK: ReqdWorkGroupSizeAttr
269+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
270+
// CHECK-NEXT: value: Int 4
271+
// CHECK-NEXT: IntegerLiteral{{.*}}4{{$}}
272+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
273+
// CHECK-NEXT: value: Int 4
274+
// CHECK-NEXT: IntegerLiteral{{.*}}4{{$}}
275+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
276+
// CHECK-NEXT: value: Int 4
277+
// CHECK-NEXT: IntegerLiteral{{.*}}4{{$}}
278+
// CHECK: SYCLIntelMaxGlobalWorkDimAttr
279+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
280+
// CHECK-NEXT: value: Int 1
281+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
282+
283+
h.single_task<class test_kernel10>(TRIFuncObjGood8());
284+
// CHECK-LABEL: FunctionDecl {{.*}}test_kernel10
285+
// CHECK: SYCLIntelMaxWorkGroupSizeAttr
286+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
287+
// CHECK-NEXT: value: Int 8
288+
// CHECK-NEXT: IntegerLiteral{{.*}}8{{$}}
289+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
290+
// CHECK-NEXT: value: Int 1
291+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
292+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
293+
// CHECK-NEXT: value: Int 1
294+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
295+
// CHECK: SYCLIntelMaxGlobalWorkDimAttr
296+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
297+
// CHECK-NEXT: value: Int 3
298+
// CHECK-NEXT: IntegerLiteral{{.*}}3{{$}}
299+
300+
h.single_task<class test_kernel11>(TRIFuncObjGood9());
301+
// CHECK-LABEL: FunctionDecl {{.*}}test_kernel11
302+
// CHECK: SYCLIntelMaxWorkGroupSizeAttr
303+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
304+
// CHECK-NEXT: value: Int 4
305+
// CHECK-NEXT: IntegerLiteral{{.*}}4{{$}}
306+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
307+
// CHECK-NEXT: value: Int 4
308+
// CHECK-NEXT: IntegerLiteral{{.*}}4{{$}}
309+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
310+
// CHECK-NEXT: value: Int 4
311+
// CHECK-NEXT: IntegerLiteral{{.*}}4{{$}}
312+
// CHECK: SYCLIntelMaxGlobalWorkDimAttr
313+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
314+
// CHECK-NEXT: value: Int 1
315+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
316+
317+
// Ignore duplicate attribute with same argument value.
318+
h.single_task<class test_kernell2>(
319+
// CHECK-LABEL: FunctionDecl {{.*}}test_kernell2 'void ()'
320+
// CHECK: SYCLIntelMaxGlobalWorkDimAttr
321+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
322+
// CHECK-NEXT: value: Int 3
323+
// CHECK-NEXT: IntegerLiteral{{.*}}3{{$}}
324+
// CHECK-NOT: SYCLIntelMaxGlobalWorkDimAttr
325+
[]() [[intel::max_global_work_dim(3),
326+
intel::max_global_work_dim(3)]] {}); // Ok
327+
});
328+
return 0;
329+
}

0 commit comments

Comments
 (0)