FmhaBwdCommonKargs Struct Reference

FmhaBwdCommonKargs Struct Reference#

Composable Kernel: ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs Struct Reference
Public Attributes | List of all members
ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs Struct Reference

#include <fmha_bwd_kernel.hpp>

Inheritance diagram for ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs:
ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdBatchModeKargs ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdGroupModeKargs

Public Attributes

const void * q_ptr
const void * k_ptr
const void * v_ptr
const void * lse_ptr
const void * do_ptr
const void * d_ptr
void * dq_acc_ptr
void * dk_ptr
void * dv_ptr
ck_tile::index_t seqlen_q
ck_tile::index_t seqlen_k
ck_tile::index_t hdim_q
ck_tile::index_t hdim_v
ck_tile::index_t num_head_q
ck_tile::index_t nhead_ratio_qk
float raw_scale
float scale
ck_tile::index_t stride_q
ck_tile::index_t stride_k
ck_tile::index_t stride_v
ck_tile::index_t stride_do
ck_tile::index_t stride_dq_acc
ck_tile::index_t stride_dk
ck_tile::index_t stride_dv
ck_tile::index_t nhead_stride_q
ck_tile::index_t nhead_stride_k
ck_tile::index_t nhead_stride_v
ck_tile::index_t nhead_stride_do
ck_tile::index_t nhead_stride_lsed
ck_tile::index_t nhead_stride_dq_acc
ck_tile::index_t nhead_stride_dk
ck_tile::index_t nhead_stride_dv

Member Data Documentation

◆ d_ptr

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
const void* ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::d_ptr

◆ dk_ptr

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
void* ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::dk_ptr

◆ do_ptr

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
const void* ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::do_ptr

◆ dq_acc_ptr

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
void* ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::dq_acc_ptr

◆ dv_ptr

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
void* ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::dv_ptr

◆ hdim_q

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
ck_tile::index_t ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::hdim_q

◆ hdim_v

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
ck_tile::index_t ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::hdim_v

◆ k_ptr

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
const void* ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::k_ptr

◆ lse_ptr

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
const void* ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::lse_ptr

◆ nhead_ratio_qk

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
ck_tile::index_t ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::nhead_ratio_qk

◆ nhead_stride_dk

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
ck_tile::index_t ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::nhead_stride_dk

◆ nhead_stride_do

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
ck_tile::index_t ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::nhead_stride_do

◆ nhead_stride_dq_acc

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
ck_tile::index_t ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::nhead_stride_dq_acc

◆ nhead_stride_dv

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
ck_tile::index_t ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::nhead_stride_dv

◆ nhead_stride_k

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
ck_tile::index_t ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::nhead_stride_k

◆ nhead_stride_lsed

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
ck_tile::index_t ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::nhead_stride_lsed

◆ nhead_stride_q

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
ck_tile::index_t ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::nhead_stride_q

◆ nhead_stride_v

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
ck_tile::index_t ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::nhead_stride_v

◆ num_head_q

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
ck_tile::index_t ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::num_head_q

◆ q_ptr

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
const void* ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::q_ptr

◆ raw_scale

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
float ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::raw_scale

◆ scale

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
float ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::scale

◆ seqlen_k

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
ck_tile::index_t ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::seqlen_k

◆ seqlen_q

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
ck_tile::index_t ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::seqlen_q

◆ stride_dk

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
ck_tile::index_t ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::stride_dk

◆ stride_do

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
ck_tile::index_t ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::stride_do

◆ stride_dq_acc

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
ck_tile::index_t ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::stride_dq_acc

◆ stride_dv

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
ck_tile::index_t ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::stride_dv

◆ stride_k

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
ck_tile::index_t ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::stride_k

◆ stride_q

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
ck_tile::index_t ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::stride_q

◆ stride_v

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
ck_tile::index_t ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::stride_v

◆ v_ptr

template<typename FmhaPipeline_, typename KGradEpiloguePipeline_, typename VGradEpiloguePipeline_, typename QGradEpiloguePipeline_ = void>
const void* ck_tile::FmhaBwdDQDKDVKernel< FmhaPipeline_, KGradEpiloguePipeline_, VGradEpiloguePipeline_, QGradEpiloguePipeline_ >::FmhaBwdCommonKargs::v_ptr
The documentation for this struct was generated from the following file: