TileFmhaFwdSplitKVTraits< kPadSeqLenQ_, kPadSeqLenK_, kPadHeadDimQ_, kPadHeadDimV_, kHasLogitsSoftCap_, BiasEnum_, kHasBiasGrad_, kStoreLSE_, kDoFp8StaticQuant_, kIsPagedKV_, kHasUnevenSplits_, kMergeNumHeadGroupsSeqLenQ_, kBlockPerCu_ > Struct Template Reference

TileFmhaFwdSplitKVTraits&lt; kPadSeqLenQ_, kPadSeqLenK_, kPadHeadDimQ_, kPadHeadDimV_, kHasLogitsSoftCap_, BiasEnum_, kHasBiasGrad_, kStoreLSE_, kDoFp8StaticQuant_, kIsPagedKV_, kHasUnevenSplits_, kMergeNumHeadGroupsSeqLenQ_, kBlockPerCu_ &gt; Struct Template Reference#

Composable Kernel: ck_tile::TileFmhaFwdSplitKVTraits< kPadSeqLenQ_, kPadSeqLenK_, kPadHeadDimQ_, kPadHeadDimV_, kHasLogitsSoftCap_, BiasEnum_, kHasBiasGrad_, kStoreLSE_, kDoFp8StaticQuant_, kIsPagedKV_, kHasUnevenSplits_, kMergeNumHeadGroupsSeqLenQ_, kBlockPerCu_ > Struct Template Reference
Static Public Attributes | List of all members
ck_tile::TileFmhaFwdSplitKVTraits< kPadSeqLenQ_, kPadSeqLenK_, kPadHeadDimQ_, kPadHeadDimV_, kHasLogitsSoftCap_, BiasEnum_, kHasBiasGrad_, kStoreLSE_, kDoFp8StaticQuant_, kIsPagedKV_, kHasUnevenSplits_, kMergeNumHeadGroupsSeqLenQ_, kBlockPerCu_ > Struct Template Reference

#include <tile_fmha_traits.hpp>

Static Public Attributes

static constexpr bool kPadSeqLenQ = kPadSeqLenQ_
static constexpr bool kPadSeqLenK = kPadSeqLenK_
static constexpr bool kPadHeadDimQ = kPadHeadDimQ_
static constexpr bool kPadHeadDimV = kPadHeadDimV_
static constexpr bool kHasLogitsSoftCap = kHasLogitsSoftCap_
static constexpr auto BiasEnum = BiasEnum_
static constexpr bool kHasBiasGrad = kHasBiasGrad_
static constexpr bool kStoreLSE = kStoreLSE_
static constexpr bool kDoFp8StaticQuant = kDoFp8StaticQuant_
static constexpr bool kIsPagedKV = kIsPagedKV_
static constexpr bool kHasUnevenSplits = kHasUnevenSplits_
static constexpr bool kMergeNumHeadGroupsSeqLenQ = kMergeNumHeadGroupsSeqLenQ_
static constexpr index_t kBlockPerCu = kBlockPerCu_

Member Data Documentation

◆ BiasEnum

template<bool kPadSeqLenQ_, bool kPadSeqLenK_, bool kPadHeadDimQ_, bool kPadHeadDimV_, bool kHasLogitsSoftCap_, BlockAttentionBiasEnum BiasEnum_, bool kHasBiasGrad_, bool kStoreLSE_, bool kDoFp8StaticQuant_, bool kIsPagedKV_, bool kHasUnevenSplits_, bool kMergeNumHeadGroupsSeqLenQ_ = false, index_t kBlockPerCu_ = -1>
auto ck_tile::TileFmhaFwdSplitKVTraits< kPadSeqLenQ_, kPadSeqLenK_, kPadHeadDimQ_, kPadHeadDimV_, kHasLogitsSoftCap_, BiasEnum_, kHasBiasGrad_, kStoreLSE_, kDoFp8StaticQuant_, kIsPagedKV_, kHasUnevenSplits_, kMergeNumHeadGroupsSeqLenQ_, kBlockPerCu_ >::BiasEnum = BiasEnum_
staticconstexpr

◆ kBlockPerCu

template<bool kPadSeqLenQ_, bool kPadSeqLenK_, bool kPadHeadDimQ_, bool kPadHeadDimV_, bool kHasLogitsSoftCap_, BlockAttentionBiasEnum BiasEnum_, bool kHasBiasGrad_, bool kStoreLSE_, bool kDoFp8StaticQuant_, bool kIsPagedKV_, bool kHasUnevenSplits_, bool kMergeNumHeadGroupsSeqLenQ_ = false, index_t kBlockPerCu_ = -1>
index_t ck_tile::TileFmhaFwdSplitKVTraits< kPadSeqLenQ_, kPadSeqLenK_, kPadHeadDimQ_, kPadHeadDimV_, kHasLogitsSoftCap_, BiasEnum_, kHasBiasGrad_, kStoreLSE_, kDoFp8StaticQuant_, kIsPagedKV_, kHasUnevenSplits_, kMergeNumHeadGroupsSeqLenQ_, kBlockPerCu_ >::kBlockPerCu = kBlockPerCu_
staticconstexpr

◆ kDoFp8StaticQuant

template<bool kPadSeqLenQ_, bool kPadSeqLenK_, bool kPadHeadDimQ_, bool kPadHeadDimV_, bool kHasLogitsSoftCap_, BlockAttentionBiasEnum BiasEnum_, bool kHasBiasGrad_, bool kStoreLSE_, bool kDoFp8StaticQuant_, bool kIsPagedKV_, bool kHasUnevenSplits_, bool kMergeNumHeadGroupsSeqLenQ_ = false, index_t kBlockPerCu_ = -1>
bool ck_tile::TileFmhaFwdSplitKVTraits< kPadSeqLenQ_, kPadSeqLenK_, kPadHeadDimQ_, kPadHeadDimV_, kHasLogitsSoftCap_, BiasEnum_, kHasBiasGrad_, kStoreLSE_, kDoFp8StaticQuant_, kIsPagedKV_, kHasUnevenSplits_, kMergeNumHeadGroupsSeqLenQ_, kBlockPerCu_ >::kDoFp8StaticQuant = kDoFp8StaticQuant_
staticconstexpr

◆ kHasBiasGrad

template<bool kPadSeqLenQ_, bool kPadSeqLenK_, bool kPadHeadDimQ_, bool kPadHeadDimV_, bool kHasLogitsSoftCap_, BlockAttentionBiasEnum BiasEnum_, bool kHasBiasGrad_, bool kStoreLSE_, bool kDoFp8StaticQuant_, bool kIsPagedKV_, bool kHasUnevenSplits_, bool kMergeNumHeadGroupsSeqLenQ_ = false, index_t kBlockPerCu_ = -1>
bool ck_tile::TileFmhaFwdSplitKVTraits< kPadSeqLenQ_, kPadSeqLenK_, kPadHeadDimQ_, kPadHeadDimV_, kHasLogitsSoftCap_, BiasEnum_, kHasBiasGrad_, kStoreLSE_, kDoFp8StaticQuant_, kIsPagedKV_, kHasUnevenSplits_, kMergeNumHeadGroupsSeqLenQ_, kBlockPerCu_ >::kHasBiasGrad = kHasBiasGrad_
staticconstexpr

◆ kHasLogitsSoftCap

template<bool kPadSeqLenQ_, bool kPadSeqLenK_, bool kPadHeadDimQ_, bool kPadHeadDimV_, bool kHasLogitsSoftCap_, BlockAttentionBiasEnum BiasEnum_, bool kHasBiasGrad_, bool kStoreLSE_, bool kDoFp8StaticQuant_, bool kIsPagedKV_, bool kHasUnevenSplits_, bool kMergeNumHeadGroupsSeqLenQ_ = false, index_t kBlockPerCu_ = -1>
bool ck_tile::TileFmhaFwdSplitKVTraits< kPadSeqLenQ_, kPadSeqLenK_, kPadHeadDimQ_, kPadHeadDimV_, kHasLogitsSoftCap_, BiasEnum_, kHasBiasGrad_, kStoreLSE_, kDoFp8StaticQuant_, kIsPagedKV_, kHasUnevenSplits_, kMergeNumHeadGroupsSeqLenQ_, kBlockPerCu_ >::kHasLogitsSoftCap = kHasLogitsSoftCap_
staticconstexpr

◆ kHasUnevenSplits

template<bool kPadSeqLenQ_, bool kPadSeqLenK_, bool kPadHeadDimQ_, bool kPadHeadDimV_, bool kHasLogitsSoftCap_, BlockAttentionBiasEnum BiasEnum_, bool kHasBiasGrad_, bool kStoreLSE_, bool kDoFp8StaticQuant_, bool kIsPagedKV_, bool kHasUnevenSplits_, bool kMergeNumHeadGroupsSeqLenQ_ = false, index_t kBlockPerCu_ = -1>
bool ck_tile::TileFmhaFwdSplitKVTraits< kPadSeqLenQ_, kPadSeqLenK_, kPadHeadDimQ_, kPadHeadDimV_, kHasLogitsSoftCap_, BiasEnum_, kHasBiasGrad_, kStoreLSE_, kDoFp8StaticQuant_, kIsPagedKV_, kHasUnevenSplits_, kMergeNumHeadGroupsSeqLenQ_, kBlockPerCu_ >::kHasUnevenSplits = kHasUnevenSplits_
staticconstexpr

◆ kIsPagedKV

template<bool kPadSeqLenQ_, bool kPadSeqLenK_, bool kPadHeadDimQ_, bool kPadHeadDimV_, bool kHasLogitsSoftCap_, BlockAttentionBiasEnum BiasEnum_, bool kHasBiasGrad_, bool kStoreLSE_, bool kDoFp8StaticQuant_, bool kIsPagedKV_, bool kHasUnevenSplits_, bool kMergeNumHeadGroupsSeqLenQ_ = false, index_t kBlockPerCu_ = -1>
bool ck_tile::TileFmhaFwdSplitKVTraits< kPadSeqLenQ_, kPadSeqLenK_, kPadHeadDimQ_, kPadHeadDimV_, kHasLogitsSoftCap_, BiasEnum_, kHasBiasGrad_, kStoreLSE_, kDoFp8StaticQuant_, kIsPagedKV_, kHasUnevenSplits_, kMergeNumHeadGroupsSeqLenQ_, kBlockPerCu_ >::kIsPagedKV = kIsPagedKV_
staticconstexpr

◆ kMergeNumHeadGroupsSeqLenQ

template<bool kPadSeqLenQ_, bool kPadSeqLenK_, bool kPadHeadDimQ_, bool kPadHeadDimV_, bool kHasLogitsSoftCap_, BlockAttentionBiasEnum BiasEnum_, bool kHasBiasGrad_, bool kStoreLSE_, bool kDoFp8StaticQuant_, bool kIsPagedKV_, bool kHasUnevenSplits_, bool kMergeNumHeadGroupsSeqLenQ_ = false, index_t kBlockPerCu_ = -1>
bool ck_tile::TileFmhaFwdSplitKVTraits< kPadSeqLenQ_, kPadSeqLenK_, kPadHeadDimQ_, kPadHeadDimV_, kHasLogitsSoftCap_, BiasEnum_, kHasBiasGrad_, kStoreLSE_, kDoFp8StaticQuant_, kIsPagedKV_, kHasUnevenSplits_, kMergeNumHeadGroupsSeqLenQ_, kBlockPerCu_ >::kMergeNumHeadGroupsSeqLenQ = kMergeNumHeadGroupsSeqLenQ_
staticconstexpr

◆ kPadHeadDimQ

template<bool kPadSeqLenQ_, bool kPadSeqLenK_, bool kPadHeadDimQ_, bool kPadHeadDimV_, bool kHasLogitsSoftCap_, BlockAttentionBiasEnum BiasEnum_, bool kHasBiasGrad_, bool kStoreLSE_, bool kDoFp8StaticQuant_, bool kIsPagedKV_, bool kHasUnevenSplits_, bool kMergeNumHeadGroupsSeqLenQ_ = false, index_t kBlockPerCu_ = -1>
bool ck_tile::TileFmhaFwdSplitKVTraits< kPadSeqLenQ_, kPadSeqLenK_, kPadHeadDimQ_, kPadHeadDimV_, kHasLogitsSoftCap_, BiasEnum_, kHasBiasGrad_, kStoreLSE_, kDoFp8StaticQuant_, kIsPagedKV_, kHasUnevenSplits_, kMergeNumHeadGroupsSeqLenQ_, kBlockPerCu_ >::kPadHeadDimQ = kPadHeadDimQ_
staticconstexpr

◆ kPadHeadDimV

template<bool kPadSeqLenQ_, bool kPadSeqLenK_, bool kPadHeadDimQ_, bool kPadHeadDimV_, bool kHasLogitsSoftCap_, BlockAttentionBiasEnum BiasEnum_, bool kHasBiasGrad_, bool kStoreLSE_, bool kDoFp8StaticQuant_, bool kIsPagedKV_, bool kHasUnevenSplits_, bool kMergeNumHeadGroupsSeqLenQ_ = false, index_t kBlockPerCu_ = -1>
bool ck_tile::TileFmhaFwdSplitKVTraits< kPadSeqLenQ_, kPadSeqLenK_, kPadHeadDimQ_, kPadHeadDimV_, kHasLogitsSoftCap_, BiasEnum_, kHasBiasGrad_, kStoreLSE_, kDoFp8StaticQuant_, kIsPagedKV_, kHasUnevenSplits_, kMergeNumHeadGroupsSeqLenQ_, kBlockPerCu_ >::kPadHeadDimV = kPadHeadDimV_
staticconstexpr

◆ kPadSeqLenK

template<bool kPadSeqLenQ_, bool kPadSeqLenK_, bool kPadHeadDimQ_, bool kPadHeadDimV_, bool kHasLogitsSoftCap_, BlockAttentionBiasEnum BiasEnum_, bool kHasBiasGrad_, bool kStoreLSE_, bool kDoFp8StaticQuant_, bool kIsPagedKV_, bool kHasUnevenSplits_, bool kMergeNumHeadGroupsSeqLenQ_ = false, index_t kBlockPerCu_ = -1>
bool ck_tile::TileFmhaFwdSplitKVTraits< kPadSeqLenQ_, kPadSeqLenK_, kPadHeadDimQ_, kPadHeadDimV_, kHasLogitsSoftCap_, BiasEnum_, kHasBiasGrad_, kStoreLSE_, kDoFp8StaticQuant_, kIsPagedKV_, kHasUnevenSplits_, kMergeNumHeadGroupsSeqLenQ_, kBlockPerCu_ >::kPadSeqLenK = kPadSeqLenK_
staticconstexpr

◆ kPadSeqLenQ

template<bool kPadSeqLenQ_, bool kPadSeqLenK_, bool kPadHeadDimQ_, bool kPadHeadDimV_, bool kHasLogitsSoftCap_, BlockAttentionBiasEnum BiasEnum_, bool kHasBiasGrad_, bool kStoreLSE_, bool kDoFp8StaticQuant_, bool kIsPagedKV_, bool kHasUnevenSplits_, bool kMergeNumHeadGroupsSeqLenQ_ = false, index_t kBlockPerCu_ = -1>
bool ck_tile::TileFmhaFwdSplitKVTraits< kPadSeqLenQ_, kPadSeqLenK_, kPadHeadDimQ_, kPadHeadDimV_, kHasLogitsSoftCap_, BiasEnum_, kHasBiasGrad_, kStoreLSE_, kDoFp8StaticQuant_, kIsPagedKV_, kHasUnevenSplits_, kMergeNumHeadGroupsSeqLenQ_, kBlockPerCu_ >::kPadSeqLenQ = kPadSeqLenQ_
staticconstexpr

◆ kStoreLSE

template<bool kPadSeqLenQ_, bool kPadSeqLenK_, bool kPadHeadDimQ_, bool kPadHeadDimV_, bool kHasLogitsSoftCap_, BlockAttentionBiasEnum BiasEnum_, bool kHasBiasGrad_, bool kStoreLSE_, bool kDoFp8StaticQuant_, bool kIsPagedKV_, bool kHasUnevenSplits_, bool kMergeNumHeadGroupsSeqLenQ_ = false, index_t kBlockPerCu_ = -1>
bool ck_tile::TileFmhaFwdSplitKVTraits< kPadSeqLenQ_, kPadSeqLenK_, kPadHeadDimQ_, kPadHeadDimV_, kHasLogitsSoftCap_, BiasEnum_, kHasBiasGrad_, kStoreLSE_, kDoFp8StaticQuant_, kIsPagedKV_, kHasUnevenSplits_, kMergeNumHeadGroupsSeqLenQ_, kBlockPerCu_ >::kStoreLSE = kStoreLSE_
staticconstexpr
The documentation for this struct was generated from the following file: