Struct MetraleRegistry 

pub struct MetraleRegistry { /* private fields */ }

2026-09-25: The kernel modules of one loaded model.

The blob set is the model’s own (the caller passes the modules for this checkpoint). Obtain one with MetraleRegistry::load and pass the Arc<MetraleRegistry> along; there is no global accessor. Dropping the last handle drops the registry’s module handles.

Implementations

impl MetraleRegistry

pub fn load( ordinal: usize, kernel_blobs: &[(&'static str, &'static [u8])], ) -> Result<Arc<Self>>

2026-09-25: Load this model’s kernel modules into the process CUDA context.

Each call produces a fresh, independent module set; it shares only the context and stream with a previously loaded model.

pub fn host(&self) -> &Arc<CudaHost>

2026-09-25: The process CUDA host this registry’s modules live in.

pub fn ctx(&self) -> &Arc<CudaContext>

pub fn stream(&self) -> &Arc<CudaStream>

pub fn module_names(&self) -> impl Iterator<Item = &'static str> + '_

2026-09-25: Module names this registry loaded.

pub fn function( &self, module_name: &str, func_name: &str, ) -> Result<CudaFunction>

2026-09-25: Look up func_name in module_name through cudarc. Not memoized; see Self::function_cached.

pub fn function_cached( &self, cache: &OnceLock<CudaFunction>, module_name: &str, func_name: &str, ) -> Result<CudaFunction>

2026-09-25: Self::function, memoized in the caller’s OnceLock.

pub fn has_module(&self, module_name: &str) -> bool

2026-09-25: Whether a module of this name was loaded for this run.

pub fn raw_function_cached( &self, cache: &OnceLock<RawCudaFunc>, module_name: &str, func_name: &str, ) -> Result<RawCudaFunc>

2026-09-25: Look up func_name through the raw driver API (cuModuleGetFunction), memoized in the caller’s OnceLock.

pub fn raw_stream(&self) -> u64

2026-09-25: The raw CUstream of the process stream held by CudaHost.

pub fn device_symbol( &self, module_name: &str, symbol: &str, ) -> Result<(u64, usize)>

2026-09-25: Resolve a __device__ symbol of a loaded module to its device pointer and byte length, for drivers that read or write device globals without a kernel (the InnerQ calibration state). symbol is the linker-visible name, so C++ namespace symbols are Itanium-mangled (_ZN7tq_plus14d_innerq_scaleE).

pub unsafe fn copy_h2d_async( &self, dst: u64, src: *const c_void, bytes: usize, stream: u64, ) -> Result<()>

2026-09-25: Async H2D copy into a device pointer.

Safety

dst must be a valid device pointer, and the bytes at src must outlive the copy (keep the host buffer until the next sync on stream).

pub unsafe fn copy_d2h_async( &self, dst: *mut c_void, src: u64, bytes: usize, stream: u64, ) -> Result<()>

2026-09-25: Async D2H copy from a device pointer.

Safety

dst must stay alive until stream is synchronised.

pub fn stream_synchronize(&self, stream: u64) -> Result<()>

2026-09-25: Block the calling thread until all prior work on stream completes.

pub unsafe fn launch_on_stream( &self, raw_func: RawCudaFunc, cfg: LaunchConfig, stream_ptr: u64, kernel_params: &mut [*mut c_void], ) -> Result<()>

2026-09-25: Launch a kernel on the raw CUDA stream stream_ptr.

Safety

• kernel_params must contain valid pointers to arguments matching the kernel signature.
• stream_ptr must be a valid CUstream handle, or 0 for the legacy default stream.
• raw_func must be a valid CUfunction obtained from raw_function_cached.

Trait Implementations

impl Drop for MetraleRegistry

fn drop(&mut self)

2026-09-25: Drains both module maps (see unload_raw); cudarc unloads each module when its last Arc<CudaModule> goes.

impl Send for MetraleRegistry

impl Sync for MetraleRegistry