Feature/refactor program #845
Conversation
|
View rendered docs @ https://intelpython.github.io/dpctl/pulls/845/index.html |
oleksandr-pavlyk
force-pushed
the
feature/refactor-program
branch
2 times, most recently
from
709c97b to
5746028
Compare
Removed uses of `sycl::program` throughout the code. In SyclInterface replace DPCTLSyclProgramRef with DPCTLSyclKernelBundleRef which is reference to `sycl::kernel_bundle<sycl::bundle_state::executable>`. Functions `DPCTLProgram_*` were replaced with `DCPTLKernelBundle_*`. Functions to create program now take both context and device. Tests were modified. dpctl.SyclProgram stays with this name, but it now encapsulates DPCTLSyclKernelBundleRef instead of removed DPCTLSyclProgramRef. OpenCL functions are no longer directly used (instead of used via loader, like in the case of level-zero backend), hence removed linkage to OpenCL library.
Make sure to exercise addressof_ref() methods of SyclProgram and SyclKernel. Add an example of invalid source code to trip throwing of an exception.
|
@diptorupd As far as I can see this change show not require companion changes in numba-dpex. |
With this change, removed use of -Wno-deprecated-declarations, since code no longer uses any.
oleksandr-pavlyk
force-pushed
the
feature/refactor-program
branch
from
e9fd855 to
b4d2d8b
Compare
Added a test for non-supported backend (host). This only runs if SYCL_ENABLE_HOST_DEVICE environment variable is set.
Also parametrized test for compiling from source to run on OCL gpu and OCL cpu if found.
|
Functions Since interoperability kernel_bundle does not expose kernel ids for kernels in it, For each of this the backend-appropriate function is used to look-up the kernel by its name. SYCL does not guarantee uniqueness of the kernel with the given name in a Kernel bundles constructed from OCL source or from a single SPV are guaranteed to not have duplicates. |
There was a problem hiding this comment.
@oleksandr-pavlyk Overall it looks good. There may be a way to make the dpctl_sycl_kernel_bundle_interface.cpp more modular. E.g. having an interface and then each implementation of the interface OCL, ZE implementing the interface. But, we may soon find ourselves reinventing the DPC++ PI. So thanks, but no thanks!
Other things, I mostly glanced through so might have missed few things. If you are satisfied please feel free to merge.
Fixed typo in comment
|
Deleted rendered PR docs from intelpython.github.com/dpctl, latest should be updated shortly. 🤞 |
This change does away with use of deprecated
sycl::programclass replacing it with use ofsycl::kernel_bundle<sycl::bundle_state::executable>instead.DPCTLSyclProgramRefhas been removed, butDPCTLSyclKernelBundleRefintroduced, being an opaque reference to the executable kernel bundle as stated above.Functions
DPCTLProgram_*are renamed toDPCTLKernelBundle_*.Free functions
DPCTLKernelBundle_Create*now take context and device references (previously only tookDPCTLSyclContextRefargument).Implementation of
dpctl_sycl_program_interfaceno longer directly uses OpenCL symbols, uses dynamic_Lib helper class to load them dynamically. Thus SyclInterface library no longer links with OpenCL library.