Function oneapi_enqueue_kernel

src/kernel/device/oneapi/kernel.cpp:329–753

}

bool oneapi_enqueue_kernel(KernelContext *kernel_context,
                           const int kernel,
                           const size_t global_size,
                           const size_t local_size,
                           const uint kernel_features,
                           bool use_hardware_raytracing,
                           void **args)
{
  bool success = true;
  ::DeviceKernel device_kernel = (::DeviceKernel)kernel;
  KernelGlobalsGPU *kg = (KernelGlobalsGPU *)kernel_context->kernel_globals;
  sycl::queue *queue = reinterpret_cast<sycl::queue *>(kernel_context->queue);
  assert(queue);
  if (!queue) {
    return false;
  }

  /* Let the compiler throw an error if there are any kernels missing in this implementation. */
#  if defined(_WIN32)
#    pragma warning(error : 4062)
#  elif defined(__GNUC__)
#    pragma GCC diagnostic push
#    pragma GCC diagnostic error "-Wswitch"
#  endif

  int max_shaders = 0;

  if (device_kernel == DEVICE_KERNEL_INTEGRATOR_SORT_BUCKET_PASS ||
      device_kernel == DEVICE_KERNEL_INTEGRATOR_SORT_WRITE_PASS)
  {
    max_shaders = (kernel_context->scene_max_shaders);
  }

  try {
    queue->submit([&](sycl::handler &cgh) {
#  ifdef WITH_EMBREE_GPU
      /* Spec says it has no effect if the called kernel doesn't support the below specialization
       * constant but it can still trigger a recompilation, so we set it only if needed. */
      if (device_kernel_has_intersection(device_kernel)) {
        const RTCFeatureFlags embree_features = use_hardware_raytracing ?
                                                    oneapi_embree_features_from_kernel_features(
                                                        kernel_features) :
                                                    RTC_FEATURE_FLAG_NONE;
        cgh.set_specialization_constant<ONEAPIKernelContext::oneapi_embree_features>(
            embree_features);
      }
#  else
      (void)kernel_features;
#  endif
      switch (device_kernel) {
        case DEVICE_KERNEL_INTEGRATOR_RESET: {
          oneapi_call(kg, cgh, global_size, local_size, args, oneapi_kernel_integrator_reset);
          break;
        }
        case DEVICE_KERNEL_INTEGRATOR_INIT_FROM_CAMERA: {
          oneapi_call(
              kg, cgh, global_size, local_size, args, oneapi_kernel_integrator_init_from_camera);
          break;