Skip to content

Instantly share code, notes, and snippets.

@al42and
Created November 3, 2021 12:10
Show Gist options
  • Star 0 You must be signed in to star a gist
  • Fork 0 You must be signed in to fork a gist
  • Save al42and/b1a93f23922a1a5de563064451c85588 to your computer and use it in GitHub Desktop.
Save al42and/b1a93f23922a1a5de563064451c85588 to your computer and use it in GitHub Desktop.
#include <iostream>
#include <vector>
#include <CL/sycl.hpp>
using namespace cl;
int main() {
sycl::device dev{sycl::gpu_selector{}};
sycl::queue q{dev, sycl::property_list{sycl::property::queue::in_order{}}};
sycl::queue q1(q);
sycl::queue q2(q);
sycl::queue q3(q);
std::size_t test_size = 4096;
int *d_buf = sycl::malloc_device<int>(test_size, q);
std::vector<int> h_buf(test_size, 0);
for (int i = 0; i < 20; i++) {
q1.submit([&](sycl::handler &cgh) {
cgh.memcpy(d_buf, h_buf.data(), test_size * sizeof(int));
});
q2.submit([&](sycl::handler &cgh) {
cgh.parallel_for<class Kernel>(
sycl::range<1>{test_size}, [=](sycl::id<1> id) {
int i = id.get(0);
d_buf[i] = i + d_buf[(i + 1) % test_size];
i = i * i * i;
i = i % test_size;
d_buf[i]++;
});
});
q3.submit([&](sycl::handler &cgh) {
cgh.memcpy(h_buf.data(), d_buf, test_size * sizeof(int));
});
}
q.wait_and_throw();
sycl::free(d_buf, q);
return 0;
}
~
Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment