tilelang.cpu.transform

CPU-specific transformation frontends.

Functions

LowerCPUKernelLaunch()	Prepare grid markers and per-launch thread counts before MaterializeKernelLaunch.
LowerCPUAtomics()	Lower tl.atomic_*_elem_op intrinsics to serial read-modify-write.
MarkCPUAtomics()	Mark atomic kernels before LowerTileOp to prevent parallel RMW races.
MaterializeCPUParallelGrid()	Convert the annotated CPU grid loop nest to OpenMP parallel loops.

Package Contents

tilelang.cpu.transform.LowerCPUKernelLaunch()

Prepare grid markers and per-launch thread counts before MaterializeKernelLaunch.

tilelang.cpu.transform.LowerCPUAtomics()

Lower tl.atomic_*_elem_op intrinsics to serial read-modify-write.

CPU targets only: the pass rewrites scalar-path atomic intrinsics into plain BufferLoad/BufferStore RMW so that both CPU codegens (c and llvm) can consume them. Tile-region atomics are lowered separately by the cpu.AtomicAdd/cpu.AtomicReduce impls inside LowerTileOp.

Returns:

fpass – The result pass

Return type:

tvm.transform.Pass

tilelang.cpu.transform.MarkCPUAtomics()

Mark atomic kernels before LowerTileOp to prevent parallel RMW races.

tilelang.cpu.transform.MaterializeCPUParallelGrid()

Convert the annotated CPU grid loop nest to OpenMP parallel loops.

CPU targets only; runs at the tail of the CPU pipeline (only inserted when the tl.cpu_parallel pass config is enabled). Consumes the tl.cpu_grid_dim annotations added by LowerCPUKernelLaunch, converts the grid loops to ForKind::kParallel (all dims on the c target for collapse(n); the first non-unit dim on llvm), and sinks function-scope allocations whose uses all lie inside the parallel region into the parallel loop body for per-worker privacy.

Returns:

fpass – The result pass

Return type:

tvm.transform.Pass