Skip to content

Commit 75197db

Browse files
committed
[SYCL] Add basic tests of reduction properties
Signed-off-by: John Pennycook <[email protected]>
1 parent b94cb3f commit 75197db

File tree

2 files changed

+80
-0
lines changed

2 files changed

+80
-0
lines changed
Lines changed: 43 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,43 @@
1+
#include <cstdio>
2+
#include <cstdlib>
3+
#include <random>
4+
5+
#include <sycl/sycl.hpp>
6+
7+
namespace syclex = sycl::ext::oneapi::experimental;
8+
9+
float sum(sycl::queue q, float *input, size_t N) {
10+
11+
float result = 0;
12+
{
13+
sycl::buffer<float> buf{&result, 1};
14+
15+
q.submit([&](sycl::handler &h) {
16+
auto reduction = sycl::reduction(
17+
buf, h, sycl::plus<>(), syclex::properties(syclex::deterministic));
18+
h.parallel_for(N, reduction,
19+
[=](size_t i, auto &reducer) { reducer += input[i]; });
20+
});
21+
}
22+
return result;
23+
}
24+
25+
int main(int argc, char *argv[]) {
26+
27+
constexpr size_t N = 1024;
28+
float *array = new float[N];
29+
30+
std::random_device rd;
31+
std::mt19937 gen(rd());
32+
std::uniform_real_distribution<float> dist(0.0f, 1.0f);
33+
std::generate(array, array + N, [&]() { return dist(gen); });
34+
35+
sycl::queue q;
36+
float x = sum(q, array, N);
37+
float y = sum(q, array, N);
38+
39+
// NB: determinism guarantees bitwise reproducible reductions for floats
40+
assert(sycl::bit_cast<unsigned int>(x) == sycl::bit_cast<unsigned int>(y));
41+
42+
delete[] array;
43+
}
Lines changed: 37 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,37 @@
1+
#include <cstdio>
2+
#include <cstdlib>
3+
#include <numeric>
4+
5+
#include <sycl/sycl.hpp>
6+
7+
namespace syclex = sycl::ext::oneapi::experimental;
8+
9+
int sum(sycl::queue q, int *input, size_t N) {
10+
11+
int result = 42;
12+
{
13+
sycl::buffer<int> buf{&result, 1};
14+
15+
q.submit([&](sycl::handler &h) {
16+
auto reduction =
17+
sycl::reduction(buf, h, sycl::plus<>(),
18+
syclex::properties(syclex::initialize_to_identity));
19+
h.parallel_for(N, reduction,
20+
[=](size_t i, auto &reducer) { reducer += input[i]; });
21+
});
22+
}
23+
return result;
24+
}
25+
26+
int main(int argc, char *argv[]) {
27+
28+
constexpr size_t N = 32;
29+
int *array = new int[N];
30+
std::iota(array, array + N, 1);
31+
32+
sycl::queue q;
33+
int x = sum(q, array, N);
34+
assert(x == 528);
35+
36+
delete[] array;
37+
}

0 commit comments

Comments
 (0)