MLIR
18.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/Matchers.h"
#include "mlir/IR/OpImplementation.h"
#include "mlir/IR/PatternMatch.h"
#include "mlir/IR/SymbolTable.h"
#include "mlir/IR/TypeUtilities.h"
#include "mlir/Interfaces/FunctionImplementation.h"
#include "mlir/Interfaces/SideEffectInterfaces.h"
#include "mlir/Support/LogicalResult.h"
#include "mlir/Transforms/InliningUtils.h"
#include "llvm/ADT/STLExtras.h"
#include "llvm/ADT/TypeSwitch.h"
#include "llvm/Support/CommandLine.h"
#include "llvm/Support/ErrorHandling.h"
#include "llvm/Support/StringSaver.h"
#include <cassert>
#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"
#include "mlir/Dialect/GPU/IR/CompilationAttrInterfaces.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 |
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 LogicalResult | 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 | parseLaunchDimType (OpAsmParser &parser, Type &dimTy, std::optional< OpAsmParser::UnresolvedOperand > clusterValue, Type &clusterXTy, Type &clusterYTy, Type &clusterZTy) |
static void | printLaunchDimType (OpAsmPrinter &printer, Operation *op, Type dimTy, Value clusterValue, Type clusterXTy, Type clusterYTy, Type clusterZTy) |
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 ParseResult | parseOffloadingHandler (OpAsmParser &parser, Attribute &offloadingHandler) |
static void | printOffloadingHandler (OpAsmPrinter &printer, Operation *op, Attribute offloadingHandler) |
#define GET_ATTRDEF_CLASSES |
Definition at line 2194 of file GPUDialect.cpp.
#define GET_ATTRDEF_LIST |
#define GET_OP_CLASSES |
Definition at line 2197 of file GPUDialect.cpp.
#define GET_OP_LIST |
|
static |
Definition at line 544 of file GPUDialect.cpp.
|
static |
Definition at line 1484 of file GPUDialect.cpp.
|
static |
Definition at line 1441 of file GPUDialect.cpp.
|
static |
Definition at line 219 of file GPUDialect.cpp.
|
static |
Definition at line 566 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 403 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 442 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 1277 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 1131 of file GPUDialect.cpp.
References mlir::failure(), mlir::get(), mlir::AsmParser::getContext(), mlir::AsmParser::parseOptionalColon(), mlir::AsmParser::parseType(), mlir::succeeded(), and mlir::success().
|
static |
Definition at line 1153 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 1745 of file GPUDialect.cpp.
References mlir::failure(), mlir::Builder::getAttr(), mlir::AsmParser::getBuilder(), mlir::AsmParser::parseAttribute(), mlir::AsmParser::parseGreater(), mlir::AsmParser::parseOptionalLess(), mlir::succeeded(), and mlir::success().
|
static |
Definition at line 799 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 579 of file GPUDialect.cpp.
|
static |
Prints optional async dependencies with its leading keyword.
(async
)? ([
ssa-id-list ]
)?
Definition at line 419 of file GPUDialect.cpp.
|
static |
Prints a GPU function memory attribution.
Definition at line 453 of file GPUDialect.cpp.
References mlir::Value::getType().
|
static |
Definition at line 1393 of file GPUDialect.cpp.
References mlir::detail::enumerate(), mlir::Value::getType(), and mlir::OpAsmPrinter::printOptionalAttrDict().
|
static |
Definition at line 1146 of file GPUDialect.cpp.
|
static |
Definition at line 1169 of file GPUDialect.cpp.
References mlir::OpAsmPrinter::printOperand(), and mlir::AsmPrinter::printType().
|
static |
Definition at line 1758 of file GPUDialect.cpp.
|
static |
Definition at line 756 of file GPUDialect.cpp.
References mlir::gpu::KernelDim3::x, mlir::gpu::KernelDim3::y, and mlir::gpu::KernelDim3::z.
|
static |
Definition at line 1508 of file GPUDialect.cpp.
|
static |
Definition at line 1457 of file GPUDialect.cpp.
|
static |
Verifies a GPU function memory attribution.
Definition at line 465 of file GPUDialect.cpp.
|
static |
Definition at line 1596 of file GPUDialect.cpp.
|
static |
Definition at line 491 of file GPUDialect.cpp.
References mlir::failure(), and mlir::success().