Skip to content
This repository was archived by the owner on Mar 28, 2023. It is now read-only.

Commit 3c9884e

Browse files
[SYCL] Add a test with reduction kernel accepting sycl::item (#1403)
Complementary to intel/llvm#7478.
1 parent 1192df5 commit 3c9884e

File tree

1 file changed

+52
-0
lines changed

1 file changed

+52
-0
lines changed
Lines changed: 52 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,52 @@
1+
// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out
2+
// RUN: %CPU_RUN_PLACEHOLDER %t.out
3+
// RUN: %GPU_RUN_PLACEHOLDER %t.out
4+
// RUN: %ACC_RUN_PLACEHOLDER %t.out
5+
6+
#include <sycl/sycl.hpp>
7+
8+
using namespace sycl;
9+
10+
int main() {
11+
queue q;
12+
auto *RedMem = malloc_shared<int>(1, q);
13+
auto *Success = malloc_shared<bool>(1, q);
14+
*Success = true;
15+
16+
*RedMem = 0;
17+
q.parallel_for(range<1>{7}, reduction(RedMem, std::plus<int>{}),
18+
[=](item<1> Item, auto &Red) {
19+
Red += 1;
20+
if (Item.get_range(0) != 7)
21+
*Success = false;
22+
if (Item.get_id(0) == 7)
23+
*Success = false;
24+
})
25+
.wait();
26+
27+
assert(*RedMem == 7);
28+
assert(*Success);
29+
30+
*RedMem = 0;
31+
q.parallel_for(range<2>{1030, 7}, reduction(RedMem, std::plus<int>{}),
32+
[=](item<2> Item, auto &Red) {
33+
Red += 1;
34+
if (Item.get_range(0) != 1030)
35+
*Success = false;
36+
if (Item.get_range(1) != 7)
37+
*Success = false;
38+
39+
if (Item.get_id(0) == 1030)
40+
*Success = false;
41+
if (Item.get_id(1) == 7)
42+
*Success = false;
43+
})
44+
.wait();
45+
46+
assert(*RedMem == 1030 * 7);
47+
assert(*Success);
48+
49+
free(RedMem, q);
50+
free(Success, q);
51+
return 0;
52+
}

0 commit comments

Comments
 (0)