1
2
3
4
5
6
7
8
9
10
11
12
13
14
15
16
17
18
19
20
21
22
23
24
25
26
27
28
29
30
31
32
33
34
35
36
37
38
39
40
41
42
43
44
45
46
47
48
49
50
51
52
53
54
55
56
57
58
59
60
61
62
63
64
65
66
67
68
69
|
#include <forstio/test/suite.hpp>
#include "../c++/remote.hpp"
namespace {
namespace schema {
using namespace saw::schema;
using Calculator = Interface<
Member<
Function<Tuple<Int64, Int64>, Int64>, "add"
>
, Member<
Function<Tuple<Int64, Int64>, Int64>, "multiply"
>
>;
}
SAW_TEST("SYCL Interface Calculator"){
using namespace saw;
cl::sycl::queue cmd_queue;
interface<schema::Calculator, encode::Sycl<encode::Native>, cl::sycl::queue*> cl_iface {
[](data<schema::Tuple<schema::Int64, schema::Int64>, encode::Sycl<encode::Native>>& in, cl::sycl::queue* cmd) -> data<schema::Int64> {
std::array<int64_t,2> h_xy{in.get<0>().get(), in.get<1>().get()};
int64_t res{};
cl::sycl::buffer<int64_t,1> d_xy { h_xy.data(), h_xy.size() };
cl::sycl::buffer<int64_t,1> d_z { &res, 1u };
cmd->submit([&](cl::sycl::handler& h){
auto a_xy = d_xy.get_access<cl::sycl::access::mode::read>(h);
auto a_z = d_z.get_access<cl::sycl::access::mode::write>(h);
h.parallel_for(cl::sycl::range<1>(1u), [=] (cl::sycl::id<1> it){
a_z[0] = a_xy[0] + a_xy[1];
});
});
cmd->wait();
return data<schema::Int64>{res};
},
[](data<schema::Tuple<schema::Int64, schema::Int64>,encode::Sycl<encode::Native>>& in, cl::sycl::queue* cmd) -> data<schema::Int64> {
return data<schema::Int64>{in.get<0>().get() * in.get<1>().get()};
}
};
int64_t x = 1;
int64_t y = -2;
int64_t sum = x + y;
int64_t mult = x * y;
data<schema::Tuple<schema::Int64, schema::Int64>> input;
input.template get<0>().set(x);
input.template get<1>().set(y);
{
auto eov = cl_iface.template call<"add">(input, &cmd_queue);
SAW_EXPECT(eov.is_value(), "Returned error on add");
SAW_EXPECT(eov.get_value().get() == sum, "Addition was incorrect");
}
{
auto eov = cl_iface.template call<"multiply">(input, &cmd_queue);
SAW_EXPECT(eov.is_value(), "Returned error on add");
SAW_EXPECT(eov.get_value().get() == mult, "Addition was incorrect");
}
}
}
|
