Add basic Rust launch binding

Author: szymon-zadworny

oneapi-rs-sys/src/queue.cpp

@@ -55,7 +55,7 @@ std::unique_ptr<Event> barrier(std::unique_ptr<Queue> &queue,
 
 void wait(std::unique_ptr<Queue> &queue) { queue->wait(); }
 
-std::unique_ptr<Event> launch(std::unique_ptr<Queue> queue, Kernel const &kernel, rust::Slice<rust::slice<std::uint8_t const> const> args) {
+std::unique_ptr<Event> launch(std::unique_ptr<Queue> &queue, Kernel const &kernel, rust::Slice<rust::slice<std::uint8_t const> const> args) {
   return std::make_unique<Event>(queue->submit([&](sycl::handler &cgh) {
     for (std::size_t i = 0; i < args.size(); ++i)
       cgh.set_arg(i, syclexp::raw_kernel_arg(args[i].data(), args[i].size()));

oneapi-rs/src/buffer.rs

@@ -141,9 +141,11 @@ impl<T, A: UsmAlloc> IntoFuture for EnqueuedBuffer<T, A> {
 
 unsafe impl<T: Pod, A: UsmAlloc> KernelArgument for Buffer<T, A> {
     unsafe fn as_raw_arg(&self) -> &[u8] {
+        let data_ptr: *const NonNull<_> = &self.data;
+        let cast_ptr = data_ptr as *const u8;
         unsafe {
             slice::from_raw_parts(
-                self.data.as_ptr() as *mut u8, 
+                cast_ptr, 
                 std::mem::size_of::<*mut u8>()
             )
         }

oneapi-rs/src/kernel_bundle.rs

@@ -54,3 +54,7 @@ unsafe impl<T: Pod> KernelArgument for T {
         bytemuck::bytes_of(self)
     }
 }
+
+pub unsafe trait KernelArgumentList<const ARGC: usize> {
+    unsafe fn as_raw_arg_list(&self) -> [&[u8]; ARGC];
+}

oneapi-rs/src/queue.rs

@@ -10,7 +10,7 @@ use bytemuck::Pod;
 use oneapi_rs_sys::{queue::ffi, types::ffi::EventPtr};
 
 use crate::{
-    buffer::{Buffer, EnqueuedBuffer}, context::Context, device::Device, event::Event, usm::{HostAllocator, SharedAllocator, UsmAlloc, UsmAllocator},
+    buffer::{Buffer, EnqueuedBuffer}, context::Context, device::Device, event::Event, kernel_bundle::{Kernel, KernelArgumentList}, usm::{HostAllocator, SharedAllocator, UsmAlloc, UsmAllocator},
 };
 
 /// The `Queue` connects a host program to a single device. Programs submit tasks to a device via the
@@ -127,6 +127,10 @@ impl Queue {
     pub fn wait(&mut self) {
         ffi::wait(&mut self.0);
     }
+
+    pub unsafe fn launch<const ARGC: usize>(&mut self, kernel: &Kernel, args: impl KernelArgumentList<ARGC>) -> Event {
+        unsafe { ffi::launch(&mut self.0, &kernel.0, &args.as_raw_arg_list()) }.into()
+    }
 }
 
 impl From<&Device> for Queue {