/home/docs/checkouts/readthedocs.org/user_builds/advanced-micro-devices-composable-kernel/checkouts/develop/include/ck_tile/ops/rmsnorm2d/pipeline/rmsnorm2d_fwd_pipeline_problem.hpp Source File

• develop
• include
• ck_tile
• ops
• rmsnorm2d
• pipeline

rmsnorm2d_fwd_pipeline_problem.hpp

Go to the documentation of this file.

1 // SPDX-License-Identifier: MIT

2 // Copyright (c) 2018-2025, Advanced Micro Devices, Inc. All rights reserved.

3 

4 #pragma once

5 

6 #include "ck_tile/core/utility/type_traits.hpp"

7 

8 namespace ck_tile {

9 

10 template <typename XDataType_,

11  typename GammaDataType_,

12  typename ComputeDataType_,

13  typename YDataType_,

14  typename InvRmsDataType_,

15  typename UnquantYDataType_,

16  typename SmoothScaleDataType_,

17  typename YScaleDataType_,

18  typename BlockShape_,

19  typename Traits_>

20 struct Rmsnorm2dFwdPipelineProblem

21 {

22  using XDataType = remove_cvref_t<XDataType_>;

23  using GammaDataType = remove_cvref_t<GammaDataType_>;

24  using ComputeDataType = remove_cvref_t<ComputeDataType_>;

25  using YDataType = remove_cvref_t<YDataType_>;

26  using InvRmsDataType = remove_cvref_t<InvRmsDataType_>;

27  using UnquantYDataType = remove_cvref_t<UnquantYDataType_>;

28  using SmoothScaleDataType = remove_cvref_t<SmoothScaleDataType_>;

29  using YScaleDataType = remove_cvref_t<YScaleDataType_>;

30  using BlockShape = remove_cvref_t<BlockShape_>;

31 

32  static constexpr bool kNeedCrossLaneSync = BlockShape::ThreadPerWarp_N > 1;

33  static constexpr bool kNeedCrossWarpSync = BlockShape::WarpPerBlock_N > 1;

34 

35  using Traits = remove_cvref_t<Traits_>;

36 };

37 

38 } // namespace ck_tile

ck_tile

Definition: cluster_descriptor.hpp:13

ck_tile::remove_cvref_t

remove_cv_t< std::remove_reference_t< T > > remove_cvref_t

Definition: type_traits.hpp:21

ck_tile::Rmsnorm2dFwdPipelineProblem

Definition: rmsnorm2d_fwd_pipeline_problem.hpp:21

ck_tile::Rmsnorm2dFwdPipelineProblem::GammaDataType

remove_cvref_t< GammaDataType_ > GammaDataType

Definition: rmsnorm2d_fwd_pipeline_problem.hpp:23

ck_tile::Rmsnorm2dFwdPipelineProblem::XDataType

remove_cvref_t< XDataType_ > XDataType

Definition: rmsnorm2d_fwd_pipeline_problem.hpp:22

ck_tile::Rmsnorm2dFwdPipelineProblem::UnquantYDataType

remove_cvref_t< UnquantYDataType_ > UnquantYDataType

Definition: rmsnorm2d_fwd_pipeline_problem.hpp:27

ck_tile::Rmsnorm2dFwdPipelineProblem::kNeedCrossLaneSync

static constexpr bool kNeedCrossLaneSync

Definition: rmsnorm2d_fwd_pipeline_problem.hpp:32

ck_tile::Rmsnorm2dFwdPipelineProblem::Traits

remove_cvref_t< Traits_ > Traits

Definition: rmsnorm2d_fwd_pipeline_problem.hpp:35

ck_tile::Rmsnorm2dFwdPipelineProblem::InvRmsDataType

remove_cvref_t< InvRmsDataType_ > InvRmsDataType

Definition: rmsnorm2d_fwd_pipeline_problem.hpp:26

ck_tile::Rmsnorm2dFwdPipelineProblem::YScaleDataType

remove_cvref_t< YScaleDataType_ > YScaleDataType

Definition: rmsnorm2d_fwd_pipeline_problem.hpp:29

ck_tile::Rmsnorm2dFwdPipelineProblem::kNeedCrossWarpSync

static constexpr bool kNeedCrossWarpSync

Definition: rmsnorm2d_fwd_pipeline_problem.hpp:33

ck_tile::Rmsnorm2dFwdPipelineProblem::ComputeDataType

remove_cvref_t< ComputeDataType_ > ComputeDataType

Definition: rmsnorm2d_fwd_pipeline_problem.hpp:24

ck_tile::Rmsnorm2dFwdPipelineProblem::BlockShape

remove_cvref_t< BlockShape_ > BlockShape

Definition: rmsnorm2d_fwd_pipeline_problem.hpp:30

ck_tile::Rmsnorm2dFwdPipelineProblem::YDataType

remove_cvref_t< YDataType_ > YDataType

Definition: rmsnorm2d_fwd_pipeline_problem.hpp:25

ck_tile::Rmsnorm2dFwdPipelineProblem::SmoothScaleDataType

remove_cvref_t< SmoothScaleDataType_ > SmoothScaleDataType

Definition: rmsnorm2d_fwd_pipeline_problem.hpp:28

type_traits.hpp