Skip to content

Commit

Permalink
[SYCL][ext][NFC] Fix example in codeplay_enqueue_native_command (#15265)
Browse files Browse the repository at this point in the history
Another solution would be to change the spec to bring it closer to
how AdaptiveCpp does things. But here we make the example
align with the extension spec, host_task spec, and the implementation.

Fixes #15254
  • Loading branch information
al42and authored Sep 4, 2024
1 parent e2c0fcb commit 3b6c0d9
Showing 1 changed file with 7 additions and 7 deletions.
Original file line number Diff line number Diff line change
Expand Up @@ -86,19 +86,19 @@ execution once all of its dependent events have completed.

In this example:

```
```c++
q.submit([&](sycl::handler &cgh) {
cgh.depends_on(dep_event);
cgh.ext_codeplay_enqueue_native_command([=](sycl::interop_handle &h) {
cgh.ext_codeplay_enqueue_native_command([=](sycl::interop_handle h) {
printf("This will print before dep_event has completed.\n");
// This stream has been synchronized with dep_event's underlying
// hipEvent_t
hipStream_t stream = h.get_native_queue<sycl::backend::hip>();
hipStream_t stream = h.get_native_queue<sycl::backend::ext_oneapi_hip>();
hipMemcpyAsync(target_ptr, native_mem, test_size * sizeof(int),
hipMemcpyDeviceToHost, stream);
});
});
q.wait()
q.wait();
```

The print statement may print before `dep_event` has completed. However, the
Expand Down Expand Up @@ -189,11 +189,11 @@ sycl::queue q;
q.submit([&](sycl::handler &cgh) {
sycl::accessor acc{buf, cgh};

cgh.ext_codeplay_enqueue_native_command([=](sycl::interop_handle &h) {
cgh.ext_codeplay_enqueue_native_command([=](sycl::interop_handle h) {
// Can extract device pointers from accessors
void *native_mem = h.get_native_mem<sycl::backend::hip>(acc);
void *native_mem = h.get_native_mem<sycl::backend::ext_oneapi_hip>(acc);
// Can extract stream
hipStream_t stream = h.get_native_queue<sycl::backend::hip>();
hipStream_t stream = h.get_native_queue<sycl::backend::ext_oneapi_hip>();

// Can enqueue arbitrary backend operations. This could also be a kernel
// launch or call to a library that enqueues operations on the stream etc
Expand Down

0 comments on commit 3b6c0d9

Please sign in to comment.