14 #ifndef MLIR_DIALECT_LLVMIR_NVVMDIALECT_H_
15 #define MLIR_DIALECT_LLVMIR_NVVMDIALECT_H_
28 #include "llvm/IR/IntrinsicsNVPTX.h"
30 #include "mlir/Dialect/LLVMIR/NVVMOpsEnums.h.inc"
35 inline bool operator==(
unsigned as, NVVMMemorySpace memSpace) {
36 return as ==
static_cast<unsigned>(memSpace);
38 inline bool operator==(NVVMMemorySpace memSpace,
unsigned as) {
39 return static_cast<unsigned>(memSpace) == as;
41 inline bool operator!=(
unsigned as, NVVMMemorySpace memSpace) {
42 return as !=
static_cast<unsigned>(memSpace);
44 inline bool operator!=(NVVMMemorySpace memSpace,
unsigned as) {
45 return static_cast<unsigned>(memSpace) != as;
54 std::pair<llvm::Intrinsic::ID, llvm::SmallVector<llvm::Value *>>;
59 std::pair<mlir::Type, unsigned>
inferMMAType(mlir::NVVM::MMATypes type,
60 mlir::NVVM::MMAFrag frag,
int nRow,
67 #define GET_ATTRDEF_CLASSES
68 #include "mlir/Dialect/LLVMIR/NVVMOpsAttributes.h.inc"
70 #define GET_OP_CLASSES
71 #include "mlir/Dialect/LLVMIR/NVVMOps.h.inc"
73 #include "mlir/Dialect/LLVMIR/NVVMOpsDialect.h.inc"
MLIRContext is the top-level object for a collection of MLIR operations.
std::pair< llvm::Intrinsic::ID, llvm::SmallVector< llvm::Value * > > IDArgPair
A pair type of LLVM's Intrinsic ID and args (which are llvm values).
constexpr int kSharedMemoryAlignmentBit
bool operator==(unsigned as, NVVMMemorySpace memSpace)
Utility functions to compare NVVMMemorySpace with unsigned values.
bool operator!=(unsigned as, NVVMMemorySpace memSpace)
std::pair< mlir::Type, unsigned > inferMMAType(mlir::NVVM::MMATypes type, mlir::NVVM::MMAFrag frag, int nRow, int nCol, mlir::MLIRContext *context)
Return the element type and number of elements associated with a wmma matrix of given chracteristics.
Include the generated interface declarations.