-
Notifications
You must be signed in to change notification settings - Fork 795
[SYCL][NFC] Move KernelWrapper
outside the sycl::handler
#19511
New issue
Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.
By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.
Already on GitHub? Sign in to your account
base: sycl
Are you sure you want to change the base?
Conversation
sycl/include/sycl/kernel_helper.hpp
Outdated
|
||
processProperties<detail::isKernelESIMD<KernelName>()>( | ||
KernelFunc.get(ext::oneapi::experimental::properties_tag{})); | ||
h->processLaunchProperties( |
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
I would consider removing the properties parsing from the KernelWrapper class - we could avoid passing the handler to the wrap function. Also, I feel that it might be cleaner, if the KernelWrapper is responsible just for wrapping of the kernel. I think the utility functions for property processing may still be in this header, but maybe as separate functions. Does it make sense?
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
I'll make that change in a follow up PR.
sycl/include/sycl/kernel_helper.hpp
Outdated
/// the same, thus unnecessary increasing compilation time. | ||
template <bool IsESIMDKernel, | ||
typename PropertiesT = ext::oneapi::experimental::empty_properties_t> | ||
void processProperties([[maybe_unused]] PropertiesT Props) { |
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
This function is now only doing the validation, right? Maybe it can be called "validateProperties" or something similar?
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
I'll be decoupling property processing from KernelWrapper in a subsequent PR, so I'll rename this method then.
This change looks great, thanks Udit! |
KernelWrapper
outside the sycl::handler
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
Pull Request Overview
This PR moves the KernelWrapper
struct and related kernel launch functions outside the sycl::handler
class into the detail::
namespace to support their use in no-handler submission paths. The refactoring creates a new kernel_launch_helper.hpp
header containing the moved code while updating references throughout the codebase.
- Moved
KernelWrapper
andKernelWrapperSingletonFuncs
fromhandler
class todetail::
namespace - Created new header
sycl/kernel_launch_helper.hpp
for kernel launch utilities - Updated template parameter types and function calls to accommodate the new namespace structure
Reviewed Changes
Copilot reviewed 6 out of 6 changed files in this pull request and generated 5 comments.
File | Description |
---|---|
sycl/kernel_launch_helper.hpp |
New header containing moved KernelWrapper and kernel launch functions |
sycl/handler.hpp |
Removed kernel wrapper code and updated references to use detail:: namespace |
Test files | Updated expected error locations from handler.hpp to kernel_launch_helper.hpp |
sycl_detail_core.hpp.cpp |
Updated include dependency order due to header reorganization |
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
esimd test change lgtm
sycl/include/sycl/handler.hpp
Outdated
template <WrapAs, typename, typename, typename, typename, typename, typename> | ||
friend struct detail::KernelWrapper; |
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
How much do you need? Would it be reasonable to use HandlerAccess
above instead?
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
If I understand your question correctly, the idea is to reuse KernelWrapper
for no-handler submission path, so it would be better to keep HandlerAccess
and KernelWrapper
separate.
KernelWrapper
is also used in the no-handler submission path, so moving this struct outside the handler class into the detail:: namespace. This PR also moves functions likekernel_single_task
that are called byKernelWrapper
outside thesycl::handler
class.