Skip to content

Conversation

@EwanC
Copy link
Contributor

@EwanC EwanC commented Oct 15, 2024

Implement Dynamic Command-Group feature specified in PR [SYCL][Graph] Add specification for kernel binary updates. 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.

@EwanC EwanC force-pushed the ewan/dyn_cgf_impl branch from 9ddbe49 to 266d60c Compare October 15, 2024 15:25
@EwanC EwanC changed the title [SYCL][Graph] Dynamic command-groups [SYCL][Graph] Implement dynamic command-groups Oct 16, 2024
@EwanC EwanC force-pushed the ewan/dyn_cgf_impl branch from 266d60c to 70652e0 Compare October 16, 2024 12:40
@EwanC EwanC force-pushed the ewan/dyn_cgf_impl branch from 70652e0 to 6f38259 Compare October 17, 2024 08:25
@EwanC EwanC force-pushed the ewan/dyn_cgf_impl branch from 6f38259 to 62c50d3 Compare October 17, 2024 14:59
@EwanC EwanC force-pushed the ewan/dyn_cgf_impl branch from 62c50d3 to c1dfee9 Compare October 18, 2024 12:12
@EwanC EwanC force-pushed the ewan/dyn_cgf_impl branch from c1dfee9 to 1c16b5b Compare October 18, 2024 16:09
@EwanC EwanC force-pushed the ewan/dyn_cgf_impl branch from 1c16b5b to bd67c82 Compare October 21, 2024 21:06
@EwanC EwanC force-pushed the ewan/dyn_cgf_impl branch from bd67c82 to 54db889 Compare October 22, 2024 09:59
@EwanC EwanC force-pushed the ewan/dyn_cgf_impl branch from 54db889 to b8bbdd6 Compare October 22, 2024 20:48
Ewan Crawford added 3 commits October 31, 2024 15:55
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.
Create helper function for getting UR details out of CG.
@EwanC EwanC force-pushed the ewan/dyn_cgf_impl branch from 331d19b to 3737c98 Compare October 31, 2024 19:58
@EwanC EwanC marked this pull request as ready for review November 1, 2024 09:42
@EwanC EwanC requested review from a team as code owners November 1, 2024 09:42
@EwanC EwanC requested a review from sergey-semenov November 1, 2024 09:42
@EwanC
Copy link
Contributor Author

EwanC commented Nov 6, 2024

@intel/llvm-reviewers-runtime could you review this please

@EwanC
Copy link
Contributor Author

EwanC commented Nov 8, 2024

@intel/llvm-gatekeepers This is ready to merge, thanks

@martygrant martygrant merged commit 5c73a20 into intel:sycl Nov 8, 2024
14 checks passed
sommerlukas pushed a commit that referenced this pull request Dec 6, 2024
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:
* #16154
* #15700

---------

Co-authored-by: Ewan Crawford <[email protected]>
Co-authored-by: Ben Tracy <[email protected]>
Co-authored-by: Pablo Reble <[email protected]>
Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment

Labels

None yet

Projects

None yet

Development

Successfully merging this pull request may close these issues.

4 participants