llvm/tools/mlir/include/mlir/Dialect/LLVMIR/ROCDLOps.h.inc

/*===- TableGen'erated file -------------------------------------*- C++ -*-===*\
|*                                                                            *|
|* Op Declarations                                                            *|
|*                                                                            *|
|* Automatically generated file, do not edit!                                 *|
|* From: ROCDLOps.td                                                          *|
|*                                                                            *|
\*===----------------------------------------------------------------------===*/

namespace mlir {
namespace ROCDL {
class BallotOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class BarrierOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class BarrierSignalOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class BarrierWaitOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class BlockDimXOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class BlockDimYOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class BlockDimZOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class BlockIdXOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class BlockIdYOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class BlockIdZOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class CvtF32Bf8Op;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class CvtF32Fp8Op;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class CvtPkBf8F32Op;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class CvtPkFp8F32Op;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class CvtPkRtz;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class CvtSrBf8F32Op;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class CvtSrFp8F32Op;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class DPPUpdateOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class DsBpermuteOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class DsSwizzleOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class GridDimXOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class GridDimYOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class GridDimZOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class MakeBufferRsrcOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class MbcntHiOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class MbcntLoOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class RawBufferAtomicCmpSwap;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class RawBufferAtomicFAddOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class RawBufferAtomicFMaxOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class RawBufferAtomicSMaxOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class RawBufferAtomicUMinOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class RawBufferLoadOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class RawBufferStoreOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class RawPtrBufferAtomicCmpSwap;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class RawPtrBufferAtomicFaddOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class RawPtrBufferAtomicFmaxOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class RawPtrBufferAtomicSmaxOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class RawPtrBufferAtomicUminOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class RawPtrBufferLoadOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class RawPtrBufferStoreOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class SBarrierOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class SchedBarrier;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class SetPrioOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class ThreadIdXOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class ThreadIdYOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class ThreadIdZOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class WaitDscntOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class WaitcntOp;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_16x16x16bf16_1k;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_16x16x16f16;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_16x16x1f32;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_16x16x2bf16;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_16x16x32_bf8_bf8;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_16x16x32_bf8_fp8;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_16x16x32_fp8_bf8;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_16x16x32_fp8_fp8;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_16x16x4bf16_1k;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_16x16x4f16;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_16x16x4f32;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_16x16x8_xf32;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_16x16x8bf16;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_32x32x16_bf8_bf8;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_32x32x16_bf8_fp8;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_32x32x16_fp8_bf8;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_32x32x16_fp8_fp8;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_32x32x1f32;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_32x32x2bf16;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_32x32x2f32;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_32x32x4_xf32;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_32x32x4bf16;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_32x32x4bf16_1k;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_32x32x4f16;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_32x32x8bf16_1k;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_32x32x8f16;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_4x4x1f32;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_4x4x2bf16;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_4x4x4bf16_1k;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f32_4x4x4f16;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f64_16x16x4f64;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_f64_4x4x4f64;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_i32_16x16x16i8;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_i32_16x16x32_i8;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_i32_16x16x4i8;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_i32_32x32x16_i8;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_i32_32x32x4i8;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_i32_32x32x8i8;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class mfma_i32_4x4x4i8;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class wmma_bf16_16x16x16_bf16;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class wmma_f16_16x16x16_f16;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class wmma_f32_16x16x16_bf16;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class wmma_f32_16x16x16_bf8;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class wmma_f32_16x16x16_f16;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class wmma_f32_16x16x16_fp8;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class wmma_i32_16x16x16_iu4;
} // namespace ROCDL
} // namespace mlir
namespace mlir {
namespace ROCDL {
class wmma_i32_16x16x16_iu8;
} // namespace ROCDL
} // namespace mlir
#ifdef GET_OP_CLASSES
#undef GET_OP_CLASSES

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::BallotOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class BallotOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class BallotOpGenericAdaptor : public detail::BallotOpGenericAdaptorBase { … };
class BallotOpAdaptor : public BallotOpGenericAdaptor<::mlir::ValueRange> { … };
class BallotOp : public ::mlir::Op<BallotOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::OneOperand, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::BallotOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::BarrierOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class BarrierOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class BarrierOpGenericAdaptor : public detail::BarrierOpGenericAdaptorBase { … };
class BarrierOpAdaptor : public BarrierOpGenericAdaptor<::mlir::ValueRange> { … };
class BarrierOp : public ::mlir::Op<BarrierOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::ZeroResults, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::ZeroOperands, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::BarrierOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::BarrierSignalOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class BarrierSignalOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class BarrierSignalOpGenericAdaptor : public detail::BarrierSignalOpGenericAdaptorBase { … };
class BarrierSignalOpAdaptor : public BarrierSignalOpGenericAdaptor<::mlir::ValueRange> { … };
class BarrierSignalOp : public ::mlir::Op<BarrierSignalOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::ZeroResults, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::ZeroOperands, ::mlir::OpTrait::OpInvariants, ::mlir::BytecodeOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::BarrierSignalOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::BarrierWaitOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class BarrierWaitOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class BarrierWaitOpGenericAdaptor : public detail::BarrierWaitOpGenericAdaptorBase { … };
class BarrierWaitOpAdaptor : public BarrierWaitOpGenericAdaptor<::mlir::ValueRange> { … };
class BarrierWaitOp : public ::mlir::Op<BarrierWaitOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::ZeroResults, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::ZeroOperands, ::mlir::OpTrait::OpInvariants, ::mlir::BytecodeOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::BarrierWaitOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::BlockDimXOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class BlockDimXOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class BlockDimXOpGenericAdaptor : public detail::BlockDimXOpGenericAdaptorBase { … };
class BlockDimXOpAdaptor : public BlockDimXOpGenericAdaptor<::mlir::ValueRange> { … };
class BlockDimXOp : public ::mlir::Op<BlockDimXOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::ZeroOperands, ::mlir::OpTrait::OpInvariants, ::mlir::BytecodeOpInterface::Trait, ::mlir::ConditionallySpeculatable::Trait, ::mlir::OpTrait::AlwaysSpeculatableImplTrait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::BlockDimXOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::BlockDimYOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class BlockDimYOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class BlockDimYOpGenericAdaptor : public detail::BlockDimYOpGenericAdaptorBase { … };
class BlockDimYOpAdaptor : public BlockDimYOpGenericAdaptor<::mlir::ValueRange> { … };
class BlockDimYOp : public ::mlir::Op<BlockDimYOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::ZeroOperands, ::mlir::OpTrait::OpInvariants, ::mlir::BytecodeOpInterface::Trait, ::mlir::ConditionallySpeculatable::Trait, ::mlir::OpTrait::AlwaysSpeculatableImplTrait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::BlockDimYOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::BlockDimZOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class BlockDimZOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class BlockDimZOpGenericAdaptor : public detail::BlockDimZOpGenericAdaptorBase { … };
class BlockDimZOpAdaptor : public BlockDimZOpGenericAdaptor<::mlir::ValueRange> { … };
class BlockDimZOp : public ::mlir::Op<BlockDimZOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::ZeroOperands, ::mlir::OpTrait::OpInvariants, ::mlir::BytecodeOpInterface::Trait, ::mlir::ConditionallySpeculatable::Trait, ::mlir::OpTrait::AlwaysSpeculatableImplTrait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::BlockDimZOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::BlockIdXOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class BlockIdXOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class BlockIdXOpGenericAdaptor : public detail::BlockIdXOpGenericAdaptorBase { … };
class BlockIdXOpAdaptor : public BlockIdXOpGenericAdaptor<::mlir::ValueRange> { … };
class BlockIdXOp : public ::mlir::Op<BlockIdXOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::ZeroOperands, ::mlir::OpTrait::OpInvariants, ::mlir::BytecodeOpInterface::Trait, ::mlir::ConditionallySpeculatable::Trait, ::mlir::OpTrait::AlwaysSpeculatableImplTrait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::BlockIdXOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::BlockIdYOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class BlockIdYOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class BlockIdYOpGenericAdaptor : public detail::BlockIdYOpGenericAdaptorBase { … };
class BlockIdYOpAdaptor : public BlockIdYOpGenericAdaptor<::mlir::ValueRange> { … };
class BlockIdYOp : public ::mlir::Op<BlockIdYOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::ZeroOperands, ::mlir::OpTrait::OpInvariants, ::mlir::BytecodeOpInterface::Trait, ::mlir::ConditionallySpeculatable::Trait, ::mlir::OpTrait::AlwaysSpeculatableImplTrait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::BlockIdYOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::BlockIdZOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class BlockIdZOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class BlockIdZOpGenericAdaptor : public detail::BlockIdZOpGenericAdaptorBase { … };
class BlockIdZOpAdaptor : public BlockIdZOpGenericAdaptor<::mlir::ValueRange> { … };
class BlockIdZOp : public ::mlir::Op<BlockIdZOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::ZeroOperands, ::mlir::OpTrait::OpInvariants, ::mlir::BytecodeOpInterface::Trait, ::mlir::ConditionallySpeculatable::Trait, ::mlir::OpTrait::AlwaysSpeculatableImplTrait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::BlockIdZOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::CvtF32Bf8Op declarations
//===----------------------------------------------------------------------===//

namespace detail {
class CvtF32Bf8OpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class CvtF32Bf8OpGenericAdaptor : public detail::CvtF32Bf8OpGenericAdaptorBase { … };
class CvtF32Bf8OpAdaptor : public CvtF32Bf8OpGenericAdaptor<::mlir::ValueRange> { … };
class CvtF32Bf8Op : public ::mlir::Op<CvtF32Bf8Op, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<2>::Impl, ::mlir::OpTrait::OpInvariants, ::mlir::ConditionallySpeculatable::Trait, ::mlir::OpTrait::AlwaysSpeculatableImplTrait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::CvtF32Bf8Op)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::CvtF32Fp8Op declarations
//===----------------------------------------------------------------------===//

namespace detail {
class CvtF32Fp8OpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class CvtF32Fp8OpGenericAdaptor : public detail::CvtF32Fp8OpGenericAdaptorBase { … };
class CvtF32Fp8OpAdaptor : public CvtF32Fp8OpGenericAdaptor<::mlir::ValueRange> { … };
class CvtF32Fp8Op : public ::mlir::Op<CvtF32Fp8Op, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<2>::Impl, ::mlir::OpTrait::OpInvariants, ::mlir::ConditionallySpeculatable::Trait, ::mlir::OpTrait::AlwaysSpeculatableImplTrait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::CvtF32Fp8Op)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::CvtPkBf8F32Op declarations
//===----------------------------------------------------------------------===//

namespace detail {
class CvtPkBf8F32OpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class CvtPkBf8F32OpGenericAdaptor : public detail::CvtPkBf8F32OpGenericAdaptorBase { … };
class CvtPkBf8F32OpAdaptor : public CvtPkBf8F32OpGenericAdaptor<::mlir::ValueRange> { … };
class CvtPkBf8F32Op : public ::mlir::Op<CvtPkBf8F32Op, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<4>::Impl, ::mlir::OpTrait::OpInvariants, ::mlir::ConditionallySpeculatable::Trait, ::mlir::OpTrait::AlwaysSpeculatableImplTrait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::CvtPkBf8F32Op)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::CvtPkFp8F32Op declarations
//===----------------------------------------------------------------------===//

namespace detail {
class CvtPkFp8F32OpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class CvtPkFp8F32OpGenericAdaptor : public detail::CvtPkFp8F32OpGenericAdaptorBase { … };
class CvtPkFp8F32OpAdaptor : public CvtPkFp8F32OpGenericAdaptor<::mlir::ValueRange> { … };
class CvtPkFp8F32Op : public ::mlir::Op<CvtPkFp8F32Op, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<4>::Impl, ::mlir::OpTrait::OpInvariants, ::mlir::ConditionallySpeculatable::Trait, ::mlir::OpTrait::AlwaysSpeculatableImplTrait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::CvtPkFp8F32Op)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::CvtPkRtz declarations
//===----------------------------------------------------------------------===//

namespace detail {
class CvtPkRtzGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class CvtPkRtzGenericAdaptor : public detail::CvtPkRtzGenericAdaptorBase { … };
class CvtPkRtzAdaptor : public CvtPkRtzGenericAdaptor<::mlir::ValueRange> { … };
class CvtPkRtz : public ::mlir::Op<CvtPkRtz, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<2>::Impl, ::mlir::OpTrait::OpInvariants, ::mlir::ConditionallySpeculatable::Trait, ::mlir::OpTrait::AlwaysSpeculatableImplTrait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::CvtPkRtz)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::CvtSrBf8F32Op declarations
//===----------------------------------------------------------------------===//

namespace detail {
class CvtSrBf8F32OpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class CvtSrBf8F32OpGenericAdaptor : public detail::CvtSrBf8F32OpGenericAdaptorBase { … };
class CvtSrBf8F32OpAdaptor : public CvtSrBf8F32OpGenericAdaptor<::mlir::ValueRange> { … };
class CvtSrBf8F32Op : public ::mlir::Op<CvtSrBf8F32Op, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<4>::Impl, ::mlir::OpTrait::OpInvariants, ::mlir::ConditionallySpeculatable::Trait, ::mlir::OpTrait::AlwaysSpeculatableImplTrait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::CvtSrBf8F32Op)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::CvtSrFp8F32Op declarations
//===----------------------------------------------------------------------===//

namespace detail {
class CvtSrFp8F32OpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class CvtSrFp8F32OpGenericAdaptor : public detail::CvtSrFp8F32OpGenericAdaptorBase { … };
class CvtSrFp8F32OpAdaptor : public CvtSrFp8F32OpGenericAdaptor<::mlir::ValueRange> { … };
class CvtSrFp8F32Op : public ::mlir::Op<CvtSrFp8F32Op, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<4>::Impl, ::mlir::OpTrait::OpInvariants, ::mlir::ConditionallySpeculatable::Trait, ::mlir::OpTrait::AlwaysSpeculatableImplTrait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::CvtSrFp8F32Op)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::DPPUpdateOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class DPPUpdateOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class DPPUpdateOpGenericAdaptor : public detail::DPPUpdateOpGenericAdaptorBase { … };
class DPPUpdateOpAdaptor : public DPPUpdateOpGenericAdaptor<::mlir::ValueRange> { … };
class DPPUpdateOp : public ::mlir::Op<DPPUpdateOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<2>::Impl, ::mlir::OpTrait::OpInvariants, ::mlir::BytecodeOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::DPPUpdateOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::DsBpermuteOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class DsBpermuteOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class DsBpermuteOpGenericAdaptor : public detail::DsBpermuteOpGenericAdaptorBase { … };
class DsBpermuteOpAdaptor : public DsBpermuteOpGenericAdaptor<::mlir::ValueRange> { … };
class DsBpermuteOp : public ::mlir::Op<DsBpermuteOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::IntegerType>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<2>::Impl, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::DsBpermuteOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::DsSwizzleOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class DsSwizzleOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class DsSwizzleOpGenericAdaptor : public detail::DsSwizzleOpGenericAdaptorBase { … };
class DsSwizzleOpAdaptor : public DsSwizzleOpGenericAdaptor<::mlir::ValueRange> { … };
class DsSwizzleOp : public ::mlir::Op<DsSwizzleOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::IntegerType>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<2>::Impl, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::DsSwizzleOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::GridDimXOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class GridDimXOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class GridDimXOpGenericAdaptor : public detail::GridDimXOpGenericAdaptorBase { … };
class GridDimXOpAdaptor : public GridDimXOpGenericAdaptor<::mlir::ValueRange> { … };
class GridDimXOp : public ::mlir::Op<GridDimXOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::ZeroOperands, ::mlir::OpTrait::OpInvariants, ::mlir::BytecodeOpInterface::Trait, ::mlir::ConditionallySpeculatable::Trait, ::mlir::OpTrait::AlwaysSpeculatableImplTrait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::GridDimXOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::GridDimYOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class GridDimYOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class GridDimYOpGenericAdaptor : public detail::GridDimYOpGenericAdaptorBase { … };
class GridDimYOpAdaptor : public GridDimYOpGenericAdaptor<::mlir::ValueRange> { … };
class GridDimYOp : public ::mlir::Op<GridDimYOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::ZeroOperands, ::mlir::OpTrait::OpInvariants, ::mlir::BytecodeOpInterface::Trait, ::mlir::ConditionallySpeculatable::Trait, ::mlir::OpTrait::AlwaysSpeculatableImplTrait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::GridDimYOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::GridDimZOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class GridDimZOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class GridDimZOpGenericAdaptor : public detail::GridDimZOpGenericAdaptorBase { … };
class GridDimZOpAdaptor : public GridDimZOpGenericAdaptor<::mlir::ValueRange> { … };
class GridDimZOp : public ::mlir::Op<GridDimZOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::ZeroOperands, ::mlir::OpTrait::OpInvariants, ::mlir::BytecodeOpInterface::Trait, ::mlir::ConditionallySpeculatable::Trait, ::mlir::OpTrait::AlwaysSpeculatableImplTrait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::GridDimZOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::MakeBufferRsrcOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class MakeBufferRsrcOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class MakeBufferRsrcOpGenericAdaptor : public detail::MakeBufferRsrcOpGenericAdaptorBase { … };
class MakeBufferRsrcOpAdaptor : public MakeBufferRsrcOpGenericAdaptor<::mlir::ValueRange> { … };
class MakeBufferRsrcOp : public ::mlir::Op<MakeBufferRsrcOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::LLVM::LLVMPointerType>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<4>::Impl, ::mlir::OpTrait::OpInvariants, ::mlir::ConditionallySpeculatable::Trait, ::mlir::OpTrait::AlwaysSpeculatableImplTrait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::MakeBufferRsrcOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::MbcntHiOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class MbcntHiOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class MbcntHiOpGenericAdaptor : public detail::MbcntHiOpGenericAdaptorBase { … };
class MbcntHiOpAdaptor : public MbcntHiOpGenericAdaptor<::mlir::ValueRange> { … };
class MbcntHiOp : public ::mlir::Op<MbcntHiOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<2>::Impl, ::mlir::OpTrait::OpInvariants, ::mlir::ConditionallySpeculatable::Trait, ::mlir::OpTrait::AlwaysSpeculatableImplTrait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::MbcntHiOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::MbcntLoOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class MbcntLoOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class MbcntLoOpGenericAdaptor : public detail::MbcntLoOpGenericAdaptorBase { … };
class MbcntLoOpAdaptor : public MbcntLoOpGenericAdaptor<::mlir::ValueRange> { … };
class MbcntLoOp : public ::mlir::Op<MbcntLoOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<2>::Impl, ::mlir::OpTrait::OpInvariants, ::mlir::ConditionallySpeculatable::Trait, ::mlir::OpTrait::AlwaysSpeculatableImplTrait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::MbcntLoOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::RawBufferAtomicCmpSwap declarations
//===----------------------------------------------------------------------===//

namespace detail {
class RawBufferAtomicCmpSwapGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class RawBufferAtomicCmpSwapGenericAdaptor : public detail::RawBufferAtomicCmpSwapGenericAdaptorBase { … };
class RawBufferAtomicCmpSwapAdaptor : public RawBufferAtomicCmpSwapGenericAdaptor<::mlir::ValueRange> { … };
class RawBufferAtomicCmpSwap : public ::mlir::Op<RawBufferAtomicCmpSwap, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<6>::Impl, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::RawBufferAtomicCmpSwap)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::RawBufferAtomicFAddOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class RawBufferAtomicFAddOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class RawBufferAtomicFAddOpGenericAdaptor : public detail::RawBufferAtomicFAddOpGenericAdaptorBase { … };
class RawBufferAtomicFAddOpAdaptor : public RawBufferAtomicFAddOpGenericAdaptor<::mlir::ValueRange> { … };
class RawBufferAtomicFAddOp : public ::mlir::Op<RawBufferAtomicFAddOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::ZeroResults, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<5>::Impl, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::RawBufferAtomicFAddOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::RawBufferAtomicFMaxOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class RawBufferAtomicFMaxOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class RawBufferAtomicFMaxOpGenericAdaptor : public detail::RawBufferAtomicFMaxOpGenericAdaptorBase { … };
class RawBufferAtomicFMaxOpAdaptor : public RawBufferAtomicFMaxOpGenericAdaptor<::mlir::ValueRange> { … };
class RawBufferAtomicFMaxOp : public ::mlir::Op<RawBufferAtomicFMaxOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::ZeroResults, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<5>::Impl, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::RawBufferAtomicFMaxOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::RawBufferAtomicSMaxOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class RawBufferAtomicSMaxOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class RawBufferAtomicSMaxOpGenericAdaptor : public detail::RawBufferAtomicSMaxOpGenericAdaptorBase { … };
class RawBufferAtomicSMaxOpAdaptor : public RawBufferAtomicSMaxOpGenericAdaptor<::mlir::ValueRange> { … };
class RawBufferAtomicSMaxOp : public ::mlir::Op<RawBufferAtomicSMaxOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::ZeroResults, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<5>::Impl, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::RawBufferAtomicSMaxOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::RawBufferAtomicUMinOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class RawBufferAtomicUMinOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class RawBufferAtomicUMinOpGenericAdaptor : public detail::RawBufferAtomicUMinOpGenericAdaptorBase { … };
class RawBufferAtomicUMinOpAdaptor : public RawBufferAtomicUMinOpGenericAdaptor<::mlir::ValueRange> { … };
class RawBufferAtomicUMinOp : public ::mlir::Op<RawBufferAtomicUMinOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::ZeroResults, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<5>::Impl, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::RawBufferAtomicUMinOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::RawBufferLoadOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class RawBufferLoadOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class RawBufferLoadOpGenericAdaptor : public detail::RawBufferLoadOpGenericAdaptorBase { … };
class RawBufferLoadOpAdaptor : public RawBufferLoadOpGenericAdaptor<::mlir::ValueRange> { … };
class RawBufferLoadOp : public ::mlir::Op<RawBufferLoadOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<4>::Impl, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::RawBufferLoadOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::RawBufferStoreOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class RawBufferStoreOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class RawBufferStoreOpGenericAdaptor : public detail::RawBufferStoreOpGenericAdaptorBase { … };
class RawBufferStoreOpAdaptor : public RawBufferStoreOpGenericAdaptor<::mlir::ValueRange> { … };
class RawBufferStoreOp : public ::mlir::Op<RawBufferStoreOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::ZeroResults, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<5>::Impl, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::RawBufferStoreOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::RawPtrBufferAtomicCmpSwap declarations
//===----------------------------------------------------------------------===//

namespace detail {
class RawPtrBufferAtomicCmpSwapGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class RawPtrBufferAtomicCmpSwapGenericAdaptor : public detail::RawPtrBufferAtomicCmpSwapGenericAdaptorBase { … };
class RawPtrBufferAtomicCmpSwapAdaptor : public RawPtrBufferAtomicCmpSwapGenericAdaptor<::mlir::ValueRange> { … };
class RawPtrBufferAtomicCmpSwap : public ::mlir::Op<RawPtrBufferAtomicCmpSwap, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<6>::Impl, ::mlir::OpTrait::OpInvariants, ::mlir::BytecodeOpInterface::Trait, ::mlir::LLVM::AliasAnalysisOpInterface::Trait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::RawPtrBufferAtomicCmpSwap)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::RawPtrBufferAtomicFaddOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class RawPtrBufferAtomicFaddOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class RawPtrBufferAtomicFaddOpGenericAdaptor : public detail::RawPtrBufferAtomicFaddOpGenericAdaptorBase { … };
class RawPtrBufferAtomicFaddOpAdaptor : public RawPtrBufferAtomicFaddOpGenericAdaptor<::mlir::ValueRange> { … };
class RawPtrBufferAtomicFaddOp : public ::mlir::Op<RawPtrBufferAtomicFaddOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::ZeroResults, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<5>::Impl, ::mlir::OpTrait::OpInvariants, ::mlir::BytecodeOpInterface::Trait, ::mlir::LLVM::AliasAnalysisOpInterface::Trait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::RawPtrBufferAtomicFaddOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::RawPtrBufferAtomicFmaxOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class RawPtrBufferAtomicFmaxOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class RawPtrBufferAtomicFmaxOpGenericAdaptor : public detail::RawPtrBufferAtomicFmaxOpGenericAdaptorBase { … };
class RawPtrBufferAtomicFmaxOpAdaptor : public RawPtrBufferAtomicFmaxOpGenericAdaptor<::mlir::ValueRange> { … };
class RawPtrBufferAtomicFmaxOp : public ::mlir::Op<RawPtrBufferAtomicFmaxOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::ZeroResults, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<5>::Impl, ::mlir::OpTrait::OpInvariants, ::mlir::BytecodeOpInterface::Trait, ::mlir::LLVM::AliasAnalysisOpInterface::Trait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::RawPtrBufferAtomicFmaxOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::RawPtrBufferAtomicSmaxOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class RawPtrBufferAtomicSmaxOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class RawPtrBufferAtomicSmaxOpGenericAdaptor : public detail::RawPtrBufferAtomicSmaxOpGenericAdaptorBase { … };
class RawPtrBufferAtomicSmaxOpAdaptor : public RawPtrBufferAtomicSmaxOpGenericAdaptor<::mlir::ValueRange> { … };
class RawPtrBufferAtomicSmaxOp : public ::mlir::Op<RawPtrBufferAtomicSmaxOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::ZeroResults, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<5>::Impl, ::mlir::OpTrait::OpInvariants, ::mlir::BytecodeOpInterface::Trait, ::mlir::LLVM::AliasAnalysisOpInterface::Trait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::RawPtrBufferAtomicSmaxOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::RawPtrBufferAtomicUminOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class RawPtrBufferAtomicUminOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class RawPtrBufferAtomicUminOpGenericAdaptor : public detail::RawPtrBufferAtomicUminOpGenericAdaptorBase { … };
class RawPtrBufferAtomicUminOpAdaptor : public RawPtrBufferAtomicUminOpGenericAdaptor<::mlir::ValueRange> { … };
class RawPtrBufferAtomicUminOp : public ::mlir::Op<RawPtrBufferAtomicUminOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::ZeroResults, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<5>::Impl, ::mlir::OpTrait::OpInvariants, ::mlir::BytecodeOpInterface::Trait, ::mlir::LLVM::AliasAnalysisOpInterface::Trait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::RawPtrBufferAtomicUminOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::RawPtrBufferLoadOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class RawPtrBufferLoadOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class RawPtrBufferLoadOpGenericAdaptor : public detail::RawPtrBufferLoadOpGenericAdaptorBase { … };
class RawPtrBufferLoadOpAdaptor : public RawPtrBufferLoadOpGenericAdaptor<::mlir::ValueRange> { … };
class RawPtrBufferLoadOp : public ::mlir::Op<RawPtrBufferLoadOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<4>::Impl, ::mlir::OpTrait::OpInvariants, ::mlir::BytecodeOpInterface::Trait, ::mlir::LLVM::AliasAnalysisOpInterface::Trait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::RawPtrBufferLoadOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::RawPtrBufferStoreOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class RawPtrBufferStoreOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class RawPtrBufferStoreOpGenericAdaptor : public detail::RawPtrBufferStoreOpGenericAdaptorBase { … };
class RawPtrBufferStoreOpAdaptor : public RawPtrBufferStoreOpGenericAdaptor<::mlir::ValueRange> { … };
class RawPtrBufferStoreOp : public ::mlir::Op<RawPtrBufferStoreOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::ZeroResults, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::NOperands<5>::Impl, ::mlir::OpTrait::OpInvariants, ::mlir::BytecodeOpInterface::Trait, ::mlir::LLVM::AliasAnalysisOpInterface::Trait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::RawPtrBufferStoreOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::SBarrierOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class SBarrierOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class SBarrierOpGenericAdaptor : public detail::SBarrierOpGenericAdaptorBase { … };
class SBarrierOpAdaptor : public SBarrierOpGenericAdaptor<::mlir::ValueRange> { … };
class SBarrierOp : public ::mlir::Op<SBarrierOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::ZeroResults, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::ZeroOperands, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::SBarrierOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::SchedBarrier declarations
//===----------------------------------------------------------------------===//

namespace detail {
class SchedBarrierGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class SchedBarrierGenericAdaptor : public detail::SchedBarrierGenericAdaptorBase { … };
class SchedBarrierAdaptor : public SchedBarrierGenericAdaptor<::mlir::ValueRange> { … };
class SchedBarrier : public ::mlir::Op<SchedBarrier, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::ZeroResults, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::ZeroOperands, ::mlir::OpTrait::OpInvariants, ::mlir::BytecodeOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::SchedBarrier)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::SetPrioOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class SetPrioOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class SetPrioOpGenericAdaptor : public detail::SetPrioOpGenericAdaptorBase { … };
class SetPrioOpAdaptor : public SetPrioOpGenericAdaptor<::mlir::ValueRange> { … };
class SetPrioOp : public ::mlir::Op<SetPrioOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::ZeroResults, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::ZeroOperands, ::mlir::OpTrait::OpInvariants, ::mlir::BytecodeOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::SetPrioOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::ThreadIdXOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class ThreadIdXOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class ThreadIdXOpGenericAdaptor : public detail::ThreadIdXOpGenericAdaptorBase { … };
class ThreadIdXOpAdaptor : public ThreadIdXOpGenericAdaptor<::mlir::ValueRange> { … };
class ThreadIdXOp : public ::mlir::Op<ThreadIdXOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::ZeroOperands, ::mlir::OpTrait::OpInvariants, ::mlir::BytecodeOpInterface::Trait, ::mlir::ConditionallySpeculatable::Trait, ::mlir::OpTrait::AlwaysSpeculatableImplTrait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::ThreadIdXOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::ThreadIdYOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class ThreadIdYOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class ThreadIdYOpGenericAdaptor : public detail::ThreadIdYOpGenericAdaptorBase { … };
class ThreadIdYOpAdaptor : public ThreadIdYOpGenericAdaptor<::mlir::ValueRange> { … };
class ThreadIdYOp : public ::mlir::Op<ThreadIdYOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::ZeroOperands, ::mlir::OpTrait::OpInvariants, ::mlir::BytecodeOpInterface::Trait, ::mlir::ConditionallySpeculatable::Trait, ::mlir::OpTrait::AlwaysSpeculatableImplTrait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::ThreadIdYOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::ThreadIdZOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class ThreadIdZOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class ThreadIdZOpGenericAdaptor : public detail::ThreadIdZOpGenericAdaptorBase { … };
class ThreadIdZOpAdaptor : public ThreadIdZOpGenericAdaptor<::mlir::ValueRange> { … };
class ThreadIdZOp : public ::mlir::Op<ThreadIdZOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::ZeroOperands, ::mlir::OpTrait::OpInvariants, ::mlir::BytecodeOpInterface::Trait, ::mlir::ConditionallySpeculatable::Trait, ::mlir::OpTrait::AlwaysSpeculatableImplTrait, ::mlir::MemoryEffectOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::ThreadIdZOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::WaitDscntOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class WaitDscntOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class WaitDscntOpGenericAdaptor : public detail::WaitDscntOpGenericAdaptorBase { … };
class WaitDscntOpAdaptor : public WaitDscntOpGenericAdaptor<::mlir::ValueRange> { … };
class WaitDscntOp : public ::mlir::Op<WaitDscntOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::ZeroResults, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::ZeroOperands, ::mlir::OpTrait::OpInvariants, ::mlir::BytecodeOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::WaitDscntOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::WaitcntOp declarations
//===----------------------------------------------------------------------===//

namespace detail {
class WaitcntOpGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class WaitcntOpGenericAdaptor : public detail::WaitcntOpGenericAdaptorBase { … };
class WaitcntOpAdaptor : public WaitcntOpGenericAdaptor<::mlir::ValueRange> { … };
class WaitcntOp : public ::mlir::Op<WaitcntOp, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::ZeroResults, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::ZeroOperands, ::mlir::OpTrait::OpInvariants, ::mlir::BytecodeOpInterface::Trait> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::WaitcntOp)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::mfma_f32_16x16x16bf16_1k declarations
//===----------------------------------------------------------------------===//

namespace detail {
class mfma_f32_16x16x16bf16_1kGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class mfma_f32_16x16x16bf16_1kGenericAdaptor : public detail::mfma_f32_16x16x16bf16_1kGenericAdaptorBase { … };
class mfma_f32_16x16x16bf16_1kAdaptor : public mfma_f32_16x16x16bf16_1kGenericAdaptor<::mlir::ValueRange> { … };
class mfma_f32_16x16x16bf16_1k : public ::mlir::Op<mfma_f32_16x16x16bf16_1k, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_16x16x16bf16_1k)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::mfma_f32_16x16x16f16 declarations
//===----------------------------------------------------------------------===//

namespace detail {
class mfma_f32_16x16x16f16GenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class mfma_f32_16x16x16f16GenericAdaptor : public detail::mfma_f32_16x16x16f16GenericAdaptorBase { … };
class mfma_f32_16x16x16f16Adaptor : public mfma_f32_16x16x16f16GenericAdaptor<::mlir::ValueRange> { … };
class mfma_f32_16x16x16f16 : public ::mlir::Op<mfma_f32_16x16x16f16, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_16x16x16f16)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::mfma_f32_16x16x1f32 declarations
//===----------------------------------------------------------------------===//

namespace detail {
class mfma_f32_16x16x1f32GenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class mfma_f32_16x16x1f32GenericAdaptor : public detail::mfma_f32_16x16x1f32GenericAdaptorBase { … };
class mfma_f32_16x16x1f32Adaptor : public mfma_f32_16x16x1f32GenericAdaptor<::mlir::ValueRange> { … };
class mfma_f32_16x16x1f32 : public ::mlir::Op<mfma_f32_16x16x1f32, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_16x16x1f32)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::mfma_f32_16x16x2bf16 declarations
//===----------------------------------------------------------------------===//

namespace detail {
class mfma_f32_16x16x2bf16GenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class mfma_f32_16x16x2bf16GenericAdaptor : public detail::mfma_f32_16x16x2bf16GenericAdaptorBase { … };
class mfma_f32_16x16x2bf16Adaptor : public mfma_f32_16x16x2bf16GenericAdaptor<::mlir::ValueRange> { … };
class mfma_f32_16x16x2bf16 : public ::mlir::Op<mfma_f32_16x16x2bf16, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_16x16x2bf16)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::mfma_f32_16x16x32_bf8_bf8 declarations
//===----------------------------------------------------------------------===//

namespace detail {
class mfma_f32_16x16x32_bf8_bf8GenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class mfma_f32_16x16x32_bf8_bf8GenericAdaptor : public detail::mfma_f32_16x16x32_bf8_bf8GenericAdaptorBase { … };
class mfma_f32_16x16x32_bf8_bf8Adaptor : public mfma_f32_16x16x32_bf8_bf8GenericAdaptor<::mlir::ValueRange> { … };
class mfma_f32_16x16x32_bf8_bf8 : public ::mlir::Op<mfma_f32_16x16x32_bf8_bf8, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_16x16x32_bf8_bf8)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::mfma_f32_16x16x32_bf8_fp8 declarations
//===----------------------------------------------------------------------===//

namespace detail {
class mfma_f32_16x16x32_bf8_fp8GenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class mfma_f32_16x16x32_bf8_fp8GenericAdaptor : public detail::mfma_f32_16x16x32_bf8_fp8GenericAdaptorBase { … };
class mfma_f32_16x16x32_bf8_fp8Adaptor : public mfma_f32_16x16x32_bf8_fp8GenericAdaptor<::mlir::ValueRange> { … };
class mfma_f32_16x16x32_bf8_fp8 : public ::mlir::Op<mfma_f32_16x16x32_bf8_fp8, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_16x16x32_bf8_fp8)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::mfma_f32_16x16x32_fp8_bf8 declarations
//===----------------------------------------------------------------------===//

namespace detail {
class mfma_f32_16x16x32_fp8_bf8GenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class mfma_f32_16x16x32_fp8_bf8GenericAdaptor : public detail::mfma_f32_16x16x32_fp8_bf8GenericAdaptorBase { … };
class mfma_f32_16x16x32_fp8_bf8Adaptor : public mfma_f32_16x16x32_fp8_bf8GenericAdaptor<::mlir::ValueRange> { … };
class mfma_f32_16x16x32_fp8_bf8 : public ::mlir::Op<mfma_f32_16x16x32_fp8_bf8, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_16x16x32_fp8_bf8)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::mfma_f32_16x16x32_fp8_fp8 declarations
//===----------------------------------------------------------------------===//

namespace detail {
class mfma_f32_16x16x32_fp8_fp8GenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class mfma_f32_16x16x32_fp8_fp8GenericAdaptor : public detail::mfma_f32_16x16x32_fp8_fp8GenericAdaptorBase { … };
class mfma_f32_16x16x32_fp8_fp8Adaptor : public mfma_f32_16x16x32_fp8_fp8GenericAdaptor<::mlir::ValueRange> { … };
class mfma_f32_16x16x32_fp8_fp8 : public ::mlir::Op<mfma_f32_16x16x32_fp8_fp8, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_16x16x32_fp8_fp8)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::mfma_f32_16x16x4bf16_1k declarations
//===----------------------------------------------------------------------===//

namespace detail {
class mfma_f32_16x16x4bf16_1kGenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class mfma_f32_16x16x4bf16_1kGenericAdaptor : public detail::mfma_f32_16x16x4bf16_1kGenericAdaptorBase { … };
class mfma_f32_16x16x4bf16_1kAdaptor : public mfma_f32_16x16x4bf16_1kGenericAdaptor<::mlir::ValueRange> { … };
class mfma_f32_16x16x4bf16_1k : public ::mlir::Op<mfma_f32_16x16x4bf16_1k, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_16x16x4bf16_1k)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::mfma_f32_16x16x4f16 declarations
//===----------------------------------------------------------------------===//

namespace detail {
class mfma_f32_16x16x4f16GenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class mfma_f32_16x16x4f16GenericAdaptor : public detail::mfma_f32_16x16x4f16GenericAdaptorBase { … };
class mfma_f32_16x16x4f16Adaptor : public mfma_f32_16x16x4f16GenericAdaptor<::mlir::ValueRange> { … };
class mfma_f32_16x16x4f16 : public ::mlir::Op<mfma_f32_16x16x4f16, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_16x16x4f16)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::mfma_f32_16x16x4f32 declarations
//===----------------------------------------------------------------------===//

namespace detail {
class mfma_f32_16x16x4f32GenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class mfma_f32_16x16x4f32GenericAdaptor : public detail::mfma_f32_16x16x4f32GenericAdaptorBase { … };
class mfma_f32_16x16x4f32Adaptor : public mfma_f32_16x16x4f32GenericAdaptor<::mlir::ValueRange> { … };
class mfma_f32_16x16x4f32 : public ::mlir::Op<mfma_f32_16x16x4f32, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_16x16x4f32)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::mfma_f32_16x16x8_xf32 declarations
//===----------------------------------------------------------------------===//

namespace detail {
class mfma_f32_16x16x8_xf32GenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class mfma_f32_16x16x8_xf32GenericAdaptor : public detail::mfma_f32_16x16x8_xf32GenericAdaptorBase { … };
class mfma_f32_16x16x8_xf32Adaptor : public mfma_f32_16x16x8_xf32GenericAdaptor<::mlir::ValueRange> { … };
class mfma_f32_16x16x8_xf32 : public ::mlir::Op<mfma_f32_16x16x8_xf32, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_16x16x8_xf32)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::mfma_f32_16x16x8bf16 declarations
//===----------------------------------------------------------------------===//

namespace detail {
class mfma_f32_16x16x8bf16GenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class mfma_f32_16x16x8bf16GenericAdaptor : public detail::mfma_f32_16x16x8bf16GenericAdaptorBase { … };
class mfma_f32_16x16x8bf16Adaptor : public mfma_f32_16x16x8bf16GenericAdaptor<::mlir::ValueRange> { … };
class mfma_f32_16x16x8bf16 : public ::mlir::Op<mfma_f32_16x16x8bf16, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_16x16x8bf16)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::mfma_f32_32x32x16_bf8_bf8 declarations
//===----------------------------------------------------------------------===//

namespace detail {
class mfma_f32_32x32x16_bf8_bf8GenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class mfma_f32_32x32x16_bf8_bf8GenericAdaptor : public detail::mfma_f32_32x32x16_bf8_bf8GenericAdaptorBase { … };
class mfma_f32_32x32x16_bf8_bf8Adaptor : public mfma_f32_32x32x16_bf8_bf8GenericAdaptor<::mlir::ValueRange> { … };
class mfma_f32_32x32x16_bf8_bf8 : public ::mlir::Op<mfma_f32_32x32x16_bf8_bf8, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_32x32x16_bf8_bf8)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::mfma_f32_32x32x16_bf8_fp8 declarations
//===----------------------------------------------------------------------===//

namespace detail {
class mfma_f32_32x32x16_bf8_fp8GenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class mfma_f32_32x32x16_bf8_fp8GenericAdaptor : public detail::mfma_f32_32x32x16_bf8_fp8GenericAdaptorBase { … };
class mfma_f32_32x32x16_bf8_fp8Adaptor : public mfma_f32_32x32x16_bf8_fp8GenericAdaptor<::mlir::ValueRange> { … };
class mfma_f32_32x32x16_bf8_fp8 : public ::mlir::Op<mfma_f32_32x32x16_bf8_fp8, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_32x32x16_bf8_fp8)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::mfma_f32_32x32x16_fp8_bf8 declarations
//===----------------------------------------------------------------------===//

namespace detail {
class mfma_f32_32x32x16_fp8_bf8GenericAdaptorBase { … };
} // namespace detail
template <typename RangeT>
class mfma_f32_32x32x16_fp8_bf8GenericAdaptor : public detail::mfma_f32_32x32x16_fp8_bf8GenericAdaptorBase { … };
class mfma_f32_32x32x16_fp8_bf8Adaptor : public mfma_f32_32x32x16_fp8_bf8GenericAdaptor<::mlir::ValueRange> { … };
class mfma_f32_32x32x16_fp8_bf8 : public ::mlir::Op<mfma_f32_32x32x16_fp8_bf8, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … };
} // namespace ROCDL
} // namespace mlir
MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_32x32x16_fp8_bf8)

namespace mlir {
namespace ROCDL {

//===----------------------------------------------------------------------===//
// ::mlir::ROCDL::mfma_f32_32x32x16_fp8_fp8 declarations
//===----------------------------------------------------------------------===//

namespace detail {
class mfma_f32_32x32x16_fp8_fp8GenericAdaptorBase { … }template <typename RangeT>
class mfma_f32_32x32x16_fp8_fp8GenericAdaptor : public detail::mfma_f32_32x32x16_fp8_fp8GenericAdaptorBase { … }class mfma_f32_32x32x16_fp8_fp8Adaptor : public mfma_f32_32x32x16_fp8_fp8GenericAdaptor<::mlir::ValueRange> { … }class mfma_f32_32x32x16_fp8_fp8 : public ::mlir::Op<mfma_f32_32x32x16_fp8_fp8, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_32x32x16_fp8_fp8)class mfma_f32_32x32x1f32GenericAdaptorBase { … }template <typename RangeT>
class mfma_f32_32x32x1f32GenericAdaptor : public detail::mfma_f32_32x32x1f32GenericAdaptorBase { … }class mfma_f32_32x32x1f32Adaptor : public mfma_f32_32x32x1f32GenericAdaptor<::mlir::ValueRange> { … }class mfma_f32_32x32x1f32 : public ::mlir::Op<mfma_f32_32x32x1f32, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_32x32x1f32)class mfma_f32_32x32x2bf16GenericAdaptorBase { … }template <typename RangeT>
class mfma_f32_32x32x2bf16GenericAdaptor : public detail::mfma_f32_32x32x2bf16GenericAdaptorBase { … }class mfma_f32_32x32x2bf16Adaptor : public mfma_f32_32x32x2bf16GenericAdaptor<::mlir::ValueRange> { … }class mfma_f32_32x32x2bf16 : public ::mlir::Op<mfma_f32_32x32x2bf16, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_32x32x2bf16)class mfma_f32_32x32x2f32GenericAdaptorBase { … }template <typename RangeT>
class mfma_f32_32x32x2f32GenericAdaptor : public detail::mfma_f32_32x32x2f32GenericAdaptorBase { … }class mfma_f32_32x32x2f32Adaptor : public mfma_f32_32x32x2f32GenericAdaptor<::mlir::ValueRange> { … }class mfma_f32_32x32x2f32 : public ::mlir::Op<mfma_f32_32x32x2f32, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_32x32x2f32)class mfma_f32_32x32x4_xf32GenericAdaptorBase { … }template <typename RangeT>
class mfma_f32_32x32x4_xf32GenericAdaptor : public detail::mfma_f32_32x32x4_xf32GenericAdaptorBase { … }class mfma_f32_32x32x4_xf32Adaptor : public mfma_f32_32x32x4_xf32GenericAdaptor<::mlir::ValueRange> { … }class mfma_f32_32x32x4_xf32 : public ::mlir::Op<mfma_f32_32x32x4_xf32, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_32x32x4_xf32)class mfma_f32_32x32x4bf16GenericAdaptorBase { … }template <typename RangeT>
class mfma_f32_32x32x4bf16GenericAdaptor : public detail::mfma_f32_32x32x4bf16GenericAdaptorBase { … }class mfma_f32_32x32x4bf16Adaptor : public mfma_f32_32x32x4bf16GenericAdaptor<::mlir::ValueRange> { … }class mfma_f32_32x32x4bf16 : public ::mlir::Op<mfma_f32_32x32x4bf16, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_32x32x4bf16)class mfma_f32_32x32x4bf16_1kGenericAdaptorBase { … }template <typename RangeT>
class mfma_f32_32x32x4bf16_1kGenericAdaptor : public detail::mfma_f32_32x32x4bf16_1kGenericAdaptorBase { … }class mfma_f32_32x32x4bf16_1kAdaptor : public mfma_f32_32x32x4bf16_1kGenericAdaptor<::mlir::ValueRange> { … }class mfma_f32_32x32x4bf16_1k : public ::mlir::Op<mfma_f32_32x32x4bf16_1k, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_32x32x4bf16_1k)class mfma_f32_32x32x4f16GenericAdaptorBase { … }template <typename RangeT>
class mfma_f32_32x32x4f16GenericAdaptor : public detail::mfma_f32_32x32x4f16GenericAdaptorBase { … }class mfma_f32_32x32x4f16Adaptor : public mfma_f32_32x32x4f16GenericAdaptor<::mlir::ValueRange> { … }class mfma_f32_32x32x4f16 : public ::mlir::Op<mfma_f32_32x32x4f16, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_32x32x4f16)class mfma_f32_32x32x8bf16_1kGenericAdaptorBase { … }template <typename RangeT>
class mfma_f32_32x32x8bf16_1kGenericAdaptor : public detail::mfma_f32_32x32x8bf16_1kGenericAdaptorBase { … }class mfma_f32_32x32x8bf16_1kAdaptor : public mfma_f32_32x32x8bf16_1kGenericAdaptor<::mlir::ValueRange> { … }class mfma_f32_32x32x8bf16_1k : public ::mlir::Op<mfma_f32_32x32x8bf16_1k, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_32x32x8bf16_1k)class mfma_f32_32x32x8f16GenericAdaptorBase { … }template <typename RangeT>
class mfma_f32_32x32x8f16GenericAdaptor : public detail::mfma_f32_32x32x8f16GenericAdaptorBase { … }class mfma_f32_32x32x8f16Adaptor : public mfma_f32_32x32x8f16GenericAdaptor<::mlir::ValueRange> { … }class mfma_f32_32x32x8f16 : public ::mlir::Op<mfma_f32_32x32x8f16, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_32x32x8f16)class mfma_f32_4x4x1f32GenericAdaptorBase { … }template <typename RangeT>
class mfma_f32_4x4x1f32GenericAdaptor : public detail::mfma_f32_4x4x1f32GenericAdaptorBase { … }class mfma_f32_4x4x1f32Adaptor : public mfma_f32_4x4x1f32GenericAdaptor<::mlir::ValueRange> { … }class mfma_f32_4x4x1f32 : public ::mlir::Op<mfma_f32_4x4x1f32, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_4x4x1f32)class mfma_f32_4x4x2bf16GenericAdaptorBase { … }template <typename RangeT>
class mfma_f32_4x4x2bf16GenericAdaptor : public detail::mfma_f32_4x4x2bf16GenericAdaptorBase { … }class mfma_f32_4x4x2bf16Adaptor : public mfma_f32_4x4x2bf16GenericAdaptor<::mlir::ValueRange> { … }class mfma_f32_4x4x2bf16 : public ::mlir::Op<mfma_f32_4x4x2bf16, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_4x4x2bf16)class mfma_f32_4x4x4bf16_1kGenericAdaptorBase { … }template <typename RangeT>
class mfma_f32_4x4x4bf16_1kGenericAdaptor : public detail::mfma_f32_4x4x4bf16_1kGenericAdaptorBase { … }class mfma_f32_4x4x4bf16_1kAdaptor : public mfma_f32_4x4x4bf16_1kGenericAdaptor<::mlir::ValueRange> { … }class mfma_f32_4x4x4bf16_1k : public ::mlir::Op<mfma_f32_4x4x4bf16_1k, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_4x4x4bf16_1k)class mfma_f32_4x4x4f16GenericAdaptorBase { … }template <typename RangeT>
class mfma_f32_4x4x4f16GenericAdaptor : public detail::mfma_f32_4x4x4f16GenericAdaptorBase { … }class mfma_f32_4x4x4f16Adaptor : public mfma_f32_4x4x4f16GenericAdaptor<::mlir::ValueRange> { … }class mfma_f32_4x4x4f16 : public ::mlir::Op<mfma_f32_4x4x4f16, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f32_4x4x4f16)class mfma_f64_16x16x4f64GenericAdaptorBase { … }template <typename RangeT>
class mfma_f64_16x16x4f64GenericAdaptor : public detail::mfma_f64_16x16x4f64GenericAdaptorBase { … }class mfma_f64_16x16x4f64Adaptor : public mfma_f64_16x16x4f64GenericAdaptor<::mlir::ValueRange> { … }class mfma_f64_16x16x4f64 : public ::mlir::Op<mfma_f64_16x16x4f64, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f64_16x16x4f64)class mfma_f64_4x4x4f64GenericAdaptorBase { … }template <typename RangeT>
class mfma_f64_4x4x4f64GenericAdaptor : public detail::mfma_f64_4x4x4f64GenericAdaptorBase { … }class mfma_f64_4x4x4f64Adaptor : public mfma_f64_4x4x4f64GenericAdaptor<::mlir::ValueRange> { … }class mfma_f64_4x4x4f64 : public ::mlir::Op<mfma_f64_4x4x4f64, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_f64_4x4x4f64)class mfma_i32_16x16x16i8GenericAdaptorBase { … }template <typename RangeT>
class mfma_i32_16x16x16i8GenericAdaptor : public detail::mfma_i32_16x16x16i8GenericAdaptorBase { … }class mfma_i32_16x16x16i8Adaptor : public mfma_i32_16x16x16i8GenericAdaptor<::mlir::ValueRange> { … }class mfma_i32_16x16x16i8 : public ::mlir::Op<mfma_i32_16x16x16i8, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_i32_16x16x16i8)class mfma_i32_16x16x32_i8GenericAdaptorBase { … }template <typename RangeT>
class mfma_i32_16x16x32_i8GenericAdaptor : public detail::mfma_i32_16x16x32_i8GenericAdaptorBase { … }class mfma_i32_16x16x32_i8Adaptor : public mfma_i32_16x16x32_i8GenericAdaptor<::mlir::ValueRange> { … }class mfma_i32_16x16x32_i8 : public ::mlir::Op<mfma_i32_16x16x32_i8, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_i32_16x16x32_i8)class mfma_i32_16x16x4i8GenericAdaptorBase { … }template <typename RangeT>
class mfma_i32_16x16x4i8GenericAdaptor : public detail::mfma_i32_16x16x4i8GenericAdaptorBase { … }class mfma_i32_16x16x4i8Adaptor : public mfma_i32_16x16x4i8GenericAdaptor<::mlir::ValueRange> { … }class mfma_i32_16x16x4i8 : public ::mlir::Op<mfma_i32_16x16x4i8, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_i32_16x16x4i8)class mfma_i32_32x32x16_i8GenericAdaptorBase { … }template <typename RangeT>
class mfma_i32_32x32x16_i8GenericAdaptor : public detail::mfma_i32_32x32x16_i8GenericAdaptorBase { … }class mfma_i32_32x32x16_i8Adaptor : public mfma_i32_32x32x16_i8GenericAdaptor<::mlir::ValueRange> { … }class mfma_i32_32x32x16_i8 : public ::mlir::Op<mfma_i32_32x32x16_i8, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_i32_32x32x16_i8)class mfma_i32_32x32x4i8GenericAdaptorBase { … }template <typename RangeT>
class mfma_i32_32x32x4i8GenericAdaptor : public detail::mfma_i32_32x32x4i8GenericAdaptorBase { … }class mfma_i32_32x32x4i8Adaptor : public mfma_i32_32x32x4i8GenericAdaptor<::mlir::ValueRange> { … }class mfma_i32_32x32x4i8 : public ::mlir::Op<mfma_i32_32x32x4i8, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_i32_32x32x4i8)class mfma_i32_32x32x8i8GenericAdaptorBase { … }template <typename RangeT>
class mfma_i32_32x32x8i8GenericAdaptor : public detail::mfma_i32_32x32x8i8GenericAdaptorBase { … }class mfma_i32_32x32x8i8Adaptor : public mfma_i32_32x32x8i8GenericAdaptor<::mlir::ValueRange> { … }class mfma_i32_32x32x8i8 : public ::mlir::Op<mfma_i32_32x32x8i8, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_i32_32x32x8i8)class mfma_i32_4x4x4i8GenericAdaptorBase { … }template <typename RangeT>
class mfma_i32_4x4x4i8GenericAdaptor : public detail::mfma_i32_4x4x4i8GenericAdaptorBase { … }class mfma_i32_4x4x4i8Adaptor : public mfma_i32_4x4x4i8GenericAdaptor<::mlir::ValueRange> { … }class mfma_i32_4x4x4i8 : public ::mlir::Op<mfma_i32_4x4x4i8, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::mfma_i32_4x4x4i8)class wmma_bf16_16x16x16_bf16GenericAdaptorBase { … }template <typename RangeT>
class wmma_bf16_16x16x16_bf16GenericAdaptor : public detail::wmma_bf16_16x16x16_bf16GenericAdaptorBase { … }class wmma_bf16_16x16x16_bf16Adaptor : public wmma_bf16_16x16x16_bf16GenericAdaptor<::mlir::ValueRange> { … }class wmma_bf16_16x16x16_bf16 : public ::mlir::Op<wmma_bf16_16x16x16_bf16, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::wmma_bf16_16x16x16_bf16)class wmma_f16_16x16x16_f16GenericAdaptorBase { … }template <typename RangeT>
class wmma_f16_16x16x16_f16GenericAdaptor : public detail::wmma_f16_16x16x16_f16GenericAdaptorBase { … }class wmma_f16_16x16x16_f16Adaptor : public wmma_f16_16x16x16_f16GenericAdaptor<::mlir::ValueRange> { … }class wmma_f16_16x16x16_f16 : public ::mlir::Op<wmma_f16_16x16x16_f16, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::wmma_f16_16x16x16_f16)class wmma_f32_16x16x16_bf16GenericAdaptorBase { … }template <typename RangeT>
class wmma_f32_16x16x16_bf16GenericAdaptor : public detail::wmma_f32_16x16x16_bf16GenericAdaptorBase { … }class wmma_f32_16x16x16_bf16Adaptor : public wmma_f32_16x16x16_bf16GenericAdaptor<::mlir::ValueRange> { … }class wmma_f32_16x16x16_bf16 : public ::mlir::Op<wmma_f32_16x16x16_bf16, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::wmma_f32_16x16x16_bf16)class wmma_f32_16x16x16_bf8GenericAdaptorBase { … }template <typename RangeT>
class wmma_f32_16x16x16_bf8GenericAdaptor : public detail::wmma_f32_16x16x16_bf8GenericAdaptorBase { … }class wmma_f32_16x16x16_bf8Adaptor : public wmma_f32_16x16x16_bf8GenericAdaptor<::mlir::ValueRange> { … }class wmma_f32_16x16x16_bf8 : public ::mlir::Op<wmma_f32_16x16x16_bf8, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::wmma_f32_16x16x16_bf8)class wmma_f32_16x16x16_f16GenericAdaptorBase { … }template <typename RangeT>
class wmma_f32_16x16x16_f16GenericAdaptor : public detail::wmma_f32_16x16x16_f16GenericAdaptorBase { … }class wmma_f32_16x16x16_f16Adaptor : public wmma_f32_16x16x16_f16GenericAdaptor<::mlir::ValueRange> { … }class wmma_f32_16x16x16_f16 : public ::mlir::Op<wmma_f32_16x16x16_f16, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::wmma_f32_16x16x16_f16)class wmma_f32_16x16x16_fp8GenericAdaptorBase { … }template <typename RangeT>
class wmma_f32_16x16x16_fp8GenericAdaptor : public detail::wmma_f32_16x16x16_fp8GenericAdaptorBase { … }class wmma_f32_16x16x16_fp8Adaptor : public wmma_f32_16x16x16_fp8GenericAdaptor<::mlir::ValueRange> { … }class wmma_f32_16x16x16_fp8 : public ::mlir::Op<wmma_f32_16x16x16_fp8, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::wmma_f32_16x16x16_fp8)class wmma_i32_16x16x16_iu4GenericAdaptorBase { … }template <typename RangeT>
class wmma_i32_16x16x16_iu4GenericAdaptor : public detail::wmma_i32_16x16x16_iu4GenericAdaptorBase { … }class wmma_i32_16x16x16_iu4Adaptor : public wmma_i32_16x16x16_iu4GenericAdaptor<::mlir::ValueRange> { … }class wmma_i32_16x16x16_iu4 : public ::mlir::Op<wmma_i32_16x16x16_iu4, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::wmma_i32_16x16x16_iu4)class wmma_i32_16x16x16_iu8GenericAdaptorBase { … }template <typename RangeT>
class wmma_i32_16x16x16_iu8GenericAdaptor : public detail::wmma_i32_16x16x16_iu8GenericAdaptorBase { … }class wmma_i32_16x16x16_iu8Adaptor : public wmma_i32_16x16x16_iu8GenericAdaptor<::mlir::ValueRange> { … }class wmma_i32_16x16x16_iu8 : public ::mlir::Op<wmma_i32_16x16x16_iu8, ::mlir::OpTrait::ZeroRegions, ::mlir::OpTrait::OneResult, ::mlir::OpTrait::OneTypedResult<::mlir::Type>::Impl, ::mlir::OpTrait::ZeroSuccessors, ::mlir::OpTrait::VariadicOperands, ::mlir::OpTrait::OpInvariants> { … }MLIR_DECLARE_EXPLICIT_TYPE_ID(::mlir::ROCDL::wmma_i32_16x16x16_iu8)#endif  // GET_OP_CLASSES