AddRmsnorm2dRdquantFwd< Pipeline_ > Struct Template Reference

AddRmsnorm2dRdquantFwd&lt; Pipeline_ &gt; Struct Template Reference#

Composable Kernel: ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ > Struct Template Reference
Classes | Public Types | Public Member Functions | Static Public Member Functions | Static Public Attributes | List of all members
ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ > Struct Template Reference

#include <add_rmsnorm2d_rdquant_fwd_kernel.hpp>

Classes

struct  Kargs
struct  t2s
struct  t2s< float >
struct  t2s< ck_tile::fp16_t >
struct  t2s< ck_tile::bf16_t >
struct  t2s< ck_tile::fp8_t >
struct  t2s< ck_tile::bf8_t >

Public Types

using Pipeline = remove_cvref_t<Pipeline_>
using Problem = typename Pipeline::Problem
using ADataType = remove_cvref_t<typename Problem::ADataType>
using BDataType = remove_cvref_t<typename Problem::BDataType>
using GammaDataType = remove_cvref_t<typename Problem::GammaDataType>
using ComputeDataType = remove_cvref_t<typename Problem::ComputeDataType>
using XDataType = remove_cvref_t<typename Problem::XDataType>
using YScaleDataType = remove_cvref_t<typename Problem::YScaleDataType>
using QYDataType = remove_cvref_t<typename Problem::QYDataType>
using Hargs = AddRmsnorm2dRdquantFwdHostArgs

Public Member Functions

CK_TILE_DEVICE void operator() (Kargs kargs) const

Static Public Member Functions

static CK_TILE_HOST constexpr Kargs MakeKargs (const Hargs &hargs)
static CK_TILE_HOST constexpr auto GridSize (const Hargs &hargs)
static CK_TILE_HOST constexpr auto BlockSize ()
static CK_TILE_HOST_DEVICE constexpr index_t GetSmemSize ()
static CK_TILE_HOST std::string GetName ()

Static Public Attributes

static constexpr bool kSaveX = Problem::kSaveX
static constexpr index_t Block_M = Problem::BlockShape::Block_M
static constexpr index_t Block_N = Problem::BlockShape::Block_N
static constexpr bool kPadM = false
static constexpr bool kPadN = Problem::kPadN
static constexpr bool kThreePass = Problem::kThreePass
static constexpr index_t ThreadPerWarp_N = Problem::BlockShape::ThreadPerWarp_N
static constexpr index_t Vector_N = Problem::BlockShape::Vector_N
static constexpr index_t Repeat_N = Problem::BlockShape::Repeat_N
static constexpr index_t kBlockSize = Problem::BlockShape::BlockSize
static constexpr auto I0 = number<0>{}
static constexpr auto I1 = number<1>{}

Member Typedef Documentation

◆ ADataType

template<typename Pipeline_>
using ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::ADataType = remove_cvref_t<typename Problem::ADataType>

◆ BDataType

template<typename Pipeline_>
using ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::BDataType = remove_cvref_t<typename Problem::BDataType>

◆ ComputeDataType

template<typename Pipeline_>
using ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::ComputeDataType = remove_cvref_t<typename Problem::ComputeDataType>

◆ GammaDataType

template<typename Pipeline_>
using ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::GammaDataType = remove_cvref_t<typename Problem::GammaDataType>

◆ Hargs

template<typename Pipeline_>
using ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::Hargs = AddRmsnorm2dRdquantFwdHostArgs

◆ Pipeline

template<typename Pipeline_>
using ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::Pipeline = remove_cvref_t<Pipeline_>

◆ Problem

template<typename Pipeline_>
using ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::Problem = typename Pipeline::Problem

◆ QYDataType

template<typename Pipeline_>
using ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::QYDataType = remove_cvref_t<typename Problem::QYDataType>

◆ XDataType

template<typename Pipeline_>
using ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::XDataType = remove_cvref_t<typename Problem::XDataType>

◆ YScaleDataType

template<typename Pipeline_>
using ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::YScaleDataType = remove_cvref_t<typename Problem::YScaleDataType>

Member Function Documentation

◆ BlockSize()

template<typename Pipeline_>
CK_TILE_HOST constexpr auto ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::BlockSize ( )
inlinestaticconstexpr

◆ GetName()

template<typename Pipeline_>
CK_TILE_HOST std::string ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::GetName ( )
inlinestatic

◆ GetSmemSize()

template<typename Pipeline_>
CK_TILE_HOST_DEVICE constexpr index_t ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::GetSmemSize ( )
inlinestaticconstexpr

◆ GridSize()

template<typename Pipeline_>
CK_TILE_HOST constexpr auto ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::GridSize ( const Hargs & hargs)
inlinestaticconstexpr

◆ MakeKargs()

template<typename Pipeline_>
CK_TILE_HOST constexpr Kargs ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::MakeKargs ( const Hargs & hargs)
inlinestaticconstexpr

◆ operator()()

template<typename Pipeline_>
CK_TILE_DEVICE void ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::operator() ( Kargs kargs) const
inline

Member Data Documentation

◆ Block_M

template<typename Pipeline_>
index_t ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::Block_M = Problem::BlockShape::Block_M
staticconstexpr

◆ Block_N

template<typename Pipeline_>
index_t ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::Block_N = Problem::BlockShape::Block_N
staticconstexpr

◆ I0

template<typename Pipeline_>
auto ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::I0 = number<0>{}
staticconstexpr

◆ I1

template<typename Pipeline_>
auto ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::I1 = number<1>{}
staticconstexpr

◆ kBlockSize

template<typename Pipeline_>
index_t ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::kBlockSize = Problem::BlockShape::BlockSize
staticconstexpr

◆ kPadM

template<typename Pipeline_>
bool ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::kPadM = false
staticconstexpr

◆ kPadN

template<typename Pipeline_>
bool ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::kPadN = Problem::kPadN
staticconstexpr

◆ kSaveX

template<typename Pipeline_>
bool ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::kSaveX = Problem::kSaveX
staticconstexpr

◆ kThreePass

template<typename Pipeline_>
bool ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::kThreePass = Problem::kThreePass
staticconstexpr

◆ Repeat_N

template<typename Pipeline_>
index_t ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::Repeat_N = Problem::BlockShape::Repeat_N
staticconstexpr

◆ ThreadPerWarp_N

template<typename Pipeline_>
index_t ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::ThreadPerWarp_N = Problem::BlockShape::ThreadPerWarp_N
staticconstexpr

◆ Vector_N

template<typename Pipeline_>
index_t ck_tile::AddRmsnorm2dRdquantFwd< Pipeline_ >::Vector_N = Problem::BlockShape::Vector_N
staticconstexpr
The documentation for this struct was generated from the following file: