[SYCL][Graph] Add specification for kernel binary updates#14896
[SYCL][Graph] Add specification for kernel binary updates#14896sommerlukas merged 7 commits intointel:syclfrom
Conversation
4637510 to
fcf08a6
Compare
Adds the kernel binary update feature to the sycl graph specification. This introduces a new dynamic_command_group class which can be used to update the command-group function of a kernel nodes in graphs.
fcf08a6 to
119ba28
Compare
Co-authored-by: Pablo Reble <pablo.reble@intel.com>
Co-authored-by: Pablo Reble <pablo.reble@intel.com>
Implement Dynamic Command-Group feature specified in PR [[SYCL][Graph] Add specification for kernel binary updates](intel#14896) This feature enables updating `ur_kernel_handle_t` objects in graph nodes between executions as well as parameters and execution range of nodes. This functionality is currently supported on CUDA & HIP which are used for testing in the new E2E tests. Level Zero support will follow shortly, resulting in the removal of the `XFAIL` labels from the E2E tests. The code for adding nodes to a graph has been refactored to split out verification of edges, and marking memory objects used in a node, as separate helper functions. This allows path for adding a command-group node to do this functions over each CG in the list before creating the node itself. The `dynamic_parameter_impl` code has also been refactored so the code is shared for updating a dynamic parameter used in both a regular kernel node and a dynamic command-group node. See the addition to the design doc for further details on the implementation.
Implement Dynamic Command-Group feature specified in PR [[SYCL][Graph] Add specification for kernel binary updates](#14896). This feature enables updating `ur_kernel_handle_t` objects in graph nodes between executions as well as parameters and execution range of nodes. Points to note in this change: * The functionality is currently supported on CUDA & HIP which are used for testing in the new E2E tests. Level Zero support will follow shortly, resulting in the removal of the `XFAIL` labels with tracker number from the E2E tests. * The code for adding nodes to a graph has been refactored to split out verification of edges, and marking memory objects used in a node, as separate helper functions. This allows path for adding a command-group node to do this functions over each CG in the list before creating the node itself. * The `dynamic_parameter_impl` code has also been refactored so the code is shared for updating a dynamic parameter used in both a regular kernel node and a dynamic command-group node. * There is now no need for the `handler::setNDRangeUsed()` API now that graph kernel nodes can update between kernels using `sycl::nd_range` and `sycl::range`. The functionality in this method has be turned into a no-op, however removing the method is an ABI breaking change, so it remains guarded by the `__INTEL_PREVIEW_BREAKING_CHANGES` macro. See the addition to the design doc for further details on the implementation.
Bensuo
left a comment
There was a problem hiding this comment.
LGTM, just some minor comments about wording.
sycl/doc/extensions/experimental/sycl_ext_oneapi_graph.asciidoc
Outdated
Show resolved
Hide resolved
sycl/doc/extensions/experimental/sycl_ext_oneapi_graph.asciidoc
Outdated
Show resolved
Hide resolved
sycl/doc/extensions/experimental/sycl_ext_oneapi_graph.asciidoc
Outdated
Show resolved
Hide resolved
hdelan
left a comment
There was a problem hiding this comment.
LGTM apart from a few questions.
sycl/doc/extensions/experimental/sycl_ext_oneapi_graph.asciidoc
Outdated
Show resolved
Hide resolved
Also refine spec wording based on PR feedback
33ae1f4 to
3edf870
Compare
|
ping @intel/llvm-reviewers-runtime for required approval, also @gmlueck to review spec language |
|
@intel/llvm-gatekeepers Can this be merged please |
Merging would associate Fabio's private mail with this commit, which is against policy:
@fabiomestre would need to update their Github profile to use the correct mail address for commits. |
After further consideration, our policy was mainly to avoid commits using the anonymous I'm merging this PR. |
Adds the kernel binary update feature to the sycl graph specification. This introduces a new dynamic_command_group class which can be used to update the command-group function of a kernel nodes in graphs.
Implemented in: