Skip to content
Closed
Show file tree
Hide file tree
Changes from 1 commit
Commits
File filter

Filter by extension

Filter by extension

Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
Original file line number Diff line number Diff line change
@@ -1,4 +1,4 @@
= sycl_ext_codeplay_enqueue_native_command
= sycl_ext_oneapi_enqueue_native_command

:source-highlighter: coderay
:coderay-linenums-mode: table
Expand Down Expand Up @@ -66,7 +66,7 @@ This extension is derived from the experimental AdaptiveCpp extension,
`enqueue_custom_operation` which is documented
https://github.com/AdaptiveCpp/AdaptiveCpp/blob/develop/doc/enqueue-custom-operation.md[here].

The goal of `ext_codeplay_enqueue_native_command` is to integrate interop
The goal of `ext_oneapi_enqueue_native_command` is to integrate interop
work within the SYCL runtime's creation of the asynchronous SYCL DAG. As such,
the user defined lambda must only enqueue asynchronous, as opposed to
synchronous, backend work within the user lambda. Asynchronous work must only
Expand All @@ -75,11 +75,11 @@ be submitted to the native queue obtained from

=== Differences with `host_task`

A callable submitted to `ext_codeplay_enqueue_native_command` won't wait
A callable submitted to `ext_oneapi_enqueue_native_command` won't wait
on its dependent events to execute. The dependencies passed to an
`ext_codeplay_enqueue_native_command` submission will result in dependencies being
`ext_oneapi_enqueue_native_command` submission will result in dependencies being
implicitly handled in the backend API, using the native queue object associated
with the SYCL queue that the `sycl_ext_codeplay_enqueue_native_command` is
with the SYCL queue that the `sycl_ext_oneapi_enqueue_native_command` is
submitted to. This gives different synchronization guarantees from normal SYCL
`host_task` s, which guarantee that the `host_task` callable will only begin
execution once all of its dependent events have completed.
Expand All @@ -89,7 +89,7 @@ In this example:
```
q.submit([&](sycl::handler &cgh) {
cgh.depends_on(dep_event);
cgh.ext_codeplay_enqueue_native_command([=](sycl::interop_handle &h) {
cgh.ext_oneapi_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
Expand All @@ -112,7 +112,7 @@ will only happen once the host task's dependent events are observed to be
complete on the host.

A SYCL event returned by a submission of a
`ext_codeplay_enqueue_native_command` command is only complete once the
`ext_oneapi_enqueue_native_command` command is only complete once the
asynchronous work enqueued to the native queue obtained through
`interop_handle::get_native_queue()` has completed.

Expand Down Expand Up @@ -147,7 +147,7 @@ class:
```c++
class handler {
template <typename Func>
void ext_codeplay_enqueue_native_command(Func&& interopCallable);
void ext_oneapi_enqueue_native_command(Func&& interopCallable);
};
```

Expand All @@ -156,7 +156,7 @@ parameter of type `interop_handle`.

_Effects_: The `interopCallable` object is called exactly once, and this call
may be made asynchronously even after the calling thread returns from
`ext_codeplay_enqueue_native_command`.
`ext_oneapi_enqueue_native_command`.

The call to `interopCallable` may submit one or more asynchronous tasks to the
native backend object obtained from `interop_handle::get_native_queue`, and
Expand Down Expand Up @@ -189,7 +189,7 @@ 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_oneapi_enqueue_native_command([=](sycl::interop_handle &h) {
// Can extract device pointers from accessors
void *native_mem = h.get_native_mem<sycl::backend::hip>(acc);
// Can extract stream
Expand All @@ -211,7 +211,7 @@ q.wait();

=== sycl_ext_oneapi_graph

`ext_codeplay_enqueue_native_command`
`ext_oneapi_enqueue_native_command`
cannot be used in graph nodes. A synchronous exception will be thrown with error
code `invalid` if a user tries to add them to a graph.

Original file line number Diff line number Diff line change
Expand Up @@ -1943,10 +1943,10 @@ The kernels loaded using
link:../experimental/sycl_ext_oneapi_kernel_compiler_spirv.asciidoc[sycl_ext_oneapi_kernel_compiler_spirv]
behave as normal when used in graph nodes.

==== sycl_ext_codeplay_enqueue_native_command
==== sycl_ext_oneapi_enqueue_native_command

`ext_codeplay_enqueue_native_command`, defined in
link:../experimental/sycl_ext_codeplay_enqueue_native_command.asciidoc[sycl_ext_codeplay_enqueue_native_command]
`ext_oneapi_enqueue_native_command`, defined in
link:../experimental/sycl_ext_oneapi_enqueue_native_command.asciidoc[sycl_ext_oneapi_enqueue_native_command]
cannot be used in graph nodes. A synchronous exception will be thrown with error
code `invalid` if a user tries to add them to a graph.

Expand Down
6 changes: 3 additions & 3 deletions sycl/include/sycl/ext/oneapi/experimental/graph.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -55,7 +55,7 @@ enum class UnsupportedGraphFeatures {
sycl_ext_oneapi_device_global = 6,
sycl_ext_oneapi_bindless_images = 7,
sycl_ext_oneapi_experimental_cuda_cluster_launch = 8,
sycl_ext_codeplay_enqueue_native_command = 9
sycl_ext_oneapi_enqueue_native_command = 9
};

inline const char *
Expand All @@ -80,8 +80,8 @@ UnsupportedFeatureToString(UnsupportedGraphFeatures Feature) {
return "sycl_ext_oneapi_bindless_images";
case UGF::sycl_ext_oneapi_experimental_cuda_cluster_launch:
return "sycl_ext_oneapi_experimental_cuda_cluster_launch";
case UGF::sycl_ext_codeplay_enqueue_native_command:
return "sycl_ext_codeplay_enqueue_native_command";
case UGF::sycl_ext_oneapi_enqueue_native_command:
return "sycl_ext_oneapi_enqueue_native_command";
}

assert(false && "Unhandled graphs feature");
Expand Down
8 changes: 4 additions & 4 deletions sycl/include/sycl/handler.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -1867,7 +1867,7 @@ class __SYCL_EXPORT handler {
template <typename FuncT>
std::enable_if_t<detail::check_fn_signature<std::remove_reference_t<FuncT>,
void(interop_handle)>::value>
ext_codeplay_enqueue_native_command_impl(FuncT &&Func) {
ext_oneapi_enqueue_native_command_impl(FuncT &&Func) {
throwIfActionIsCreated();

// Need to copy these rather than move so that we can check associated
Expand Down Expand Up @@ -2090,11 +2090,11 @@ class __SYCL_EXPORT handler {
template <typename FuncT>
std::enable_if_t<detail::check_fn_signature<std::remove_reference_t<FuncT>,
void(interop_handle)>::value>
ext_codeplay_enqueue_native_command(FuncT &&Func) {
ext_oneapi_enqueue_native_command(FuncT &&Func) {
throwIfGraphAssociated<
ext::oneapi::experimental::detail::UnsupportedGraphFeatures::
sycl_ext_codeplay_enqueue_native_command>();
ext_codeplay_enqueue_native_command_impl(Func);
sycl_ext_oneapi_enqueue_native_command>();
ext_oneapi_enqueue_native_command_impl(Func);
}

/// Defines and invokes a SYCL kernel function for the specified range and
Expand Down
2 changes: 1 addition & 1 deletion sycl/source/detail/scheduler/commands.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -3196,7 +3196,7 @@ ur_result_t ExecCGCommand::enqueueImpQueue() {
detail::getSyclObjImpl(MQueue->get_device())->getHandleRef(),
UR_DEVICE_INFO_ENQUEUE_NATIVE_COMMAND_SUPPORT_EXP,
sizeof(NativeCommandSupport), &NativeCommandSupport, nullptr);
assert(NativeCommandSupport && "ext_codeplay_enqueue_native_command is not "
assert(NativeCommandSupport && "ext_oneapi_enqueue_native_command is not "
"supported on this device");
MQueue->getPlugin()->call(urEnqueueNativeCommandExp, MQueue->getHandleRef(),
InteropFreeFunc, &CustomOpData, ReqMems.size(),
Expand Down
4 changes: 2 additions & 2 deletions sycl/test-e2e/EnqueueNativeCommand/custom-command-cuda.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -51,7 +51,7 @@ void copy(buffer<DataT, 1> &Src, buffer<DataT, 1> &Dst, queue &Q) {
if (Q.get_backend() != IH.get_backend())
throw;
};
CGH.ext_codeplay_enqueue_native_command(Func);
CGH.ext_oneapi_enqueue_native_command(Func);
});
}

Expand Down Expand Up @@ -85,7 +85,7 @@ void test_ht_buffer(queue &Q) {
Q.submit([&](handler &CGH) {
auto Acc = Buffer.get_access<mode::write>(CGH);
auto Func = [=](interop_handle IH) { /*A no-op */ };
CGH.ext_codeplay_enqueue_native_command(Func);
CGH.ext_oneapi_enqueue_native_command(Func);
});
}

Expand Down
4 changes: 2 additions & 2 deletions sycl/test-e2e/EnqueueNativeCommand/custom-command-hip.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -57,7 +57,7 @@ void copy(buffer<DataT, 1> &Src, buffer<DataT, 1> &Dst, queue &Q) {
if (Q.get_backend() != IH.get_backend())
throw;
};
CGH.ext_codeplay_enqueue_native_command(Func);
CGH.ext_oneapi_enqueue_native_command(Func);
});
}

Expand Down Expand Up @@ -91,7 +91,7 @@ void test_ht_buffer(queue &Q) {
Q.submit([&](handler &CGH) {
auto Acc = Buffer.get_access<mode::write>(CGH);
auto Func = [=](interop_handle IH) { /*A no-op */ };
CGH.ext_codeplay_enqueue_native_command(Func);
CGH.ext_oneapi_enqueue_native_command(Func);
});
}

Expand Down
Original file line number Diff line number Diff line change
Expand Up @@ -35,7 +35,7 @@ int main() {

Q.submit([&](handler &CGH) {
accessor Acc{Buf, CGH, read_write};
CGH.ext_codeplay_enqueue_native_command([=](interop_handle IH) {
CGH.ext_oneapi_enqueue_native_command([=](interop_handle IH) {
auto Ptr = IH.get_native_mem<backend::ext_oneapi_cuda>(Acc);
auto Stream = IH.get_native_queue<backend::ext_oneapi_cuda>();
int Tmp = 0;
Expand Down
4 changes: 2 additions & 2 deletions sycl/unittests/Extensions/CommandGraph/Exceptions.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -406,12 +406,12 @@ TEST_F(CommandGraphTest, BindlessExceptionCheck) {
sycl::free(ImgMemUSM, Ctxt);
}

// ext_codeplay_enqueue_native_command isn't supported with SYCL graphs
// ext_oneapi_enqueue_native_command isn't supported with SYCL graphs
TEST_F(CommandGraphTest, EnqueueCustomCommandCheck) {
std::error_code ExceptionCode = make_error_code(sycl::errc::success);
try {
Graph.add([&](sycl::handler &CGH) {
CGH.ext_codeplay_enqueue_native_command([=](sycl::interop_handle IH) {});
CGH.ext_oneapi_enqueue_native_command([=](sycl::interop_handle IH) {});
});
} catch (exception &Exception) {
ExceptionCode = Exception.code();
Expand Down