StreamKKernelArgs Struct Reference

• ck_tile
• StreamKKernel
• StreamKKernelArgs

Public Attributes | List of all members

ck_tile::StreamKKernel< TilePartitioner_, GemmPipeline_, EpiloguePipeline_ >::StreamKKernelArgs Struct Reference

ALayout and ADataType are expected to be scalars, not a tuple. More...

#include <streamk_gemm_kernel.hpp>

Inheritance diagram for ck_tile::StreamKKernel< TilePartitioner_, GemmPipeline_, EpiloguePipeline_ >::StreamKKernelArgs:

Public Attributes
StreamKReductionStrategy	reduction_strategy
	The strategy used by work groups to compute final results in C tensor.
uint32_t	num_sk_blocks
	The number of stream k blocks.
void *	workspace_ptr
	A pointer to a buffer in device memory for accumulating partial via reduction strategy.
TilePartitioner	tile_partitioner
	An instance of the TilePartioner class for assisting with mapping workgroups to the C tensor.
Public Attributes inherited from ck_tile::UniversalGemmKernelArgs< NumATensor, NumBTensor, NumDTensor >
const std::array< const void *, NumATensor >	as_ptr
	The As input tensor's pointer to device memory.
const std::array< const void *, NumBTensor >	bs_ptr
	The Bs input tensor's pointer to device memory.
const std::array< const void *, NumDTensor >	ds_ptr
	The Ds input tensor's pointer to device memory.
void *	e_ptr
	The E output tensor's pointer to device memory.
index_t	M
	GEMM's M dimension size.
index_t	N
	GEMM's N dimension size.
index_t	K
	GEMM's K dimension size.
std::array< index_t, NumATensor >	stride_As
	The distance between consecutive elements of non-contiguous dimension (in memory) of As tensor.
std::array< index_t, NumBTensor >	stride_Bs
	The distance between consecutive elements of non-contiguous dimension (in memory) of Bs tensor.
std::array< index_t, NumDTensor >	stride_Ds
	The distance between consecutive elements of non-contiguous dimension (in memory) of Ds tensor.
index_t	stride_E
	The distance between consecutive elements of non-contiguous dimension (in memory) of E tensor.
index_t	k_batch

Detailed Description

template<typename TilePartitioner_, typename GemmPipeline_, typename EpiloguePipeline_>
struct ck_tile::StreamKKernel< TilePartitioner_, GemmPipeline_, EpiloguePipeline_ >::StreamKKernelArgs

ALayout and ADataType are expected to be scalars, not a tuple.

BLayout and BDataType are expected to be scalars, not a tuple.

CLayout and CDataType are expected to be scalars, not a tuple.

Member Data Documentation

◆ num_sk_blocks

template<typename TilePartitioner_, typename GemmPipeline_, typename EpiloguePipeline_>

uint32_t ck_tile::StreamKKernel< TilePartitioner_, GemmPipeline_, EpiloguePipeline_ >::StreamKKernelArgs::num_sk_blocks

The number of stream k blocks.

◆ reduction_strategy

template<typename TilePartitioner_, typename GemmPipeline_, typename EpiloguePipeline_>

StreamKReductionStrategy ck_tile::StreamKKernel< TilePartitioner_, GemmPipeline_, EpiloguePipeline_ >::StreamKKernelArgs::reduction_strategy

The strategy used by work groups to compute final results in C tensor.

◆ tile_partitioner

template<typename TilePartitioner_, typename GemmPipeline_, typename EpiloguePipeline_>

TilePartitioner ck_tile::StreamKKernel< TilePartitioner_, GemmPipeline_, EpiloguePipeline_ >::StreamKKernelArgs::tile_partitioner

An instance of the TilePartioner class for assisting with mapping workgroups to the C tensor.

◆ workspace_ptr

template<typename TilePartitioner_, typename GemmPipeline_, typename EpiloguePipeline_>

void* ck_tile::StreamKKernel< TilePartitioner_, GemmPipeline_, EpiloguePipeline_ >::StreamKKernelArgs::workspace_ptr

A pointer to a buffer in device memory for accumulating partial via reduction strategy.

---

The documentation for this struct was generated from the following file:

• streamk_gemm_kernel.hpp