BlockFmhaBwdOGradDotO< Problem, Policy > Struct Template Reference

BlockFmhaBwdOGradDotO&lt; Problem, Policy &gt; Struct Template Reference#

Composable Kernel: ck_tile::BlockFmhaBwdOGradDotO< Problem, Policy > Struct Template Reference
Public Types | Public Member Functions | Static Public Member Functions | Static Public Attributes | List of all members
ck_tile::BlockFmhaBwdOGradDotO< Problem, Policy > Struct Template Reference

#include <block_fmha_bwd_dot_do_o.hpp>

Public Types

using ODataType = remove_cvref_t<typename Problem::ODataType>
using OGradDataType = remove_cvref_t<typename Problem::OGradDataType>
using DDataType = remove_cvref_t<typename Problem::DDataType>

Public Member Functions

template<typename ODramBlockWindowTmp, typename OGradDramBlockWindowTmp, typename DDramBlockWindowTmp>
CK_TILE_HOST_DEVICE void operator() (const ODramBlockWindowTmp &o_dram_block_window_tmp, const OGradDramBlockWindowTmp &do_dram_block_window_tmp, DDramBlockWindowTmp &d_dram_block_window_tmp, float p_undrop) const

Static Public Member Functions

static CK_TILE_HOST_DEVICE constexpr ck_tile::index_t GetSmemSize ()

Static Public Attributes

static constexpr index_t kBlockPerCu = Problem::kBlockPerCu
static constexpr index_t kBlockSize = Problem::kBlockSize
static constexpr index_t kVHeaddim = Problem::kVHeaddim
static constexpr bool kIsGroupMode = Problem::kIsGroupMode
static constexpr bool kPadSeqLenQ = Problem::kPadSeqLenQ
static constexpr bool kPadHeadDimV = Problem::kPadHeadDimV
static constexpr index_t kAlignmentO
static constexpr index_t kAlignmentOGrad

Member Typedef Documentation

◆ DDataType

template<typename Problem, typename Policy = BlockFmhaBwdPipelineDefaultPolicy>
using ck_tile::BlockFmhaBwdOGradDotO< Problem, Policy >::DDataType = remove_cvref_t<typename Problem::DDataType>

◆ ODataType

template<typename Problem, typename Policy = BlockFmhaBwdPipelineDefaultPolicy>
using ck_tile::BlockFmhaBwdOGradDotO< Problem, Policy >::ODataType = remove_cvref_t<typename Problem::ODataType>

◆ OGradDataType

template<typename Problem, typename Policy = BlockFmhaBwdPipelineDefaultPolicy>
using ck_tile::BlockFmhaBwdOGradDotO< Problem, Policy >::OGradDataType = remove_cvref_t<typename Problem::OGradDataType>

Member Function Documentation

◆ GetSmemSize()

template<typename Problem, typename Policy = BlockFmhaBwdPipelineDefaultPolicy>
CK_TILE_HOST_DEVICE constexpr ck_tile::index_t ck_tile::BlockFmhaBwdOGradDotO< Problem, Policy >::GetSmemSize ( )
inlinestaticconstexpr

◆ operator()()

template<typename Problem, typename Policy = BlockFmhaBwdPipelineDefaultPolicy>
template<typename ODramBlockWindowTmp, typename OGradDramBlockWindowTmp, typename DDramBlockWindowTmp>
CK_TILE_HOST_DEVICE void ck_tile::BlockFmhaBwdOGradDotO< Problem, Policy >::operator() ( const ODramBlockWindowTmp & o_dram_block_window_tmp,
const OGradDramBlockWindowTmp & do_dram_block_window_tmp,
DDramBlockWindowTmp & d_dram_block_window_tmp,
float p_undrop ) const
inline

Member Data Documentation

◆ kAlignmentO

template<typename Problem, typename Policy = BlockFmhaBwdPipelineDefaultPolicy>
index_t ck_tile::BlockFmhaBwdOGradDotO< Problem, Policy >::kAlignmentO
staticconstexpr
Initial value:
=
kPadHeadDimV ? 1 : Policy::template GetAlignmentO<Problem>()
ck_tile::BlockFmhaBwdOGradDotO::kPadHeadDimV
static constexpr bool kPadHeadDimV
Definition block_fmha_bwd_dot_do_o.hpp:24

◆ kAlignmentOGrad

template<typename Problem, typename Policy = BlockFmhaBwdPipelineDefaultPolicy>
index_t ck_tile::BlockFmhaBwdOGradDotO< Problem, Policy >::kAlignmentOGrad
staticconstexpr
Initial value:
=
kPadHeadDimV ? 1 : Policy::template GetAlignmentO<Problem>()

◆ kBlockPerCu

template<typename Problem, typename Policy = BlockFmhaBwdPipelineDefaultPolicy>
index_t ck_tile::BlockFmhaBwdOGradDotO< Problem, Policy >::kBlockPerCu = Problem::kBlockPerCu
staticconstexpr

◆ kBlockSize

template<typename Problem, typename Policy = BlockFmhaBwdPipelineDefaultPolicy>
index_t ck_tile::BlockFmhaBwdOGradDotO< Problem, Policy >::kBlockSize = Problem::kBlockSize
staticconstexpr

◆ kIsGroupMode

template<typename Problem, typename Policy = BlockFmhaBwdPipelineDefaultPolicy>
bool ck_tile::BlockFmhaBwdOGradDotO< Problem, Policy >::kIsGroupMode = Problem::kIsGroupMode
staticconstexpr

◆ kPadHeadDimV

template<typename Problem, typename Policy = BlockFmhaBwdPipelineDefaultPolicy>
bool ck_tile::BlockFmhaBwdOGradDotO< Problem, Policy >::kPadHeadDimV = Problem::kPadHeadDimV
staticconstexpr

◆ kPadSeqLenQ

template<typename Problem, typename Policy = BlockFmhaBwdPipelineDefaultPolicy>
bool ck_tile::BlockFmhaBwdOGradDotO< Problem, Policy >::kPadSeqLenQ = Problem::kPadSeqLenQ
staticconstexpr

◆ kVHeaddim

template<typename Problem, typename Policy = BlockFmhaBwdPipelineDefaultPolicy>
index_t ck_tile::BlockFmhaBwdOGradDotO< Problem, Policy >::kVHeaddim = Problem::kVHeaddim
staticconstexpr
The documentation for this struct was generated from the following file: