rusticl/kernel: implement clEnqueueTask

Signed-off-by: Karol Herbst <kherbst@redhat.com>
Acked-by: Alyssa Rosenzweig <alyssa.rosenzweig@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/15439>

diff --git a/src/gallium/frontends/rusticl/api/icd.rs b/src/gallium/frontends/rusticl/api/icd.rs
index 5c1b38e..2c38be3 100644 (file)
--- a/src/gallium/frontends/rusticl/api/icd.rs
+++ b/src/gallium/frontends/rusticl/api/icd.rs
@@ -82,7 +82,7 @@ pub static DISPATCH: cl_icd_dispatch = cl_icd_dispatch {
     clEnqueueMapImage: Some(cl_enqueue_map_image),
     clEnqueueUnmapMemObject: Some(cl_enqueue_unmap_mem_object),
     clEnqueueNDRangeKernel: Some(cl_enqueue_ndrange_kernel),
-    clEnqueueTask: None,
+    clEnqueueTask: Some(cl_enqueue_task),
     clEnqueueNativeKernel: None,
     clEnqueueMarker: None,
     clEnqueueWaitForEvents: None,
@@ -1105,6 +1105,22 @@ extern "C" fn cl_enqueue_ndrange_kernel(
     ))
 }
 
+extern "C" fn cl_enqueue_task(
+    command_queue: cl_command_queue,
+    kernel: cl_kernel,
+    num_events_in_wait_list: cl_uint,
+    event_wait_list: *const cl_event,
+    event: *mut cl_event,
+) -> cl_int {
+    match_err!(enqueue_task(
+        command_queue,
+        kernel,
+        num_events_in_wait_list,
+        event_wait_list,
+        event
+    ))
+}
+
 extern "C" fn cl_get_extension_function_address(
     function_name: *const ::std::os::raw::c_char,
 ) -> *mut ::std::ffi::c_void {

diff --git a/src/gallium/frontends/rusticl/api/kernel.rs b/src/gallium/frontends/rusticl/api/kernel.rs
index abc5771..f29eaf6 100644 (file)
--- a/src/gallium/frontends/rusticl/api/kernel.rs
+++ b/src/gallium/frontends/rusticl/api/kernel.rs
@@ -14,6 +14,7 @@ use self::mesa_rust_util::string::*;
 use self::rusticl_opencl_gen::*;
 
 use std::collections::HashSet;
+use std::ptr;
 use std::slice;
 use std::sync::Arc;
 
@@ -385,3 +386,26 @@ pub fn enqueue_ndrange_kernel(
     //• CL_MEM_OBJECT_ALLOCATION_FAILURE if there is a failure to allocate memory for data store associated with image or buffer objects specified as arguments to kernel.
     //• CL_INVALID_OPERATION if SVM pointers are passed as arguments to a kernel and the device does not support SVM or if system pointers are passed as arguments to a kernel and/or stored inside SVM allocations passed as kernel arguments and the device does not support fine grain system SVM allocations.
 }
+
+pub fn enqueue_task(
+    command_queue: cl_command_queue,
+    kernel: cl_kernel,
+    num_events_in_wait_list: cl_uint,
+    event_wait_list: *const cl_event,
+    event: *mut cl_event,
+) -> CLResult<()> {
+    // clEnqueueTask is equivalent to calling clEnqueueNDRangeKernel with work_dim set to 1,
+    // global_work_offset set to NULL, global_work_size[0] set to 1, and local_work_size[0] set to
+    // 1.
+    enqueue_ndrange_kernel(
+        command_queue,
+        kernel,
+        1,
+        ptr::null(),
+        [1, 0, 0].as_ptr(),
+        [1, 0, 0].as_ptr(),
+        num_events_in_wait_list,
+        event_wait_list,
+        event,
+    )
+}