Add sycl_khr_free_function_commands extension #644
Conversation
This extension provides an alternative mechanism for submitting commands to a device via free-functions that require developers to opt-in to the creation of event objects. It also proposes alternative names for several commands (e.g., launch) and simplifies some concepts (e.g., by removing the need for the nd_range class).
Previous "0 or more" wording only made sense when reductions could be optionally provided to functions like parallel_for; now that there are dedicated *_reduce functions, at least one reduction is required.
"is" is more consistent with ISO C++ wording.
Renaming sycl::nd_item is not a necessary part of the API redesign for submitting work, so it should be moved to its own extension. This will also give us more time to consider the design and naming of any proposed replacement(s), including how they should interact with new functionality proposed in other KHRs.
There are currently no backends that define interop for reductions, so we can remove these functions for now. If we decide later that these functions are necessary, we can release a revision of the KHR.
| This extension provides an alternative mechanism for submitting commands to a | ||
| device via free-functions that require developers to opt-in to the creation of | ||
| [code]#event# objects. |
There was a problem hiding this comment.
I suggest adding a {note} to clarify the proposed use of the new functions and when they should be preferred over the standard queue/handler member functions.
Currently this document, with all the "Effect: equivalent to...", can look like it's adding new functions for the sake of adding new functions (except the barriers). The "Usage example" is not helping. The brief mention of "opt-in to the creation of event objects" is hard to interpret unless one knows that event creation can incur overhead on certain platforms / with certain backend design.
There was a problem hiding this comment.
Agreed. Text in the Overview section is already non-normative, so it's not necessary to format it as a {note}.
| * All commands previously submitted to this queue have completed; and | ||
| * All commands associated with this command's dependencies (e.g., via | ||
| [code]#handler::depends_on#) have completed. | ||
|
|
There was a problem hiding this comment.
Since event_barrier has a clarification for "no-op" case, shall we add one here too?
{note}Overload (1) is no-op when submitted to an in-order queue with no other dependencies (e.g., specified by [code]#handler::depends_on#). Overload (2) is no-op if [code]#queue# is in-order.{endnote}
There was a problem hiding this comment.
I've added a similar note in 165d07e, but weakened from "is" to "may be" (for both notes) in be56bb7.
Even though this is non-normative, upon reflection it seems a bad idea to guarantee that either of these things is a no-op, since that implies calling these functions does absolutely nothing and/or has no overhead. That might be true of some implementations, some of the time, but there will probably still be corner cases where something has to happen (e.g., because the user requested an event, because the implementation missed an optimization opportunity, etc).
Spec: KhronosGroup/SYCL-Docs#644 In general this is a copy-paste of test_plans/oneapi/enqueue_functions.asciidoc
Co-authored-by: Nikita Kornev <[email protected]>
The convention used elsewhere is to omit sycl:: unless it's required for clarity (e.g., to distinguish between sycl:: and std:: classes and/or functions with the same name).
I had incorrectly assumed this would be covered by the "Equivalent to" wording, but we never say "Equivalent to" a queue:: shortcut, always to a handler:: function.
| ---- | ||
| _Effects (1)_: Equivalent to [code]#h.memset(ptr, value, numBytes)#. | ||
|
|
||
| _Effects (2)_: Equivalent to [code]#q.submit([&](handler& h) { memset(h, value, |
There was a problem hiding this comment.
I think we are lacking ptr as parameter. Corrected version: q.submit([&](handler& h) { memset(h, ptr, value, numBytes); });
There was a problem hiding this comment.
Good catch, thanks! Fixed in 19440ec.
This extension provides an alternative mechanism for submitting commands to a device via free-functions that require developers to opt-in to the creation of event objects.
It also proposes alternative names for several commands (e.g.,
launch) and simplifies some concepts (e.g., by removing the need for thend_rangeclass).