Struct CutlassI8Gemm 

pub struct CutlassI8Gemm {
    pub tb_m: usize,
    pub tb_n: usize,
    pub tb_k: usize,
    pub warp_m: usize,
    pub warp_n: usize,
    pub warp_k: usize,
    pub inst_m: usize,
    pub inst_n: usize,
    pub inst_k: usize,
    pub stages: usize,
}

A CUTLASS int8 tensorop-GEMM config. cuBLAS’s own int8 GEMM kernel IS a CUTLASS kernel — disassembly of cublasGemmEx (CUDA_R_8I, TN, 2048^3, sm_86) shows it dispatch cutlass_80_tensorop_i16832gemm_s8_256x128_64x3_tn_align16. Because CUTLASS is open source, our compiler can emit the same kernel and reach cuBLAS parity (measured 99.6-100.0% head-to-head), instead of approximating it with a hand-written kernel that tops out at ~90% (ptxas will not schedule our from-scratch source into the perfect pointer-increment addressing CUTLASS’s templates coax out of it).

Fields

tb_m: usize

tb_n: usize

tb_k: usize

warp_m: usize

warp_n: usize

warp_k: usize

inst_m: usize

inst_n: usize

inst_k: usize

stages: usize

Implementations

impl CutlassI8Gemm

pub fn cublas_reference() -> Self

The exact config cuBLAS dispatches (i16832gemm_s8_256x128_64x3_tn): 256x128x64 threadblock, 64x64x64 warp, m16n8k32 IMMA, 3-stage. Verified 99.6-100.0% parity — cuBLAS IS this kernel.

pub fn is_legal(&self) -> bool

Legality on Ampere: warp tile tiles the threadblock; instruction tiles the warp; shared fits.

Trait Implementations

impl Clone for CutlassI8Gemm

fn clone(&self) -> CutlassI8Gemm

Returns a duplicate of the value.

fn clone_from(&mut self, source: &Self)

Performs copy-assignment from source.

impl Debug for CutlassI8Gemm

fn fmt(&self, f: &mut Formatter<'_>) -> Result

Formats the value using the given formatter.

impl PartialEq for CutlassI8Gemm

fn eq(&self, other: &CutlassI8Gemm) -> bool

Tests for self and other values to be equal, and is used by ==.

fn ne(&self, other: &Rhs) -> bool

Tests for !=. The default implementation is almost always sufficient, and should not be overridden without very good reason.

impl Copy for CutlassI8Gemm

impl Eq for CutlassI8Gemm

impl StructuralPartialEq for CutlassI8Gemm