NVVMDialect.cpp File Reference

#include "mlir/Dialect/LLVMIR/NVVMDialect.h"
#include "mlir/Conversion/ConvertToLLVM/ToLLVMInterface.h"
#include "mlir/Dialect/GPU/IR/CompilationInterfaces.h"
#include "mlir/Dialect/Utils/StaticValueUtils.h"
#include "mlir/IR/Builders.h"
#include "mlir/IR/BuiltinAttributes.h"
#include "mlir/IR/BuiltinTypes.h"
#include "mlir/IR/Diagnostics.h"
#include "mlir/IR/DialectImplementation.h"
#include "mlir/IR/MLIRContext.h"
#include "mlir/IR/Operation.h"
#include "mlir/IR/OperationSupport.h"
#include "mlir/IR/Types.h"
#include "mlir/Support/LogicalResult.h"
#include "llvm/ADT/STLExtras.h"
#include "llvm/ADT/TypeSwitch.h"
#include "llvm/AsmParser/Parser.h"
#include "llvm/IR/Attributes.h"
#include "llvm/IR/Function.h"
#include "llvm/IR/Type.h"
#include "llvm/Support/Casting.h"
#include "llvm/Support/SourceMgr.h"
#include "llvm/Support/raw_ostream.h"
#include <cassert>
#include <optional>
#include <string>
#include "mlir/Dialect/LLVMIR/NVVMOpsDialect.cpp.inc"
#include "mlir/Dialect/LLVMIR/NVVMOpsEnums.cpp.inc"
#include "mlir/Dialect/LLVMIR/NVVMOps.cpp.inc"
#include "mlir/Dialect/LLVMIR/NVVMOpsAttributes.cpp.inc"

Go to the source code of this file.

Macros
#define	GET_OP_LIST
#define	GET_ATTRDEF_LIST
#define	GET_OP_CLASSES
#define	GET_ATTRDEF_CLASSES

Functions
static void	printNVVMIntrinsicOp (OpAsmPrinter &p, Operation *op)
static bool	isInt4PtxType (MMATypes type)
static bool	isInt8PtxType (MMATypes type)
static bool	isIntegerPtxType (MMATypes type)
static std::pair< mlir::Type, unsigned >	inferMMATypeFromMNK (NVVM::MMATypes type, NVVM::MMAFrag frag, int m, int n, int k, MLIRContext *context)
FailureOr< int >	getAllowedSizeK (NVVM::WGMMATypes typeA)
LogicalResult	isAllowedWGMMADataType (NVVM::WGMMATypes typeD, NVVM::WGMMATypes typeA, NVVM::WGMMATypes typeB)
LogicalResult	isAllowedSizeN (int sizeN, NVVM::WGMMATypes typeA)

Macro Definition Documentation

GET_ATTRDEF_CLASSES

#define GET_ATTRDEF_CLASSES

Definition at line 1142 of file NVVMDialect.cpp.

GET_ATTRDEF_LIST

#define GET_ATTRDEF_LIST

GET_OP_CLASSES

#define GET_OP_CLASSES

Definition at line 1139 of file NVVMDialect.cpp.

GET_OP_LIST

#define GET_OP_LIST

Function Documentation

getAllowedSizeK()

FailureOr<int> getAllowedSizeK

(

NVVM::WGMMATypes

typeA

)

Definition at line 743 of file NVVMDialect.cpp.

References mlir::failure().

inferMMATypeFromMNK()

static std::pair<mlir::Type, unsigned> inferMMATypeFromMNK ( NVVM::MMATypes  type, NVVM::MMAFrag  frag, int  m, int  n, int  k, MLIRContext *  context  )

static

Definition at line 617 of file NVVMDialect.cpp.

References mlir::NVVM::inferMMAType().

isAllowedSizeN()

LogicalResult isAllowedSizeN	(	int	sizeN,
		NVVM::WGMMATypes	typeA
	)

Definition at line 799 of file NVVMDialect.cpp.

References mlir::failure(), and mlir::success().

isAllowedWGMMADataType()

LogicalResult isAllowedWGMMADataType	(	NVVM::WGMMATypes	typeD,
		NVVM::WGMMATypes	typeA,
		NVVM::WGMMATypes	typeB
	)

Definition at line 757 of file NVVMDialect.cpp.

References mlir::failure(), and mlir::success().

isInt4PtxType()

static bool isInt4PtxType ( MMATypes  type )

static

Definition at line 142 of file NVVMDialect.cpp.

Referenced by isIntegerPtxType().

isInt8PtxType()

static bool isInt8PtxType ( MMATypes  type )

static

Definition at line 146 of file NVVMDialect.cpp.

Referenced by isIntegerPtxType().

isIntegerPtxType()

static bool isIntegerPtxType ( MMATypes  type )

static

Definition at line 150 of file NVVMDialect.cpp.

References isInt4PtxType(), and isInt8PtxType().

printNVVMIntrinsicOp()

static void printNVVMIntrinsicOp ( OpAsmPrinter &  p, Operation *  op  )

static

Definition at line 55 of file NVVMDialect.cpp.