BlockFmhaFwdSplitKVCombinePipelineDefaultPolicy Struct Reference

BlockFmhaFwdSplitKVCombinePipelineDefaultPolicy Struct Reference#

Composable Kernel: ck_tile::BlockFmhaFwdSplitKVCombinePipelineDefaultPolicy Struct Reference
ck_tile::BlockFmhaFwdSplitKVCombinePipelineDefaultPolicy Struct Reference

#include <block_fmha_fwd_splitkv_combine_pipeline_default_policy.hpp>

Static Public Member Functions

template<index_t NumWarps, index_t M, index_t N, typename DataType>
static CK_TILE_HOST_DEVICE constexpr auto GetMaxNumWarpsForTile ()
template<index_t NumWarps, index_t M, index_t N, typename DataType>
static CK_TILE_HOST_DEVICE constexpr auto GetVectorSizeForTile ()
template<typename Problem>
static CK_TILE_HOST_DEVICE constexpr auto GetAlignmentLSE ()
template<typename Problem>
static CK_TILE_HOST_DEVICE constexpr auto GetAlignmentOacc ()
template<typename Problem>
static CK_TILE_HOST_DEVICE constexpr auto GetAlignmentO ()
template<typename Problem>
static CK_TILE_HOST_DEVICE constexpr ck_tile::index_t GetSmemSizeLSEacc ()
template<typename Problem>
static CK_TILE_HOST_DEVICE constexpr ck_tile::index_t GetSmemSizeOacc ()
template<typename Problem>
static CK_TILE_HOST_DEVICE constexpr ck_tile::index_t GetSmemSize ()
template<typename Problem>
static CK_TILE_HOST_DEVICE constexpr auto MakeLSEaccDramTileDistribution ()
template<typename Problem>
static CK_TILE_HOST_DEVICE constexpr auto MakeLSEaccLdsStoreBlockDescriptor ()
template<typename Problem>
static CK_TILE_HOST_DEVICE constexpr auto MakeLSEaccLdsBlockDescriptor ()
template<typename Problem>
static CK_TILE_HOST_DEVICE constexpr auto MakeOaccLdsBlockDescriptor ()
template<typename Problem>
static CK_TILE_HOST_DEVICE constexpr auto MakeLSEaccRegTileDistribution ()
template<typename Problem>
static CK_TILE_HOST_DEVICE constexpr auto MakeOaccDramTileDistribution ()
template<typename Problem>
static CK_TILE_HOST_DEVICE constexpr auto MakeOaccResultDramTileDistribution ()

Member Function Documentation

◆ GetAlignmentLSE()

template<typename Problem>
CK_TILE_HOST_DEVICE constexpr auto ck_tile::BlockFmhaFwdSplitKVCombinePipelineDefaultPolicy::GetAlignmentLSE ( )
inlinestaticconstexpr

◆ GetAlignmentO()

template<typename Problem>
CK_TILE_HOST_DEVICE constexpr auto ck_tile::BlockFmhaFwdSplitKVCombinePipelineDefaultPolicy::GetAlignmentO ( )
inlinestaticconstexpr

◆ GetAlignmentOacc()

template<typename Problem>
CK_TILE_HOST_DEVICE constexpr auto ck_tile::BlockFmhaFwdSplitKVCombinePipelineDefaultPolicy::GetAlignmentOacc ( )
inlinestaticconstexpr

◆ GetMaxNumWarpsForTile()

template<index_t NumWarps, index_t M, index_t N, typename DataType>
CK_TILE_HOST_DEVICE constexpr auto ck_tile::BlockFmhaFwdSplitKVCombinePipelineDefaultPolicy::GetMaxNumWarpsForTile ( )
inlinestaticconstexpr

◆ GetSmemSize()

template<typename Problem>
CK_TILE_HOST_DEVICE constexpr ck_tile::index_t ck_tile::BlockFmhaFwdSplitKVCombinePipelineDefaultPolicy::GetSmemSize ( )
inlinestaticconstexpr

◆ GetSmemSizeLSEacc()

template<typename Problem>
CK_TILE_HOST_DEVICE constexpr ck_tile::index_t ck_tile::BlockFmhaFwdSplitKVCombinePipelineDefaultPolicy::GetSmemSizeLSEacc ( )
inlinestaticconstexpr

◆ GetSmemSizeOacc()

template<typename Problem>
CK_TILE_HOST_DEVICE constexpr ck_tile::index_t ck_tile::BlockFmhaFwdSplitKVCombinePipelineDefaultPolicy::GetSmemSizeOacc ( )
inlinestaticconstexpr

◆ GetVectorSizeForTile()

template<index_t NumWarps, index_t M, index_t N, typename DataType>
CK_TILE_HOST_DEVICE constexpr auto ck_tile::BlockFmhaFwdSplitKVCombinePipelineDefaultPolicy::GetVectorSizeForTile ( )
inlinestaticconstexpr

◆ MakeLSEaccDramTileDistribution()

template<typename Problem>
CK_TILE_HOST_DEVICE constexpr auto ck_tile::BlockFmhaFwdSplitKVCombinePipelineDefaultPolicy::MakeLSEaccDramTileDistribution ( )
inlinestaticconstexpr

◆ MakeLSEaccLdsBlockDescriptor()

template<typename Problem>
CK_TILE_HOST_DEVICE constexpr auto ck_tile::BlockFmhaFwdSplitKVCombinePipelineDefaultPolicy::MakeLSEaccLdsBlockDescriptor ( )
inlinestaticconstexpr

◆ MakeLSEaccLdsStoreBlockDescriptor()

template<typename Problem>
CK_TILE_HOST_DEVICE constexpr auto ck_tile::BlockFmhaFwdSplitKVCombinePipelineDefaultPolicy::MakeLSEaccLdsStoreBlockDescriptor ( )
inlinestaticconstexpr

◆ MakeLSEaccRegTileDistribution()

template<typename Problem>
CK_TILE_HOST_DEVICE constexpr auto ck_tile::BlockFmhaFwdSplitKVCombinePipelineDefaultPolicy::MakeLSEaccRegTileDistribution ( )
inlinestaticconstexpr

◆ MakeOaccDramTileDistribution()

template<typename Problem>
CK_TILE_HOST_DEVICE constexpr auto ck_tile::BlockFmhaFwdSplitKVCombinePipelineDefaultPolicy::MakeOaccDramTileDistribution ( )
inlinestaticconstexpr

◆ MakeOaccLdsBlockDescriptor()

template<typename Problem>
CK_TILE_HOST_DEVICE constexpr auto ck_tile::BlockFmhaFwdSplitKVCombinePipelineDefaultPolicy::MakeOaccLdsBlockDescriptor ( )
inlinestaticconstexpr

◆ MakeOaccResultDramTileDistribution()

template<typename Problem>
CK_TILE_HOST_DEVICE constexpr auto ck_tile::BlockFmhaFwdSplitKVCombinePipelineDefaultPolicy::MakeOaccResultDramTileDistribution ( )
inlinestaticconstexpr

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