AddRmsnorm2dRdquantFwdPipelineProblem< ADataType_, BDataType_, GammaDataType_, ComputeDataType_, XDataType_, YScaleDataType_, QYDataType_, BlockShape_, kPadN_, kSaveX_, kThreePass_ > Struct Template Reference

AddRmsnorm2dRdquantFwdPipelineProblem&lt; ADataType_, BDataType_, GammaDataType_, ComputeDataType_, XDataType_, YScaleDataType_, QYDataType_, BlockShape_, kPadN_, kSaveX_, kThreePass_ &gt; Struct Template Reference#

Composable Kernel: ck_tile::AddRmsnorm2dRdquantFwdPipelineProblem< ADataType_, BDataType_, GammaDataType_, ComputeDataType_, XDataType_, YScaleDataType_, QYDataType_, BlockShape_, kPadN_, kSaveX_, kThreePass_ > Struct Template Reference
Public Types | Static Public Attributes | List of all members
ck_tile::AddRmsnorm2dRdquantFwdPipelineProblem< ADataType_, BDataType_, GammaDataType_, ComputeDataType_, XDataType_, YScaleDataType_, QYDataType_, BlockShape_, kPadN_, kSaveX_, kThreePass_ > Struct Template Reference

#include <add_rmsnorm2d_rdquant_fwd_pipeline_problem.hpp>

Public Types

using ADataType = remove_cvref_t<ADataType_>
using BDataType = remove_cvref_t<BDataType_>
using GammaDataType = remove_cvref_t<GammaDataType_>
using ComputeDataType = remove_cvref_t<ComputeDataType_>
using XDataType = remove_cvref_t<XDataType_>
using YScaleDataType = remove_cvref_t<YScaleDataType_>
using QYDataType = remove_cvref_t<QYDataType_>
using BlockShape = remove_cvref_t<BlockShape_>

Static Public Attributes

static constexpr bool kNeedCrossLaneSync = BlockShape::ThreadPerWarp_N > 1
static constexpr bool kNeedCrossWarpSync = BlockShape::WarpPerBlock_N > 1
static constexpr bool kPadN = kPadN_
static constexpr bool kSaveX = kSaveX_
static constexpr bool kThreePass = kThreePass_

Member Typedef Documentation

◆ ADataType

template<typename ADataType_, typename BDataType_, typename GammaDataType_, typename ComputeDataType_, typename XDataType_, typename YScaleDataType_, typename QYDataType_, typename BlockShape_, bool kPadN_, bool kSaveX_, bool kThreePass_>
using ck_tile::AddRmsnorm2dRdquantFwdPipelineProblem< ADataType_, BDataType_, GammaDataType_, ComputeDataType_, XDataType_, YScaleDataType_, QYDataType_, BlockShape_, kPadN_, kSaveX_, kThreePass_ >::ADataType = remove_cvref_t<ADataType_>

◆ BDataType

template<typename ADataType_, typename BDataType_, typename GammaDataType_, typename ComputeDataType_, typename XDataType_, typename YScaleDataType_, typename QYDataType_, typename BlockShape_, bool kPadN_, bool kSaveX_, bool kThreePass_>
using ck_tile::AddRmsnorm2dRdquantFwdPipelineProblem< ADataType_, BDataType_, GammaDataType_, ComputeDataType_, XDataType_, YScaleDataType_, QYDataType_, BlockShape_, kPadN_, kSaveX_, kThreePass_ >::BDataType = remove_cvref_t<BDataType_>

◆ BlockShape

template<typename ADataType_, typename BDataType_, typename GammaDataType_, typename ComputeDataType_, typename XDataType_, typename YScaleDataType_, typename QYDataType_, typename BlockShape_, bool kPadN_, bool kSaveX_, bool kThreePass_>
using ck_tile::AddRmsnorm2dRdquantFwdPipelineProblem< ADataType_, BDataType_, GammaDataType_, ComputeDataType_, XDataType_, YScaleDataType_, QYDataType_, BlockShape_, kPadN_, kSaveX_, kThreePass_ >::BlockShape = remove_cvref_t<BlockShape_>

◆ ComputeDataType

template<typename ADataType_, typename BDataType_, typename GammaDataType_, typename ComputeDataType_, typename XDataType_, typename YScaleDataType_, typename QYDataType_, typename BlockShape_, bool kPadN_, bool kSaveX_, bool kThreePass_>
using ck_tile::AddRmsnorm2dRdquantFwdPipelineProblem< ADataType_, BDataType_, GammaDataType_, ComputeDataType_, XDataType_, YScaleDataType_, QYDataType_, BlockShape_, kPadN_, kSaveX_, kThreePass_ >::ComputeDataType = remove_cvref_t<ComputeDataType_>

◆ GammaDataType

template<typename ADataType_, typename BDataType_, typename GammaDataType_, typename ComputeDataType_, typename XDataType_, typename YScaleDataType_, typename QYDataType_, typename BlockShape_, bool kPadN_, bool kSaveX_, bool kThreePass_>
using ck_tile::AddRmsnorm2dRdquantFwdPipelineProblem< ADataType_, BDataType_, GammaDataType_, ComputeDataType_, XDataType_, YScaleDataType_, QYDataType_, BlockShape_, kPadN_, kSaveX_, kThreePass_ >::GammaDataType = remove_cvref_t<GammaDataType_>

◆ QYDataType

template<typename ADataType_, typename BDataType_, typename GammaDataType_, typename ComputeDataType_, typename XDataType_, typename YScaleDataType_, typename QYDataType_, typename BlockShape_, bool kPadN_, bool kSaveX_, bool kThreePass_>
using ck_tile::AddRmsnorm2dRdquantFwdPipelineProblem< ADataType_, BDataType_, GammaDataType_, ComputeDataType_, XDataType_, YScaleDataType_, QYDataType_, BlockShape_, kPadN_, kSaveX_, kThreePass_ >::QYDataType = remove_cvref_t<QYDataType_>

◆ XDataType

template<typename ADataType_, typename BDataType_, typename GammaDataType_, typename ComputeDataType_, typename XDataType_, typename YScaleDataType_, typename QYDataType_, typename BlockShape_, bool kPadN_, bool kSaveX_, bool kThreePass_>
using ck_tile::AddRmsnorm2dRdquantFwdPipelineProblem< ADataType_, BDataType_, GammaDataType_, ComputeDataType_, XDataType_, YScaleDataType_, QYDataType_, BlockShape_, kPadN_, kSaveX_, kThreePass_ >::XDataType = remove_cvref_t<XDataType_>

◆ YScaleDataType

template<typename ADataType_, typename BDataType_, typename GammaDataType_, typename ComputeDataType_, typename XDataType_, typename YScaleDataType_, typename QYDataType_, typename BlockShape_, bool kPadN_, bool kSaveX_, bool kThreePass_>
using ck_tile::AddRmsnorm2dRdquantFwdPipelineProblem< ADataType_, BDataType_, GammaDataType_, ComputeDataType_, XDataType_, YScaleDataType_, QYDataType_, BlockShape_, kPadN_, kSaveX_, kThreePass_ >::YScaleDataType = remove_cvref_t<YScaleDataType_>

Member Data Documentation

◆ kNeedCrossLaneSync

template<typename ADataType_, typename BDataType_, typename GammaDataType_, typename ComputeDataType_, typename XDataType_, typename YScaleDataType_, typename QYDataType_, typename BlockShape_, bool kPadN_, bool kSaveX_, bool kThreePass_>
bool ck_tile::AddRmsnorm2dRdquantFwdPipelineProblem< ADataType_, BDataType_, GammaDataType_, ComputeDataType_, XDataType_, YScaleDataType_, QYDataType_, BlockShape_, kPadN_, kSaveX_, kThreePass_ >::kNeedCrossLaneSync = BlockShape::ThreadPerWarp_N > 1
staticconstexpr

◆ kNeedCrossWarpSync

template<typename ADataType_, typename BDataType_, typename GammaDataType_, typename ComputeDataType_, typename XDataType_, typename YScaleDataType_, typename QYDataType_, typename BlockShape_, bool kPadN_, bool kSaveX_, bool kThreePass_>
bool ck_tile::AddRmsnorm2dRdquantFwdPipelineProblem< ADataType_, BDataType_, GammaDataType_, ComputeDataType_, XDataType_, YScaleDataType_, QYDataType_, BlockShape_, kPadN_, kSaveX_, kThreePass_ >::kNeedCrossWarpSync = BlockShape::WarpPerBlock_N > 1
staticconstexpr

◆ kPadN

template<typename ADataType_, typename BDataType_, typename GammaDataType_, typename ComputeDataType_, typename XDataType_, typename YScaleDataType_, typename QYDataType_, typename BlockShape_, bool kPadN_, bool kSaveX_, bool kThreePass_>
bool ck_tile::AddRmsnorm2dRdquantFwdPipelineProblem< ADataType_, BDataType_, GammaDataType_, ComputeDataType_, XDataType_, YScaleDataType_, QYDataType_, BlockShape_, kPadN_, kSaveX_, kThreePass_ >::kPadN = kPadN_
staticconstexpr

◆ kSaveX

template<typename ADataType_, typename BDataType_, typename GammaDataType_, typename ComputeDataType_, typename XDataType_, typename YScaleDataType_, typename QYDataType_, typename BlockShape_, bool kPadN_, bool kSaveX_, bool kThreePass_>
bool ck_tile::AddRmsnorm2dRdquantFwdPipelineProblem< ADataType_, BDataType_, GammaDataType_, ComputeDataType_, XDataType_, YScaleDataType_, QYDataType_, BlockShape_, kPadN_, kSaveX_, kThreePass_ >::kSaveX = kSaveX_
staticconstexpr

◆ kThreePass

template<typename ADataType_, typename BDataType_, typename GammaDataType_, typename ComputeDataType_, typename XDataType_, typename YScaleDataType_, typename QYDataType_, typename BlockShape_, bool kPadN_, bool kSaveX_, bool kThreePass_>
bool ck_tile::AddRmsnorm2dRdquantFwdPipelineProblem< ADataType_, BDataType_, GammaDataType_, ComputeDataType_, XDataType_, YScaleDataType_, QYDataType_, BlockShape_, kPadN_, kSaveX_, kThreePass_ >::kThreePass = kThreePass_
staticconstexpr
The documentation for this struct was generated from the following file: