From 932ed87ebb55aa538adab2bb4be8b847dd547ec8 Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Fri, 17 Jul 2026 14:59:38 +0000 Subject: [PATCH 01/27] Add Context module --- oneapi-rs-sys/build.rs | 3 +++ oneapi-rs-sys/include/context.hpp | 25 +++++++++++++++++++++++++ oneapi-rs-sys/include/types.hpp | 2 ++ oneapi-rs-sys/src/context-sys.rs | 25 +++++++++++++++++++++++++ oneapi-rs-sys/src/context.cpp | 16 ++++++++++++++++ oneapi-rs-sys/src/lib.rs | 3 +++ oneapi-rs-sys/src/types-sys.rs | 6 ++++++ 7 files changed, 80 insertions(+) create mode 100644 oneapi-rs-sys/include/context.hpp create mode 100644 oneapi-rs-sys/src/context-sys.rs create mode 100644 oneapi-rs-sys/src/context.cpp diff --git a/oneapi-rs-sys/build.rs b/oneapi-rs-sys/build.rs index 853dffc..38d4e37 100644 --- a/oneapi-rs-sys/build.rs +++ b/oneapi-rs-sys/build.rs @@ -20,6 +20,7 @@ fn main() { "src/queue-sys.rs", "src/usm-sys.rs", "src/event-sys.rs", + "src/context-sys.rs", ]; let cpp_sources = [ @@ -28,6 +29,7 @@ fn main() { "src/queue.cpp", "src/usm.cpp", "src/event.cpp", + "src/context.cpp", ]; let cpp_headers = [ @@ -37,6 +39,7 @@ fn main() { "include/queue.hpp", "include/usm.hpp", "include/event.hpp", + "include/context.hpp", ]; cxx_build::bridges(&rust_sources) diff --git a/oneapi-rs-sys/include/context.hpp b/oneapi-rs-sys/include/context.hpp new file mode 100644 index 0000000..a436608 --- /dev/null +++ b/oneapi-rs-sys/include/context.hpp @@ -0,0 +1,25 @@ +// +// Copyright (C) 2026 Intel Corporation +// +// Under the MIT License or the Apache License v2.0. +// See LICENSE-MIT and LICENSE-APACHE for license information. +// SPDX-License-Identifier: MIT OR Apache-2.0 +// + +#pragma once + +#include "rust/cxx.h" +#include "oneapi-rs-sys/include/types.hpp" + +#include + +#include + +namespace sycl_shims { +struct DevicePtr; +struct ContextPtr; +} // namespace sycl_shims + +namespace sycl_shims::context { +std::unique_ptr new_context(rust::Vec); +} // namespace sycl_shims::context diff --git a/oneapi-rs-sys/include/types.hpp b/oneapi-rs-sys/include/types.hpp index db147b1..bb975e8 100644 --- a/oneapi-rs-sys/include/types.hpp +++ b/oneapi-rs-sys/include/types.hpp @@ -15,4 +15,6 @@ using Device = sycl::device; using Platform = sycl::platform; using Queue = sycl::queue; using Event = sycl::event; +using Context = sycl::context; } // namespace sycl_shims + diff --git a/oneapi-rs-sys/src/context-sys.rs b/oneapi-rs-sys/src/context-sys.rs new file mode 100644 index 0000000..e5f9093 --- /dev/null +++ b/oneapi-rs-sys/src/context-sys.rs @@ -0,0 +1,25 @@ +// +// Copyright (C) 2026 Intel Corporation +// +// Under the MIT License or the Apache License v2.0. +// See LICENSE-MIT and LICENSE-APACHE for license information. +// SPDX-License-Identifier: MIT OR Apache-2.0 +// + +#[cxx::bridge(namespace = "sycl_shims::context")] +pub mod ffi { + #[namespace = "sycl_shims"] + extern "C++" { + include!("oneapi-rs-sys/src/types-sys.rs.h"); + type DevicePtr = crate::types::ffi::DevicePtr; + } + + unsafe extern "C++" { + include!("oneapi-rs-sys/include/context.hpp"); + + #[namespace = "sycl_shims"] + type Context = crate::types::ffi::Context; + + fn new_context(devices: Vec) -> UniquePtr; + } +} diff --git a/oneapi-rs-sys/src/context.cpp b/oneapi-rs-sys/src/context.cpp new file mode 100644 index 0000000..9468652 --- /dev/null +++ b/oneapi-rs-sys/src/context.cpp @@ -0,0 +1,16 @@ +// +// Copyright (C) 2026 Intel Corporation +// +// Under the MIT License or the Apache License v2.0. +// See LICENSE-MIT and LICENSE-APACHE for license information. +// SPDX-License-Identifier: MIT OR Apache-2.0 +// + +#include "oneapi-rs-sys/include/context.hpp" +#include "oneapi-rs-sys/src/context-sys.rs.h" + +namespace sycl_shims::context { +std::unique_ptr new_context(rust::Vec devices) { + return std::make_unique(); +} +} // namespace sycl_shims::context diff --git a/oneapi-rs-sys/src/lib.rs b/oneapi-rs-sys/src/lib.rs index f2f723d..9b5fed2 100644 --- a/oneapi-rs-sys/src/lib.rs +++ b/oneapi-rs-sys/src/lib.rs @@ -23,3 +23,6 @@ pub mod usm; #[path = "event-sys.rs"] pub mod event; + +#[path = "context-sys.rs"] +pub mod context; diff --git a/oneapi-rs-sys/src/types-sys.rs b/oneapi-rs-sys/src/types-sys.rs index 2c9a49e..b1e5574 100644 --- a/oneapi-rs-sys/src/types-sys.rs +++ b/oneapi-rs-sys/src/types-sys.rs @@ -37,6 +37,7 @@ pub mod ffi { type Platform; type Queue; type Event; + type Context; } // This is a workaround - cxx currently doesn't support passing @@ -55,6 +56,10 @@ pub mod ffi { ptr: UniquePtr, } + struct ContextPtr { + ptr: UniquePtr + } + #[derive(Debug)] enum DeviceType { Cpu, @@ -78,6 +83,7 @@ pub mod ffi { impl UniquePtr {} impl UniquePtr {} impl UniquePtr {} + impl UniquePtr {} impl Vec {} impl Vec {} From 47891a48230e5728b9407d686455dd583ecb554a Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Fri, 17 Jul 2026 15:06:13 +0000 Subject: [PATCH 02/27] Add C++ context device ctor binding --- oneapi-rs-sys/src/context.cpp | 5 ++++- 1 file changed, 4 insertions(+), 1 deletion(-) diff --git a/oneapi-rs-sys/src/context.cpp b/oneapi-rs-sys/src/context.cpp index 9468652..4bd6c10 100644 --- a/oneapi-rs-sys/src/context.cpp +++ b/oneapi-rs-sys/src/context.cpp @@ -11,6 +11,9 @@ namespace sycl_shims::context { std::unique_ptr new_context(rust::Vec devices) { - return std::make_unique(); + std::vector raw_devices; + for (auto&& d: devices) + raw_devices.push_back(std::move(*d.ptr.release())); + return std::make_unique(raw_devices); } } // namespace sycl_shims::context From dab4d5dc08e167f04295b346cfb168201ebf61fe Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Fri, 17 Jul 2026 15:17:46 +0000 Subject: [PATCH 03/27] Add Clone impl for Device --- oneapi-rs-sys/include/device.hpp | 1 + oneapi-rs-sys/src/device-sys.rs | 1 + oneapi-rs-sys/src/device.cpp | 4 ++++ oneapi-rs/src/device.rs | 12 ++++++++++++ 4 files changed, 18 insertions(+) diff --git a/oneapi-rs-sys/include/device.hpp b/oneapi-rs-sys/include/device.hpp index 0cbbf1a..bbed562 100644 --- a/oneapi-rs-sys/include/device.hpp +++ b/oneapi-rs-sys/include/device.hpp @@ -25,4 +25,5 @@ DeviceType get_device_type(Device const &); rust::String get_version(Device const &); rust::String get_name(Device const &); std::unique_ptr get_platform(Device const &); +std::unique_ptr clone(Device const &); } // namespace sycl_shims::device diff --git a/oneapi-rs-sys/src/device-sys.rs b/oneapi-rs-sys/src/device-sys.rs index f272172..9c460ea 100644 --- a/oneapi-rs-sys/src/device-sys.rs +++ b/oneapi-rs-sys/src/device-sys.rs @@ -31,5 +31,6 @@ pub mod ffi { fn get_device_type(device: &Device) -> DeviceType; fn get_version(device: &Device) -> String; fn get_name(device: &Device) -> String; + fn clone(device: &Device) -> UniquePtr; } } diff --git a/oneapi-rs-sys/src/device.cpp b/oneapi-rs-sys/src/device.cpp index 6d87359..64d287c 100644 --- a/oneapi-rs-sys/src/device.cpp +++ b/oneapi-rs-sys/src/device.cpp @@ -53,4 +53,8 @@ rust::String get_name(Device const &device) { std::unique_ptr get_platform(Device const &device) { return std::make_unique(device.get_platform()); } + +std::unique_ptr clone(Device const & device) { + return std::make_unique(sycl::device(device)); +} } // namespace sycl_shims::device diff --git a/oneapi-rs/src/device.rs b/oneapi-rs/src/device.rs index cee1aac..8560062 100644 --- a/oneapi-rs/src/device.rs +++ b/oneapi-rs/src/device.rs @@ -15,6 +15,12 @@ use crate::{info::device::DeviceInfo, platform::Platform}; /// The `Device` struct provides the common reference semantics. pub struct Device(pub(crate) cxx::UniquePtr); +impl From> for Device { + fn from(value: cxx::UniquePtr) -> Self { + Self(value) + } +} + impl Device { /// Returns a [`Vec`] containing all the root devices from all SYCL backends /// available in the system which have the device type encapsulated by [`DeviceType`](crate::info::DeviceType). @@ -39,3 +45,9 @@ impl Device { Platform(raw_platform) } } + +impl Clone for Device { + fn clone(&self) -> Self { + ffi::clone(&self.0).into() + } +} From 03e29765e5820171c3d8aaa47d8f1205274df3e2 Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Fri, 17 Jul 2026 15:18:07 +0000 Subject: [PATCH 04/27] Add Rust Context binding --- oneapi-rs/src/context.rs | 32 ++++++++++++++++++++++++++++++++ oneapi-rs/src/lib.rs | 2 ++ 2 files changed, 34 insertions(+) create mode 100644 oneapi-rs/src/context.rs diff --git a/oneapi-rs/src/context.rs b/oneapi-rs/src/context.rs new file mode 100644 index 0000000..0b0cf82 --- /dev/null +++ b/oneapi-rs/src/context.rs @@ -0,0 +1,32 @@ +// +// Copyright (C) 2026 Intel Corporation +// +// Under the MIT License or the Apache License v2.0. +// See LICENSE-MIT and LICENSE-APACHE for license information. +// SPDX-License-Identifier: MIT OR Apache-2.0 +// + +use oneapi_rs_sys::{context::ffi, types::ffi::DevicePtr}; + +use crate::device::Device; + +/// A context represents the runtime data structures and state required by a SYCL backend API +/// to interact with a group of devices associated with a platform. +pub struct Context(pub(crate) cxx::UniquePtr); + +impl From> for Context { + fn from(value: cxx::UniquePtr) -> Self { + Self(value) + } +} + +impl Context { + pub fn new(devices: &[&Device]) -> Self { + let devices = devices + .iter() + .map(|d| DevicePtr { ptr: (*d).clone().0 }) + .collect::>(); + + ffi::new_context(devices).into() + } +} diff --git a/oneapi-rs/src/lib.rs b/oneapi-rs/src/lib.rs index 922b460..5c8f4fb 100644 --- a/oneapi-rs/src/lib.rs +++ b/oneapi-rs/src/lib.rs @@ -13,3 +13,5 @@ pub mod info; pub mod platform; pub mod queue; pub mod usm; +pub mod context; + From 4bfb3976b6220b62e985d6a0d90b638a81ad2afb Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Wed, 22 Jul 2026 08:43:28 +0000 Subject: [PATCH 05/27] Add C++ queue::get_context binding --- oneapi-rs-sys/include/queue.hpp | 1 + oneapi-rs-sys/src/queue-sys.rs | 3 +++ oneapi-rs-sys/src/queue.cpp | 4 ++++ 3 files changed, 8 insertions(+) diff --git a/oneapi-rs-sys/include/queue.hpp b/oneapi-rs-sys/include/queue.hpp index c8fa997..72448e8 100644 --- a/oneapi-rs-sys/include/queue.hpp +++ b/oneapi-rs-sys/include/queue.hpp @@ -21,6 +21,7 @@ namespace sycl_shims::queue { std::unique_ptr new_queue(); std::unique_ptr new_queue_immediate(); std::unique_ptr new_queue_from_device(Device const &); +std::unique_ptr get_context(Queue const &); std::unique_ptr clone(Queue const &); std::unique_ptr memset(std::unique_ptr &, std::uint8_t *ptr, int value, std::size_t num_bytes, diff --git a/oneapi-rs-sys/src/queue-sys.rs b/oneapi-rs-sys/src/queue-sys.rs index f79c6f5..a20236b 100644 --- a/oneapi-rs-sys/src/queue-sys.rs +++ b/oneapi-rs-sys/src/queue-sys.rs @@ -22,11 +22,14 @@ pub mod ffi { #[namespace = "sycl_shims"] type Device = crate::types::ffi::Device; #[namespace = "sycl_shims"] + type Context = crate::types::ffi::Context; + #[namespace = "sycl_shims"] type Event = crate::types::ffi::Event; fn new_queue() -> UniquePtr; fn new_queue_immediate() -> UniquePtr; fn new_queue_from_device(device: &Device) -> UniquePtr; + fn get_context(queue: &Queue) -> UniquePtr; fn clone(queue: &Queue) -> UniquePtr; unsafe fn memset( queue: &mut UniquePtr, diff --git a/oneapi-rs-sys/src/queue.cpp b/oneapi-rs-sys/src/queue.cpp index 3209cd1..7da9fc6 100644 --- a/oneapi-rs-sys/src/queue.cpp +++ b/oneapi-rs-sys/src/queue.cpp @@ -26,6 +26,10 @@ std::unique_ptr new_queue_from_device(Device const &device) { return std::make_unique(sycl::queue(device, {in_order()})); } +std::unique_ptr get_context(Queue const &queue) { + return std::make_unique(queue.get_context()); +} + std::unique_ptr clone(Queue const &queue) { return std::make_unique(sycl::queue(queue)); } From d7599ec1a688a3bd707bd88dff5b72b285057b81 Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Wed, 22 Jul 2026 08:46:38 +0000 Subject: [PATCH 06/27] Add Rust Queue::get_context binding --- oneapi-rs/src/queue.rs | 10 ++++++---- 1 file changed, 6 insertions(+), 4 deletions(-) diff --git a/oneapi-rs/src/queue.rs b/oneapi-rs/src/queue.rs index 70ba4fc..fa542b3 100644 --- a/oneapi-rs/src/queue.rs +++ b/oneapi-rs/src/queue.rs @@ -10,10 +10,7 @@ use bytemuck::Pod; use oneapi_rs_sys::{queue::ffi, types::ffi::EventPtr}; use crate::{ - buffer::{Buffer, EnqueuedBuffer}, - device::Device, - event::Event, - usm::{HostAllocator, SharedAllocator, UsmAlloc, UsmAllocator}, + buffer::{Buffer, EnqueuedBuffer}, context::Context, device::Device, event::Event, usm::{HostAllocator, SharedAllocator, UsmAlloc, UsmAllocator}, }; /// The `Queue` connects a host program to a single device. Programs submit tasks to a device via the @@ -32,6 +29,11 @@ impl Queue { Self(ffi::new_queue_immediate()) } + /// Returns the SYCL queue’s context. + pub fn get_context(&self) -> Context { + ffi::get_context(&self.0).into() + } + /// Allocates zeroed memory and creates a host-side [`Buffer`] that can store an array of T. pub fn alloc_host( &mut self, From 337fca0a863cd86923ea62c846e5c3890871514f Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Wed, 22 Jul 2026 09:07:23 +0000 Subject: [PATCH 07/27] Add C++ syclexp::create_kernel_bundle_from_source binding --- oneapi-rs-sys/include/context.hpp | 2 ++ oneapi-rs-sys/include/types.hpp | 1 + oneapi-rs-sys/src/context-sys.rs | 5 +++++ oneapi-rs-sys/src/context.cpp | 11 +++++++++++ oneapi-rs-sys/src/types-sys.rs | 6 ++---- 5 files changed, 21 insertions(+), 4 deletions(-) diff --git a/oneapi-rs-sys/include/context.hpp b/oneapi-rs-sys/include/context.hpp index a436608..d0d5ff8 100644 --- a/oneapi-rs-sys/include/context.hpp +++ b/oneapi-rs-sys/include/context.hpp @@ -22,4 +22,6 @@ struct ContextPtr; namespace sycl_shims::context { std::unique_ptr new_context(rust::Vec); +std::unique_ptr create_kernel_bundle_from_source(Context const &ctxt, + rust::Str source); } // namespace sycl_shims::context diff --git a/oneapi-rs-sys/include/types.hpp b/oneapi-rs-sys/include/types.hpp index bb975e8..fcfa37c 100644 --- a/oneapi-rs-sys/include/types.hpp +++ b/oneapi-rs-sys/include/types.hpp @@ -16,5 +16,6 @@ using Platform = sycl::platform; using Queue = sycl::queue; using Event = sycl::event; using Context = sycl::context; +using SourceKernelBundle = sycl::kernel_bundle; } // namespace sycl_shims diff --git a/oneapi-rs-sys/src/context-sys.rs b/oneapi-rs-sys/src/context-sys.rs index e5f9093..114861f 100644 --- a/oneapi-rs-sys/src/context-sys.rs +++ b/oneapi-rs-sys/src/context-sys.rs @@ -20,6 +20,11 @@ pub mod ffi { #[namespace = "sycl_shims"] type Context = crate::types::ffi::Context; + #[namespace = "sycl_shims"] + type SourceKernelBundle = crate::types::ffi::SourceKernelBundle; + fn new_context(devices: Vec) -> UniquePtr; + fn create_kernel_bundle_from_source(ctxt: &Context, source: &str) + -> UniquePtr; } } diff --git a/oneapi-rs-sys/src/context.cpp b/oneapi-rs-sys/src/context.cpp index 4bd6c10..6602104 100644 --- a/oneapi-rs-sys/src/context.cpp +++ b/oneapi-rs-sys/src/context.cpp @@ -9,6 +9,8 @@ #include "oneapi-rs-sys/include/context.hpp" #include "oneapi-rs-sys/src/context-sys.rs.h" +namespace syclexp = sycl::ext::oneapi::experimental; + namespace sycl_shims::context { std::unique_ptr new_context(rust::Vec devices) { std::vector raw_devices; @@ -16,4 +18,13 @@ std::unique_ptr new_context(rust::Vec devices) { raw_devices.push_back(std::move(*d.ptr.release())); return std::make_unique(raw_devices); } + +std::unique_ptr create_kernel_bundle_from_source(Context const &ctxt, + rust::Str source) { + return std::make_unique(syclexp::create_kernel_bundle_from_source( + ctxt, + syclexp::source_language::sycl, + std::string(source) + )); +} } // namespace sycl_shims::context diff --git a/oneapi-rs-sys/src/types-sys.rs b/oneapi-rs-sys/src/types-sys.rs index b1e5574..d8ffb29 100644 --- a/oneapi-rs-sys/src/types-sys.rs +++ b/oneapi-rs-sys/src/types-sys.rs @@ -38,6 +38,7 @@ pub mod ffi { type Queue; type Event; type Context; + type SourceKernelBundle; } // This is a workaround - cxx currently doesn't support passing @@ -56,10 +57,6 @@ pub mod ffi { ptr: UniquePtr, } - struct ContextPtr { - ptr: UniquePtr - } - #[derive(Debug)] enum DeviceType { Cpu, @@ -84,6 +81,7 @@ pub mod ffi { impl UniquePtr {} impl UniquePtr {} impl UniquePtr {} + impl UniquePtr {} impl Vec {} impl Vec {} From 166f3e2e89a869c2c893a780a5a91e54bc07a361 Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Wed, 22 Jul 2026 09:13:33 +0000 Subject: [PATCH 08/27] Add Rust SourceKernelBundle binding --- oneapi-rs/src/kernel_bundle.rs | 17 +++++++++++++++++ oneapi-rs/src/lib.rs | 1 + 2 files changed, 18 insertions(+) create mode 100644 oneapi-rs/src/kernel_bundle.rs diff --git a/oneapi-rs/src/kernel_bundle.rs b/oneapi-rs/src/kernel_bundle.rs new file mode 100644 index 0000000..e157650 --- /dev/null +++ b/oneapi-rs/src/kernel_bundle.rs @@ -0,0 +1,17 @@ +// +// Copyright (C) 2026 Intel Corporation +// +// Under the MIT License or the Apache License v2.0. +// See LICENSE-MIT and LICENSE-APACHE for license information. +// SPDX-License-Identifier: MIT OR Apache-2.0 +// + +use oneapi_rs_sys::{context::ffi, types}; + +pub struct SourceKernelBundle(pub(crate) cxx::UniquePtr); + +impl From> for SourceKernelBundle { + fn from(value: cxx::UniquePtr) -> Self { + Self(value) + } +} diff --git a/oneapi-rs/src/lib.rs b/oneapi-rs/src/lib.rs index 5c8f4fb..e2a7fc6 100644 --- a/oneapi-rs/src/lib.rs +++ b/oneapi-rs/src/lib.rs @@ -14,4 +14,5 @@ pub mod platform; pub mod queue; pub mod usm; pub mod context; +pub mod kernel_bundle; From 3764d0d4202fbf256c2846dccaad5dfe9fd50b02 Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Wed, 22 Jul 2026 09:16:01 +0000 Subject: [PATCH 09/27] Add Rust Context::create_kernel_bundle_from_source binding --- oneapi-rs/src/context.rs | 6 +++++- 1 file changed, 5 insertions(+), 1 deletion(-) diff --git a/oneapi-rs/src/context.rs b/oneapi-rs/src/context.rs index 0b0cf82..ec6b998 100644 --- a/oneapi-rs/src/context.rs +++ b/oneapi-rs/src/context.rs @@ -8,7 +8,7 @@ use oneapi_rs_sys::{context::ffi, types::ffi::DevicePtr}; -use crate::device::Device; +use crate::{device::Device, kernel_bundle::SourceKernelBundle}; /// A context represents the runtime data structures and state required by a SYCL backend API /// to interact with a group of devices associated with a platform. @@ -29,4 +29,8 @@ impl Context { ffi::new_context(devices).into() } + + pub fn create_kernel_bundle_from_source(&self, source: &str) -> SourceKernelBundle { + ffi::create_kernel_bundle_from_source(&self.0, source).into() + } } From 0129df67ce4cb960a33bf9b80319f8412d52187c Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Wed, 22 Jul 2026 09:33:19 +0000 Subject: [PATCH 10/27] Add kernel bundle module --- oneapi-rs-sys/build.rs | 3 +++ oneapi-rs-sys/include/context.hpp | 9 +++------ oneapi-rs-sys/include/kernel-bundle.hpp | 21 +++++++++++++++++++++ oneapi-rs-sys/src/context-sys.rs | 5 ----- oneapi-rs-sys/src/context.cpp | 9 --------- oneapi-rs-sys/src/kernel-bundle-sys.rs | 22 ++++++++++++++++++++++ oneapi-rs-sys/src/kernel-bundle.cpp | 23 +++++++++++++++++++++++ oneapi-rs-sys/src/lib.rs | 4 ++++ oneapi-rs/src/context.rs | 4 ++-- oneapi-rs/src/kernel_bundle.rs | 2 +- 10 files changed, 79 insertions(+), 23 deletions(-) create mode 100644 oneapi-rs-sys/include/kernel-bundle.hpp create mode 100644 oneapi-rs-sys/src/kernel-bundle-sys.rs create mode 100644 oneapi-rs-sys/src/kernel-bundle.cpp diff --git a/oneapi-rs-sys/build.rs b/oneapi-rs-sys/build.rs index 38d4e37..854c5e4 100644 --- a/oneapi-rs-sys/build.rs +++ b/oneapi-rs-sys/build.rs @@ -21,6 +21,7 @@ fn main() { "src/usm-sys.rs", "src/event-sys.rs", "src/context-sys.rs", + "src/kernel-bundle-sys.rs", ]; let cpp_sources = [ @@ -30,6 +31,7 @@ fn main() { "src/usm.cpp", "src/event.cpp", "src/context.cpp", + "src/kernel-bundle.cpp", ]; let cpp_headers = [ @@ -40,6 +42,7 @@ fn main() { "include/usm.hpp", "include/event.hpp", "include/context.hpp", + "include/kernel-bundle.hpp", ]; cxx_build::bridges(&rust_sources) diff --git a/oneapi-rs-sys/include/context.hpp b/oneapi-rs-sys/include/context.hpp index d0d5ff8..1d8f608 100644 --- a/oneapi-rs-sys/include/context.hpp +++ b/oneapi-rs-sys/include/context.hpp @@ -8,20 +8,17 @@ #pragma once -#include "rust/cxx.h" -#include "oneapi-rs-sys/include/types.hpp" +#include #include -#include +#include "rust/cxx.h" +#include "oneapi-rs-sys/include/types.hpp" namespace sycl_shims { struct DevicePtr; -struct ContextPtr; } // namespace sycl_shims namespace sycl_shims::context { std::unique_ptr new_context(rust::Vec); -std::unique_ptr create_kernel_bundle_from_source(Context const &ctxt, - rust::Str source); } // namespace sycl_shims::context diff --git a/oneapi-rs-sys/include/kernel-bundle.hpp b/oneapi-rs-sys/include/kernel-bundle.hpp new file mode 100644 index 0000000..eecac90 --- /dev/null +++ b/oneapi-rs-sys/include/kernel-bundle.hpp @@ -0,0 +1,21 @@ +// +// Copyright (C) 2026 Intel Corporation +// +// Under the MIT License or the Apache License v2.0. +// See LICENSE-MIT and LICENSE-APACHE for license information. +// SPDX-License-Identifier: MIT OR Apache-2.0 +// + +#pragma once + +#include + +#include + +#include "rust/cxx.h" +#include "oneapi-rs-sys/include/types.hpp" + +namespace sycl_shims::kernel_bundle { +std::unique_ptr create_kernel_bundle_from_source(Context const &ctxt, + rust::Str source); +} // namespace sycl_shims::kernel_bundle diff --git a/oneapi-rs-sys/src/context-sys.rs b/oneapi-rs-sys/src/context-sys.rs index 114861f..e5f9093 100644 --- a/oneapi-rs-sys/src/context-sys.rs +++ b/oneapi-rs-sys/src/context-sys.rs @@ -20,11 +20,6 @@ pub mod ffi { #[namespace = "sycl_shims"] type Context = crate::types::ffi::Context; - #[namespace = "sycl_shims"] - type SourceKernelBundle = crate::types::ffi::SourceKernelBundle; - fn new_context(devices: Vec) -> UniquePtr; - fn create_kernel_bundle_from_source(ctxt: &Context, source: &str) - -> UniquePtr; } } diff --git a/oneapi-rs-sys/src/context.cpp b/oneapi-rs-sys/src/context.cpp index 6602104..a530deb 100644 --- a/oneapi-rs-sys/src/context.cpp +++ b/oneapi-rs-sys/src/context.cpp @@ -18,13 +18,4 @@ std::unique_ptr new_context(rust::Vec devices) { raw_devices.push_back(std::move(*d.ptr.release())); return std::make_unique(raw_devices); } - -std::unique_ptr create_kernel_bundle_from_source(Context const &ctxt, - rust::Str source) { - return std::make_unique(syclexp::create_kernel_bundle_from_source( - ctxt, - syclexp::source_language::sycl, - std::string(source) - )); -} } // namespace sycl_shims::context diff --git a/oneapi-rs-sys/src/kernel-bundle-sys.rs b/oneapi-rs-sys/src/kernel-bundle-sys.rs new file mode 100644 index 0000000..4b00edc --- /dev/null +++ b/oneapi-rs-sys/src/kernel-bundle-sys.rs @@ -0,0 +1,22 @@ +// +// Copyright (C) 2026 Intel Corporation +// +// Under the MIT License or the Apache License v2.0. +// See LICENSE-MIT and LICENSE-APACHE for license information. +// SPDX-License-Identifier: MIT OR Apache-2.0 +// + +#[cxx::bridge(namespace = "sycl_shims::kernel_bundle")] +pub mod ffi { + unsafe extern "C++" { + include!("oneapi-rs-sys/include/kernel-bundle.hpp"); + + #[namespace = "sycl_shims"] + type Context = crate::types::ffi::Context; + + #[namespace = "sycl_shims"] + type SourceKernelBundle = crate::types::ffi::SourceKernelBundle; + fn create_kernel_bundle_from_source(ctxt: &Context, source: &str) + -> UniquePtr; + } +} diff --git a/oneapi-rs-sys/src/kernel-bundle.cpp b/oneapi-rs-sys/src/kernel-bundle.cpp new file mode 100644 index 0000000..fb9a58f --- /dev/null +++ b/oneapi-rs-sys/src/kernel-bundle.cpp @@ -0,0 +1,23 @@ +// +// Copyright (C) 2026 Intel Corporation +// +// Under the MIT License or the Apache License v2.0. +// See LICENSE-MIT and LICENSE-APACHE for license information. +// SPDX-License-Identifier: MIT OR Apache-2.0 +// + +#include "oneapi-rs-sys/include/kernel-bundle.hpp" +#include "oneapi-rs-sys/src/kernel-bundle-sys.rs.h" + +namespace syclexp = sycl::ext::oneapi::experimental; + +namespace sycl_shims::kernel_bundle { +std::unique_ptr create_kernel_bundle_from_source(Context const &ctxt, + rust::Str source) { + return std::make_unique(syclexp::create_kernel_bundle_from_source( + ctxt, + syclexp::source_language::sycl, + std::string(source) + )); +} +} // namespace sycl_shims::kernel_bundle diff --git a/oneapi-rs-sys/src/lib.rs b/oneapi-rs-sys/src/lib.rs index 9b5fed2..5945819 100644 --- a/oneapi-rs-sys/src/lib.rs +++ b/oneapi-rs-sys/src/lib.rs @@ -26,3 +26,7 @@ pub mod event; #[path = "context-sys.rs"] pub mod context; + +#[path = "kernel-bundle-sys.rs"] +pub mod kernel_bundle; + diff --git a/oneapi-rs/src/context.rs b/oneapi-rs/src/context.rs index ec6b998..36e78f7 100644 --- a/oneapi-rs/src/context.rs +++ b/oneapi-rs/src/context.rs @@ -6,7 +6,7 @@ // SPDX-License-Identifier: MIT OR Apache-2.0 // -use oneapi_rs_sys::{context::ffi, types::ffi::DevicePtr}; +use oneapi_rs_sys::{kernel_bundle, context::ffi, types::ffi::DevicePtr}; use crate::{device::Device, kernel_bundle::SourceKernelBundle}; @@ -31,6 +31,6 @@ impl Context { } pub fn create_kernel_bundle_from_source(&self, source: &str) -> SourceKernelBundle { - ffi::create_kernel_bundle_from_source(&self.0, source).into() + kernel_bundle::ffi::create_kernel_bundle_from_source(&self.0, source).into() } } diff --git a/oneapi-rs/src/kernel_bundle.rs b/oneapi-rs/src/kernel_bundle.rs index e157650..bdc4b08 100644 --- a/oneapi-rs/src/kernel_bundle.rs +++ b/oneapi-rs/src/kernel_bundle.rs @@ -6,7 +6,7 @@ // SPDX-License-Identifier: MIT OR Apache-2.0 // -use oneapi_rs_sys::{context::ffi, types}; +use oneapi_rs_sys::types; pub struct SourceKernelBundle(pub(crate) cxx::UniquePtr); From 43426071c66871653a79643b431a387b56d13cdf Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Wed, 22 Jul 2026 09:43:38 +0000 Subject: [PATCH 11/27] Add C++ syclexp::build binding --- oneapi-rs-sys/include/kernel-bundle.hpp | 2 ++ oneapi-rs-sys/include/types.hpp | 1 + oneapi-rs-sys/src/context.cpp | 2 -- oneapi-rs-sys/src/kernel-bundle-sys.rs | 6 ++++++ oneapi-rs-sys/src/kernel-bundle.cpp | 4 ++++ oneapi-rs-sys/src/types-sys.rs | 2 ++ 6 files changed, 15 insertions(+), 2 deletions(-) diff --git a/oneapi-rs-sys/include/kernel-bundle.hpp b/oneapi-rs-sys/include/kernel-bundle.hpp index eecac90..7666c56 100644 --- a/oneapi-rs-sys/include/kernel-bundle.hpp +++ b/oneapi-rs-sys/include/kernel-bundle.hpp @@ -18,4 +18,6 @@ namespace sycl_shims::kernel_bundle { std::unique_ptr create_kernel_bundle_from_source(Context const &ctxt, rust::Str source); + +std::unique_ptr build(std::unique_ptr &source); } // namespace sycl_shims::kernel_bundle diff --git a/oneapi-rs-sys/include/types.hpp b/oneapi-rs-sys/include/types.hpp index fcfa37c..a15269e 100644 --- a/oneapi-rs-sys/include/types.hpp +++ b/oneapi-rs-sys/include/types.hpp @@ -17,5 +17,6 @@ using Queue = sycl::queue; using Event = sycl::event; using Context = sycl::context; using SourceKernelBundle = sycl::kernel_bundle; +using ExecutableKernelBundle = sycl::kernel_bundle; } // namespace sycl_shims diff --git a/oneapi-rs-sys/src/context.cpp b/oneapi-rs-sys/src/context.cpp index a530deb..4bd6c10 100644 --- a/oneapi-rs-sys/src/context.cpp +++ b/oneapi-rs-sys/src/context.cpp @@ -9,8 +9,6 @@ #include "oneapi-rs-sys/include/context.hpp" #include "oneapi-rs-sys/src/context-sys.rs.h" -namespace syclexp = sycl::ext::oneapi::experimental; - namespace sycl_shims::context { std::unique_ptr new_context(rust::Vec devices) { std::vector raw_devices; diff --git a/oneapi-rs-sys/src/kernel-bundle-sys.rs b/oneapi-rs-sys/src/kernel-bundle-sys.rs index 4b00edc..51a48e1 100644 --- a/oneapi-rs-sys/src/kernel-bundle-sys.rs +++ b/oneapi-rs-sys/src/kernel-bundle-sys.rs @@ -16,7 +16,13 @@ pub mod ffi { #[namespace = "sycl_shims"] type SourceKernelBundle = crate::types::ffi::SourceKernelBundle; + + #[namespace = "sycl_shims"] + type ExecutableKernelBundle = crate::types::ffi::ExecutableKernelBundle; + fn create_kernel_bundle_from_source(ctxt: &Context, source: &str) -> UniquePtr; + + fn build(source: &mut UniquePtr) -> UniquePtr; } } diff --git a/oneapi-rs-sys/src/kernel-bundle.cpp b/oneapi-rs-sys/src/kernel-bundle.cpp index fb9a58f..78d4c73 100644 --- a/oneapi-rs-sys/src/kernel-bundle.cpp +++ b/oneapi-rs-sys/src/kernel-bundle.cpp @@ -20,4 +20,8 @@ std::unique_ptr create_kernel_bundle_from_source(Context con std::string(source) )); } + +std::unique_ptr build(std::unique_ptr &source) { + return std::make_unique(syclexp::build(*source)); +} } // namespace sycl_shims::kernel_bundle diff --git a/oneapi-rs-sys/src/types-sys.rs b/oneapi-rs-sys/src/types-sys.rs index d8ffb29..308cbaa 100644 --- a/oneapi-rs-sys/src/types-sys.rs +++ b/oneapi-rs-sys/src/types-sys.rs @@ -39,6 +39,7 @@ pub mod ffi { type Event; type Context; type SourceKernelBundle; + type ExecutableKernelBundle; } // This is a workaround - cxx currently doesn't support passing @@ -82,6 +83,7 @@ pub mod ffi { impl UniquePtr {} impl UniquePtr {} impl UniquePtr {} + impl UniquePtr {} impl Vec {} impl Vec {} From 24b63f575710612fa485ad2dff11858f8da7401a Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Wed, 22 Jul 2026 09:47:32 +0000 Subject: [PATCH 12/27] Add Rust SourceKernelBundle::build binding --- oneapi-rs/src/kernel_bundle.rs | 16 +++++++++++++++- 1 file changed, 15 insertions(+), 1 deletion(-) diff --git a/oneapi-rs/src/kernel_bundle.rs b/oneapi-rs/src/kernel_bundle.rs index bdc4b08..093cbbc 100644 --- a/oneapi-rs/src/kernel_bundle.rs +++ b/oneapi-rs/src/kernel_bundle.rs @@ -6,7 +6,7 @@ // SPDX-License-Identifier: MIT OR Apache-2.0 // -use oneapi_rs_sys::types; +use oneapi_rs_sys::{kernel_bundle::ffi, types}; pub struct SourceKernelBundle(pub(crate) cxx::UniquePtr); @@ -15,3 +15,17 @@ impl From> for SourceKernelBundle Self(value) } } + +impl SourceKernelBundle { + pub fn build(&mut self) -> ExecutableKernelBundle { + ffi::build(&mut self.0).into() + } +} + +pub struct ExecutableKernelBundle(pub(crate) cxx::UniquePtr); + +impl From> for ExecutableKernelBundle { + fn from(value: cxx::UniquePtr) -> Self { + Self(value) + } +} From 7c5896150e27a8f1ae14d8577f4ef4979b4cd648 Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Wed, 22 Jul 2026 10:49:28 +0000 Subject: [PATCH 13/27] Add kernel argument trait --- oneapi-rs/src/buffer.rs | 15 +++++++++++++-- oneapi-rs/src/kernel_bundle.rs | 11 +++++++++++ 2 files changed, 24 insertions(+), 2 deletions(-) diff --git a/oneapi-rs/src/buffer.rs b/oneapi-rs/src/buffer.rs index 32d2a44..048fc7c 100644 --- a/oneapi-rs/src/buffer.rs +++ b/oneapi-rs/src/buffer.rs @@ -15,11 +15,11 @@ use std::{ task::{Context, Poll}, }; +use bytemuck::Pod; use pin_project::pin_project; use crate::{ - event::{Event, EventFuture}, - usm::UsmAlloc, + event::{Event, EventFuture}, kernel_bundle::KernelArgument, usm::UsmAlloc, }; /// The Buffer struct defines a shared array of one, two or three dimensions that can be used @@ -138,3 +138,14 @@ impl IntoFuture for EnqueuedBuffer { } } } + +unsafe impl KernelArgument for Buffer { + unsafe fn as_raw_arg(&self) -> &[u8] { + unsafe { + slice::from_raw_parts( + self.data.as_ptr() as *mut u8, + std::mem::size_of::<*mut u8>() + ) + } + } +} diff --git a/oneapi-rs/src/kernel_bundle.rs b/oneapi-rs/src/kernel_bundle.rs index 093cbbc..7f44130 100644 --- a/oneapi-rs/src/kernel_bundle.rs +++ b/oneapi-rs/src/kernel_bundle.rs @@ -6,6 +6,7 @@ // SPDX-License-Identifier: MIT OR Apache-2.0 // +use bytemuck::Pod; use oneapi_rs_sys::{kernel_bundle::ffi, types}; pub struct SourceKernelBundle(pub(crate) cxx::UniquePtr); @@ -29,3 +30,13 @@ impl From> for ExecutableKern Self(value) } } + +pub unsafe trait KernelArgument { + unsafe fn as_raw_arg(&self) -> &[u8]; +} + +unsafe impl KernelArgument for T { + unsafe fn as_raw_arg(&self) -> &[u8] { + bytemuck::bytes_of(self) + } +} From 8259982095a3fd741324b031301b268816dfab6d Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Thu, 23 Jul 2026 09:50:43 +0000 Subject: [PATCH 14/27] Add C++ sycl::kernel type binding --- oneapi-rs-sys/include/types.hpp | 1 + oneapi-rs-sys/src/queue-sys.rs | 2 ++ oneapi-rs-sys/src/types-sys.rs | 1 + 3 files changed, 4 insertions(+) diff --git a/oneapi-rs-sys/include/types.hpp b/oneapi-rs-sys/include/types.hpp index a15269e..edec193 100644 --- a/oneapi-rs-sys/include/types.hpp +++ b/oneapi-rs-sys/include/types.hpp @@ -16,6 +16,7 @@ using Platform = sycl::platform; using Queue = sycl::queue; using Event = sycl::event; using Context = sycl::context; +using Kernel = sycl::kernel; using SourceKernelBundle = sycl::kernel_bundle; using ExecutableKernelBundle = sycl::kernel_bundle; } // namespace sycl_shims diff --git a/oneapi-rs-sys/src/queue-sys.rs b/oneapi-rs-sys/src/queue-sys.rs index a20236b..8a1f442 100644 --- a/oneapi-rs-sys/src/queue-sys.rs +++ b/oneapi-rs-sys/src/queue-sys.rs @@ -25,6 +25,8 @@ pub mod ffi { type Context = crate::types::ffi::Context; #[namespace = "sycl_shims"] type Event = crate::types::ffi::Event; + #[namespace = "sycl_shims"] + type Kernel = crate::types::ffi::Kernel; fn new_queue() -> UniquePtr; fn new_queue_immediate() -> UniquePtr; diff --git a/oneapi-rs-sys/src/types-sys.rs b/oneapi-rs-sys/src/types-sys.rs index 308cbaa..a577254 100644 --- a/oneapi-rs-sys/src/types-sys.rs +++ b/oneapi-rs-sys/src/types-sys.rs @@ -38,6 +38,7 @@ pub mod ffi { type Queue; type Event; type Context; + type Kernel; type SourceKernelBundle; type ExecutableKernelBundle; } From dfebe00f9d63166337bcdad4191c0132f518c980 Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Thu, 23 Jul 2026 09:52:19 +0000 Subject: [PATCH 15/27] Add basic C++ nd_launch implementation --- oneapi-rs-sys/include/queue.hpp | 1 + oneapi-rs-sys/src/queue-sys.rs | 1 + oneapi-rs-sys/src/queue.cpp | 11 +++++++++++ 3 files changed, 13 insertions(+) diff --git a/oneapi-rs-sys/include/queue.hpp b/oneapi-rs-sys/include/queue.hpp index 72448e8..ffc8288 100644 --- a/oneapi-rs-sys/include/queue.hpp +++ b/oneapi-rs-sys/include/queue.hpp @@ -28,4 +28,5 @@ std::unique_ptr memset(std::unique_ptr &, std::uint8_t *ptr, rust::Vec); std::unique_ptr barrier(std::unique_ptr &, rust::Vec); void wait(std::unique_ptr &); +std::unique_ptr launch(std::unique_ptr &, Kernel const &, rust::Slice const> args); } // namespace sycl_shims::queue diff --git a/oneapi-rs-sys/src/queue-sys.rs b/oneapi-rs-sys/src/queue-sys.rs index 8a1f442..938524a 100644 --- a/oneapi-rs-sys/src/queue-sys.rs +++ b/oneapi-rs-sys/src/queue-sys.rs @@ -42,5 +42,6 @@ pub mod ffi { ) -> UniquePtr; fn barrier(queue: &mut UniquePtr, dep_events: Vec) -> UniquePtr; fn wait(queue: &mut UniquePtr); + unsafe fn launch(queue: &mut UniquePtr, kernel: &Kernel, args: &[&[u8]]) -> UniquePtr; } } diff --git a/oneapi-rs-sys/src/queue.cpp b/oneapi-rs-sys/src/queue.cpp index 7da9fc6..9ac374a 100644 --- a/oneapi-rs-sys/src/queue.cpp +++ b/oneapi-rs-sys/src/queue.cpp @@ -12,6 +12,8 @@ using sycl::ext::intel::property::queue::immediate_command_list; using sycl::property::queue::in_order; +namespace syclexp = sycl::ext::oneapi::experimental; + namespace sycl_shims::queue { std::unique_ptr new_queue() { return std::make_unique(sycl::queue({in_order()})); @@ -52,4 +54,13 @@ std::unique_ptr barrier(std::unique_ptr &queue, } void wait(std::unique_ptr &queue) { queue->wait(); } + +std::unique_ptr launch(std::unique_ptr queue, Kernel const &kernel, rust::Slice const> args) { + return std::make_unique(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())); + + cgh.parallel_for(sycl::nd_range{{1024}, {16}}, kernel); + })); +} } // namespace sycl_shims::queue From 430cd9243c4844b17afad0afadb6a64a6106ae0d Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Thu, 23 Jul 2026 10:06:06 +0000 Subject: [PATCH 16/27] Add C++ get_kernel binding --- oneapi-rs-sys/include/kernel-bundle.hpp | 2 +- oneapi-rs-sys/src/kernel-bundle-sys.rs | 5 ++++- oneapi-rs-sys/src/kernel-bundle.cpp | 4 ++++ oneapi-rs-sys/src/types-sys.rs | 1 + 4 files changed, 10 insertions(+), 2 deletions(-) diff --git a/oneapi-rs-sys/include/kernel-bundle.hpp b/oneapi-rs-sys/include/kernel-bundle.hpp index 7666c56..e03fc76 100644 --- a/oneapi-rs-sys/include/kernel-bundle.hpp +++ b/oneapi-rs-sys/include/kernel-bundle.hpp @@ -18,6 +18,6 @@ namespace sycl_shims::kernel_bundle { std::unique_ptr create_kernel_bundle_from_source(Context const &ctxt, rust::Str source); - std::unique_ptr build(std::unique_ptr &source); +std::unique_ptr get_kernel(std::unique_ptr &, rust::Str); } // namespace sycl_shims::kernel_bundle diff --git a/oneapi-rs-sys/src/kernel-bundle-sys.rs b/oneapi-rs-sys/src/kernel-bundle-sys.rs index 51a48e1..412b382 100644 --- a/oneapi-rs-sys/src/kernel-bundle-sys.rs +++ b/oneapi-rs-sys/src/kernel-bundle-sys.rs @@ -20,9 +20,12 @@ pub mod ffi { #[namespace = "sycl_shims"] type ExecutableKernelBundle = crate::types::ffi::ExecutableKernelBundle; + #[namespace = "sycl_shims"] + type Kernel = crate::types::ffi::Kernel; + fn create_kernel_bundle_from_source(ctxt: &Context, source: &str) -> UniquePtr; - fn build(source: &mut UniquePtr) -> UniquePtr; + fn get_kernel(bundle: &mut UniquePtr, name: &str) -> UniquePtr; } } diff --git a/oneapi-rs-sys/src/kernel-bundle.cpp b/oneapi-rs-sys/src/kernel-bundle.cpp index 78d4c73..39027c9 100644 --- a/oneapi-rs-sys/src/kernel-bundle.cpp +++ b/oneapi-rs-sys/src/kernel-bundle.cpp @@ -24,4 +24,8 @@ std::unique_ptr create_kernel_bundle_from_source(Context con std::unique_ptr build(std::unique_ptr &source) { return std::make_unique(syclexp::build(*source)); } + +std::unique_ptr get_kernel(std::unique_ptr &bundle, rust::Str name) { + return std::make_unique(bundle->ext_oneapi_get_kernel(static_cast(name))); +} } // namespace sycl_shims::kernel_bundle diff --git a/oneapi-rs-sys/src/types-sys.rs b/oneapi-rs-sys/src/types-sys.rs index a577254..5174d4b 100644 --- a/oneapi-rs-sys/src/types-sys.rs +++ b/oneapi-rs-sys/src/types-sys.rs @@ -85,6 +85,7 @@ pub mod ffi { impl UniquePtr {} impl UniquePtr {} impl UniquePtr {} + impl UniquePtr {} impl Vec {} impl Vec {} From 08f79aebc5eb5cc1acd575db036feb60894a2a1c Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Thu, 23 Jul 2026 10:08:34 +0000 Subject: [PATCH 17/27] Add Rust get_kernel binding --- oneapi-rs/src/kernel_bundle.rs | 14 ++++++++++++++ 1 file changed, 14 insertions(+) diff --git a/oneapi-rs/src/kernel_bundle.rs b/oneapi-rs/src/kernel_bundle.rs index 7f44130..7c261d1 100644 --- a/oneapi-rs/src/kernel_bundle.rs +++ b/oneapi-rs/src/kernel_bundle.rs @@ -31,6 +31,20 @@ impl From> for ExecutableKern } } +impl ExecutableKernelBundle { + pub fn get_kernel(&mut self, name: &str) -> Kernel { + ffi::get_kernel(&mut self.0, name).into() + } +} + +pub struct Kernel(pub(crate) cxx::UniquePtr); + +impl From> for Kernel { + fn from(value: cxx::UniquePtr) -> Self { + Self(value) + } +} + pub unsafe trait KernelArgument { unsafe fn as_raw_arg(&self) -> &[u8]; } From 7529b424846ef8dd480b1d382754fd3853676ea1 Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Thu, 23 Jul 2026 10:13:02 +0000 Subject: [PATCH 18/27] Add basic Rust launch binding --- oneapi-rs-sys/src/queue.cpp | 2 +- oneapi-rs/src/buffer.rs | 4 +++- oneapi-rs/src/kernel_bundle.rs | 4 ++++ oneapi-rs/src/queue.rs | 6 +++++- 4 files changed, 13 insertions(+), 3 deletions(-) diff --git a/oneapi-rs-sys/src/queue.cpp b/oneapi-rs-sys/src/queue.cpp index 9ac374a..9f1a56f 100644 --- a/oneapi-rs-sys/src/queue.cpp +++ b/oneapi-rs-sys/src/queue.cpp @@ -55,7 +55,7 @@ std::unique_ptr barrier(std::unique_ptr &queue, void wait(std::unique_ptr &queue) { queue->wait(); } -std::unique_ptr launch(std::unique_ptr queue, Kernel const &kernel, rust::Slice const> args) { +std::unique_ptr launch(std::unique_ptr &queue, Kernel const &kernel, rust::Slice const> args) { return std::make_unique(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())); diff --git a/oneapi-rs/src/buffer.rs b/oneapi-rs/src/buffer.rs index 048fc7c..ac2ebc9 100644 --- a/oneapi-rs/src/buffer.rs +++ b/oneapi-rs/src/buffer.rs @@ -141,9 +141,11 @@ impl IntoFuture for EnqueuedBuffer { unsafe impl KernelArgument for Buffer { 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>() ) } diff --git a/oneapi-rs/src/kernel_bundle.rs b/oneapi-rs/src/kernel_bundle.rs index 7c261d1..c29fff1 100644 --- a/oneapi-rs/src/kernel_bundle.rs +++ b/oneapi-rs/src/kernel_bundle.rs @@ -54,3 +54,7 @@ unsafe impl KernelArgument for T { bytemuck::bytes_of(self) } } + +pub unsafe trait KernelArgumentList { + unsafe fn as_raw_arg_list(&self) -> [&[u8]; ARGC]; +} diff --git a/oneapi-rs/src/queue.rs b/oneapi-rs/src/queue.rs index fa542b3..9a41201 100644 --- a/oneapi-rs/src/queue.rs +++ b/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(&mut self, kernel: &Kernel, args: impl KernelArgumentList) -> Event { + unsafe { ffi::launch(&mut self.0, &kernel.0, &args.as_raw_arg_list()) }.into() + } } impl From<&Device> for Queue { From 446563e6898b8d4d47496607cc19f39827f6223d Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Thu, 23 Jul 2026 13:19:00 +0000 Subject: [PATCH 19/27] Add kernel launch example --- oneapi-rs-sys/build.rs | 1 + oneapi-rs/examples/kernel_launch.rs | 53 +++++++++++++++++++++++++++++ 2 files changed, 54 insertions(+) create mode 100644 oneapi-rs/examples/kernel_launch.rs diff --git a/oneapi-rs-sys/build.rs b/oneapi-rs-sys/build.rs index 854c5e4..121cc68 100644 --- a/oneapi-rs-sys/build.rs +++ b/oneapi-rs-sys/build.rs @@ -53,6 +53,7 @@ fn main() { .compile("oneapi-shim"); println!("cargo::rustc-link-lib=sycl"); + println!("cargo::rustc-link-lib=ze_loader"); println!("cargo::rustc-link-lib=intlc"); for source in cpp_sources { diff --git a/oneapi-rs/examples/kernel_launch.rs b/oneapi-rs/examples/kernel_launch.rs new file mode 100644 index 0000000..8240d7e --- /dev/null +++ b/oneapi-rs/examples/kernel_launch.rs @@ -0,0 +1,53 @@ +// +// Copyright (C) 2026 Intel Corporation +// +// Under the MIT License or the Apache License v2.0. +// See LICENSE-MIT and LICENSE-APACHE for license information. +// SPDX-License-Identifier: MIT OR Apache-2.0 +// + +use oneapi_rs::{buffer::Buffer, kernel_bundle::{KernelArgument, KernelArgumentList}, queue::Queue, usm::{SharedAllocator, UsmAllocator}}; + +static IOTA_SRC: &str = r#" +#include +namespace syclext = sycl::ext::oneapi; +namespace syclexp = sycl::ext::oneapi::experimental; + +extern "C" +SYCL_EXT_ONEAPI_FUNCTION_PROPERTY((syclexp::nd_range_kernel<1>)) +void iota(float start, float *ptr) { + size_t id = syclext::this_work_item::get_nd_item<1>().get_global_linear_id(); + ptr[id] = start + static_cast(id); +} +"#; + +struct IotaArgs<'a> { + start: f32, + buffer: &'a mut Buffer> +} + +unsafe impl<'a> KernelArgumentList<2> for IotaArgs<'a> { + unsafe fn as_raw_arg_list(&self) -> [&[u8]; 2] { + return [ + unsafe { self.start.as_raw_arg() }, + unsafe { self.buffer.as_raw_arg() } + ] + } +} + +fn main() { + let mut queue = Queue::new(); + let mut buffer = queue.alloc_shared::(1024).wait(); + + let kernel = queue.get_context() + .create_kernel_bundle_from_source(IOTA_SRC) + .build() + .get_kernel("iota"); + + unsafe { queue.launch(&kernel, IotaArgs { start: 3.14, buffer: &mut buffer }) }.wait(); + + for e in buffer.iter() { + print!("{e} "); + } + println!(); +} From dc08aab213b46717b7f492b7713591c5f1a8b6c3 Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Thu, 23 Jul 2026 14:27:15 +0000 Subject: [PATCH 20/27] Add C++ multidimensional launch bindings --- oneapi-rs-sys/include/queue.hpp | 15 ++++++++++++- oneapi-rs-sys/src/queue-sys.rs | 29 ++++++++++++++++++++++++- oneapi-rs-sys/src/queue.cpp | 38 ++++++++++++++++++++++++++++++--- oneapi-rs-sys/src/types-sys.rs | 11 ++++++++++ 4 files changed, 88 insertions(+), 5 deletions(-) diff --git a/oneapi-rs-sys/include/queue.hpp b/oneapi-rs-sys/include/queue.hpp index ffc8288..588941f 100644 --- a/oneapi-rs-sys/include/queue.hpp +++ b/oneapi-rs-sys/include/queue.hpp @@ -15,6 +15,8 @@ namespace sycl_shims { struct EventPtr; +struct Range2; +struct Range3; } // namespace sycl_shims namespace sycl_shims::queue { @@ -28,5 +30,16 @@ std::unique_ptr memset(std::unique_ptr &, std::uint8_t *ptr, rust::Vec); std::unique_ptr barrier(std::unique_ptr &, rust::Vec); void wait(std::unique_ptr &); -std::unique_ptr launch(std::unique_ptr &, Kernel const &, rust::Slice const> args); + +std::unique_ptr +launch_1d(std::unique_ptr &, unsigned long global_size, unsigned long local_size, + Kernel const &, rust::Slice const> args); + +std::unique_ptr +launch_2d(std::unique_ptr &, Range2 global_size, Range2 local_size, + Kernel const &, rust::Slice const> args); + +std::unique_ptr +launch_3d(std::unique_ptr &, Range3 global_size, Range3 local_size, + Kernel const &, rust::Slice const> args); } // namespace sycl_shims::queue diff --git a/oneapi-rs-sys/src/queue-sys.rs b/oneapi-rs-sys/src/queue-sys.rs index 938524a..ea16687 100644 --- a/oneapi-rs-sys/src/queue-sys.rs +++ b/oneapi-rs-sys/src/queue-sys.rs @@ -12,6 +12,8 @@ pub mod ffi { extern "C++" { include!("oneapi-rs-sys/src/types-sys.rs.h"); type EventPtr = crate::types::ffi::EventPtr; + type Range2 = crate::types::ffi::Range2; + type Range3 = crate::types::ffi::Range3; } unsafe extern "C++" { @@ -33,6 +35,7 @@ pub mod ffi { fn new_queue_from_device(device: &Device) -> UniquePtr; fn get_context(queue: &Queue) -> UniquePtr; fn clone(queue: &Queue) -> UniquePtr; + unsafe fn memset( queue: &mut UniquePtr, ptr: *mut u8, @@ -40,8 +43,32 @@ pub mod ffi { num_bytes: usize, dep_events: Vec, ) -> UniquePtr; + fn barrier(queue: &mut UniquePtr, dep_events: Vec) -> UniquePtr; fn wait(queue: &mut UniquePtr); - unsafe fn launch(queue: &mut UniquePtr, kernel: &Kernel, args: &[&[u8]]) -> UniquePtr; + + unsafe fn launch_1d( + queue: &mut UniquePtr, + global_size: u64, + local_size: u64, + kernel: &Kernel, + args: &[&[u8]], + ) -> UniquePtr; + + unsafe fn launch_2d( + queue: &mut UniquePtr, + global_size: Range2, + local_size: Range2, + kernel: &Kernel, + args: &[&[u8]], + ) -> UniquePtr; + + unsafe fn launch_3d( + queue: &mut UniquePtr, + global_size: Range3, + local_size: Range3, + kernel: &Kernel, + args: &[&[u8]], + ) -> UniquePtr; } } diff --git a/oneapi-rs-sys/src/queue.cpp b/oneapi-rs-sys/src/queue.cpp index 9f1a56f..2ce405c 100644 --- a/oneapi-rs-sys/src/queue.cpp +++ b/oneapi-rs-sys/src/queue.cpp @@ -55,12 +55,44 @@ std::unique_ptr barrier(std::unique_ptr &queue, void wait(std::unique_ptr &queue) { queue->wait(); } -std::unique_ptr launch(std::unique_ptr &queue, Kernel const &kernel, rust::Slice const> args) { +template +std::unique_ptr +launch(std::unique_ptr &queue, sycl::nd_range nd_range, + Kernel const &kernel, + rust::Slice const> args) { return std::make_unique(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())); - - cgh.parallel_for(sycl::nd_range{{1024}, {16}}, kernel); + + cgh.parallel_for(nd_range, kernel); })); } + +std::unique_ptr +launch_1d(std::unique_ptr &queue, unsigned long global_size, + unsigned long local_size, Kernel const &kernel, + rust::Slice const> args) { + return launch(queue, sycl::nd_range<1>{{global_size}, {local_size}}, kernel, + args); +} + +std::unique_ptr +launch_2d(std::unique_ptr &queue, Range2 global_size, Range2 local_size, + Kernel const &kernel, + rust::Slice const> args) { + return launch(queue, + sycl::nd_range<2>{{global_size.x, global_size.y}, + {local_size.x, local_size.y}}, + kernel, args); +} + +std::unique_ptr +launch_3d(std::unique_ptr &queue, Range3 global_size, Range3 local_size, + Kernel const &kernel, + rust::Slice const> args) { + return launch(queue, + sycl::nd_range<3>{{global_size.x, global_size.y, global_size.z}, + {local_size.x, local_size.y, local_size.z}}, + kernel, args); +} } // namespace sycl_shims::queue diff --git a/oneapi-rs-sys/src/types-sys.rs b/oneapi-rs-sys/src/types-sys.rs index 5174d4b..20c91f4 100644 --- a/oneapi-rs-sys/src/types-sys.rs +++ b/oneapi-rs-sys/src/types-sys.rs @@ -77,6 +77,17 @@ pub mod ffi { Complete, Unknown, } + + struct Range2 { + x: u64, + y: u64, + } + + struct Range3 { + x: u64, + y: u64, + z: u64, + } impl UniquePtr {} impl UniquePtr {} From 9e3b755eef2e658bfa5e96858b5e6c6d51a86eb9 Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Thu, 23 Jul 2026 14:56:55 +0000 Subject: [PATCH 21/27] Add Rust NdRange support --- oneapi-rs/examples/kernel_launch.rs | 4 +- oneapi-rs/src/kernel_bundle.rs | 81 +++++++++++++++++++++++++++++ oneapi-rs/src/queue.rs | 19 +++++-- 3 files changed, 99 insertions(+), 5 deletions(-) diff --git a/oneapi-rs/examples/kernel_launch.rs b/oneapi-rs/examples/kernel_launch.rs index 8240d7e..8f944f7 100644 --- a/oneapi-rs/examples/kernel_launch.rs +++ b/oneapi-rs/examples/kernel_launch.rs @@ -6,7 +6,7 @@ // SPDX-License-Identifier: MIT OR Apache-2.0 // -use oneapi_rs::{buffer::Buffer, kernel_bundle::{KernelArgument, KernelArgumentList}, queue::Queue, usm::{SharedAllocator, UsmAllocator}}; +use oneapi_rs::{buffer::Buffer, kernel_bundle::{KernelArgument, KernelArgumentList, NdRange, Range}, queue::Queue, usm::{SharedAllocator, UsmAllocator}}; static IOTA_SRC: &str = r#" #include @@ -44,7 +44,7 @@ fn main() { .build() .get_kernel("iota"); - unsafe { queue.launch(&kernel, IotaArgs { start: 3.14, buffer: &mut buffer }) }.wait(); + unsafe { queue.launch(NdRange::new([1024], [16]), &kernel, IotaArgs { start: 3.14, buffer: &mut buffer }) }.wait(); for e in buffer.iter() { print!("{e} "); diff --git a/oneapi-rs/src/kernel_bundle.rs b/oneapi-rs/src/kernel_bundle.rs index c29fff1..32caee5 100644 --- a/oneapi-rs/src/kernel_bundle.rs +++ b/oneapi-rs/src/kernel_bundle.rs @@ -6,9 +6,13 @@ // SPDX-License-Identifier: MIT OR Apache-2.0 // +use std::marker::PhantomData; + use bytemuck::Pod; use oneapi_rs_sys::{kernel_bundle::ffi, types}; +use crate::{event::Event, queue::Queue}; + pub struct SourceKernelBundle(pub(crate) cxx::UniquePtr); impl From> for SourceKernelBundle { @@ -58,3 +62,80 @@ unsafe impl KernelArgument for T { pub unsafe trait KernelArgumentList { unsafe fn as_raw_arg_list(&self) -> [&[u8]; ARGC]; } + + +pub type Range = [u64; DIMENSIONS]; + +pub struct NdRange { + pub group_size: Range, + pub local_size: Range +} + +impl NdRange { + pub fn new(group_size: Range, local_size: Range) -> Self { + Self { + group_size, + local_size + } + } +} + +pub trait ValidDimension { + unsafe fn launch(&self, queue: &mut Queue, kernel: &Kernel, args: impl KernelArgumentList) -> Event; +} + +impl ValidDimension for NdRange<1> { + unsafe fn launch(&self, queue: &mut Queue, kernel: &Kernel, args: impl KernelArgumentList) -> Event { + unsafe { + oneapi_rs_sys::queue::ffi::launch_1d( + &mut queue.0, + self.group_size[0], + self.local_size[0], + &kernel.0, + &args.as_raw_arg_list() + ) + }.into() + } +} + +impl ValidDimension for NdRange<2> { + unsafe fn launch(&self, queue: &mut Queue, kernel: &Kernel, args: impl KernelArgumentList) -> Event { + unsafe { + oneapi_rs_sys::queue::ffi::launch_2d( + &mut queue.0, + types::ffi::Range2 { + x: self.group_size[0], + y: self.group_size[1], + }, + types::ffi::Range2 { + x: self.local_size[0], + y: self.local_size[1], + }, + &kernel.0, + &args.as_raw_arg_list() + ) + }.into() + } +} + +impl ValidDimension for NdRange<3> { + unsafe fn launch(&self, queue: &mut Queue, kernel: &Kernel, args: impl KernelArgumentList) -> Event { + unsafe { + oneapi_rs_sys::queue::ffi::launch_3d( + &mut queue.0, + types::ffi::Range3 { + x: self.group_size[0], + y: self.group_size[1], + z: self.group_size[2], + }, + types::ffi::Range3 { + x: self.local_size[0], + y: self.local_size[1], + z: self.local_size[2], + }, + &kernel.0, + &args.as_raw_arg_list() + ) + }.into() + } +} diff --git a/oneapi-rs/src/queue.rs b/oneapi-rs/src/queue.rs index 9a41201..5160a23 100644 --- a/oneapi-rs/src/queue.rs +++ b/oneapi-rs/src/queue.rs @@ -10,7 +10,12 @@ use bytemuck::Pod; use oneapi_rs_sys::{queue::ffi, types::ffi::EventPtr}; use crate::{ - buffer::{Buffer, EnqueuedBuffer}, context::Context, device::Device, event::Event, kernel_bundle::{Kernel, KernelArgumentList}, usm::{HostAllocator, SharedAllocator, UsmAlloc, UsmAllocator}, + buffer::{Buffer, EnqueuedBuffer}, + context::Context, + device::Device, + event::Event, + kernel_bundle::{Kernel, KernelArgumentList, NdRange, ValidDimension}, + usm::{HostAllocator, SharedAllocator, UsmAlloc, UsmAllocator}, }; /// The `Queue` connects a host program to a single device. Programs submit tasks to a device via the @@ -128,8 +133,16 @@ impl Queue { ffi::wait(&mut self.0); } - pub unsafe fn launch(&mut self, kernel: &Kernel, args: impl KernelArgumentList) -> Event { - unsafe { ffi::launch(&mut self.0, &kernel.0, &args.as_raw_arg_list()) }.into() + pub unsafe fn launch( + &mut self, + nd_range: NdRange, + kernel: &Kernel, + args: impl KernelArgumentList, + ) -> Event + where + NdRange: ValidDimension, + { + unsafe { nd_range.launch(self, kernel, args) } } } From 1021932bba2c7c00733e4b0055cc81a20d49bad9 Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Thu, 23 Jul 2026 15:18:35 +0000 Subject: [PATCH 22/27] Reuse Rust dimension arrays --- oneapi-rs-sys/include/queue.hpp | 12 ++++++++---- oneapi-rs-sys/src/queue-sys.rs | 5 +++-- oneapi-rs-sys/src/queue.cpp | 23 +++++++++++++---------- oneapi-rs-sys/src/types-sys.rs | 14 ++++++++------ oneapi-rs/src/kernel_bundle.rs | 28 ++++++---------------------- 5 files changed, 38 insertions(+), 44 deletions(-) diff --git a/oneapi-rs-sys/include/queue.hpp b/oneapi-rs-sys/include/queue.hpp index 588941f..8ac204a 100644 --- a/oneapi-rs-sys/include/queue.hpp +++ b/oneapi-rs-sys/include/queue.hpp @@ -15,6 +15,7 @@ namespace sycl_shims { struct EventPtr; +struct Range1; struct Range2; struct Range3; } // namespace sycl_shims @@ -32,14 +33,17 @@ std::unique_ptr barrier(std::unique_ptr &, rust::Vec); void wait(std::unique_ptr &); std::unique_ptr -launch_1d(std::unique_ptr &, unsigned long global_size, unsigned long local_size, - Kernel const &, rust::Slice const> args); +launch_1d(std::unique_ptr &, Range1 global_size, Range1 local_size, + Kernel const &, + rust::Slice const> args); std::unique_ptr launch_2d(std::unique_ptr &, Range2 global_size, Range2 local_size, - Kernel const &, rust::Slice const> args); + Kernel const &, + rust::Slice const> args); std::unique_ptr launch_3d(std::unique_ptr &, Range3 global_size, Range3 local_size, - Kernel const &, rust::Slice const> args); + Kernel const &, + rust::Slice const> args); } // namespace sycl_shims::queue diff --git a/oneapi-rs-sys/src/queue-sys.rs b/oneapi-rs-sys/src/queue-sys.rs index ea16687..d66e620 100644 --- a/oneapi-rs-sys/src/queue-sys.rs +++ b/oneapi-rs-sys/src/queue-sys.rs @@ -12,6 +12,7 @@ pub mod ffi { extern "C++" { include!("oneapi-rs-sys/src/types-sys.rs.h"); type EventPtr = crate::types::ffi::EventPtr; + type Range1 = crate::types::ffi::Range1; type Range2 = crate::types::ffi::Range2; type Range3 = crate::types::ffi::Range3; } @@ -49,8 +50,8 @@ pub mod ffi { unsafe fn launch_1d( queue: &mut UniquePtr, - global_size: u64, - local_size: u64, + global_size: Range1, + local_size: Range1, kernel: &Kernel, args: &[&[u8]], ) -> UniquePtr; diff --git a/oneapi-rs-sys/src/queue.cpp b/oneapi-rs-sys/src/queue.cpp index 2ce405c..afef17b 100644 --- a/oneapi-rs-sys/src/queue.cpp +++ b/oneapi-rs-sys/src/queue.cpp @@ -69,11 +69,12 @@ launch(std::unique_ptr &queue, sycl::nd_range nd_range, } std::unique_ptr -launch_1d(std::unique_ptr &queue, unsigned long global_size, - unsigned long local_size, Kernel const &kernel, +launch_1d(std::unique_ptr &queue, Range1 global_size, Range1 local_size, + Kernel const &kernel, rust::Slice const> args) { - return launch(queue, sycl::nd_range<1>{{global_size}, {local_size}}, kernel, - args); + return launch(queue, + sycl::nd_range<1>{{global_size.data[0]}, {local_size.data[0]}}, + kernel, args); } std::unique_ptr @@ -81,8 +82,8 @@ launch_2d(std::unique_ptr &queue, Range2 global_size, Range2 local_size, Kernel const &kernel, rust::Slice const> args) { return launch(queue, - sycl::nd_range<2>{{global_size.x, global_size.y}, - {local_size.x, local_size.y}}, + sycl::nd_range<2>{{global_size.data[0], global_size.data[1]}, + {local_size.data[0], local_size.data[1]}}, kernel, args); } @@ -90,9 +91,11 @@ std::unique_ptr launch_3d(std::unique_ptr &queue, Range3 global_size, Range3 local_size, Kernel const &kernel, rust::Slice const> args) { - return launch(queue, - sycl::nd_range<3>{{global_size.x, global_size.y, global_size.z}, - {local_size.x, local_size.y, local_size.z}}, - kernel, args); + return launch( + queue, + sycl::nd_range<3>{ + {global_size.data[0], global_size.data[1], global_size.data[2]}, + {local_size.data[0], local_size.data[1], local_size.data[2]}}, + kernel, args); } } // namespace sycl_shims::queue diff --git a/oneapi-rs-sys/src/types-sys.rs b/oneapi-rs-sys/src/types-sys.rs index 20c91f4..77f4cb9 100644 --- a/oneapi-rs-sys/src/types-sys.rs +++ b/oneapi-rs-sys/src/types-sys.rs @@ -77,16 +77,18 @@ pub mod ffi { Complete, Unknown, } - + + // cxx doesn't support const generic parameters + struct Range1 { + data: [u64; 1] + } + struct Range2 { - x: u64, - y: u64, + data: [u64; 2] } struct Range3 { - x: u64, - y: u64, - z: u64, + data: [u64; 3] } impl UniquePtr {} diff --git a/oneapi-rs/src/kernel_bundle.rs b/oneapi-rs/src/kernel_bundle.rs index 32caee5..d926c76 100644 --- a/oneapi-rs/src/kernel_bundle.rs +++ b/oneapi-rs/src/kernel_bundle.rs @@ -6,8 +6,6 @@ // SPDX-License-Identifier: MIT OR Apache-2.0 // -use std::marker::PhantomData; - use bytemuck::Pod; use oneapi_rs_sys::{kernel_bundle::ffi, types}; @@ -89,8 +87,8 @@ impl ValidDimension for NdRange<1> { unsafe { oneapi_rs_sys::queue::ffi::launch_1d( &mut queue.0, - self.group_size[0], - self.local_size[0], + types::ffi::Range1 { data: self.group_size }, + types::ffi::Range1 { data: self.local_size }, &kernel.0, &args.as_raw_arg_list() ) @@ -103,14 +101,8 @@ impl ValidDimension for NdRange<2> { unsafe { oneapi_rs_sys::queue::ffi::launch_2d( &mut queue.0, - types::ffi::Range2 { - x: self.group_size[0], - y: self.group_size[1], - }, - types::ffi::Range2 { - x: self.local_size[0], - y: self.local_size[1], - }, + types::ffi::Range2 { data: self.group_size }, + types::ffi::Range2 { data: self.local_size }, &kernel.0, &args.as_raw_arg_list() ) @@ -123,16 +115,8 @@ impl ValidDimension for NdRange<3> { unsafe { oneapi_rs_sys::queue::ffi::launch_3d( &mut queue.0, - types::ffi::Range3 { - x: self.group_size[0], - y: self.group_size[1], - z: self.group_size[2], - }, - types::ffi::Range3 { - x: self.local_size[0], - y: self.local_size[1], - z: self.local_size[2], - }, + types::ffi::Range3 { data: self.group_size }, + types::ffi::Range3 { data: self.local_size }, &kernel.0, &args.as_raw_arg_list() ) From 2fa3f903c7899ba023f6e1f5c96420cef6df540c Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Thu, 23 Jul 2026 15:19:37 +0000 Subject: [PATCH 23/27] cargo fmt --- oneapi-rs-sys/src/kernel-bundle-sys.rs | 11 ++-- oneapi-rs-sys/src/lib.rs | 1 - oneapi-rs-sys/src/types-sys.rs | 6 +-- oneapi-rs/examples/kernel_launch.rs | 31 ++++++++--- oneapi-rs/src/buffer.rs | 11 ++-- oneapi-rs/src/context.rs | 6 ++- oneapi-rs/src/kernel_bundle.rs | 72 +++++++++++++++++++------- oneapi-rs/src/lib.rs | 5 +- 8 files changed, 97 insertions(+), 46 deletions(-) diff --git a/oneapi-rs-sys/src/kernel-bundle-sys.rs b/oneapi-rs-sys/src/kernel-bundle-sys.rs index 412b382..2e53ab4 100644 --- a/oneapi-rs-sys/src/kernel-bundle-sys.rs +++ b/oneapi-rs-sys/src/kernel-bundle-sys.rs @@ -23,9 +23,14 @@ pub mod ffi { #[namespace = "sycl_shims"] type Kernel = crate::types::ffi::Kernel; - fn create_kernel_bundle_from_source(ctxt: &Context, source: &str) - -> UniquePtr; + fn create_kernel_bundle_from_source( + ctxt: &Context, + source: &str, + ) -> UniquePtr; fn build(source: &mut UniquePtr) -> UniquePtr; - fn get_kernel(bundle: &mut UniquePtr, name: &str) -> UniquePtr; + fn get_kernel( + bundle: &mut UniquePtr, + name: &str, + ) -> UniquePtr; } } diff --git a/oneapi-rs-sys/src/lib.rs b/oneapi-rs-sys/src/lib.rs index 5945819..514e0d8 100644 --- a/oneapi-rs-sys/src/lib.rs +++ b/oneapi-rs-sys/src/lib.rs @@ -29,4 +29,3 @@ pub mod context; #[path = "kernel-bundle-sys.rs"] pub mod kernel_bundle; - diff --git a/oneapi-rs-sys/src/types-sys.rs b/oneapi-rs-sys/src/types-sys.rs index 77f4cb9..c675165 100644 --- a/oneapi-rs-sys/src/types-sys.rs +++ b/oneapi-rs-sys/src/types-sys.rs @@ -80,15 +80,15 @@ pub mod ffi { // cxx doesn't support const generic parameters struct Range1 { - data: [u64; 1] + data: [u64; 1], } struct Range2 { - data: [u64; 2] + data: [u64; 2], } struct Range3 { - data: [u64; 3] + data: [u64; 3], } impl UniquePtr {} diff --git a/oneapi-rs/examples/kernel_launch.rs b/oneapi-rs/examples/kernel_launch.rs index 8f944f7..7376abb 100644 --- a/oneapi-rs/examples/kernel_launch.rs +++ b/oneapi-rs/examples/kernel_launch.rs @@ -6,7 +6,12 @@ // SPDX-License-Identifier: MIT OR Apache-2.0 // -use oneapi_rs::{buffer::Buffer, kernel_bundle::{KernelArgument, KernelArgumentList, NdRange, Range}, queue::Queue, usm::{SharedAllocator, UsmAllocator}}; +use oneapi_rs::{ + buffer::Buffer, + kernel_bundle::{KernelArgument, KernelArgumentList, NdRange}, + queue::Queue, + usm::{SharedAllocator, UsmAllocator}, +}; static IOTA_SRC: &str = r#" #include @@ -23,15 +28,14 @@ void iota(float start, float *ptr) { struct IotaArgs<'a> { start: f32, - buffer: &'a mut Buffer> + buffer: &'a mut Buffer>, } unsafe impl<'a> KernelArgumentList<2> for IotaArgs<'a> { unsafe fn as_raw_arg_list(&self) -> [&[u8]; 2] { - return [ - unsafe { self.start.as_raw_arg() }, - unsafe { self.buffer.as_raw_arg() } - ] + return [unsafe { self.start.as_raw_arg() }, unsafe { + self.buffer.as_raw_arg() + }]; } } @@ -39,12 +43,23 @@ fn main() { let mut queue = Queue::new(); let mut buffer = queue.alloc_shared::(1024).wait(); - let kernel = queue.get_context() + let kernel = queue + .get_context() .create_kernel_bundle_from_source(IOTA_SRC) .build() .get_kernel("iota"); - unsafe { queue.launch(NdRange::new([1024], [16]), &kernel, IotaArgs { start: 3.14, buffer: &mut buffer }) }.wait(); + unsafe { + queue.launch( + NdRange::new([1024], [16]), + &kernel, + IotaArgs { + start: 3.14, + buffer: &mut buffer, + }, + ) + } + .wait(); for e in buffer.iter() { print!("{e} "); diff --git a/oneapi-rs/src/buffer.rs b/oneapi-rs/src/buffer.rs index ac2ebc9..d54e1f7 100644 --- a/oneapi-rs/src/buffer.rs +++ b/oneapi-rs/src/buffer.rs @@ -19,7 +19,9 @@ use bytemuck::Pod; use pin_project::pin_project; use crate::{ - event::{Event, EventFuture}, kernel_bundle::KernelArgument, usm::UsmAlloc, + event::{Event, EventFuture}, + kernel_bundle::KernelArgument, + usm::UsmAlloc, }; /// The Buffer struct defines a shared array of one, two or three dimensions that can be used @@ -143,11 +145,6 @@ unsafe impl KernelArgument for Buffer { 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( - cast_ptr, - std::mem::size_of::<*mut u8>() - ) - } + unsafe { slice::from_raw_parts(cast_ptr, std::mem::size_of::<*mut u8>()) } } } diff --git a/oneapi-rs/src/context.rs b/oneapi-rs/src/context.rs index 36e78f7..661d3d6 100644 --- a/oneapi-rs/src/context.rs +++ b/oneapi-rs/src/context.rs @@ -6,7 +6,7 @@ // SPDX-License-Identifier: MIT OR Apache-2.0 // -use oneapi_rs_sys::{kernel_bundle, context::ffi, types::ffi::DevicePtr}; +use oneapi_rs_sys::{context::ffi, kernel_bundle, types::ffi::DevicePtr}; use crate::{device::Device, kernel_bundle::SourceKernelBundle}; @@ -24,7 +24,9 @@ impl Context { pub fn new(devices: &[&Device]) -> Self { let devices = devices .iter() - .map(|d| DevicePtr { ptr: (*d).clone().0 }) + .map(|d| DevicePtr { + ptr: (*d).clone().0, + }) .collect::>(); ffi::new_context(devices).into() diff --git a/oneapi-rs/src/kernel_bundle.rs b/oneapi-rs/src/kernel_bundle.rs index d926c76..22ec3ba 100644 --- a/oneapi-rs/src/kernel_bundle.rs +++ b/oneapi-rs/src/kernel_bundle.rs @@ -61,65 +61,99 @@ pub unsafe trait KernelArgumentList { unsafe fn as_raw_arg_list(&self) -> [&[u8]; ARGC]; } - pub type Range = [u64; DIMENSIONS]; pub struct NdRange { pub group_size: Range, - pub local_size: Range + pub local_size: Range, } impl NdRange { pub fn new(group_size: Range, local_size: Range) -> Self { Self { group_size, - local_size + local_size, } } } pub trait ValidDimension { - unsafe fn launch(&self, queue: &mut Queue, kernel: &Kernel, args: impl KernelArgumentList) -> Event; + unsafe fn launch( + &self, + queue: &mut Queue, + kernel: &Kernel, + args: impl KernelArgumentList, + ) -> Event; } impl ValidDimension for NdRange<1> { - unsafe fn launch(&self, queue: &mut Queue, kernel: &Kernel, args: impl KernelArgumentList) -> Event { + unsafe fn launch( + &self, + queue: &mut Queue, + kernel: &Kernel, + args: impl KernelArgumentList, + ) -> Event { unsafe { oneapi_rs_sys::queue::ffi::launch_1d( &mut queue.0, - types::ffi::Range1 { data: self.group_size }, - types::ffi::Range1 { data: self.local_size }, + types::ffi::Range1 { + data: self.group_size, + }, + types::ffi::Range1 { + data: self.local_size, + }, &kernel.0, - &args.as_raw_arg_list() + &args.as_raw_arg_list(), ) - }.into() + } + .into() } } impl ValidDimension for NdRange<2> { - unsafe fn launch(&self, queue: &mut Queue, kernel: &Kernel, args: impl KernelArgumentList) -> Event { + unsafe fn launch( + &self, + queue: &mut Queue, + kernel: &Kernel, + args: impl KernelArgumentList, + ) -> Event { unsafe { oneapi_rs_sys::queue::ffi::launch_2d( &mut queue.0, - types::ffi::Range2 { data: self.group_size }, - types::ffi::Range2 { data: self.local_size }, + types::ffi::Range2 { + data: self.group_size, + }, + types::ffi::Range2 { + data: self.local_size, + }, &kernel.0, - &args.as_raw_arg_list() + &args.as_raw_arg_list(), ) - }.into() + } + .into() } } impl ValidDimension for NdRange<3> { - unsafe fn launch(&self, queue: &mut Queue, kernel: &Kernel, args: impl KernelArgumentList) -> Event { + unsafe fn launch( + &self, + queue: &mut Queue, + kernel: &Kernel, + args: impl KernelArgumentList, + ) -> Event { unsafe { oneapi_rs_sys::queue::ffi::launch_3d( &mut queue.0, - types::ffi::Range3 { data: self.group_size }, - types::ffi::Range3 { data: self.local_size }, + types::ffi::Range3 { + data: self.group_size, + }, + types::ffi::Range3 { + data: self.local_size, + }, &kernel.0, - &args.as_raw_arg_list() + &args.as_raw_arg_list(), ) - }.into() + } + .into() } } diff --git a/oneapi-rs/src/lib.rs b/oneapi-rs/src/lib.rs index e2a7fc6..e8593e9 100644 --- a/oneapi-rs/src/lib.rs +++ b/oneapi-rs/src/lib.rs @@ -7,12 +7,11 @@ // pub mod buffer; +pub mod context; pub mod device; pub mod event; pub mod info; +pub mod kernel_bundle; pub mod platform; pub mod queue; pub mod usm; -pub mod context; -pub mod kernel_bundle; - From b9c1a63724b44310328ff61356496ef9f3bfc873 Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Thu, 23 Jul 2026 15:25:35 +0000 Subject: [PATCH 24/27] Rename kernel_bundle module to kernel --- oneapi-rs/examples/kernel_launch.rs | 2 +- oneapi-rs/src/buffer.rs | 2 +- oneapi-rs/src/context.rs | 2 +- oneapi-rs/src/{kernel_bundle.rs => kernel.rs} | 0 oneapi-rs/src/lib.rs | 2 +- oneapi-rs/src/queue.rs | 2 +- 6 files changed, 5 insertions(+), 5 deletions(-) rename oneapi-rs/src/{kernel_bundle.rs => kernel.rs} (100%) diff --git a/oneapi-rs/examples/kernel_launch.rs b/oneapi-rs/examples/kernel_launch.rs index 7376abb..8eacb05 100644 --- a/oneapi-rs/examples/kernel_launch.rs +++ b/oneapi-rs/examples/kernel_launch.rs @@ -8,7 +8,7 @@ use oneapi_rs::{ buffer::Buffer, - kernel_bundle::{KernelArgument, KernelArgumentList, NdRange}, + kernel::{KernelArgument, KernelArgumentList, NdRange}, queue::Queue, usm::{SharedAllocator, UsmAllocator}, }; diff --git a/oneapi-rs/src/buffer.rs b/oneapi-rs/src/buffer.rs index d54e1f7..4ef71c2 100644 --- a/oneapi-rs/src/buffer.rs +++ b/oneapi-rs/src/buffer.rs @@ -20,7 +20,7 @@ use pin_project::pin_project; use crate::{ event::{Event, EventFuture}, - kernel_bundle::KernelArgument, + kernel::KernelArgument, usm::UsmAlloc, }; diff --git a/oneapi-rs/src/context.rs b/oneapi-rs/src/context.rs index 661d3d6..0dfff64 100644 --- a/oneapi-rs/src/context.rs +++ b/oneapi-rs/src/context.rs @@ -8,7 +8,7 @@ use oneapi_rs_sys::{context::ffi, kernel_bundle, types::ffi::DevicePtr}; -use crate::{device::Device, kernel_bundle::SourceKernelBundle}; +use crate::{device::Device, kernel::SourceKernelBundle}; /// A context represents the runtime data structures and state required by a SYCL backend API /// to interact with a group of devices associated with a platform. diff --git a/oneapi-rs/src/kernel_bundle.rs b/oneapi-rs/src/kernel.rs similarity index 100% rename from oneapi-rs/src/kernel_bundle.rs rename to oneapi-rs/src/kernel.rs diff --git a/oneapi-rs/src/lib.rs b/oneapi-rs/src/lib.rs index e8593e9..ee970bb 100644 --- a/oneapi-rs/src/lib.rs +++ b/oneapi-rs/src/lib.rs @@ -11,7 +11,7 @@ pub mod context; pub mod device; pub mod event; pub mod info; -pub mod kernel_bundle; +pub mod kernel; pub mod platform; pub mod queue; pub mod usm; diff --git a/oneapi-rs/src/queue.rs b/oneapi-rs/src/queue.rs index 5160a23..cb8050a 100644 --- a/oneapi-rs/src/queue.rs +++ b/oneapi-rs/src/queue.rs @@ -14,7 +14,7 @@ use crate::{ context::Context, device::Device, event::Event, - kernel_bundle::{Kernel, KernelArgumentList, NdRange, ValidDimension}, + kernel::{Kernel, KernelArgumentList, NdRange, ValidDimension}, usm::{HostAllocator, SharedAllocator, UsmAlloc, UsmAllocator}, }; From 02ef136fb6cb5bbc7ae07ac7c7c85fe60526cc68 Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Thu, 23 Jul 2026 15:29:06 +0000 Subject: [PATCH 25/27] Move ranges to range module --- oneapi-rs/examples/kernel_launch.rs | 3 +- oneapi-rs/src/kernel.rs | 99 ------------------------ oneapi-rs/src/lib.rs | 1 + oneapi-rs/src/queue.rs | 3 +- oneapi-rs/src/range.rs | 112 ++++++++++++++++++++++++++++ 5 files changed, 117 insertions(+), 101 deletions(-) create mode 100644 oneapi-rs/src/range.rs diff --git a/oneapi-rs/examples/kernel_launch.rs b/oneapi-rs/examples/kernel_launch.rs index 8eacb05..6224b27 100644 --- a/oneapi-rs/examples/kernel_launch.rs +++ b/oneapi-rs/examples/kernel_launch.rs @@ -8,8 +8,9 @@ use oneapi_rs::{ buffer::Buffer, - kernel::{KernelArgument, KernelArgumentList, NdRange}, + kernel::{KernelArgument, KernelArgumentList}, queue::Queue, + range::NdRange, usm::{SharedAllocator, UsmAllocator}, }; diff --git a/oneapi-rs/src/kernel.rs b/oneapi-rs/src/kernel.rs index 22ec3ba..c29fff1 100644 --- a/oneapi-rs/src/kernel.rs +++ b/oneapi-rs/src/kernel.rs @@ -9,8 +9,6 @@ use bytemuck::Pod; use oneapi_rs_sys::{kernel_bundle::ffi, types}; -use crate::{event::Event, queue::Queue}; - pub struct SourceKernelBundle(pub(crate) cxx::UniquePtr); impl From> for SourceKernelBundle { @@ -60,100 +58,3 @@ unsafe impl KernelArgument for T { pub unsafe trait KernelArgumentList { unsafe fn as_raw_arg_list(&self) -> [&[u8]; ARGC]; } - -pub type Range = [u64; DIMENSIONS]; - -pub struct NdRange { - pub group_size: Range, - pub local_size: Range, -} - -impl NdRange { - pub fn new(group_size: Range, local_size: Range) -> Self { - Self { - group_size, - local_size, - } - } -} - -pub trait ValidDimension { - unsafe fn launch( - &self, - queue: &mut Queue, - kernel: &Kernel, - args: impl KernelArgumentList, - ) -> Event; -} - -impl ValidDimension for NdRange<1> { - unsafe fn launch( - &self, - queue: &mut Queue, - kernel: &Kernel, - args: impl KernelArgumentList, - ) -> Event { - unsafe { - oneapi_rs_sys::queue::ffi::launch_1d( - &mut queue.0, - types::ffi::Range1 { - data: self.group_size, - }, - types::ffi::Range1 { - data: self.local_size, - }, - &kernel.0, - &args.as_raw_arg_list(), - ) - } - .into() - } -} - -impl ValidDimension for NdRange<2> { - unsafe fn launch( - &self, - queue: &mut Queue, - kernel: &Kernel, - args: impl KernelArgumentList, - ) -> Event { - unsafe { - oneapi_rs_sys::queue::ffi::launch_2d( - &mut queue.0, - types::ffi::Range2 { - data: self.group_size, - }, - types::ffi::Range2 { - data: self.local_size, - }, - &kernel.0, - &args.as_raw_arg_list(), - ) - } - .into() - } -} - -impl ValidDimension for NdRange<3> { - unsafe fn launch( - &self, - queue: &mut Queue, - kernel: &Kernel, - args: impl KernelArgumentList, - ) -> Event { - unsafe { - oneapi_rs_sys::queue::ffi::launch_3d( - &mut queue.0, - types::ffi::Range3 { - data: self.group_size, - }, - types::ffi::Range3 { - data: self.local_size, - }, - &kernel.0, - &args.as_raw_arg_list(), - ) - } - .into() - } -} diff --git a/oneapi-rs/src/lib.rs b/oneapi-rs/src/lib.rs index ee970bb..3e274b2 100644 --- a/oneapi-rs/src/lib.rs +++ b/oneapi-rs/src/lib.rs @@ -14,4 +14,5 @@ pub mod info; pub mod kernel; pub mod platform; pub mod queue; +pub mod range; pub mod usm; diff --git a/oneapi-rs/src/queue.rs b/oneapi-rs/src/queue.rs index cb8050a..c961d81 100644 --- a/oneapi-rs/src/queue.rs +++ b/oneapi-rs/src/queue.rs @@ -14,7 +14,8 @@ use crate::{ context::Context, device::Device, event::Event, - kernel::{Kernel, KernelArgumentList, NdRange, ValidDimension}, + kernel::{Kernel, KernelArgumentList}, + range::{NdRange, ValidDimension}, usm::{HostAllocator, SharedAllocator, UsmAlloc, UsmAllocator}, }; diff --git a/oneapi-rs/src/range.rs b/oneapi-rs/src/range.rs new file mode 100644 index 0000000..d4515f7 --- /dev/null +++ b/oneapi-rs/src/range.rs @@ -0,0 +1,112 @@ +// +// Copyright (C) 2026 Intel Corporation +// +// Under the MIT License or the Apache License v2.0. +// See LICENSE-MIT and LICENSE-APACHE for license information. +// SPDX-License-Identifier: MIT OR Apache-2.0 +// + +use oneapi_rs_sys::types; + +use crate::{ + event::Event, + kernel::{Kernel, KernelArgumentList}, + queue::Queue, +}; + +pub type Range = [u64; DIMENSIONS]; + +pub struct NdRange { + pub group_size: Range, + pub local_size: Range, +} + +impl NdRange { + pub fn new(group_size: Range, local_size: Range) -> Self { + Self { + group_size, + local_size, + } + } +} + +pub trait ValidDimension { + unsafe fn launch( + &self, + queue: &mut Queue, + kernel: &Kernel, + args: impl KernelArgumentList, + ) -> Event; +} + +impl ValidDimension for NdRange<1> { + unsafe fn launch( + &self, + queue: &mut Queue, + kernel: &Kernel, + args: impl KernelArgumentList, + ) -> Event { + unsafe { + oneapi_rs_sys::queue::ffi::launch_1d( + &mut queue.0, + types::ffi::Range1 { + data: self.group_size, + }, + types::ffi::Range1 { + data: self.local_size, + }, + &kernel.0, + &args.as_raw_arg_list(), + ) + } + .into() + } +} + +impl ValidDimension for NdRange<2> { + unsafe fn launch( + &self, + queue: &mut Queue, + kernel: &Kernel, + args: impl KernelArgumentList, + ) -> Event { + unsafe { + oneapi_rs_sys::queue::ffi::launch_2d( + &mut queue.0, + types::ffi::Range2 { + data: self.group_size, + }, + types::ffi::Range2 { + data: self.local_size, + }, + &kernel.0, + &args.as_raw_arg_list(), + ) + } + .into() + } +} + +impl ValidDimension for NdRange<3> { + unsafe fn launch( + &self, + queue: &mut Queue, + kernel: &Kernel, + args: impl KernelArgumentList, + ) -> Event { + unsafe { + oneapi_rs_sys::queue::ffi::launch_3d( + &mut queue.0, + types::ffi::Range3 { + data: self.group_size, + }, + types::ffi::Range3 { + data: self.local_size, + }, + &kernel.0, + &args.as_raw_arg_list(), + ) + } + .into() + } +} From 6140a1b96c31acfdf39b8e0be417ee681a4be847 Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Fri, 24 Jul 2026 15:27:28 +0000 Subject: [PATCH 26/27] Add documentation --- oneapi-rs/src/kernel.rs | 5 +++++ oneapi-rs/src/queue.rs | 2 ++ oneapi-rs/src/range.rs | 8 ++++++++ 3 files changed, 15 insertions(+) diff --git a/oneapi-rs/src/kernel.rs b/oneapi-rs/src/kernel.rs index c29fff1..0e5d6a4 100644 --- a/oneapi-rs/src/kernel.rs +++ b/oneapi-rs/src/kernel.rs @@ -9,6 +9,7 @@ use bytemuck::Pod; use oneapi_rs_sys::{kernel_bundle::ffi, types}; +/// A kernel bundle which stores loaded SYCL source code. pub struct SourceKernelBundle(pub(crate) cxx::UniquePtr); impl From> for SourceKernelBundle { @@ -23,6 +24,7 @@ impl SourceKernelBundle { } } +/// A kernel bundle which stores compiled SYCL kernels. pub struct ExecutableKernelBundle(pub(crate) cxx::UniquePtr); impl From> for ExecutableKernelBundle { @@ -37,6 +39,7 @@ impl ExecutableKernelBundle { } } +/// An executable SYCL kernel. pub struct Kernel(pub(crate) cxx::UniquePtr); impl From> for Kernel { @@ -45,6 +48,7 @@ impl From> for Kernel { } } +/// Types which can be passed as SYCL kernel arguments. pub unsafe trait KernelArgument { unsafe fn as_raw_arg(&self) -> &[u8]; } @@ -55,6 +59,7 @@ unsafe impl KernelArgument for T { } } +/// Types which describe an argument list for a SYCL kernel. pub unsafe trait KernelArgumentList { unsafe fn as_raw_arg_list(&self) -> [&[u8]; ARGC]; } diff --git a/oneapi-rs/src/queue.rs b/oneapi-rs/src/queue.rs index c961d81..2f8e351 100644 --- a/oneapi-rs/src/queue.rs +++ b/oneapi-rs/src/queue.rs @@ -134,6 +134,8 @@ impl Queue { ffi::wait(&mut self.0); } + /// Enqueues a kernel object to the queue an an ND-range kernel, using the number of work-items + /// specified by the [`NdRange`] nd_range. pub unsafe fn launch( &mut self, nd_range: NdRange, diff --git a/oneapi-rs/src/range.rs b/oneapi-rs/src/range.rs index d4515f7..8c11aca 100644 --- a/oneapi-rs/src/range.rs +++ b/oneapi-rs/src/range.rs @@ -14,8 +14,15 @@ use crate::{ queue::Queue, }; +/// `Range` is a 1D, 2D or 3D vector that defines the iteration domain of either a single work-group +/// in a parallel dispatch, or the overall dimensions of the dispatch. pub type Range = [u64; DIMENSIONS]; +/// The `NdRange` struct defines the iteration domain of both the work-groups and the overall +/// dispatch. +/// +/// An `NdRange` comprises two [`Range`] parameters: the whole range over which the kernel is to be +/// executed and the range of each work group. pub struct NdRange { pub group_size: Range, pub local_size: Range, @@ -30,6 +37,7 @@ impl NdRange { } } +/// [`NdRange`] types which are limited to 1, 2 or 3 dimensions. pub trait ValidDimension { unsafe fn launch( &self, From 3e95505e64b3984b56be0f88bc5eb6eb4622f733 Mon Sep 17 00:00:00 2001 From: Szymon Zadworny Date: Fri, 24 Jul 2026 15:29:34 +0000 Subject: [PATCH 27/27] clang-format --- oneapi-rs-sys/include/context.hpp | 2 +- oneapi-rs-sys/include/kernel-bundle.hpp | 12 +++++++----- oneapi-rs-sys/include/types.hpp | 7 ++++--- oneapi-rs-sys/src/context.cpp | 2 +- oneapi-rs-sys/src/device.cpp | 2 +- oneapi-rs-sys/src/kernel-bundle.cpp | 21 +++++++++++---------- 6 files changed, 25 insertions(+), 21 deletions(-) diff --git a/oneapi-rs-sys/include/context.hpp b/oneapi-rs-sys/include/context.hpp index 1d8f608..ffc53b8 100644 --- a/oneapi-rs-sys/include/context.hpp +++ b/oneapi-rs-sys/include/context.hpp @@ -12,8 +12,8 @@ #include -#include "rust/cxx.h" #include "oneapi-rs-sys/include/types.hpp" +#include "rust/cxx.h" namespace sycl_shims { struct DevicePtr; diff --git a/oneapi-rs-sys/include/kernel-bundle.hpp b/oneapi-rs-sys/include/kernel-bundle.hpp index e03fc76..cc7dbe7 100644 --- a/oneapi-rs-sys/include/kernel-bundle.hpp +++ b/oneapi-rs-sys/include/kernel-bundle.hpp @@ -12,12 +12,14 @@ #include -#include "rust/cxx.h" #include "oneapi-rs-sys/include/types.hpp" +#include "rust/cxx.h" namespace sycl_shims::kernel_bundle { -std::unique_ptr create_kernel_bundle_from_source(Context const &ctxt, - rust::Str source); -std::unique_ptr build(std::unique_ptr &source); -std::unique_ptr get_kernel(std::unique_ptr &, rust::Str); +std::unique_ptr +create_kernel_bundle_from_source(Context const &ctxt, rust::Str source); +std::unique_ptr +build(std::unique_ptr &source); +std::unique_ptr get_kernel(std::unique_ptr &, + rust::Str); } // namespace sycl_shims::kernel_bundle diff --git a/oneapi-rs-sys/include/types.hpp b/oneapi-rs-sys/include/types.hpp index edec193..567da69 100644 --- a/oneapi-rs-sys/include/types.hpp +++ b/oneapi-rs-sys/include/types.hpp @@ -17,7 +17,8 @@ using Queue = sycl::queue; using Event = sycl::event; using Context = sycl::context; using Kernel = sycl::kernel; -using SourceKernelBundle = sycl::kernel_bundle; -using ExecutableKernelBundle = sycl::kernel_bundle; +using SourceKernelBundle = + sycl::kernel_bundle; +using ExecutableKernelBundle = + sycl::kernel_bundle; } // namespace sycl_shims - diff --git a/oneapi-rs-sys/src/context.cpp b/oneapi-rs-sys/src/context.cpp index 4bd6c10..dd91592 100644 --- a/oneapi-rs-sys/src/context.cpp +++ b/oneapi-rs-sys/src/context.cpp @@ -12,7 +12,7 @@ namespace sycl_shims::context { std::unique_ptr new_context(rust::Vec devices) { std::vector raw_devices; - for (auto&& d: devices) + for (auto &&d : devices) raw_devices.push_back(std::move(*d.ptr.release())); return std::make_unique(raw_devices); } diff --git a/oneapi-rs-sys/src/device.cpp b/oneapi-rs-sys/src/device.cpp index 64d287c..1063f0f 100644 --- a/oneapi-rs-sys/src/device.cpp +++ b/oneapi-rs-sys/src/device.cpp @@ -54,7 +54,7 @@ std::unique_ptr get_platform(Device const &device) { return std::make_unique(device.get_platform()); } -std::unique_ptr clone(Device const & device) { +std::unique_ptr clone(Device const &device) { return std::make_unique(sycl::device(device)); } } // namespace sycl_shims::device diff --git a/oneapi-rs-sys/src/kernel-bundle.cpp b/oneapi-rs-sys/src/kernel-bundle.cpp index 39027c9..27fe928 100644 --- a/oneapi-rs-sys/src/kernel-bundle.cpp +++ b/oneapi-rs-sys/src/kernel-bundle.cpp @@ -12,20 +12,21 @@ namespace syclexp = sycl::ext::oneapi::experimental; namespace sycl_shims::kernel_bundle { -std::unique_ptr create_kernel_bundle_from_source(Context const &ctxt, - rust::Str source) { - return std::make_unique(syclexp::create_kernel_bundle_from_source( - ctxt, - syclexp::source_language::sycl, - std::string(source) - )); +std::unique_ptr +create_kernel_bundle_from_source(Context const &ctxt, rust::Str source) { + return std::make_unique( + syclexp::create_kernel_bundle_from_source( + ctxt, syclexp::source_language::sycl, std::string(source))); } -std::unique_ptr build(std::unique_ptr &source) { +std::unique_ptr +build(std::unique_ptr &source) { return std::make_unique(syclexp::build(*source)); } -std::unique_ptr get_kernel(std::unique_ptr &bundle, rust::Str name) { - return std::make_unique(bundle->ext_oneapi_get_kernel(static_cast(name))); +std::unique_ptr +get_kernel(std::unique_ptr &bundle, rust::Str name) { + return std::make_unique( + bundle->ext_oneapi_get_kernel(static_cast(name))); } } // namespace sycl_shims::kernel_bundle