Skip to content

Commit e74d3a7

Browse files
sergey-semenovsergey-semenov
authored
[SYCL][NFC] Split unit test kernels based on their args (#19247)
Most of the unit test kernels don't have any arguments, and those that do typically use a single accessor. Prior to this change, kernel size was passed as a template parameter of the kernel function type, while still sharing the same kernel name. This causes problems for the upcoming change to kernel name based caches, which will be tied to both types and strings. Since currently there are only 2 cases for such test kernels: no arguments and a single accessor, this patch simply adds a separate kernel for the latter case.
1 parent 3d6647e commit e74d3a7

39 files changed

+310
-315
lines changed

‎sycl/unittests/Extensions/CommandGraph/Barrier.cpp

Lines changed: 59 additions & 59 deletions
Original file line numberDiff line numberDiff line change
@@ -14,19 +14,19 @@ using namespace sycl::ext::oneapi;
1414
TEST_F(CommandGraphTest, EnqueueBarrier) {
1515
Graph.begin_recording(Queue);
1616
auto Node1Graph = Queue.submit(
17-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
17+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
1818
auto Node2Graph = Queue.submit(
19-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
19+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
2020
auto Node3Graph = Queue.submit(
21-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
21+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
2222

2323
auto Barrier =
2424
Queue.submit([&](sycl::handler &cgh) { cgh.ext_oneapi_barrier(); });
2525

2626
auto Node4Graph = Queue.submit(
27-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
27+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
2828
auto Node5Graph = Queue.submit(
29-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
29+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
3030
Graph.end_recording(Queue);
3131

3232
experimental::detail::graph_impl &GraphImpl = *getSyclObjImpl(Graph);
@@ -55,19 +55,19 @@ TEST_F(CommandGraphTest, EnqueueBarrierMultipleQueues) {
5555
sycl::queue Queue2{Queue.get_context(), Dev};
5656
Graph.begin_recording({Queue, Queue2});
5757
auto Node1Graph = Queue.submit(
58-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
58+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
5959
auto Node2Graph = Queue.submit(
60-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
60+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
6161
auto Node3Graph = Queue.submit(
62-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
62+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
6363

6464
auto Barrier = Queue2.submit(
6565
[&](sycl::handler &cgh) { cgh.ext_oneapi_barrier({Node2Graph}); });
6666

6767
auto Node4Graph = Queue2.submit(
68-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
68+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
6969
auto Node5Graph = Queue2.submit(
70-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
70+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
7171
Graph.end_recording();
7272

7373
experimental::detail::graph_impl &GraphImpl = *getSyclObjImpl(Graph);
@@ -118,21 +118,21 @@ TEST_F(CommandGraphTest, EnqueueBarrierMultipleQueues) {
118118
TEST_F(CommandGraphTest, EnqueueBarrierWaitList) {
119119
Graph.begin_recording(Queue);
120120
auto Node1Graph = Queue.submit(
121-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
121+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
122122
auto Node2Graph = Queue.submit(
123-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
123+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
124124
auto Node3Graph = Queue.submit(
125-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
125+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
126126

127127
auto Barrier = Queue.submit([&](sycl::handler &cgh) {
128128
cgh.ext_oneapi_barrier({Node1Graph, Node2Graph});
129129
});
130130

131131
auto Node4Graph = Queue.submit(
132-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
132+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
133133
auto Node5Graph = Queue.submit([&](sycl::handler &cgh) {
134134
cgh.depends_on(Node3Graph);
135-
cgh.single_task<TestKernel<>>([]() {});
135+
cgh.single_task<TestKernel>([]() {});
136136
});
137137

138138
Graph.end_recording(Queue);
@@ -167,22 +167,22 @@ TEST_F(CommandGraphTest, EnqueueBarrierWaitListMultipleQueues) {
167167
sycl::queue Queue2{Queue.get_context(), Dev};
168168
Graph.begin_recording({Queue, Queue2});
169169
auto Node1Graph = Queue.submit(
170-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
170+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
171171
auto Node2Graph = Queue2.submit(
172-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
172+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
173173
auto Node3Graph = Queue2.submit(
174-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
174+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
175175

176176
// Node1Graph comes from Queue, and Node2Graph comes from Queue2
177177
auto Barrier = Queue.submit([&](sycl::handler &cgh) {
178178
cgh.ext_oneapi_barrier({Node1Graph, Node2Graph});
179179
});
180180

181181
auto Node4Graph = Queue.submit(
182-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
182+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
183183
auto Node5Graph = Queue.submit([&](sycl::handler &cgh) {
184184
cgh.depends_on(Node3Graph);
185-
cgh.single_task<TestKernel<>>([]() {});
185+
cgh.single_task<TestKernel>([]() {});
186186
});
187187

188188
auto Barrier2 = Queue2.submit([&](sycl::handler &cgh) {
@@ -223,32 +223,32 @@ TEST_F(CommandGraphTest, EnqueueBarrierWaitListMultipleQueues) {
223223
TEST_F(CommandGraphTest, EnqueueMultipleBarrier) {
224224
Graph.begin_recording(Queue);
225225
auto Node1Graph = Queue.submit(
226-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
226+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
227227
auto Node2Graph = Queue.submit(
228-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
228+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
229229
auto Node3Graph = Queue.submit(
230-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
230+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
231231

232232
auto Barrier1 = Queue.submit([&](sycl::handler &cgh) {
233233
cgh.ext_oneapi_barrier({Node1Graph, Node2Graph});
234234
});
235235

236236
auto Node4Graph = Queue.submit(
237-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
237+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
238238
auto Node5Graph = Queue.submit([&](sycl::handler &cgh) {
239239
cgh.depends_on(Node3Graph);
240-
cgh.single_task<TestKernel<>>([]() {});
240+
cgh.single_task<TestKernel>([]() {});
241241
});
242242

243243
auto Barrier2 =
244244
Queue.submit([&](sycl::handler &cgh) { cgh.ext_oneapi_barrier(); });
245245

246246
auto Node6Graph = Queue.submit(
247-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
247+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
248248
auto Node7Graph = Queue.submit(
249-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
249+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
250250
auto Node8Graph = Queue.submit(
251-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
251+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
252252

253253
Graph.end_recording(Queue);
254254

@@ -312,7 +312,7 @@ TEST_F(CommandGraphTest, InOrderQueueWithPreviousCommand) {
312312
sycl::queue InOrderQueue{Dev, Properties};
313313

314314
auto NonGraphEvent = InOrderQueue.submit(
315-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
315+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
316316

317317
Graph.begin_recording(InOrderQueue);
318318

@@ -354,10 +354,10 @@ TEST_F(CommandGraphTest, InOrderQueuesWithBarrier) {
354354
Graph.begin_recording({InOrderQueue1, InOrderQueue2, InOrderQueue3});
355355

356356
auto Node1 = InOrderQueue1.submit(
357-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
357+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
358358

359359
auto Node2 = InOrderQueue2.submit(
360-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
360+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
361361

362362
InOrderQueue3.ext_oneapi_submit_barrier({Node1});
363363

@@ -401,10 +401,10 @@ TEST_F(CommandGraphTest, InOrderQueuesWithBarrierWaitList) {
401401
Graph.begin_recording({InOrderQueue1, InOrderQueue2});
402402

403403
auto Node1 = InOrderQueue1.submit(
404-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
404+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
405405

406406
auto Node2 = InOrderQueue2.submit(
407-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
407+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
408408

409409
auto BarrierNode = InOrderQueue2.ext_oneapi_submit_barrier({Node1});
410410

@@ -442,15 +442,15 @@ TEST_F(CommandGraphTest, InOrderQueuesWithEmptyBarrierWaitList) {
442442
Graph.begin_recording({InOrderQueue1, InOrderQueue2});
443443

444444
auto Node1 = InOrderQueue1.submit(
445-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
445+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
446446

447447
auto Node2 = InOrderQueue2.submit(
448-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
448+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
449449

450450
auto BarrierNode = InOrderQueue1.ext_oneapi_submit_barrier();
451451

452452
auto Node3 = InOrderQueue2.submit(
453-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
453+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
454454

455455
Graph.end_recording();
456456

@@ -502,16 +502,16 @@ TEST_F(CommandGraphTest, BarrierMixedQueueTypes) {
502502
Graph.begin_recording({InOrderQueue, OutOfOrderQueue});
503503

504504
auto Node1 = OutOfOrderQueue.submit(
505-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
505+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
506506

507507
auto Node2 = OutOfOrderQueue.submit(
508-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
508+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
509509

510510
auto BarrierNode = InOrderQueue.ext_oneapi_submit_barrier({Node1, Node2});
511511

512512
auto Node3 = OutOfOrderQueue.submit([&](sycl::handler &cgh) {
513513
cgh.depends_on(Node2);
514-
cgh.single_task<TestKernel<>>([]() {});
514+
cgh.single_task<TestKernel>([]() {});
515515
});
516516

517517
Graph.end_recording();
@@ -563,14 +563,14 @@ TEST_F(CommandGraphTest, BarrierBetweenExplicitNodes) {
563563
InOrderQueue};
564564

565565
auto Node1 = Graph.add(
566-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
566+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
567567

568568
Graph.begin_recording(InOrderQueue);
569569
auto BarrierNode = InOrderQueue.ext_oneapi_submit_barrier();
570570
Graph.end_recording();
571571

572572
auto Node2 = Graph.add(
573-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); },
573+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); },
574574
{experimental::property::node::depends_on(Node1)});
575575

576576
// Check the graph structure
@@ -604,26 +604,26 @@ TEST_F(CommandGraphTest, BarrierMultipleOOOQueue) {
604604
Graph.begin_recording({Queue, Queue2});
605605

606606
auto Node1 = Queue.submit(
607-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
607+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
608608

609609
auto Node2 = Queue.submit(
610-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
610+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
611611

612612
auto Node3 = Queue2.submit(
613-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
613+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
614614

615615
auto Node4 = Queue2.submit(
616-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
616+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
617617

618618
auto BarrierNode = Queue.ext_oneapi_submit_barrier();
619619

620620
auto Node5 = Queue2.submit([&](sycl::handler &cgh) {
621621
cgh.depends_on({Node3, Node4});
622-
cgh.single_task<TestKernel<>>([]() {});
622+
cgh.single_task<TestKernel>([]() {});
623623
});
624624

625625
auto Node6 = Queue.submit(
626-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
626+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
627627

628628
Graph.end_recording();
629629

@@ -682,15 +682,15 @@ TEST_F(CommandGraphTest, BarrierMultipleInOrderQueue) {
682682
Graph.begin_recording({InOrderQueue1, InOrderQueue2});
683683

684684
auto Node1 = InOrderQueue1.submit(
685-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
685+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
686686

687687
auto Node2 = InOrderQueue2.submit(
688-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
688+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
689689

690690
auto BarrierNode = InOrderQueue1.ext_oneapi_submit_barrier();
691691

692692
auto Node3 = InOrderQueue2.submit(
693-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
693+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
694694

695695
Graph.end_recording();
696696

@@ -733,15 +733,15 @@ TEST_F(CommandGraphTest, BarrierMultipleMixedOrderQueues) {
733733
Graph.begin_recording({Queue, InOrderQueue});
734734

735735
auto Node1 = Queue.submit(
736-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
736+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
737737

738738
auto Node2 = InOrderQueue.submit(
739-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
739+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
740740

741741
auto BarrierNode = Queue.ext_oneapi_submit_barrier();
742742

743743
auto Node3 = InOrderQueue.submit(
744-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
744+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
745745

746746
Graph.end_recording();
747747

@@ -827,9 +827,9 @@ TEST_F(CommandGraphTest, BarrierWithInOrderCommands) {
827827

828828
Graph.begin_recording({InOrderQueue1, InOrderQueue2});
829829
auto Node1 = InOrderQueue1.submit(
830-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
830+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
831831
auto Node2 = InOrderQueue2.submit(
832-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
832+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
833833
Graph.end_recording();
834834

835835
Graph.begin_recording({InOrderQueue1, InOrderQueue2});
@@ -839,9 +839,9 @@ TEST_F(CommandGraphTest, BarrierWithInOrderCommands) {
839839

840840
Graph.begin_recording({InOrderQueue1, InOrderQueue2});
841841
auto Node3 = InOrderQueue1.submit(
842-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
842+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
843843
auto Node4 = InOrderQueue2.submit(
844-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
844+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
845845
Graph.end_recording();
846846

847847
Graph.begin_recording({InOrderQueue1, InOrderQueue2});
@@ -851,9 +851,9 @@ TEST_F(CommandGraphTest, BarrierWithInOrderCommands) {
851851

852852
Graph.begin_recording({InOrderQueue1, InOrderQueue2});
853853
auto Node5 = InOrderQueue1.submit(
854-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
854+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
855855
auto Node6 = InOrderQueue2.submit(
856-
[&](sycl::handler &cgh) { cgh.single_task<TestKernel<>>([]() {}); });
856+
[&](sycl::handler &cgh) { cgh.single_task<TestKernel>([]() {}); });
857857
Graph.end_recording();
858858

859859
Graph.begin_recording({InOrderQueue1, InOrderQueue2});

0 commit comments

Comments
 (0)