tvm_tirx.backend.cuda.layout

Contents

tvm_tirx.backend.cuda.layout#

CUDA-specific memory and instruction layout factories.

Classes#

SwizzleMode(*values)

The swizzle mode of the TMA.

Functions#

tmem_datapath_layout(datapath, rows, cols[, ...])

Return the TileLayout for a tcgen05 MMA datapath.

tmem_mma_operand_layout(operand, shape, ...)

Resolve the TMEM TileLayout for a tcgen05 MMA operand from operand role + instruction params.

mma_atom_layout(dtype, swizzle_mode)

Generate the MMA-compatible shared-memory atom layout.

mma_atom_shape(dtype, swizzle_mode[, shape])

Generate the MMA-compatible shared-memory atom shape.

mma_shared_layout(dtype, swizzle_mode, shape)

Generate the MMA-compatible shared-memory layout for shape and dtype.
Contents