Expand description
Per-actor *Dispatch traits + their *DispatchCtx bundles
(Phase 0.3).
Each kernel actor (cuBLAS, cuBLASLt, cuDNN, cuFFT, cuRAND, cuSOLVER, cuSPARSE, cuTENSOR, NCCL, NVRTC) eventually exposes a typed public API like:
blas_actor.tell(BlasMsg::gemm::<f16>(GemmRequest { ... }));Internally, the request is boxed as Box<dyn GemmDispatch> and
the actor’s handle loop calls dispatch(self, &ctx) which carries
the cuBLAS handle, stream, and completion strategy. This avoids
the N-fold (op × dtype) variant explosion in the actor’s Msg
enum without giving up typed GpuRef<T> requests on the public
API.
§Status
In Phase 0.3 only NVRTC actually adopts the pattern: see
NvrtcLaunchDispatch + NvrtcDispatchCtx and the migrated
NvrtcActor. The remaining actor traits
(GemmDispatch, BlasLtDispatch, …) ship as stub trait
declarations with *DispatchCtx<'a> placeholder structs whose
handle fields are PhantomData until the matching actor migrates
in a follow-up PR (per the migration order in the Phase 0 plan).
The shared kernel-arg traits (DevSliceArg, ScalarArg)
collapse the per-dtype KernelArg::DevSlice* / KernelArg::Scalar*
variants into a single boxed-dyn pair plus a literal Usize(usize)
variant.
Re-exports§
pub use crate::kernel::solver::SolverDispatch;
Structs§
- Blas
Dispatch Ctx - Bundle of resources every cuBLAS dispatcher needs to run an op.
- Blas
LtDispatch Ctx - Per-call context handed to a
BlasLtDispatch::dispatchimpl. - Collective
Dispatch Ctx - Per-call context handed to a
CollectiveDispatch::dispatchimpl. Carries the NCCL communicator (cudarc wraps it) plus the device state and completion strategy. - Cudnn
Dispatch Ctx CudnnDispatchis owned bykernel::cudnn(Phase 2 cuDNN). Re-exported here for symmetry with other actors’ dispatch traits. Context handed to aCudnnDispatch::dispatchcall.- FftDispatch
Ctx - Per-execution context bundle handed to every
FftDispatch::dispatch. The actor packs its current stream + completion strategy + plan handle (already resolved against the LRU cache) so dispatch impls stay lean. - Nvrtc
Dispatch Ctx - Per-launch context bundle for
NvrtcLaunchDispatch::dispatch. - Send
Sparse Handle - Phase 4 cuSPARSE handle wrapper. Raw pointer is
!Sendby default; cuSPARSE is thread-safe per-handle as long as a given handle is only touched by one stream at a time (the actor-per-handle invariant). - Sparse
Dispatch Ctx - Per-call context handed to a
SparseDispatch::dispatchimpl. - Tensor
Dispatch Ctx TensorDispatchis owned bykernel::tensor(Phase 2 cuTENSOR).- Workspace
Pool TensorDispatchis owned bykernel::tensor(Phase 2 cuTENSOR).
Enums§
- DispatchD
Type - Alias used by the NCCL CollectiveDispatch (Phase 2). Maps onto the
canonical
atomr_accel::DType. Compact discriminant forAccelDtype::KIND. - Sparse
Op - Op-kind tag a
SparseDispatchexposes.
Traits§
- Blas
L1Dispatch - Erased L1 ops: axpy, dot, nrm2, scal, asum, iamax, iamin, copy, swap, rot.
- Blas
L2Dispatch - Erased L2 ops: gemv, ger.
- Blas
L3Dispatch - Erased L3 ops other than gemm: geam, syrk, trsm.
- Blas
LtDispatch - Boxed-dispatch trait the cuBLASLt actor uses to call into a typed
MatmulRequest<T>after type-erasing it through the mailbox. - Collective
Dispatch - Boxed-dispatch trait for NCCL collectives. The
CollectiveActorhandles the message envelope; each typed request struct (e.g.AllReduceRequest<T: NcclReduceSupported>) implements this so the actor stays single-mailbox while the dtype dimension travels in the box. - Cudnn
Dispatch CudnnDispatchis owned bykernel::cudnn(Phase 2 cuDNN). Re-exported here for symmetry with other actors’ dispatch traits. Dispatch trait for typed cuDNN ops.- DevSlice
Arg - Type-erased typed-device-slice argument for an NVRTC kernel launch.
- FftDispatch
- Dispatch trait for typed cuFFT requests (
FftRequest<T>forT: FftSupported). - Gemm
Dispatch - Erased
GemmRequest<T>. Implementors live inkernel::blas::gemm. - Gemm
Strided Batched Dispatch - Erased
GemmStridedBatchedRequest<T>. - Nvrtc
Launch Dispatch - Boxed-dispatch trait for an NVRTC kernel launch request.
- RngDispatch
- Erased payload accepted by
RngActorviaRngMsg::Fill. - Scalar
Arg - Type-erased typed-host-scalar argument for an NVRTC kernel launch.
- Sparse
Dispatch - Box-erased cuSPARSE op (Phase 4).
- Tensor
Dispatch TensorDispatchis owned bykernel::tensor(Phase 2 cuTENSOR).
Functions§
Type Aliases§
- Gemm
Dispatch Ctx GemmDispatchCtxis now an alias forBlasDispatchCtx(the cuBLAS agent unified the per-op contexts since every cuBLAS dispatcher needs the same set: cuBLAS handle, stream, completion, state).