Add more documentation

sycl/sycl-rs/src/buffer.rs

@@ -121,7 +121,10 @@ impl<T, A: UsmAlloc> EnqueuedBuffer<T, A> {
 }
 
 impl<T, A: UsmAlloc> EnqueuedBuffer<T, A> {
-    /// Waits for [`Buffer`] initialization to finish.
+    /// Performs a blocking wait for the [`Buffer`] initialization to complete. Returns an error if
+    /// a synchronous SYCL exception occurs.
+    ///
+    /// Dropping an enqueued buffer does not wait for its completion.
     pub fn wait(mut self) -> Result<Buffer<T, A>> {
         self.event.wait().map(|_| self.buffer)
     }

sycl/sycl-rs/src/event.rs

@@ -21,6 +21,10 @@ use crate::{Result, info::InfoTarget, private::Sealed, queue::Queue};
 pub struct Event(pub(crate) cxx::UniquePtr<ffi::Event>);
 
 impl Event {
+    /// Performs a blocking wait for the event to complete. Returns an error if a synchronous SYCL
+    /// exception occurs.
+    ///
+    /// Dropping the event does not wait for its completion.
     pub fn wait(&mut self) -> Result<()> {
         ffi::wait(&mut self.0)
     }

sycl/sycl-rs/src/kernel.rs

@@ -50,7 +50,14 @@ impl From<cxx::UniquePtr<types::ffi::Kernel>> for Kernel {
 }
 
 /// Types which can be passed as SYCL kernel arguments.
+///
+/// Safety: a type implement this trait must mirror the representation and alignment of the
+/// corresponding SYCL kernel argument structure.
 pub unsafe trait KernelArgument {
+    /// Converts self to a raw byte representation.
+    ///
+    /// Safety: This function returns a reference to raw bytes. These bytes will be passed to FFI
+    /// functions. The caller must make sure these functions respect Rust's aliasing rules.
     unsafe fn as_raw_arg(&self) -> &[u8];
 }
 
@@ -61,7 +68,14 @@ unsafe impl<T: Pod> KernelArgument for T {
 }
 
 /// Types which describe an argument list for a SYCL kernel.
+///
+/// Safety: a type implement this trait must mirror the representation and alignment of each
+/// corresponding SYCL kernel argument inside the returned array.
 pub unsafe trait KernelArgumentList<const ARGC: usize> {
+    /// Converts each struct member to a raw byte representation.
+    ///
+    /// Safety: This function returns references to raw bytes. These bytes will be passed to FFI
+    /// functions. The caller must make sure these functions respect Rust's aliasing rules.
     unsafe fn as_raw_arg_list(&self) -> [&[u8]; ARGC];
 }
 

sycl/sycl-rs/src/lib.rs

@@ -20,6 +20,10 @@
 //! source <oneapi_install_directory>/setvars.sh
 //! ```
 //!
+//! This project was tested on oneAPI Toolkit 2026.1 and requires the Unified Runtime over Level Zero
+//! driver version 1.14.37020 or newer. For more detailed information check out the
+//! [required extensions](crate#required-extensions) section.
+//!
 //! # Getting started
 //! ### Building the crate
 //! Before building this crate you need to source the `setvars.sh` file. You can then build it as
@@ -31,57 +35,67 @@
 //! You must also source `setvars.sh` before running any SYCL program.
 //!
 //! ### Hello world
-//! 1. Create a [`Queue`](crate::queue::Queue). It's the main entry point to the SYCL API.
-//! ```rust,ignore
-//! let mut queue = Queue::new();
-//! ```
 //!
-//! 2. Create an [USM buffer](crate::buffer::Buffer) for your data.
-//! ```rust,ignore
-//! let mut device_buffer = queue.alloc_device::<f64>(1024).wait();
 //! ```
+//! # use sycl_rs::prelude::*;
 //!
-//! 3. Build a SYCL kernel.
-//! ```rust,ignore
-//! let kernel = queue
-//!     .get_context()
-//!     .create_kernel_bundle_from_source(IOTA_SRC)
-//!     .build()
-//!     .get_kernel("iota");
-//! ```
+//! # static IOTA_SRC: &str = r#"
+//! # #include <sycl/sycl.hpp>
+//! # 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<float>(id);
+//! # }
+//! # "#;
+//! #
+//! fn main() -> sycl_rs::Result<()> {
+//!     // 1. Create a Queue. It's the main entry point to the SYCL API.
+//!     let mut queue = Queue::new();
+//!     let mut device_buffer = queue.alloc_device::<f32>(1024)?.wait()?;
 //!
-//! 4. Launch your kernel.
-//! ```rust,ignore
-//! unsafe {
-//!     queue.launch(
-//!         NdRange::new([1024], [16]),
-//!         &kernel,
-//!         (3.14, &mut device_buffer),
-//!     )
-//! }
-//! .wait();
-//! ```
+//!     // 3. Build a SYCL kernel.
+//!     let kernel = queue
+//!         .get_context()
+//!         .create_kernel_bundle_from_source(IOTA_SRC)?
+//!         .build()?
+//!         .get_kernel("iota")?;
 //!
-//! 5. Copy your data to the host.
-//! ```rust,ignore
-//! let mut host_buffer = queue.alloc_host::<f64>(1024).wait();
-//! queue.copy(&device_buffer, &mut host_buffer).wait();
-//! ```
+//!     // 4. Launch your kernel.
+//!     unsafe {
+//!         queue.launch(
+//!             NdRange::new([1024], [16]),
+//!             &kernel,
+//!             (3.14_f32, &mut device_buffer),
+//!         )
+//!     }?
+//!     .wait()?;
 //!
-//! You can access your host data just like a normal Rust slice.
-//! ```rust,ignore
-//! for e in host_buffer.iter() {
-//!     print!("{e} ");
+//!     let mut host_buffer = queue.alloc_host::<f32>(1024)?.wait()?;
+//!
+//!     // 5. Copy your data to the host.
+//!     queue.copy(&device_buffer, &mut host_buffer)?.wait()?;
+//!
+//!     // You can access your host data just like a normal Rust slice.
+//!     for e in host_buffer.iter() {
+//!         print!("{e} ");
+//!     }
+//!     println!();
+//!
+//!     Ok(())
 //! }
-//! println!();
 //! ```
 //!
 //! # Safety model
 //! - USM allocations are represented by a zero-cost `Buffer` type managed through RAII.
 //!   - Note: Unlike SYCL buffers, SYCL-rs buffers do not rely on accessors.
 //! - Buffers are zero-initialized by default.
 //! - Buffers can only store types that implement [`bytemuck::Pod`].
-//! - Kernel launch is inherently unsafe.
+//! - Kernel launch is inherently unsafe. In particular, the caller must ensure that every argument
+//!   has the correct representation, layout, and alignment.
 //!
 //! # Asynchronous programming model
 //! Each queue operation returns an [`Event`](`crate::event::Event`). You can synchronously
@@ -90,6 +104,19 @@
 //! You can also synchronously call [`Queue::wait()`](crate::queue::Queue::wait) to wait for a
 //! [`Queue`](crate::queue::Queue) directly. To do the same asynchronously you have to `.await` an
 //! event returned by [`Queue::barrier()`](crate::queue::Queue::barrier).
+//!
+//! All basic SYCL wrapper types (`Queue`, `Event`, `Context`, `Platform`, `Device`) are thread safe as
+//! indicated by the provided [`Send`] and [`Sync`] trait implementations. However - Buffers are
+//! not thread-safe. If you need a thread-safe Buffer you need to wrap it in an `Arc<Mutex<T>>`.
+//!
+//! # Required extensions
+//! This project requires the following SYCL extensions to work:
+//! - [sycl_ext_oneapi_kernel_compiler](https://github.com/intel/llvm/blob/sycl/sycl/doc/extensions/experimental/sycl_ext_oneapi_kernel_compiler.asciidoc)
+//! - [sycl_ext_oneapi_raw_kernel_arg](https://github.com/intel/llvm/blob/sycl/sycl/doc/extensions/experimental/sycl_ext_oneapi_raw_kernel_arg.asciidoc)
+//!
+//! The following extensions are also required for async support:
+//! - [sycl_ext_intel_queue_immediate_command_list](https://github.com/intel/llvm/blob/sycl/sycl/doc/extensions/supported/sycl_ext_intel_queue_immediate_command_list.asciidoc)
+//! - [sycl_ext_oneapi_enqueue_barrier](https://github.com/intel/llvm/blob/sycl/sycl/doc/extensions/supported/sycl_ext_oneapi_enqueue_barrier.asciidoc)
 
 pub mod buffer;
 pub mod context;

sycl/sycl-rs/src/queue.rs

@@ -137,13 +137,19 @@ impl Queue {
         ffi::barrier(&mut self.0, dep_events).map(Into::into)
     }
 
-    /// Performs a blocking wait for the completion of all enqueued tasks in the queue.
+    /// Performs a blocking wait for the completion of all enqueued tasks in the queue. Returns an
+    /// error if a synchronous SYCL exception occurs.
+    ///
+    /// Dropping the queue does not wait for its completion.
     pub fn wait(&mut self) -> Result<()> {
         ffi::wait(&mut self.0)
     }
 
     /// Enqueues a kernel object to the queue as an ND-range kernel, using the number of work-items
     /// specified by the [`NdRange`] nd_range.
+    ///
+    /// Safety: The caller must make sure each argument matches the launched SYCL kernel's
+    /// signature, including their respective size, layout and alignment.
     pub unsafe fn launch<const ARGC: usize, const DIMENSIONS: usize>(
         &mut self,
         nd_range: NdRange<DIMENSIONS>,

sycl/sycl-rs/src/usm.rs

@@ -22,15 +22,23 @@ pub struct UsmAllocator<T: UsmAllocatorKind> {
 }
 
 /// A marker trait for USM allocators.
+///
+/// Safety: a type implementing this trait must be a valid USM allocator managed by a SYCL runtime.
 pub unsafe trait UsmAlloc: Allocator {}
 
 unsafe impl<T: UsmAllocatorKind> UsmAlloc for UsmAllocator<T> {}
 
 pub trait UsmAllocatorKind {
+    /// Allocates uninitialized memory.
+    /// Safety: the caller must not read uninitialized memory. The caller must also free this
+    /// memory manually.
     unsafe fn alloc(alignment: usize, num_bytes: usize, queue: &Queue) -> CxxResult<*mut u8>;
 }
 
 /// A marker trait for host-accessible USM allocators.
+///
+/// Safety: a type implementing this trait must be a valid USM allocator managed by a SYCL runtime,
+/// that allocates memory accessible from the host.
 pub unsafe trait HostAccessible {}
 
 impl<T: UsmAllocatorKind> From<&Queue> for UsmAllocator<T> {
@@ -61,6 +69,7 @@ unsafe impl<T: UsmAllocatorKind> Allocator for UsmAllocator<T> {
 }
 
 /// An allocator for Device-side buffers
+///
 /// Safety: memory allocated by this allocator cannot be accessed on the host side
 #[allow(dead_code)]
 pub struct DeviceAllocator;