tilelang.intrinsics.mma_layout

Attributes

shared_16x16_to_mma_32x8_layout
shared_16x16_to_mma_32x8_layout_trans

Functions

ldmatrix_32x8_to_shared_16x16_layout(thread_id, local_id)
ldmatrix_trans_32x8_to_shared_16x16_layout(thread_id, ...)
ldmatrix_16x32_to_shared_16x32_layout_a(thread_id, ...)
ldmatrix_16x32_to_shared_16x32_layout_b(thread_id, ...)
ldmatrix_32x16_to_shared_16x32_layout_a(thread_id, ...)
ldmatrix_32x16_to_shared_16x32_layout_b(thread_id, ...)
mma_store_32x8_to_shared_16x16_layout(thread_id, local_id)
shared_16x16_to_mma_32x8_layout_sr(i, j)
shared_16x16_to_mma_32x8_layout_rs(i, j)
shared_16x32_to_mma_32x16_layout(i, j)
shared_32x16_to_mma_32x16_layout(i, j)
mma_32x8_to_shared_16x16_layout(thread_id, local_id)
shared_16x16_to_mma_32x8_smoothlayout(i, j)
shared_16x32_to_mma_32x16_smoothlayout(i, j)
shared_32x16_to_mma_32x16_smoothlayout(i, j)
get_swizzle_layout(row_idx, col_idx, row_size, dtype)
make_mma_swizzle_layout(shared_buf[, is_smooth])

Module Contents

tilelang.intrinsics.mma_layout.ldmatrix_32x8_to_shared_16x16_layout(thread_id, local_id)

tilelang.intrinsics.mma_layout.ldmatrix_trans_32x8_to_shared_16x16_layout(thread_id, local_id)

tilelang.intrinsics.mma_layout.ldmatrix_16x32_to_shared_16x32_layout_a(thread_id, local_id)

tilelang.intrinsics.mma_layout.ldmatrix_16x32_to_shared_16x32_layout_b(thread_id, local_id)

tilelang.intrinsics.mma_layout.ldmatrix_32x16_to_shared_16x32_layout_a(thread_id, local_id)

tilelang.intrinsics.mma_layout.ldmatrix_32x16_to_shared_16x32_layout_b(thread_id, local_id)

tilelang.intrinsics.mma_layout.mma_store_32x8_to_shared_16x16_layout(thread_id, local_id)

tilelang.intrinsics.mma_layout.shared_16x16_to_mma_32x8_layout_sr(i, j)

tilelang.intrinsics.mma_layout.shared_16x16_to_mma_32x8_layout_rs(i, j)

tilelang.intrinsics.mma_layout.shared_16x16_to_mma_32x8_layout

tilelang.intrinsics.mma_layout.shared_16x16_to_mma_32x8_layout_trans

tilelang.intrinsics.mma_layout.shared_16x32_to_mma_32x16_layout(i, j)

tilelang.intrinsics.mma_layout.shared_32x16_to_mma_32x16_layout(i, j)

tilelang.intrinsics.mma_layout.mma_32x8_to_shared_16x16_layout(thread_id, local_id)

tilelang.intrinsics.mma_layout.shared_16x16_to_mma_32x8_smoothlayout(i, j)

tilelang.intrinsics.mma_layout.shared_16x32_to_mma_32x16_smoothlayout(i, j)

tilelang.intrinsics.mma_layout.shared_32x16_to_mma_32x16_smoothlayout(i, j)

tilelang.intrinsics.mma_layout.get_swizzle_layout(row_idx, col_idx, row_size, dtype, swizzle_bytes=None)

Parameters:

dtype (Union[tvm.DataType, str])

tilelang.intrinsics.mma_layout.make_mma_swizzle_layout(shared_buf, is_smooth=False)

Parameters:

is_smooth (bool)