DynamicQuantEpilogue< Problem_, Policy_ > Struct Template Reference

DynamicQuantEpilogue&lt; Problem_, Policy_ &gt; Struct Template Reference#

Composable Kernel: ck_tile::DynamicQuantEpilogue< Problem_, Policy_ > Struct Template Reference
Public Types | Public Member Functions | Static Public Member Functions | Static Public Attributes | List of all members
ck_tile::DynamicQuantEpilogue< Problem_, Policy_ > Struct Template Reference

#include <dynamic_quant_epilogue.hpp>

Public Types

using Problem = remove_cvref_t< Problem_ >
 
using AccDataType = remove_cvref_t< typename Problem::AccDataType >
 
using SmoothScaleDataType = remove_cvref_t< typename Problem::SmoothScaleDataType >
 
using YScaleDataType = remove_cvref_t< typename Problem::YScaleDataType >
 
using ODataType = remove_cvref_t< typename Problem::ODataType >
 
using BlockShape = remove_cvref_t< typename Problem::BlockShape >
 

Public Member Functions

template<typename ODramWindowTmp , typename YScaleWindow , typename OAccTile >
CK_TILE_DEVICE auto Impl (ODramWindowTmp &o_dram_window_tmp, YScaleWindow &y_scale_window, const OAccTile &o_acc_tile, void *smem)
 
template<typename ODramWindowTmp , typename SmoothScaleWindow , typename YScaleWindow , typename OAccTile >
CK_TILE_DEVICE auto operator() (ODramWindowTmp &o_dram_window_tmp, const SmoothScaleWindow &sm_scale_window_, YScaleWindow &y_scale_window, const OAccTile &o_acc_tile, void *smem)
 
template<typename ODramWindowTmp , typename YScaleWindow , typename OAccTile >
CK_TILE_DEVICE auto operator() (ODramWindowTmp &o_dram_window_tmp, YScaleWindow &y_scale_window, const OAccTile &o_acc_tile, void *smem)
 

Static Public Member Functions

static constexpr CK_TILE_HOST_DEVICE auto GetBlockReduce2d ()
 
static constexpr CK_TILE_HOST_DEVICE auto GetBlockReduce2dSync ()
 
static constexpr CK_TILE_HOST_DEVICE auto GetBlockReduce2dCrossWarpSync ()
 
static constexpr CK_TILE_DEVICE auto MakeSmoothInputScaleTileDistribution ()
 
static constexpr CK_TILE_HOST_DEVICE index_t GetSmemSize ()
 

Static Public Attributes

static constexpr bool kPadM = Problem::Traits::kPadM
 
static constexpr bool kPadN = Problem::Traits::kPadN
 
static constexpr bool UseRawStore = Problem::Traits::UseRawStore
 
static constexpr bool UseMax3 = Problem::Traits::UseMax3
 

Member Typedef Documentation

◆ AccDataType

template<typename Problem_ , typename Policy_ = void>
using ck_tile::DynamicQuantEpilogue< Problem_, Policy_ >::AccDataType = remove_cvref_t<typename Problem::AccDataType>

◆ BlockShape

template<typename Problem_ , typename Policy_ = void>
using ck_tile::DynamicQuantEpilogue< Problem_, Policy_ >::BlockShape = remove_cvref_t<typename Problem::BlockShape>

◆ ODataType

template<typename Problem_ , typename Policy_ = void>
using ck_tile::DynamicQuantEpilogue< Problem_, Policy_ >::ODataType = remove_cvref_t<typename Problem::ODataType>

◆ Problem

template<typename Problem_ , typename Policy_ = void>
using ck_tile::DynamicQuantEpilogue< Problem_, Policy_ >::Problem = remove_cvref_t<Problem_>

◆ SmoothScaleDataType

template<typename Problem_ , typename Policy_ = void>
using ck_tile::DynamicQuantEpilogue< Problem_, Policy_ >::SmoothScaleDataType = remove_cvref_t<typename Problem::SmoothScaleDataType>

◆ YScaleDataType

template<typename Problem_ , typename Policy_ = void>
using ck_tile::DynamicQuantEpilogue< Problem_, Policy_ >::YScaleDataType = remove_cvref_t<typename Problem::YScaleDataType>

Member Function Documentation

◆ GetBlockReduce2d()

template<typename Problem_ , typename Policy_ = void>
static constexpr CK_TILE_HOST_DEVICE auto ck_tile::DynamicQuantEpilogue< Problem_, Policy_ >::GetBlockReduce2d ( )
inlinestaticconstexpr

◆ GetBlockReduce2dCrossWarpSync()

template<typename Problem_ , typename Policy_ = void>
static constexpr CK_TILE_HOST_DEVICE auto ck_tile::DynamicQuantEpilogue< Problem_, Policy_ >::GetBlockReduce2dCrossWarpSync ( )
inlinestaticconstexpr

◆ GetBlockReduce2dSync()

template<typename Problem_ , typename Policy_ = void>
static constexpr CK_TILE_HOST_DEVICE auto ck_tile::DynamicQuantEpilogue< Problem_, Policy_ >::GetBlockReduce2dSync ( )
inlinestaticconstexpr

◆ GetSmemSize()

template<typename Problem_ , typename Policy_ = void>
static constexpr CK_TILE_HOST_DEVICE index_t ck_tile::DynamicQuantEpilogue< Problem_, Policy_ >::GetSmemSize ( )
inlinestaticconstexpr

◆ Impl()

template<typename Problem_ , typename Policy_ = void>
template<typename ODramWindowTmp , typename YScaleWindow , typename OAccTile >
CK_TILE_DEVICE auto ck_tile::DynamicQuantEpilogue< Problem_, Policy_ >::Impl ( ODramWindowTmp &  o_dram_window_tmp,
YScaleWindow &  y_scale_window,
const OAccTile &  o_acc_tile,
void *  smem 
)
inline

◆ MakeSmoothInputScaleTileDistribution()

template<typename Problem_ , typename Policy_ = void>
static constexpr CK_TILE_DEVICE auto ck_tile::DynamicQuantEpilogue< Problem_, Policy_ >::MakeSmoothInputScaleTileDistribution ( )
inlinestaticconstexpr

◆ operator()() [1/2]

template<typename Problem_ , typename Policy_ = void>
template<typename ODramWindowTmp , typename SmoothScaleWindow , typename YScaleWindow , typename OAccTile >
CK_TILE_DEVICE auto ck_tile::DynamicQuantEpilogue< Problem_, Policy_ >::operator() ( ODramWindowTmp &  o_dram_window_tmp,
const SmoothScaleWindow &  sm_scale_window_,
YScaleWindow &  y_scale_window,
const OAccTile &  o_acc_tile,
void *  smem 
)
inline

◆ operator()() [2/2]

template<typename Problem_ , typename Policy_ = void>
template<typename ODramWindowTmp , typename YScaleWindow , typename OAccTile >
CK_TILE_DEVICE auto ck_tile::DynamicQuantEpilogue< Problem_, Policy_ >::operator() ( ODramWindowTmp &  o_dram_window_tmp,
YScaleWindow &  y_scale_window,
const OAccTile &  o_acc_tile,
void *  smem 
)
inline

Member Data Documentation

◆ kPadM

template<typename Problem_ , typename Policy_ = void>
constexpr bool ck_tile::DynamicQuantEpilogue< Problem_, Policy_ >::kPadM = Problem::Traits::kPadM
staticconstexpr

◆ kPadN

template<typename Problem_ , typename Policy_ = void>
constexpr bool ck_tile::DynamicQuantEpilogue< Problem_, Policy_ >::kPadN = Problem::Traits::kPadN
staticconstexpr

◆ UseMax3

template<typename Problem_ , typename Policy_ = void>
constexpr bool ck_tile::DynamicQuantEpilogue< Problem_, Policy_ >::UseMax3 = Problem::Traits::UseMax3
staticconstexpr

◆ UseRawStore

template<typename Problem_ , typename Policy_ = void>
constexpr bool ck_tile::DynamicQuantEpilogue< Problem_, Policy_ >::UseRawStore = Problem::Traits::UseRawStore
staticconstexpr
The documentation for this struct was generated from the following file:
  • /home/docs/checkouts/readthedocs.org/user_builds/advanced-micro-devices-composable-kernel/checkouts/docs-6.4.3/include/ck_tile/ops/epilogue/dynamic_quant_epilogue.hpp