GemmMultiABDHostArgs< NumATensor, NumBTensor, NumDTensor > Struct Template Reference

GemmMultiABDHostArgs< NumATensor, NumBTensor, NumDTensor > Struct Template Reference#

Composable Kernel: ck_tile::GemmMultiABDHostArgs< NumATensor, NumBTensor, NumDTensor > Struct Template Reference

The MultiABD GEMM kernel host arguments. More...

#include <gemm_multi_abd_kernel.hpp>

Public Member Functions

CK_TILE_HOST GemmMultiABDHostArgs (const std::array< const void *, NumATensor > &as_ptr_, const std::array< const void *, NumBTensor > &bs_ptr_, const std::array< const void *, NumDTensor > &ds_ptr_, void *e_ptr_, index_t k_batch_, index_t M_, index_t N_, index_t K_, const std::array< index_t, NumATensor > &stride_As_, const std::array< index_t, NumBTensor > &stride_Bs_, const std::array< index_t, NumDTensor > &stride_Ds_, index_t stride_E_)

Public Attributes
const std::array< const void *, NumATensor >	as_ptr
const std::array< const void *, NumBTensor >	bs_ptr
const std::array< const void *, NumDTensor >	ds_ptr
union {
void * e_ptr
void * c_ptr
};
index_t	M
index_t	N
index_t	K
const std::array< index_t, NumATensor >	stride_As
const std::array< index_t, NumBTensor >	stride_Bs
const std::array< index_t, NumDTensor >	stride_Ds
union {
index_t stride_E
index_t stride_C
};
index_t	k_batch

Detailed Description

template<index_t NumATensor, index_t NumBTensor, index_t NumDTensor>
struct ck_tile::GemmMultiABDHostArgs< NumATensor, NumBTensor, NumDTensor >

The MultiABD GEMM kernel host arguments.

Overview: This structure is passed to GemmKernelMultiABD when creating kernel arguments object. It contain all necessary information required to build proper kernel argument and launch kernel on GPU. This structure defines the GEMM problem configuration by stating all required information like M,N,K sizes and respective strides. NumATensor describes the number of A tensors. The minimum number of tensors is 1(required). NumBTensor describes the number of B tensors. The minimum number of tensors is 1(required). NumDTensor describes the number of D tensors. The minimum number of tensors is 0(not required).

Constructor & Destructor Documentation

◆ GemmMultiABDHostArgs()

template<index_t NumATensor, index_t NumBTensor, index_t NumDTensor>

CK_TILE_HOST ck_tile::GemmMultiABDHostArgs< NumATensor, NumBTensor, NumDTensor >::GemmMultiABDHostArgs	(	const std::array< const void *, NumATensor > &	as_ptr_,
		const std::array< const void *, NumBTensor > &	bs_ptr_,
		const std::array< const void *, NumDTensor > &	ds_ptr_,
		void *	e_ptr_,
		index_t	k_batch_,
		index_t	M_,
		index_t	N_,
		index_t	K_,
		const std::array< index_t, NumATensor > &	stride_As_,
		const std::array< index_t, NumBTensor > &	stride_Bs_,
		const std::array< index_t, NumDTensor > &	stride_Ds_,
		index_t	stride_E_ )

inline

Member Data Documentation

◆ [union]

union { ... } ck_tile::GemmMultiABDHostArgs< NumATensor, NumBTensor, NumDTensor >

◆ [union]

union { ... } ck_tile::GemmMultiABDHostArgs< NumATensor, NumBTensor, NumDTensor >

◆ as_ptr

template<index_t NumATensor, index_t NumBTensor, index_t NumDTensor>

const std::array<const void*, NumATensor> ck_tile::GemmMultiABDHostArgs< NumATensor, NumBTensor, NumDTensor >::as_ptr

◆ bs_ptr

template<index_t NumATensor, index_t NumBTensor, index_t NumDTensor>

const std::array<const void*, NumBTensor> ck_tile::GemmMultiABDHostArgs< NumATensor, NumBTensor, NumDTensor >::bs_ptr

◆ c_ptr

template<index_t NumATensor, index_t NumBTensor, index_t NumDTensor>

void* ck_tile::GemmMultiABDHostArgs< NumATensor, NumBTensor, NumDTensor >::c_ptr

◆ ds_ptr

template<index_t NumATensor, index_t NumBTensor, index_t NumDTensor>

const std::array<const void*, NumDTensor> ck_tile::GemmMultiABDHostArgs< NumATensor, NumBTensor, NumDTensor >::ds_ptr

◆ e_ptr

template<index_t NumATensor, index_t NumBTensor, index_t NumDTensor>

void* ck_tile::GemmMultiABDHostArgs< NumATensor, NumBTensor, NumDTensor >::e_ptr

◆ K

template<index_t NumATensor, index_t NumBTensor, index_t NumDTensor>

index_t ck_tile::GemmMultiABDHostArgs< NumATensor, NumBTensor, NumDTensor >::K

◆ k_batch

template<index_t NumATensor, index_t NumBTensor, index_t NumDTensor>

index_t ck_tile::GemmMultiABDHostArgs< NumATensor, NumBTensor, NumDTensor >::k_batch

◆ M

template<index_t NumATensor, index_t NumBTensor, index_t NumDTensor>

index_t ck_tile::GemmMultiABDHostArgs< NumATensor, NumBTensor, NumDTensor >::M

◆ N

template<index_t NumATensor, index_t NumBTensor, index_t NumDTensor>

index_t ck_tile::GemmMultiABDHostArgs< NumATensor, NumBTensor, NumDTensor >::N

◆ stride_As

template<index_t NumATensor, index_t NumBTensor, index_t NumDTensor>

const std::array<index_t, NumATensor> ck_tile::GemmMultiABDHostArgs< NumATensor, NumBTensor, NumDTensor >::stride_As

◆ stride_Bs

template<index_t NumATensor, index_t NumBTensor, index_t NumDTensor>

const std::array<index_t, NumBTensor> ck_tile::GemmMultiABDHostArgs< NumATensor, NumBTensor, NumDTensor >::stride_Bs

◆ stride_C

template<index_t NumATensor, index_t NumBTensor, index_t NumDTensor>

index_t ck_tile::GemmMultiABDHostArgs< NumATensor, NumBTensor, NumDTensor >::stride_C

◆ stride_Ds

template<index_t NumATensor, index_t NumBTensor, index_t NumDTensor>

const std::array<index_t, NumDTensor> ck_tile::GemmMultiABDHostArgs< NumATensor, NumBTensor, NumDTensor >::stride_Ds

◆ stride_E

template<index_t NumATensor, index_t NumBTensor, index_t NumDTensor>

index_t ck_tile::GemmMultiABDHostArgs< NumATensor, NumBTensor, NumDTensor >::stride_E

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

gemm_multi_abd_kernel.hpp