MLIR
17.0.0git
|
#include "mlir/Dialect/GPU/IR/GPUDialect.h"
#include "mlir/Dialect/Arith/IR/Arith.h"
#include "mlir/Dialect/MemRef/IR/MemRef.h"
#include "mlir/IR/Attributes.h"
#include "mlir/IR/Builders.h"
#include "mlir/IR/BuiltinAttributes.h"
#include "mlir/IR/BuiltinOps.h"
#include "mlir/IR/BuiltinTypes.h"
#include "mlir/IR/DialectImplementation.h"
#include "mlir/IR/FunctionImplementation.h"
#include "mlir/IR/Matchers.h"
#include "mlir/IR/OpImplementation.h"
#include "mlir/IR/PatternMatch.h"
#include "mlir/IR/TypeUtilities.h"
#include "mlir/Interfaces/SideEffectInterfaces.h"
#include "mlir/Transforms/InliningUtils.h"
#include "llvm/ADT/TypeSwitch.h"
#include "mlir/Dialect/GPU/IR/GPUOpsDialect.cpp.inc"
#include "mlir/Dialect/GPU/IR/GPUOps.cpp.inc"
#include "mlir/Dialect/GPU/IR/GPUOpsAttributes.cpp.inc"
#include "mlir/Dialect/GPU/IR/GPUOpInterfaces.cpp.inc"
#include "mlir/Dialect/GPU/IR/GPUOpsEnums.cpp.inc"
Go to the source code of this file.
Classes | |
struct | FoldLaunchArguments |
Simplify the gpu.launch when the range of a thread or block ID is trivially known to be one. More... | |
Macros | |
#define | GET_OP_LIST |
#define | GET_ATTRDEF_LIST |
#define | GET_ATTRDEF_CLASSES |
#define | GET_OP_CLASSES |
Enumerations | |
enum | GPUMemorySpace { kGenericMemorySpace = 0 , kGlobalMemorySpace = 1 , kSharedMemorySpace = 3 } |
GPU memory space identifiers. More... | |
Functions | |
static std::string | getSparseHandleKeyword (SparseHandleKind kind) |
static ParseResult | parseAsyncDependencies (OpAsmParser &parser, Type &asyncTokenType, SmallVectorImpl< OpAsmParser::UnresolvedOperand > &asyncDependencies) |
Parses an optional list of async operands with an optional leading keyword. More... | |
static void | printAsyncDependencies (OpAsmPrinter &printer, Operation *op, Type asyncTokenType, OperandRange asyncDependencies) |
Prints optional async dependencies with its leading keyword. More... | |
static ParseResult | parseAttributions (OpAsmParser &parser, StringRef keyword, SmallVectorImpl< OpAsmParser::Argument > &args) |
Parses a GPU function memory attribution. More... | |
static void | printAttributions (OpAsmPrinter &p, StringRef keyword, ArrayRef< BlockArgument > values) |
Prints a GPU function memory attribution. More... | |
static LogicalResult | verifyAttributions (Operation *op, ArrayRef< BlockArgument > attributions, gpu::AddressSpace memorySpace) |
Verifies a GPU function memory attribution. More... | |
static bool | verifyReduceOpAndType (gpu::AllReduceOperation opName, Type resType) |
static bool | canMakeGroupOpUniform (Operation *op) |
static ParseResult | parseAllReduceOperation (AsmParser &parser, AllReduceOperationAttr &attr) |
static void | printAllReduceOperation (AsmPrinter &printer, Operation *op, AllReduceOperationAttr attr) |
static void | printSizeAssignment (OpAsmPrinter &p, KernelDim3 size, KernelDim3 operands, KernelDim3 ids) |
static ParseResult | parseSizeAssignment (OpAsmParser &parser, MutableArrayRef< OpAsmParser::UnresolvedOperand > sizes, MutableArrayRef< OpAsmParser::UnresolvedOperand > regionSizes, MutableArrayRef< OpAsmParser::UnresolvedOperand > indices) |
static ParseResult | parseLaunchFuncOperands (OpAsmParser &parser, SmallVectorImpl< OpAsmParser::UnresolvedOperand > &argNames, SmallVectorImpl< Type > &argTypes) |
static void | printLaunchFuncOperands (OpAsmPrinter &printer, Operation *, OperandRange operands, TypeRange types) |
static ParseResult | parseAttributions (OpAsmParser &parser, StringRef keyword, SmallVectorImpl< OpAsmParser::Argument > &args, Attribute &attributionAttrs) |
Parses a GPU function memory attribution. More... | |
static void | printAttributions (OpAsmPrinter &p, StringRef keyword, ArrayRef< BlockArgument > values, ArrayAttr attributes) |
static DictionaryAttr | getAttributionAttrs (GPUFuncOp op, unsigned index, StringAttr attrName) |
static void | setAttributionAttrs (GPUFuncOp op, unsigned index, DictionaryAttr value, StringAttr attrName) |
static Attribute | getAttributionAttr (GPUFuncOp op, unsigned index, StringAttr name, StringAttr attrsName) |
static void | setAttributionAttr (GPUFuncOp op, unsigned index, StringAttr name, Attribute value, StringAttr attrsName) |
static LogicalResult | verifyKnownLaunchSizeAttr (gpu::GPUFuncOp op, StringRef attrName) |
static bool | isLastMemrefDimUnitStride (MemRefType type) |
Return true if the last dimension of the MemRefType has unit stride. More... | |
#define GET_ATTRDEF_CLASSES |
Definition at line 1784 of file GPUDialect.cpp.
#define GET_ATTRDEF_LIST |
#define GET_OP_CLASSES |
Definition at line 1787 of file GPUDialect.cpp.
#define GET_OP_LIST |
enum GPUMemorySpace |
GPU memory space identifiers.
Enumerator | |
---|---|
kGenericMemorySpace | Generic memory space identifier. |
kGlobalMemorySpace | Global memory space identifier. |
kSharedMemorySpace | Shared memory space identifier. |
Definition at line 117 of file GPUDialect.cpp.
|
static |
Definition at line 471 of file GPUDialect.cpp.
|
static |
Definition at line 1302 of file GPUDialect.cpp.
|
static |
Definition at line 1259 of file GPUDialect.cpp.
|
static |
Definition at line 164 of file GPUDialect.cpp.
|
static |
Return true if the last dimension of the MemRefType has unit stride.
Also return true for memrefs with no strides.
Definition at line 1564 of file GPUDialect.cpp.
References mlir::failed(), and mlir::getStridesAndOffset().
|
static |
Definition at line 493 of file GPUDialect.cpp.
References mlir::AsmParser::parseOptionalKeyword().
|
static |
Parses an optional list of async operands with an optional leading keyword.
(async
)? ([
ssa-id-list ]
)?
This method is used by the tablegen assembly format for async ops as well.
Definition at line 342 of file GPUDialect.cpp.
References mlir::AsmParser::emitError(), mlir::AsmParser::getBuilder(), mlir::AsmParser::getCurrentLocation(), mlir::OpAsmParser::getNumResults(), mlir::Builder::getType(), mlir::AsmParser::OptionalSquare, mlir::OpAsmParser::parseOperandList(), mlir::AsmParser::parseOptionalKeyword(), and mlir::succeeded().
|
static |
Parses a GPU function memory attribution.
memory-attribution ::= (workgroup
(
ssa-id-and-type-list )
)? (private
(
ssa-id-and-type-list )
)?
Note that this function parses only one of the two similar parts, with the keyword provided as argument.
Definition at line 381 of file GPUDialect.cpp.
References mlir::failed(), mlir::AsmParser::Paren, mlir::OpAsmParser::parseArgumentList(), mlir::AsmParser::parseOptionalKeyword(), and mlir::success().
|
static |
Parses a GPU function memory attribution.
memory-attribution ::= (workgroup
(
ssa-id-and-type-list )
)? (private
(
ssa-id-and-type-list )
)?
Note that this function parses only one of the two similar parts, with the keyword provided as argument.
Definition at line 1095 of file GPUDialect.cpp.
References mlir::OpAsmParser::Argument::attrs, mlir::failed(), mlir::Builder::getArrayAttr(), mlir::AsmParser::getBuilder(), mlir::Builder::getDictionaryAttr(), mlir::AsmParser::Paren, mlir::OpAsmParser::parseArgumentList(), mlir::AsmParser::parseOptionalKeyword(), and mlir::success().
|
static |
Definition at line 994 of file GPUDialect.cpp.
References mlir::failure(), mlir::AsmParser::Paren, mlir::AsmParser::parseColonType(), mlir::AsmParser::parseCommaSeparatedList(), mlir::OpAsmParser::parseOperand(), mlir::AsmParser::parseOptionalKeyword(), and mlir::success().
|
static |
Definition at line 725 of file GPUDialect.cpp.
References mlir::failure(), mlir::AsmParser::Paren, mlir::AsmParser::parseComma(), mlir::AsmParser::parseEqual(), mlir::AsmParser::parseKeyword(), mlir::AsmParser::parseLParen(), mlir::OpAsmParser::parseOperand(), mlir::OpAsmParser::parseOperandList(), and mlir::AsmParser::parseRParen().
|
static |
Definition at line 506 of file GPUDialect.cpp.
|
static |
Prints optional async dependencies with its leading keyword.
(async
)? ([
ssa-id-list ]
)?
Definition at line 358 of file GPUDialect.cpp.
|
static |
Prints a GPU function memory attribution.
Definition at line 392 of file GPUDialect.cpp.
References mlir::Value::getType().
|
static |
Definition at line 1211 of file GPUDialect.cpp.
References mlir::detail::enumerate(), mlir::Value::getType(), and mlir::OpAsmPrinter::printOptionalAttrDict().
|
static |
Definition at line 1010 of file GPUDialect.cpp.
References mlir::OpAsmPrinter::printOperand(), and mlir::AsmPrinter::printType().
|
static |
Definition at line 682 of file GPUDialect.cpp.
References mlir::gpu::KernelDim3::x, mlir::gpu::KernelDim3::y, and mlir::gpu::KernelDim3::z.
|
static |
Definition at line 1326 of file GPUDialect.cpp.
|
static |
Definition at line 1275 of file GPUDialect.cpp.
|
static |
Verifies a GPU function memory attribution.
Definition at line 404 of file GPUDialect.cpp.
|
static |
Definition at line 1414 of file GPUDialect.cpp.
|
static |
Definition at line 430 of file GPUDialect.cpp.