TileGemmQuantTraits< kPadM_, kPadN_, kPadK_, PreshuffleQuant_, PreshuffleB_, ALayout_, BLayout_, CLayout_, QuantType_, AQLayout_, BQLayout_, TransposeC_, DoubleSmemBuffer_, UsePersistentKernel_ > Struct Template Reference

TileGemmQuantTraits&lt; kPadM_, kPadN_, kPadK_, PreshuffleQuant_, PreshuffleB_, ALayout_, BLayout_, CLayout_, QuantType_, AQLayout_, BQLayout_, TransposeC_, DoubleSmemBuffer_, UsePersistentKernel_ &gt; Struct Template Reference#

Composable Kernel: ck_tile::TileGemmQuantTraits< kPadM_, kPadN_, kPadK_, PreshuffleQuant_, PreshuffleB_, ALayout_, BLayout_, CLayout_, QuantType_, AQLayout_, BQLayout_, TransposeC_, DoubleSmemBuffer_, UsePersistentKernel_ > Struct Template Reference
Public Types | Static Public Attributes | List of all members
ck_tile::TileGemmQuantTraits< kPadM_, kPadN_, kPadK_, PreshuffleQuant_, PreshuffleB_, ALayout_, BLayout_, CLayout_, QuantType_, AQLayout_, BQLayout_, TransposeC_, DoubleSmemBuffer_, UsePersistentKernel_ > Struct Template Reference

#include <tile_gemm_quant_traits.hpp>

Public Types

using ALayout = ALayout_
using BLayout = BLayout_
using CLayout = CLayout_
using AQLayout = AQLayout_
using BQLayout = BQLayout_
using AsLayout = ALayout_
using BsLayout = BLayout_

Static Public Attributes

static constexpr bool kPadM = kPadM_
static constexpr bool kPadN = kPadN_
static constexpr bool kPadK = kPadK_
static constexpr QuantType kQuantType = QuantType_
static constexpr int _VectorSize = 16
static constexpr bool DoubleSmemBuffer = DoubleSmemBuffer_
static constexpr bool TransposeC = TransposeC_
static constexpr bool UseStructuredSparsity = false
static constexpr index_t NumWaveGroups = 1
static constexpr bool UsePersistentKernel = UsePersistentKernel_
static constexpr bool PreshuffleQuant = PreshuffleQuant_
static constexpr bool PreshuffleB = PreshuffleB_

Member Typedef Documentation

◆ ALayout

template<bool kPadM_, bool kPadN_, bool kPadK_, bool PreshuffleQuant_, bool PreshuffleB_, typename ALayout_, typename BLayout_, typename CLayout_, QuantType QuantType_, typename AQLayout_ = ALayout_, typename BQLayout_ = BLayout_, bool TransposeC_ = false, bool DoubleSmemBuffer_ = false, bool UsePersistentKernel_ = false>
using ck_tile::TileGemmQuantTraits< kPadM_, kPadN_, kPadK_, PreshuffleQuant_, PreshuffleB_, ALayout_, BLayout_, CLayout_, QuantType_, AQLayout_, BQLayout_, TransposeC_, DoubleSmemBuffer_, UsePersistentKernel_ >::ALayout = ALayout_

◆ AQLayout

template<bool kPadM_, bool kPadN_, bool kPadK_, bool PreshuffleQuant_, bool PreshuffleB_, typename ALayout_, typename BLayout_, typename CLayout_, QuantType QuantType_, typename AQLayout_ = ALayout_, typename BQLayout_ = BLayout_, bool TransposeC_ = false, bool DoubleSmemBuffer_ = false, bool UsePersistentKernel_ = false>
using ck_tile::TileGemmQuantTraits< kPadM_, kPadN_, kPadK_, PreshuffleQuant_, PreshuffleB_, ALayout_, BLayout_, CLayout_, QuantType_, AQLayout_, BQLayout_, TransposeC_, DoubleSmemBuffer_, UsePersistentKernel_ >::AQLayout = AQLayout_

◆ AsLayout

template<bool kPadM_, bool kPadN_, bool kPadK_, bool PreshuffleQuant_, bool PreshuffleB_, typename ALayout_, typename BLayout_, typename CLayout_, QuantType QuantType_, typename AQLayout_ = ALayout_, typename BQLayout_ = BLayout_, bool TransposeC_ = false, bool DoubleSmemBuffer_ = false, bool UsePersistentKernel_ = false>
using ck_tile::TileGemmQuantTraits< kPadM_, kPadN_, kPadK_, PreshuffleQuant_, PreshuffleB_, ALayout_, BLayout_, CLayout_, QuantType_, AQLayout_, BQLayout_, TransposeC_, DoubleSmemBuffer_, UsePersistentKernel_ >::AsLayout = ALayout_

◆ BLayout

template<bool kPadM_, bool kPadN_, bool kPadK_, bool PreshuffleQuant_, bool PreshuffleB_, typename ALayout_, typename BLayout_, typename CLayout_, QuantType QuantType_, typename AQLayout_ = ALayout_, typename BQLayout_ = BLayout_, bool TransposeC_ = false, bool DoubleSmemBuffer_ = false, bool UsePersistentKernel_ = false>
using ck_tile::TileGemmQuantTraits< kPadM_, kPadN_, kPadK_, PreshuffleQuant_, PreshuffleB_, ALayout_, BLayout_, CLayout_, QuantType_, AQLayout_, BQLayout_, TransposeC_, DoubleSmemBuffer_, UsePersistentKernel_ >::BLayout = BLayout_

◆ BQLayout

template<bool kPadM_, bool kPadN_, bool kPadK_, bool PreshuffleQuant_, bool PreshuffleB_, typename ALayout_, typename BLayout_, typename CLayout_, QuantType QuantType_, typename AQLayout_ = ALayout_, typename BQLayout_ = BLayout_, bool TransposeC_ = false, bool DoubleSmemBuffer_ = false, bool UsePersistentKernel_ = false>
using ck_tile::TileGemmQuantTraits< kPadM_, kPadN_, kPadK_, PreshuffleQuant_, PreshuffleB_, ALayout_, BLayout_, CLayout_, QuantType_, AQLayout_, BQLayout_, TransposeC_, DoubleSmemBuffer_, UsePersistentKernel_ >::BQLayout = BQLayout_

◆ BsLayout

template<bool kPadM_, bool kPadN_, bool kPadK_, bool PreshuffleQuant_, bool PreshuffleB_, typename ALayout_, typename BLayout_, typename CLayout_, QuantType QuantType_, typename AQLayout_ = ALayout_, typename BQLayout_ = BLayout_, bool TransposeC_ = false, bool DoubleSmemBuffer_ = false, bool UsePersistentKernel_ = false>
using ck_tile::TileGemmQuantTraits< kPadM_, kPadN_, kPadK_, PreshuffleQuant_, PreshuffleB_, ALayout_, BLayout_, CLayout_, QuantType_, AQLayout_, BQLayout_, TransposeC_, DoubleSmemBuffer_, UsePersistentKernel_ >::BsLayout = BLayout_

◆ CLayout

template<bool kPadM_, bool kPadN_, bool kPadK_, bool PreshuffleQuant_, bool PreshuffleB_, typename ALayout_, typename BLayout_, typename CLayout_, QuantType QuantType_, typename AQLayout_ = ALayout_, typename BQLayout_ = BLayout_, bool TransposeC_ = false, bool DoubleSmemBuffer_ = false, bool UsePersistentKernel_ = false>
using ck_tile::TileGemmQuantTraits< kPadM_, kPadN_, kPadK_, PreshuffleQuant_, PreshuffleB_, ALayout_, BLayout_, CLayout_, QuantType_, AQLayout_, BQLayout_, TransposeC_, DoubleSmemBuffer_, UsePersistentKernel_ >::CLayout = CLayout_

Member Data Documentation

◆ _VectorSize

template<bool kPadM_, bool kPadN_, bool kPadK_, bool PreshuffleQuant_, bool PreshuffleB_, typename ALayout_, typename BLayout_, typename CLayout_, QuantType QuantType_, typename AQLayout_ = ALayout_, typename BQLayout_ = BLayout_, bool TransposeC_ = false, bool DoubleSmemBuffer_ = false, bool UsePersistentKernel_ = false>
int ck_tile::TileGemmQuantTraits< kPadM_, kPadN_, kPadK_, PreshuffleQuant_, PreshuffleB_, ALayout_, BLayout_, CLayout_, QuantType_, AQLayout_, BQLayout_, TransposeC_, DoubleSmemBuffer_, UsePersistentKernel_ >::_VectorSize = 16
staticconstexpr

◆ DoubleSmemBuffer

template<bool kPadM_, bool kPadN_, bool kPadK_, bool PreshuffleQuant_, bool PreshuffleB_, typename ALayout_, typename BLayout_, typename CLayout_, QuantType QuantType_, typename AQLayout_ = ALayout_, typename BQLayout_ = BLayout_, bool TransposeC_ = false, bool DoubleSmemBuffer_ = false, bool UsePersistentKernel_ = false>
bool ck_tile::TileGemmQuantTraits< kPadM_, kPadN_, kPadK_, PreshuffleQuant_, PreshuffleB_, ALayout_, BLayout_, CLayout_, QuantType_, AQLayout_, BQLayout_, TransposeC_, DoubleSmemBuffer_, UsePersistentKernel_ >::DoubleSmemBuffer = DoubleSmemBuffer_
staticconstexpr

◆ kPadK

template<bool kPadM_, bool kPadN_, bool kPadK_, bool PreshuffleQuant_, bool PreshuffleB_, typename ALayout_, typename BLayout_, typename CLayout_, QuantType QuantType_, typename AQLayout_ = ALayout_, typename BQLayout_ = BLayout_, bool TransposeC_ = false, bool DoubleSmemBuffer_ = false, bool UsePersistentKernel_ = false>
bool ck_tile::TileGemmQuantTraits< kPadM_, kPadN_, kPadK_, PreshuffleQuant_, PreshuffleB_, ALayout_, BLayout_, CLayout_, QuantType_, AQLayout_, BQLayout_, TransposeC_, DoubleSmemBuffer_, UsePersistentKernel_ >::kPadK = kPadK_
staticconstexpr

◆ kPadM

template<bool kPadM_, bool kPadN_, bool kPadK_, bool PreshuffleQuant_, bool PreshuffleB_, typename ALayout_, typename BLayout_, typename CLayout_, QuantType QuantType_, typename AQLayout_ = ALayout_, typename BQLayout_ = BLayout_, bool TransposeC_ = false, bool DoubleSmemBuffer_ = false, bool UsePersistentKernel_ = false>
bool ck_tile::TileGemmQuantTraits< kPadM_, kPadN_, kPadK_, PreshuffleQuant_, PreshuffleB_, ALayout_, BLayout_, CLayout_, QuantType_, AQLayout_, BQLayout_, TransposeC_, DoubleSmemBuffer_, UsePersistentKernel_ >::kPadM = kPadM_
staticconstexpr

◆ kPadN

template<bool kPadM_, bool kPadN_, bool kPadK_, bool PreshuffleQuant_, bool PreshuffleB_, typename ALayout_, typename BLayout_, typename CLayout_, QuantType QuantType_, typename AQLayout_ = ALayout_, typename BQLayout_ = BLayout_, bool TransposeC_ = false, bool DoubleSmemBuffer_ = false, bool UsePersistentKernel_ = false>
bool ck_tile::TileGemmQuantTraits< kPadM_, kPadN_, kPadK_, PreshuffleQuant_, PreshuffleB_, ALayout_, BLayout_, CLayout_, QuantType_, AQLayout_, BQLayout_, TransposeC_, DoubleSmemBuffer_, UsePersistentKernel_ >::kPadN = kPadN_
staticconstexpr

◆ kQuantType

template<bool kPadM_, bool kPadN_, bool kPadK_, bool PreshuffleQuant_, bool PreshuffleB_, typename ALayout_, typename BLayout_, typename CLayout_, QuantType QuantType_, typename AQLayout_ = ALayout_, typename BQLayout_ = BLayout_, bool TransposeC_ = false, bool DoubleSmemBuffer_ = false, bool UsePersistentKernel_ = false>
QuantType ck_tile::TileGemmQuantTraits< kPadM_, kPadN_, kPadK_, PreshuffleQuant_, PreshuffleB_, ALayout_, BLayout_, CLayout_, QuantType_, AQLayout_, BQLayout_, TransposeC_, DoubleSmemBuffer_, UsePersistentKernel_ >::kQuantType = QuantType_
staticconstexpr

◆ NumWaveGroups

template<bool kPadM_, bool kPadN_, bool kPadK_, bool PreshuffleQuant_, bool PreshuffleB_, typename ALayout_, typename BLayout_, typename CLayout_, QuantType QuantType_, typename AQLayout_ = ALayout_, typename BQLayout_ = BLayout_, bool TransposeC_ = false, bool DoubleSmemBuffer_ = false, bool UsePersistentKernel_ = false>
index_t ck_tile::TileGemmQuantTraits< kPadM_, kPadN_, kPadK_, PreshuffleQuant_, PreshuffleB_, ALayout_, BLayout_, CLayout_, QuantType_, AQLayout_, BQLayout_, TransposeC_, DoubleSmemBuffer_, UsePersistentKernel_ >::NumWaveGroups = 1
staticconstexpr

◆ PreshuffleB

template<bool kPadM_, bool kPadN_, bool kPadK_, bool PreshuffleQuant_, bool PreshuffleB_, typename ALayout_, typename BLayout_, typename CLayout_, QuantType QuantType_, typename AQLayout_ = ALayout_, typename BQLayout_ = BLayout_, bool TransposeC_ = false, bool DoubleSmemBuffer_ = false, bool UsePersistentKernel_ = false>
bool ck_tile::TileGemmQuantTraits< kPadM_, kPadN_, kPadK_, PreshuffleQuant_, PreshuffleB_, ALayout_, BLayout_, CLayout_, QuantType_, AQLayout_, BQLayout_, TransposeC_, DoubleSmemBuffer_, UsePersistentKernel_ >::PreshuffleB = PreshuffleB_
staticconstexpr

◆ PreshuffleQuant

template<bool kPadM_, bool kPadN_, bool kPadK_, bool PreshuffleQuant_, bool PreshuffleB_, typename ALayout_, typename BLayout_, typename CLayout_, QuantType QuantType_, typename AQLayout_ = ALayout_, typename BQLayout_ = BLayout_, bool TransposeC_ = false, bool DoubleSmemBuffer_ = false, bool UsePersistentKernel_ = false>
bool ck_tile::TileGemmQuantTraits< kPadM_, kPadN_, kPadK_, PreshuffleQuant_, PreshuffleB_, ALayout_, BLayout_, CLayout_, QuantType_, AQLayout_, BQLayout_, TransposeC_, DoubleSmemBuffer_, UsePersistentKernel_ >::PreshuffleQuant = PreshuffleQuant_
staticconstexpr

◆ TransposeC

template<bool kPadM_, bool kPadN_, bool kPadK_, bool PreshuffleQuant_, bool PreshuffleB_, typename ALayout_, typename BLayout_, typename CLayout_, QuantType QuantType_, typename AQLayout_ = ALayout_, typename BQLayout_ = BLayout_, bool TransposeC_ = false, bool DoubleSmemBuffer_ = false, bool UsePersistentKernel_ = false>
bool ck_tile::TileGemmQuantTraits< kPadM_, kPadN_, kPadK_, PreshuffleQuant_, PreshuffleB_, ALayout_, BLayout_, CLayout_, QuantType_, AQLayout_, BQLayout_, TransposeC_, DoubleSmemBuffer_, UsePersistentKernel_ >::TransposeC = TransposeC_
staticconstexpr

◆ UsePersistentKernel

template<bool kPadM_, bool kPadN_, bool kPadK_, bool PreshuffleQuant_, bool PreshuffleB_, typename ALayout_, typename BLayout_, typename CLayout_, QuantType QuantType_, typename AQLayout_ = ALayout_, typename BQLayout_ = BLayout_, bool TransposeC_ = false, bool DoubleSmemBuffer_ = false, bool UsePersistentKernel_ = false>
bool ck_tile::TileGemmQuantTraits< kPadM_, kPadN_, kPadK_, PreshuffleQuant_, PreshuffleB_, ALayout_, BLayout_, CLayout_, QuantType_, AQLayout_, BQLayout_, TransposeC_, DoubleSmemBuffer_, UsePersistentKernel_ >::UsePersistentKernel = UsePersistentKernel_
staticconstexpr

◆ UseStructuredSparsity

template<bool kPadM_, bool kPadN_, bool kPadK_, bool PreshuffleQuant_, bool PreshuffleB_, typename ALayout_, typename BLayout_, typename CLayout_, QuantType QuantType_, typename AQLayout_ = ALayout_, typename BQLayout_ = BLayout_, bool TransposeC_ = false, bool DoubleSmemBuffer_ = false, bool UsePersistentKernel_ = false>
bool ck_tile::TileGemmQuantTraits< kPadM_, kPadN_, kPadK_, PreshuffleQuant_, PreshuffleB_, ALayout_, BLayout_, CLayout_, QuantType_, AQLayout_, BQLayout_, TransposeC_, DoubleSmemBuffer_, UsePersistentKernel_ >::UseStructuredSparsity = false
staticconstexpr
The documentation for this struct was generated from the following file: