Skip to content

Commit f196d0b

Browse files
smanna12smanna12
authored
[SYCL][FPGA][NFC] Refactor of ReqdWorkGroupSize attribute tests (#6431)
* [SYCL][FPGA][NFC] Refactor of ReqdWorkGroupSize attribute tests Signed-off-by: Soumi Manna <[email protected]>
1 parent 25c34bf commit f196d0b

7 files changed

+322
-424
lines changed

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

Lines changed: 201 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,201 @@
1+
// RUN: %clang_cc1 -fsycl-is-device -internal-isystem %S/Inputs -sycl-std=2017 -Wno-sycl-2017-compat -ast-dump %s | FileCheck %s
2+
3+
// Test for AST of reqd_work_group_size kernel attribute in SYCL 1.2.1.
4+
5+
#include "sycl.hpp"
6+
7+
using namespace cl::sycl;
8+
queue q;
9+
10+
// Test that checks template parameter support on member function of class template.
11+
template <int SIZE, int SIZE1, int SIZE2>
12+
class KernelFunctor {
13+
public:
14+
[[sycl::reqd_work_group_size(SIZE, SIZE1, SIZE2)]] void operator()() {}
15+
};
16+
17+
void test() {
18+
KernelFunctor<16, 1, 1>();
19+
}
20+
21+
// CHECK: ClassTemplateDecl {{.*}} {{.*}} KernelFunctor
22+
// CHECK: ClassTemplateSpecializationDecl {{.*}} {{.*}} class KernelFunctor definition
23+
// CHECK: CXXRecordDecl {{.*}} {{.*}} implicit class KernelFunctor
24+
// CHECK: ReqdWorkGroupSizeAttr
25+
// CHECK-NEXT: ConstantExpr{{.*}}'int'
26+
// CHECK-NEXT: value: Int 16
27+
// CHECK-NEXT: SubstNonTypeTemplateParmExpr {{.*}}
28+
// CHECK-NEXT: NonTypeTemplateParmDecl {{.*}}
29+
// CHECK-NEXT: IntegerLiteral{{.*}}16{{$}}
30+
// CHECK-NEXT: ConstantExpr{{.*}}'int'
31+
// CHECK-NEXT: value: Int 1
32+
// CHECK-NEXT: SubstNonTypeTemplateParmExpr {{.*}}
33+
// CHECK-NEXT: NonTypeTemplateParmDecl {{.*}}
34+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
35+
// CHECK-NEXT: ConstantExpr{{.*}}'int'
36+
// CHECK-NEXT: value: Int 1
37+
// CHECK-NEXT: SubstNonTypeTemplateParmExpr {{.*}}
38+
// CHECK-NEXT: NonTypeTemplateParmDecl {{.*}}
39+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
40+
41+
// Test that checks template parameter support on function.
42+
template <int N, int N1, int N2>
43+
[[sycl::reqd_work_group_size(N, N1, N2)]] void func3() {}
44+
45+
int check() {
46+
func3<8, 8, 8>();
47+
return 0;
48+
}
49+
// CHECK: FunctionTemplateDecl {{.*}} {{.*}} func3
50+
// CHECK: FunctionDecl {{.*}} {{.*}} used func3 'void ()'
51+
// CHECK: ReqdWorkGroupSizeAttr
52+
// CHECK-NEXT: ConstantExpr{{.*}}'int'
53+
// CHECK-NEXT: value: Int 8
54+
// CHECK: SubstNonTypeTemplateParmExpr {{.*}}
55+
// CHECK-NEXT: NonTypeTemplateParmDecl {{.*}}
56+
// CHECK-NEXT: IntegerLiteral{{.*}}8{{$}}
57+
// CHECK-NEXT: ConstantExpr{{.*}}'int'
58+
// CHECK-NEXT: value: Int 8
59+
// CHECK: SubstNonTypeTemplateParmExpr {{.*}}
60+
// CHECK-NEXT: NonTypeTemplateParmDecl {{.*}}
61+
// CHECK-NEXT: IntegerLiteral{{.*}}8{{$}}
62+
// CHECK-NEXT: ConstantExpr{{.*}}'int'
63+
// CHECK-NEXT: value: Int 8
64+
// CHECK: SubstNonTypeTemplateParmExpr {{.*}}
65+
// CHECK-NEXT: NonTypeTemplateParmDecl {{.*}}
66+
// CHECK-NEXT: IntegerLiteral{{.*}}8{{$}}
67+
68+
[[sycl::reqd_work_group_size(4)]] void f4x1x1() {}
69+
70+
class Functor16 {
71+
public:
72+
[[sycl::reqd_work_group_size(16)]] void operator()() const {}
73+
};
74+
75+
class Functor16x16x16 {
76+
public:
77+
[[sycl::reqd_work_group_size(16, 16, 16)]] void operator()() const {}
78+
};
79+
80+
class Functor {
81+
public:
82+
void operator()() const {
83+
f4x1x1();
84+
}
85+
};
86+
87+
class FunctorAttr {
88+
public:
89+
[[sycl::reqd_work_group_size(128, 128, 128)]] void operator()() const {}
90+
};
91+
92+
// Test of redeclaration of [[intel::max_work_group_size()]] and [[sycl::reqd_work_group_size()]].
93+
[[intel::no_global_work_offset]] void func1();
94+
[[intel::max_work_group_size(4, 4, 4)]] void func1();
95+
[[sycl::reqd_work_group_size(2, 2, 2)]] void func1() {}
96+
97+
[[sycl::reqd_work_group_size(32, 32, 32)]] void f32x32x32() {}
98+
99+
int main() {
100+
q.submit([&](handler &h) {
101+
// CHECK: FunctionDecl {{.*}} {{.*}}kernel_name1
102+
// CHECK: ReqdWorkGroupSizeAttr
103+
// CHECK-NEXT: ConstantExpr{{.*}}'int'
104+
// CHECK-NEXT: value: Int 16
105+
// CHECK-NEXT: IntegerLiteral{{.*}}16{{$}}
106+
// CHECK-NEXT: ConstantExpr{{.*}}'int'
107+
// CHECK-NEXT: value: Int 1
108+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
109+
// CHECK-NEXT: ConstantExpr{{.*}}'int'
110+
// CHECK-NEXT: value: Int 1
111+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
112+
Functor16 f16;
113+
h.single_task<class kernel_name1>(f16);
114+
115+
// CHECK: FunctionDecl {{.*}} {{.*}}kernel_name2
116+
// CHECK: ReqdWorkGroupSizeAttr
117+
// CHECK-NEXT: ConstantExpr{{.*}}'int'
118+
// CHECK-NEXT: value: Int 4
119+
// CHECK-NEXT: IntegerLiteral{{.*}}4{{$}}
120+
// CHECK-NEXT: ConstantExpr{{.*}}'int'
121+
// CHECK-NEXT: value: Int 1
122+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
123+
// CHECK-NEXT: ConstantExpr{{.*}}'int'
124+
// CHECK-NEXT: value: Int 1
125+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
126+
Functor f;
127+
h.single_task<class kernel_name2>(f);
128+
129+
// CHECK: FunctionDecl {{.*}} {{.*}}kernel_name3
130+
// CHECK: ReqdWorkGroupSizeAttr
131+
// CHECK-NEXT: ConstantExpr{{.*}}'int'
132+
// CHECK-NEXT: value: Int 16
133+
// CHECK-NEXT: IntegerLiteral{{.*}}16{{$}}
134+
// CHECK-NEXT: ConstantExpr{{.*}}'int'
135+
// CHECK-NEXT: value: Int 16
136+
// CHECK-NEXT: IntegerLiteral{{.*}}16{{$}}
137+
// CHECK-NEXT: ConstantExpr{{.*}}'int'
138+
// CHECK-NEXT: value: Int 16
139+
// CHECK-NEXT: IntegerLiteral{{.*}}16{{$}}
140+
Functor16x16x16 f16x16x16;
141+
h.single_task<class kernel_name3>(f16x16x16);
142+
143+
// CHECK: FunctionDecl {{.*}} {{.*}}kernel_name4
144+
// CHECK: ReqdWorkGroupSizeAttr
145+
// CHECK-NEXT: ConstantExpr{{.*}}'int'
146+
// CHECK-NEXT: value: Int 128
147+
// CHECK-NEXT: IntegerLiteral{{.*}}128{{$}}
148+
// CHECK-NEXT: ConstantExpr{{.*}}'int'
149+
// CHECK-NEXT: value: Int 128
150+
// CHECK-NEXT: IntegerLiteral{{.*}}128{{$}}
151+
// CHECK-NEXT: ConstantExpr{{.*}}'int'
152+
// CHECK-NEXT: value: Int 128
153+
// CHECK-NEXT: IntegerLiteral{{.*}}128{{$}}
154+
FunctorAttr fattr;
155+
h.single_task<class kernel_name4>(fattr);
156+
157+
// CHECK: FunctionDecl {{.*}} {{.*}}kernel_name5
158+
// CHECK: ReqdWorkGroupSizeAttr
159+
// CHECK-NEXT: ConstantExpr{{.*}}'int'
160+
// CHECK-NEXT: value: Int 32
161+
// CHECK-NEXT: IntegerLiteral{{.*}}32{{$}}
162+
// CHECK-NEXT: ConstantExpr{{.*}}'int'
163+
// CHECK-NEXT: value: Int 32
164+
// CHECK-NEXT: IntegerLiteral{{.*}}32{{$}}
165+
// CHECK-NEXT: ConstantExpr{{.*}}'int'
166+
// CHECK-NEXT: value: Int 32
167+
// CHECK-NEXT: IntegerLiteral{{.*}}32{{$}}
168+
h.single_task<class kernel_name5>([]() [[sycl::reqd_work_group_size(32, 32, 32)]] {
169+
f32x32x32();
170+
});
171+
172+
// CHECK: FunctionDecl {{.*}} {{.*}}kernel_name6
173+
// CHECK: SYCLIntelNoGlobalWorkOffsetAttr
174+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
175+
// CHECK-NEXT: value: Int 1
176+
// CHECK-NEXT: IntegerLiteral{{.*}}1{{$}}
177+
// CHECK: SYCLIntelMaxWorkGroupSizeAttr {{.*}} Inherited
178+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
179+
// CHECK-NEXT: value: Int 4
180+
// CHECK-NEXT: IntegerLiteral{{.*}}4{{$}}
181+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
182+
// CHECK-NEXT: value: Int 4
183+
// CHECK-NEXT: IntegerLiteral{{.*}}4{{$}}
184+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
185+
// CHECK-NEXT: value: Int 4
186+
// CHECK-NEXT: IntegerLiteral{{.*}}4{{$}}
187+
// CHECK: ReqdWorkGroupSizeAttr
188+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
189+
// CHECK-NEXT: value: Int 2
190+
// CHECK-NEXT: IntegerLiteral{{.*}}2{{$}}
191+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
192+
// CHECK-NEXT: value: Int 2
193+
// CHECK-NEXT: IntegerLiteral{{.*}}2{{$}}
194+
// CHECK-NEXT: ConstantExpr {{.*}} 'int'
195+
// CHECK-NEXT: value: Int 2
196+
// CHECK-NEXT: IntegerLiteral{{.*}}2{{$}}
197+
h.single_task<class kernel_name6>(
198+
[]() { func1(); });
199+
});
200+
return 0;
201+
}

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

Lines changed: 0 additions & 116 deletions
This file was deleted.

0 commit comments

Comments
 (0)