Cycles: oneAPI: preload kernels only when not using prebuilt binaries

sycl::build triggers compilation even if prebuilt binaries are
available, we'll have to find a better way in this case.

2 files changed, 8 insertions, 2 deletions

diff --git a/intern/cycles/kernel/CMakeLists.txt b/intern/cycles/kernel/CMakeLists.txt
index dfd21c4e675..e5d1c3d6563 100644
--- a/intern/cycles/kernel/CMakeLists.txt
+++ b/intern/cycles/kernel/CMakeLists.txt
@@ -767,6 +767,8 @@ if(WITH_CYCLES_DEVICE_ONEAPI)
   string(PREPEND CYCLES_ONEAPI_SYCL_OPTIONS_spir64_gen "-device ${CYCLES_ONEAPI_SPIR64_GEN_DEVICES} ")
 
   if (WITH_CYCLES_ONEAPI_BINARIES)
+    # AoT binaries aren't currently reused when calling sycl::build.
+    list (APPEND sycl_compiler_flags -DSYCL_SKIP_KERNELS_PRELOAD)
     # Iterate over all targest and their options
     list (JOIN CYCLES_ONEAPI_SYCL_TARGETS "," targets_string)
     list (APPEND sycl_compiler_flags -fsycl-targets=${targets_string})
diff --git a/intern/cycles/kernel/device/oneapi/kernel.cpp b/intern/cycles/kernel/device/oneapi/kernel.cpp
index d15d6509d6a..1f32d3406ea 100644
--- a/intern/cycles/kernel/device/oneapi/kernel.cpp
+++ b/intern/cycles/kernel/device/oneapi/kernel.cpp
@@ -144,6 +144,10 @@ size_t oneapi_kernel_preferred_local_size(SyclQueue *queue,
 
 bool oneapi_load_kernels(SyclQueue *queue_, const uint requested_features)
 {
+#  ifdef SYCL_SKIP_KERNELS_PRELOAD
+  (void)queue_;
+  (void)requested_features;
+#  else
   assert(queue_);
   sycl::queue *queue = reinterpret_cast<sycl::queue *>(queue_);
 
@@ -175,7 +179,7 @@ bool oneapi_load_kernels(SyclQueue *queue_, const uint requested_features)
 
       sycl::kernel_bundle<sycl::bundle_state::input> one_kernel_bundle =
           sycl::get_kernel_bundle<sycl::bundle_state::input>(queue->get_context(), {kernel_id});
-      sycl::build(one_kernel_bundle, {queue->get_device()}, sycl::property::queue::in_order());
+      sycl::build(one_kernel_bundle);
     }
   }
   catch (sycl::exception const &e) {
@@ -184,7 +188,7 @@ bool oneapi_load_kernels(SyclQueue *queue_, const uint requested_features)
     }
     return false;
   }
-
+#  endif
   return true;
 }