#include "Conversion/GPUCommon/IndexIntrinsicsOpLowering.h"
Inheritance diagram for mlir::gpu::index_lowering::OpLowering< Op, XOp, YOp, ZOp >:
Public Member Functions | |
| OpLowering (const LLVMTypeConverter &typeConverter) | |
| OpLowering (const LLVMTypeConverter &typeConverter, IndexKind indexKind, IntrType intrType) | |
| LogicalResult | matchAndRewrite (Op op, typename Op::Adaptor adaptor, ConversionPatternRewriter &rewriter) const override |
| Public Member Functions inherited from mlir::ConvertOpToLLVMPattern< Op > | |
| ConvertOpToLLVMPattern (const LLVMTypeConverter &typeConverter, PatternBenefit benefit=1) | |
| void | rewrite (Operation *op, ArrayRef< Value > operands, ConversionPatternRewriter &rewriter) const final |
| Wrappers around the RewritePattern methods that pass the derived op type. More... | |
| void | rewrite (Operation *op, ArrayRef< ValueRange > operands, ConversionPatternRewriter &rewriter) const final |
| virtual void | rewrite (Op op, OpAdaptor adaptor, ConversionPatternRewriter &rewriter) const |
| virtual void | rewrite (Op op, OneToNOpAdaptor adaptor, ConversionPatternRewriter &rewriter) const |
| LogicalResult | match (Operation *op) const final |
| Attempt to match against code rooted at the specified operation, which is the same operation code as getRootKind(). More... | |
| virtual LogicalResult | match (Op op) const |
| Rewrite and Match methods that operate on the SourceOp type. More... | |
| LogicalResult | matchAndRewrite (Operation *op, ArrayRef< Value > operands, ConversionPatternRewriter &rewriter) const final |
| Hook for derived classes to implement combined matching and rewriting. More... | |
| LogicalResult | matchAndRewrite (Operation *op, ArrayRef< ValueRange > operands, ConversionPatternRewriter &rewriter) const final |
| Hook for derived classes to implement combined matching and rewriting. More... | |
| virtual LogicalResult | matchAndRewrite (Op op, OpAdaptor adaptor, ConversionPatternRewriter &rewriter) const |
| virtual LogicalResult | matchAndRewrite (Op op, OneToNOpAdaptor adaptor, ConversionPatternRewriter &rewriter) const |
| Public Member Functions inherited from mlir::ConvertToLLVMPattern | |
| ConvertToLLVMPattern (StringRef rootOpName, MLIRContext *context, const LLVMTypeConverter &typeConverter, PatternBenefit benefit=1) | |
| Public Member Functions inherited from mlir::ConversionPattern | |
| LogicalResult | matchAndRewrite (Operation *op, PatternRewriter &rewriter) const final |
| Attempt to match and rewrite the IR root at the specified operation. More... | |
| const TypeConverter * | getTypeConverter () const |
| Return the type converter held by this pattern, or nullptr if the pattern does not require type conversion. More... | |
| template<typename ConverterTy > | |
| std::enable_if_t< std::is_base_of< TypeConverter, ConverterTy >::value, const ConverterTy * > | getTypeConverter () const |
| Public Member Functions inherited from mlir::RewritePattern | |
| virtual | ~RewritePattern ()=default |
| virtual void | rewrite (Operation *op, PatternRewriter &rewriter) const |
| Rewrite the IR rooted at the specified operation with the result of this pattern, generating any new operations with the specified builder. More... | |
| Public Member Functions inherited from mlir::Pattern | |
| ArrayRef< OperationName > | getGeneratedOps () const |
| Return a list of operations that may be generated when rewriting an operation instance with this pattern. More... | |
| std::optional< OperationName > | getRootKind () const |
| Return the root node that this pattern matches. More... | |
| std::optional< TypeID > | getRootInterfaceID () const |
| Return the interface ID used to match the root operation of this pattern. More... | |
| std::optional< TypeID > | getRootTraitID () const |
| Return the trait ID used to match the root operation of this pattern. More... | |
| PatternBenefit | getBenefit () const |
| Return the benefit (the inverse of "cost") of matching this pattern. More... | |
| bool | hasBoundedRewriteRecursion () const |
| Returns true if this pattern is known to result in recursive application, i.e. More... | |
| MLIRContext * | getContext () const |
| Return the MLIRContext used to create this pattern. More... | |
| StringRef | getDebugName () const |
| Return a readable name for this pattern. More... | |
| void | setDebugName (StringRef name) |
| Set the human readable debug name used for this pattern. More... | |
| ArrayRef< StringRef > | getDebugLabels () const |
| Return the set of debug labels attached to this pattern. More... | |
| void | addDebugLabels (ArrayRef< StringRef > labels) |
| Add the provided debug labels to this pattern. More... | |
| void | addDebugLabels (StringRef label) |
Additional Inherited Members | |
| Public Types inherited from mlir::ConvertOpToLLVMPattern< Op > | |
| using | OpAdaptor = typename SourceOp::Adaptor |
| using | OneToNOpAdaptor = typename SourceOp::template GenericAdaptor< ArrayRef< ValueRange > > |
| Static Public Member Functions inherited from mlir::RewritePattern | |
| template<typename T , typename... Args> | |
| static std::unique_ptr< T > | create (Args &&...args) |
This method provides a convenient interface for creating and initializing derived rewrite patterns of the given type T. More... | |
| Protected Member Functions inherited from mlir::ConvertToLLVMPattern | |
| LLVM::LLVMDialect & | getDialect () const |
| Returns the LLVM dialect. More... | |
| const LLVMTypeConverter * | getTypeConverter () const |
| Type | getIndexType () const |
| Gets the MLIR type wrapping the LLVM integer type whose bit width is defined by the used type converter. More... | |
| Type | getIntPtrType (unsigned addressSpace=0) const |
| Gets the MLIR type wrapping the LLVM integer type whose bit width corresponds to that of a LLVM pointer type. More... | |
| Type | getVoidType () const |
| Gets the MLIR type wrapping the LLVM void type. More... | |
| Type | getVoidPtrType () const |
| Get the MLIR type wrapping the LLVM i8* type. More... | |
| Value | getStridedElementPtr (Location loc, MemRefType type, Value memRefDesc, ValueRange indices, ConversionPatternRewriter &rewriter) const |
| bool | isConvertibleAndHasIdentityMaps (MemRefType type) const |
| Returns if the given memref has identity maps and the element type is convertible to LLVM. More... | |
| Type | getElementPtrType (MemRefType type) const |
| Returns the type of a pointer to an element of the memref. More... | |
| void | getMemRefDescriptorSizes (Location loc, MemRefType memRefType, ValueRange dynamicSizes, ConversionPatternRewriter &rewriter, SmallVectorImpl< Value > &sizes, SmallVectorImpl< Value > &strides, Value &size, bool sizeInBytes=true) const |
Computes sizes, strides and buffer size of memRefType with identity layout. More... | |
| Value | getSizeInBytes (Location loc, Type type, ConversionPatternRewriter &rewriter) const |
| Computes the size of type in bytes. More... | |
| Value | getNumElements (Location loc, MemRefType memRefType, ValueRange dynamicSizes, ConversionPatternRewriter &rewriter) const |
| Computes total number of elements for the given MemRef and dynamicSizes. More... | |
| MemRefDescriptor | createMemRefDescriptor (Location loc, MemRefType memRefType, Value allocatedPtr, Value alignedPtr, ArrayRef< Value > sizes, ArrayRef< Value > strides, ConversionPatternRewriter &rewriter) const |
| Creates and populates a canonical memref descriptor struct. More... | |
| LogicalResult | copyUnrankedDescriptors (OpBuilder &builder, Location loc, TypeRange origTypes, SmallVectorImpl< Value > &operands, bool toDynamic) const |
| Copies the memory descriptor for any operands that were unranked descriptors originally to heap-allocated memory (if toDynamic is true) or to stack-allocated memory (otherwise). More... | |
| Protected Member Functions inherited from mlir::ConversionPattern | |
| template<typename... Args> | |
| ConversionPattern (const TypeConverter &typeConverter, Args &&...args) | |
| Construct a conversion pattern with the given converter, and forward the remaining arguments to RewritePattern. More... | |
| SmallVector< Value > | getOneToOneAdaptorOperands (ArrayRef< ValueRange > operands) const |
| Given an array of value ranges, which are the inputs to a 1:N adaptor, try to extract the single value of each range to construct a the inputs for a 1:1 adaptor. More... | |
| Protected Member Functions inherited from mlir::RewritePattern | |
| Pattern (StringRef rootName, PatternBenefit benefit, MLIRContext *context, ArrayRef< StringRef > generatedNames={}) | |
Inherit the base constructors from Pattern. More... | |
| Pattern (MatchAnyOpTypeTag tag, PatternBenefit benefit, MLIRContext *context, ArrayRef< StringRef > generatedNames={}) | |
Inherit the base constructors from Pattern. More... | |
| Pattern (MatchInterfaceOpTypeTag tag, TypeID interfaceID, PatternBenefit benefit, MLIRContext *context, ArrayRef< StringRef > generatedNames={}) | |
Inherit the base constructors from Pattern. More... | |
| Pattern (MatchTraitOpTypeTag tag, TypeID traitID, PatternBenefit benefit, MLIRContext *context, ArrayRef< StringRef > generatedNames={}) | |
Inherit the base constructors from Pattern. More... | |
| Protected Member Functions inherited from mlir::Pattern | |
| Pattern (StringRef rootName, PatternBenefit benefit, MLIRContext *context, ArrayRef< StringRef > generatedNames={}) | |
| Construct a pattern with a certain benefit that matches the operation with the given root name. More... | |
| Pattern (MatchAnyOpTypeTag tag, PatternBenefit benefit, MLIRContext *context, ArrayRef< StringRef > generatedNames={}) | |
| Construct a pattern that may match any operation type. More... | |
| Pattern (MatchInterfaceOpTypeTag tag, TypeID interfaceID, PatternBenefit benefit, MLIRContext *context, ArrayRef< StringRef > generatedNames={}) | |
Construct a pattern that may match any operation that implements the interface defined by the provided interfaceID. More... | |
| Pattern (MatchTraitOpTypeTag tag, TypeID traitID, PatternBenefit benefit, MLIRContext *context, ArrayRef< StringRef > generatedNames={}) | |
Construct a pattern that may match any operation that implements the trait defined by the provided traitID. More... | |
| void | setHasBoundedRewriteRecursion (bool hasBoundedRecursionArg=true) |
| Set the flag detailing if this pattern has bounded rewrite recursion or not. More... | |
| Static Protected Member Functions inherited from mlir::ConvertToLLVMPattern | |
| static Value | createIndexAttrConstant (OpBuilder &builder, Location loc, Type resultType, int64_t value) |
Create a constant Op producing a value of resultType from an index-typed integer attribute. More... | |
| Protected Attributes inherited from mlir::ConversionPattern | |
| const TypeConverter * | typeConverter = nullptr |
| An optional type converter for use by this pattern. More... | |
Detailed Description
template<typename Op, typename XOp, typename YOp, typename ZOp>
struct mlir::gpu::index_lowering::OpLowering< Op, XOp, YOp, ZOp >
Definition at line 32 of file IndexIntrinsicsOpLowering.h.
Constructor & Destructor Documentation
◆ OpLowering() [1/2]
template<typename Op , typename XOp , typename YOp , typename ZOp >
|
inlineexplicit |
Definition at line 39 of file IndexIntrinsicsOpLowering.h.
References None, and mlir::gpu::index_lowering::Other.
◆ OpLowering() [2/2]
template<typename Op , typename XOp , typename YOp , typename ZOp >
|
inlineexplicit |
Definition at line 44 of file IndexIntrinsicsOpLowering.h.
Member Function Documentation
◆ matchAndRewrite()
template<typename Op , typename XOp , typename YOp , typename ZOp >
|
inlineoverride |
Definition at line 52 of file IndexIntrinsicsOpLowering.h.
References mlir::detail::DenseArrayAttrImpl< T >::asArrayRef(), mlir::gpu::index_lowering::Block, mlir::OpBuilder::create(), mlir::gpu::index_lowering::Dim, mlir::get(), mlir::OpState::getContext(), mlir::Builder::getContext(), mlir::OpState::getLoc(), mlir::Operation::getResult(), mlir::Operation::getResults(), mlir::gpu::index_lowering::Grid, mlir::gpu::index_lowering::Id, max(), min(), mlir::gpu::index_lowering::None, mlir::gpu::index_lowering::Other, mlir::ConversionPatternRewriter::replaceOp(), and mlir::Operation::setAttr().
The documentation for this struct was generated from the following file:
- lib/Conversion/GPUCommon/IndexIntrinsicsOpLowering.h
