Skip to content

Commit 5e06976

Browse files
MrSidimsromanovvlad
MrSidims
authored and
romanovvlad
committed
[SYCL][FPGA] Improve FPGA pipes testing
The test itself was reworked. Also were added following cases: 1. SYCL pipe constructed from template type; 2. Multiple pipes call site; 3. Tests cases for non-blocking pipes were repeated for blocking as well. Signed-off-by: Dmitry Sidorov <[email protected]>
1 parent 6710d1a commit 5e06976

File tree

1 file changed

+180
-126
lines changed

1 file changed

+180
-126
lines changed

‎sycl/test/fpga_tests/fpga_pipes.cpp

Lines changed: 180 additions & 126 deletions
Original file line numberDiff line numberDiff line change
@@ -10,171 +10,225 @@
1010
#include <CL/sycl.hpp>
1111
#include <iostream>
1212

13-
// For pipes created with namespaces set
13+
// For simple non-blocking pipes with explicit type
14+
class some_nb_pipe;
15+
16+
// For non-blocking pipes created with namespaces set
1417
namespace some {
15-
class pipe;
18+
class nb_pipe;
1619
}
1720

18-
using namespace cl::sycl;
21+
// For non-blocking template pipes
22+
template<int N>
23+
class templ_nb_pipe;
1924

20-
int main() {
21-
int data[] = {0};
25+
// For non-blocking multiple pipes
26+
template<int N>
27+
using PipeMulNb = cl::sycl::pipe<class templ_nb_pipe<N>, int>;
2228

23-
{
24-
// Test for non-blocking pipes
25-
queue Queue;
26-
using Pipe = pipe<class some_pipe, int, 1>;
27-
28-
Queue.submit([&](handler &cgh) {
29-
cgh.single_task<class foo_nb>([=]() {
30-
bool SuccessCode = false;
31-
while (!SuccessCode)
32-
Pipe::write(42, SuccessCode);
33-
});
34-
});
29+
// For simple blocking pipes with explicit type
30+
class some_bl_pipe;
3531

36-
buffer<int, 1> writeBuf(data, 1);
37-
Queue.submit([&](handler &cgh) {
38-
auto write_acc = writeBuf.get_access<access::mode::write>(cgh);
32+
// For blocking pipes created with namespaces set
33+
namespace some {
34+
class bl_pipe;
35+
}
3936

40-
cgh.single_task<class goo_nb>([=]() {
41-
bool SuccessCode = false;
42-
while (!SuccessCode)
43-
write_acc[0] = Pipe::read(SuccessCode);
44-
});
45-
});
37+
// For blocking template pipes
38+
template<int N>
39+
class templ_bl_pipe;
4640

47-
auto readHostBuffer = writeBuf.get_access<access::mode::read>();
48-
if (readHostBuffer[0] != 42) {
49-
std::cout << "Result mismatches " << readHostBuffer[0] << " Vs expected "
50-
<< 42 << std::endl;
41+
// For blocking multiple pipes
42+
template<int N>
43+
using PipeMulBl = cl::sycl::pipe<class templ_bl_pipe<N>, int>;
5144

52-
return -1;
53-
}
54-
}
45+
// Kernel names
46+
template <int TestNumber, int KernelNumber = 0>
47+
class writer;
48+
template <int TestNumber, int KernelNumber = 0>
49+
class reader;
5550

56-
{
57-
// Test for simple non-blocking pipes with explicit type
58-
queue Queue;
59-
60-
buffer<int, 1> readBuf(data, 1);
61-
Queue.submit([&](handler &cgh) {
62-
cgh.single_task<class boo_nb>([=]() {
63-
bool SuccessCode;
64-
while (!SuccessCode)
65-
pipe<class some_pipe, int, 1>::write(42, SuccessCode);
66-
});
51+
// Test for simple non-blocking pipes
52+
template<typename PipeName, int TestNumber>
53+
int test_simple_nb_pipe(cl::sycl::queue Queue) {
54+
int data[] = {0};
55+
56+
using Pipe = cl::sycl::pipe<PipeName, int>;
57+
58+
cl::sycl::buffer<int, 1> readBuf(data, 1);
59+
Queue.submit([&](cl::sycl::handler &cgh) {
60+
cgh.single_task<class writer<TestNumber>>([=]() {
61+
bool SuccessCode = false;
62+
do {
63+
Pipe::write(42, SuccessCode);
64+
} while (!SuccessCode);
6765
});
66+
});
6867

69-
buffer<int, 1> writeBuf(data, 1);
70-
Queue.submit([&](handler &cgh) {
71-
auto write_acc = writeBuf.get_access<access::mode::write>(cgh);
68+
cl::sycl::buffer<int, 1> writeBuf(data, 1);
69+
Queue.submit([&](cl::sycl::handler &cgh) {
70+
auto write_acc = writeBuf.get_access<cl::sycl::access::mode::write>(cgh);
7271

73-
cgh.single_task<class zoo_nb>([=]() {
74-
bool SuccessCode;
75-
while (!SuccessCode)
76-
write_acc[0] = pipe<class some_pipe, int, 1>::read(SuccessCode);
77-
});
72+
cgh.single_task<class reader<TestNumber>>([=]() {
73+
bool SuccessCode = false;
74+
do {
75+
write_acc[0] = Pipe::read(SuccessCode);
76+
} while (!SuccessCode);
7877
});
78+
});
7979

80-
auto readHostBuffer = writeBuf.get_access<access::mode::read>();
81-
if (readHostBuffer[0] != 42) {
82-
std::cout << "Result mismatches " << readHostBuffer[0] << " Vs expected "
83-
<< 42 << std::endl;
80+
auto readHostBuffer = writeBuf.get_access<cl::sycl::access::mode::read>();
81+
if (readHostBuffer[0] != 42) {
82+
std::cout << "Test: " << TestNumber << "\nResult mismatches "
83+
<< readHostBuffer[0] << " Vs expected " << 42 << std::endl;
8484

85-
return -1;
86-
}
85+
return -1;
8786
}
8887

89-
{
90-
// Test for simple non-blocking pipes created with namespaces set
91-
queue Queue;
92-
93-
buffer<int, 1> readBuf(data, 1);
94-
Queue.submit([&](handler &cgh) {
95-
cgh.single_task<class foo_ns>([=]() {
96-
bool SuccessCode;
97-
while (!SuccessCode)
98-
pipe<class some::pipe, int, 1>::write(42, SuccessCode);
99-
});
100-
});
88+
return 0;
89+
}
10190

102-
buffer<int, 1> writeBuf(data, 1);
103-
Queue.submit([&](handler &cgh) {
104-
auto write_acc = writeBuf.get_access<access::mode::write>(cgh);
91+
// Test for multiple non-blocking pipes
92+
template<int TestNumber>
93+
int test_multiple_nb_pipe(cl::sycl::queue Queue) {
94+
int data[] = {0};
10595

106-
cgh.single_task<class boo_ns>([=]() {
107-
bool SuccessCode;
108-
while (!SuccessCode)
109-
write_acc[0] = pipe<class some::pipe, int, 1>::read(SuccessCode);
110-
});
96+
Queue.submit([&](cl::sycl::handler &cgh) {
97+
cgh.single_task<class writer<TestNumber, /*KernelNumber*/ 1>>([=]() {
98+
bool SuccessCode = false;
99+
do {
100+
PipeMulNb<1>::write(19, SuccessCode);
101+
} while (!SuccessCode);
102+
});
103+
});
104+
105+
Queue.submit([&](cl::sycl::handler &cgh) {
106+
cgh.single_task<class writer<TestNumber, /*KernelNumber*/ 2>>([=]() {
107+
bool SuccessCode = false;
108+
do {
109+
PipeMulNb<2>::write(23, SuccessCode);
110+
} while (!SuccessCode);
111+
});
112+
});
113+
114+
cl::sycl::buffer<int, 1> writeBuf(data, 1);
115+
Queue.submit([&](cl::sycl::handler &cgh) {
116+
auto write_acc = writeBuf.get_access<cl::sycl::access::mode::write>(cgh);
117+
cgh.single_task<class reader<TestNumber>>([=]() {
118+
bool SuccessCodeA = false;
119+
int Value = 0;
120+
do {
121+
Value = PipeMulNb<1>::read(SuccessCodeA);
122+
} while (!SuccessCodeA);
123+
write_acc[0] = Value;
124+
bool SuccessCodeB = false;
125+
do {
126+
Value = PipeMulNb<2>::read(SuccessCodeB);
127+
} while (!SuccessCodeB);
128+
write_acc[0] += Value;
111129
});
130+
});
112131

113-
auto readHostBuffer = writeBuf.get_access<access::mode::read>();
114-
if (readHostBuffer[0] != 42) {
115-
std::cout << "Result mismatches " << readHostBuffer[0] << " Vs expected "
116-
<< 42 << std::endl;
132+
auto readHostBuffer = writeBuf.get_access<cl::sycl::access::mode::read>();
133+
if (readHostBuffer[0] != 42) {
134+
std::cout << "Test: " << TestNumber << "\nResult mismatches "
135+
<< readHostBuffer[0] << " Vs expected " << 42 << std::endl;
117136

118-
return -1;
119-
}
137+
return -1;
120138
}
121139

122-
{
123-
// Test for forward declared pipes
124-
queue Queue;
125-
class pipe_type_for_lambdas;
126-
127-
buffer<int, 1> readBuf(data, 1);
128-
Queue.submit([&](handler &cgh) {
129-
cgh.single_task<class foo_la>([=]() {
130-
bool SuccessCode;
131-
while (!SuccessCode)
132-
pipe<class pipe_type_for_lambdas, int>::write(42, SuccessCode);
133-
});
140+
return 0;
141+
}
142+
143+
// Test for simple blocking pipes
144+
template<typename PipeName, int TestNumber>
145+
int test_simple_bl_pipe(cl::sycl::queue Queue) {
146+
int data[] = {0};
147+
148+
using Pipe = cl::sycl::pipe<PipeName, int>;
149+
150+
cl::sycl::buffer<int, 1> readBuf(data, 1);
151+
Queue.submit([&](cl::sycl::handler &cgh) {
152+
cgh.single_task<class writer<TestNumber>>([=]() {
153+
Pipe::write(42);
134154
});
155+
});
135156

136-
buffer<int, 1> writeBuf(data, 1);
137-
Queue.submit([&](handler &cgh) {
138-
cgh.single_task<class boo_la>([=]() {
139-
bool SuccessCode;
140-
while (!SuccessCode)
141-
pipe<class pipe_type_for_lambdas, int>::read(SuccessCode);
142-
});
157+
cl::sycl::buffer<int, 1> writeBuf(data, 1);
158+
Queue.submit([&](cl::sycl::handler &cgh) {
159+
auto write_acc = writeBuf.get_access<cl::sycl::access::mode::write>(cgh);
160+
161+
cgh.single_task<class reader<TestNumber>>([=]() {
162+
write_acc[0] = Pipe::read();
143163
});
164+
});
144165

145-
auto readHostBuffer = writeBuf.get_access<access::mode::read>();
146-
if (readHostBuffer[0] != 42) {
147-
std::cout << "Result mismatches " << readHostBuffer[0] << " Vs expected "
148-
<< 42 << std::endl;
166+
auto readHostBuffer = writeBuf.get_access<cl::sycl::access::mode::read>();
167+
if (readHostBuffer[0] != 42) {
168+
std::cout << "Test: " << TestNumber << "\nResult mismatches "
169+
<< readHostBuffer[0] << " Vs expected " << 42 << std::endl;
149170

150-
return -1;
151-
}
171+
return -1;
152172
}
153173

154-
{
155-
// Test for blocking pipes
156-
queue Queue;
157-
using Pipe = pipe<class some_pipe, int, 1>;
174+
return 0;
175+
}
176+
177+
// Test for multiple blocking pipes
178+
template<int TestNumber>
179+
int test_multiple_bl_pipe(cl::sycl::queue Queue) {
180+
int data[] = {0};
158181

159-
Queue.submit([&](handler &cgh) {
160-
cgh.single_task<class foo_b>([=]() { Pipe::write(42); });
182+
Queue.submit([&](cl::sycl::handler &cgh) {
183+
cgh.single_task<class writer<TestNumber, /*KernelNumber*/ 1>>([=]() {
184+
PipeMulBl<1>::write(19);
161185
});
186+
});
162187

163-
buffer<int, 1> writeBuf(data, 1);
164-
Queue.submit([&](handler &cgh) {
165-
auto write_acc = writeBuf.get_access<access::mode::write>(cgh);
166-
167-
cgh.single_task<class goo_b>([=]() { write_acc[0] = Pipe::read(); });
188+
Queue.submit([&](cl::sycl::handler &cgh) {
189+
cgh.single_task<class writer<TestNumber, /*KernelNumber*/ 2>>([=]() {
190+
PipeMulBl<2>::write(23);
168191
});
192+
});
193+
194+
cl::sycl::buffer<int, 1> writeBuf(data, 1);
195+
Queue.submit([&](cl::sycl::handler &cgh) {
196+
auto write_acc = writeBuf.get_access<cl::sycl::access::mode::write>(cgh);
197+
cgh.single_task<class reader<TestNumber>>([=]() {
198+
write_acc[0] = PipeMulBl<1>::read();
199+
write_acc[0] += PipeMulBl<2>::read();
200+
});
201+
});
169202

170-
auto readHostBuffer = writeBuf.get_access<access::mode::read>();
171-
if (readHostBuffer[0] != 42) {
172-
std::cout << "Result mismatches " << readHostBuffer[0] << " Vs expected "
173-
<< 42 << std::endl;
203+
auto readHostBuffer = writeBuf.get_access<cl::sycl::access::mode::read>();
204+
if (readHostBuffer[0] != 42) {
205+
std::cout << "Test: " << TestNumber << "\nResult mismatches "
206+
<< readHostBuffer[0] << " Vs expected " << 42 << std::endl;
174207

175-
return -1;
176-
}
208+
return -1;
177209
}
178210

179211
return 0;
180212
}
213+
214+
int main() {
215+
cl::sycl::queue Queue;
216+
217+
// Non-blocking pipes
218+
int Result = test_simple_nb_pipe<some_nb_pipe, /*test number*/ 1>(Queue);
219+
Result &= test_simple_nb_pipe<some::nb_pipe, /*test number*/ 2>(Queue);
220+
class forward_nb_pipe;
221+
Result &= test_simple_nb_pipe<forward_nb_pipe, /*test number*/ 3>(Queue);
222+
Result &= test_simple_nb_pipe<templ_nb_pipe<0>, /*test number*/ 4>(Queue);
223+
Result &= test_multiple_nb_pipe</*test number*/ 5>(Queue);
224+
225+
// Blocking pipes
226+
Result &= test_simple_bl_pipe<some_bl_pipe, /*test number*/ 6>(Queue);
227+
Result &= test_simple_bl_pipe<some::bl_pipe, /*test number*/ 7>(Queue);
228+
class forward_bl_pipe;
229+
Result &= test_simple_bl_pipe<forward_bl_pipe, /*test number*/ 8>(Queue);
230+
Result &= test_simple_bl_pipe<templ_bl_pipe<0>, /*test number*/ 9>(Queue);
231+
Result &= test_multiple_bl_pipe</*test number*/ 10>(Queue);
232+
233+
return Result;
234+
}

0 commit comments

Comments
 (0)