Description
Describe the bug
Simple description.
After calling get_kernel_bundle(const context& ctxt, const std::vector<kernel_id>& kernelIds);, the kernel only is built on device 0, even the kernel is compatible on device 1. Launching kernel on device 1 will raise runtime error PI_ERROR_INVALID_QUEUE.
The issue was filed internally. We have got the root cause. We synced with the SYCL compiler community, and they suggested using another API as an alternative solution to fix the issue temporarily. In parallel, the SYCL compiler community will fix the issue. The purpose of the issue here is to show Intel's transparency in PyTorch community.
Problematic spot:
https://github.com/intel/llvm/blob/sycl/sycl/source/detail/program_manager/program_manager.cpp#L2508-L2513
To reproduce
#include <sycl/sycl.hpp>
class KernelName;
int main() {
sycl::platform platform;
auto devices = platform.get_devices();
std::cout << "Init devices, count: " << devices.size() << std::endl;
auto dev0 = devices[0];
auto dev1 = devices[1];
auto ctx = sycl::context({dev0, dev1});
auto q1 = sycl::queue(ctx, dev1);
sycl::kernel_id kid = sycl::get_kernel_id<KernelName>();
sycl::kernel_bundle KernelBundleExecutable =
sycl::get_kernel_bundle<sycl::bundle_state::executable>(ctx, {kid});
q1.submit([=](sycl::handler &cgh) {
cgh.use_kernel_bundle(KernelBundleExecutable);
cgh.single_task<class KernelName>([=]() {
});
});
q1.wait();
Environment
Compiler: pytorch-gpu-dev-0.5, Intel(R) oneAPI DPC+/C+ Compiler 2024.1.3 (2024.1.3.20240604)
UMD: PyTorch Prerequisites for Intel® GPUs
ii libigc-dev 1.0.16900.24-91422.04 amd64 Intel graphics co>22.04 amd64 Intel graphics co>
ii libigc-tools 1.0.16900.24-914
ii libigc1 1.0.16900.24-91422.04 amd64 Intel graphics co>22.04 amd64 Intel graphics co>
ii libigdfcl-dev 1.0.16900.24-914
ii libigdfcl1 1.0.16900.24-91422.04 amd64 Intel graphics co>22.04 amd64 Intel(R) Graphics>
ii intel-level-zero-gpu 1.3.29735.27-914
Additional context
No response