Namespaces
	detail
	Check if a Callable type can be invoked with the given set of arg types.

	edsc

	functional

	fxpmath

	gpu

	impl

	linalg

	LLVM

	loop

	matcher

	matchers

	NVVM

	ops_assertions

	OpTrait

	quant

	quantizer

	ROCDL

	spirv

	StandardAttributes

	StandardTypes

	tblgen

	vector

Classes
class	AbstractOperation

struct	AffineApplyNormalizer

class	AffineApplyOp

class	AffineBinaryOpExpr

class	AffineBound

class	AffineConstantExpr
	An integer constant appearing in affine expression. More...

struct	AffineCopyOptions
	Explicit copy / DMA generation options for mlir::affineDataCopyGenerate. More...

class	AffineDimExpr
	A dimensional identifier appearing in an affine expression. More...

class	AffineDmaStartOp

class	AffineDmaWaitOp

class	AffineExpr

class	AffineExprVisitor

class	AffineLoadOp

class	AffineMap

class	AffineMapAttr

class	AffineOpsDialect

class	AffineStoreOp

class	AffineSymbolExpr
	A symbolic identifier appearing in an affine expression. More...

class	AffineValueMap

class	AnalysisManager

class	ArrayAttr

class	Attribute

class	AttributeStorage

class	BaseMemRefType
	Base MemRef for Ranked and Unranked variants. More...

class	Block
	`Block` represents an ordered list of `Operation`s. More...

class	BlockAndValueMapping

class	BlockArgument
	Block arguments are values. More...

class	BlockOperand
	Terminator operations can have Block operands to represent successors. More...

class	BoolAttr

class	Builder

class	CallGraph

class	CallGraphNode

struct	CallInterfaceCallable

class	CallSiteLoc

struct	ClassID

class	ComplexType

struct	ComputationSliceState

class	ConstantFloatOp

class	ConstantIndexOp

class	ConstantIntOp

class	ConversionPattern

class	ConversionPatternRewriter

class	ConversionTarget
	This class describes a specific conversion target. More...

class	DenseElementsAttr

class	DenseFPElementsAttr

class	DenseIntElementsAttr

struct	DependenceComponent

struct	DependenceResult

class	Diagnostic

class	DiagnosticArgument
	A variant type that holds a single argument for a diagnostic. More...

class	DiagnosticEngine

class	Dialect

class	DialectAsmParser

class	DialectAsmPrinter

class	DialectHooks

struct	DialectHooksRegistration

class	DialectInlinerInterface

class	DialectInterface
	This class represents an interface overridden for a single dialect. More...

class	DialectInterfaceCollection

struct	DialectRegistration

class	DictionaryAttr

class	DmaStartOp

class	DmaWaitOp

class	DominanceInfo
	A class for computing basic dominance information. More...

class	ElementsAttr

struct	EmptyPipelineOptions

class	ExecutionEngine

class	FileLineColLoc

class	FilteredValueUseIterator

class	FlatAffineConstraints

class	FlatSymbolRefAttr

class	FloatAttr

class	FloatType

class	FoldingHook

class	FoldingHook< ConcreteType, isSingleResult, typename std::enable_if< isSingleResult >::type >

class	FuncOp

struct	FunctionPass

struct	FunctionTraits

struct	FunctionTraits< ReturnType(*)(Args...), false >
	Overload for non-class function types. More...

struct	FunctionTraits< ReturnType(ClassType::*)(Args...) const, false >
	Overload for class function types. More...

class	FunctionType
	Function types map from a list of inputs to a list of results. More...

class	FusedLoc

struct	FusionResult

class	GenInfo

struct	GenNameParser
	Adds command line option for each registered generator. More...

struct	GenRegistration

struct	GPUIndexIntrinsicOpLowering

class	Identifier

class	indexed_accessor_iterator

class	indexed_accessor_range

class	IndexType

class	InFlightDiagnostic

class	InlinerInterface

class	IntegerAttr

class	IntegerSet

class	IntegerSetAttr

class	IntegerType
	Integer types can have arbitrary bitwidth up to a large fixed limit. More...

class	IntegerValueSet
	An IntegerValueSet is an integer set plus its operands. More...

class	IntInfty

class	IRMultiObjectWithUseList

class	IRObjectWithUseList
	This class represents a single IR object that contains a use list. More...

class	IROperand

class	Lexer
	This class breaks up the current file into a token stream. More...

class	LinalgTypeConverter

class	Liveness

class	LivenessBlockInfo
	This class represents liveness information on block level. More...

class	LLVMOpLowering

class	LLVMTypeConverter
	Conversion from types in the Standard dialect to the LLVM IR dialect. More...

class	Location

class	LocationAttr

struct	LogicalResult

struct	LoopNestStats

struct	MemRefAccess
	Encapsulates a memref load or store access information. More...

class	MemRefDescriptor

struct	MemRefRegion

class	MemRefType

class	MLIRContext

class	MLIRContextImpl

class	ModuleAnalysisManager

class	ModuleOp

struct	ModulePass

class	ModuleTerminatorOp

struct	MutableAffineMap
	A mutable affine map. Its affine expressions are however unique. More...

struct	MutableIntegerSet
	A mutable integer set. Its affine expressions are however unique. More...

class	NamedAttributeList

class	NameLoc
	Represents an identity name attached to a child location. More...

class	NestedMatch

class	NestedPattern

class	NestedPatternContext

class	NoneType

class	Op

class	OpaqueAttr

class	OpaqueElementsAttr

class	OpaqueLoc

class	OpaqueType

class	OpAsmDialectInterface

class	OpAsmParser

class	OpAsmPrinter

class	OpBuilder

struct	OpConversionPattern

class	OperandElementTypeIterator

class	OperandRange
	This class implements the operand iterators for the Operation class. More...

class	Operation

class	OperationFolder

class	OperationName

class	OperationPass

struct	OperationPass< PassT, void >

struct	OperationState

class	OpFolderDialectInterface

class	OpFoldResult
	This class represents a single result from folding an operation. More...

class	OpInterface

class	OpOperand
	A reference to a value, suitable for use as an operand of an operation. More...

class	OpPassBase

class	OpPassManager

class	OpPrintingFlags

class	OpResult
	This is a value defined by a result of an operation. More...

struct	OpRewritePattern

class	OpState

class	OptionalParseResult

struct	OpToFuncCallLowering

class	OwningModuleRef

class	OwningRewritePatternList

class	ParallelDiagnosticHandler

class	ParseResult

class	Pass

class	PassInfo
	A structure to represent the information for a derived pass class. More...

class	PassInstrumentation

class	PassInstrumentor

class	PassManager
	The main pass manager and pipeline builder. More...

class	PassPipelineCLParser

class	PassPipelineInfo
	A structure to represent the information of a registered pass pipeline. More...

class	PassPipelineOptions

struct	PassPipelineRegistration

struct	PassPipelineRegistration< EmptyPipelineOptions >

struct	PassRegistration

class	PassRegistryEntry

class	Pattern

class	PatternBenefit

class	PatternRewriter

class	PatternState

class	PostDominanceInfo
	A class for computing basic postdominance information. More...

class	PredecessorIterator

class	RankedTensorType

class	Region

class	RegionRange

class	ResultElementTypeIterator

class	ResultRange
	This class implements the result iterators for the Operation class. More...

class	RewritePattern

class	RewritePatternMatcher

class	ScopedDiagnosticHandler

class	SDBM

class	SDBMConstantExpr
	SDBM constant expression, wraps a 64-bit integer. More...

class	SDBMDialect

class	SDBMDiffExpr

class	SDBMDimExpr

class	SDBMDirectExpr

class	SDBMExpr

class	SDBMInputExpr

class	SDBMNegExpr

class	SDBMStripeExpr

class	SDBMSumExpr
	SDBM sum expression. LHS is a term expression and RHS is a constant. More...

class	SDBMSymbolExpr

class	SDBMTermExpr

class	SDBMVaryingExpr

class	SDBMVisitor

class	ShapedType

class	SideEffectsDialectInterface

class	SideEffectsInterface

class	SimpleAffineExprFlattener

class	SimpleObjectCache
	A simple object cache following Lang's LLJITWithObjectCache example. More...

class	SourceMgrDiagnosticHandler
	This class is a utility diagnostic handler for use with llvm::SourceMgr. More...

class	SourceMgrDiagnosticVerifierHandler

class	SparseElementsAttr

class	SPIRVOpLowering
	Base class to define a conversion pattern to lower `SourceOp` into SPIR-V. More...

class	SPIRVTypeConverter

class	SplatElementsAttr

class	StandardOpsDialect

class	StorageUniquer

class	StringAttr

class	StructBuilder

class	SuccessorRange
	This class implements the successor iterators for Block. More...

class	SymbolRefAttr

class	SymbolTable

class	TensorType

class	Token
	This represents a token in the MLIR syntax. More...

struct	TranslateFromMLIRRegistration

struct	TranslateRegistration

struct	TranslateToMLIRRegistration

struct	TranslationParser

class	TupleType

class	Type

class	TypeAttr

class	TypeConverter

class	TypeStorage
	Base storage class appearing in a Type. More...

class	TypeSwitch

class	TypeSwitch< T, void >
	Specialization of TypeSwitch for void returning callables. More...

class	UnitAttr

class	UnknownLoc

class	UnrankedMemRefDescriptor

class	UnrankedMemRefType

class	UnrankedTensorType

class	Value

class	ValueRange

class	ValueTypeIterator
	This class implements iteration on the types of a given range of values. More...

class	ValueUseIterator
	An iterator over all of the uses of an IR object. More...

class	ValueUserIterator
	An iterator over all users of a ValueBase. More...

class	VectorType

class	VulkanLayoutUtils

class	WalkResult

Typedefs
using	DominanceInfoNode = llvm::DomTreeNodeBase< Block >

using	VectorizableLoopFun = std::function< bool(AffineForOp)>

using	FilterFunctionType = std::function< bool(Operation &)>

using	TransitiveFilter = std::function< bool(Operation *)>

using	OwnedCubin = std::unique_ptr< std::vector< char > >

using	CubinGenerator = std::function< OwnedCubin(const std::string &, Location, StringRef)>

using	LLVMPatternListFiller = std::function< void(LLVMTypeConverter &, OwningRewritePatternList &)>

using	LLVMTypeConverterMaker = std::function< std::unique_ptr< LLVMTypeConverter >(MLIRContext *)>

using	NamedAttribute = std::pair< Identifier, Attribute >

using	DefaultAttributeStorage = AttributeStorage

using	AttributeStorageAllocator = StorageUniquer::StorageAllocator

using	DialectConstantDecodeHook = std::function< bool(const OpaqueElementsAttr, ElementsAttr &)>

using	DialectConstantFoldHook = std::function< LogicalResult(Operation *, ArrayRef< Attribute >, SmallVectorImpl< Attribute > &)>

using	DialectExtractElementHook = std::function< Attribute(const OpaqueElementsAttr, ArrayRef< uint64_t >)>

using	DialectAllocatorFunction = std::function< void(MLIRContext *)>

using	DialectHooksSetter = std::function< void(MLIRContext *)>

template<typename OpTy >
using	OperandAdaptor = typename OpTy::OperandAdaptor

using	OpAsmSetValueNameFn = function_ref< void(Value, StringRef)>

using	PatternMatchResult = Optional< std::unique_ptr< PatternState > >

using	DefaultTypeStorage = TypeStorage

using	TypeStorageAllocator = StorageUniquer::StorageAllocator

using	OperandElementTypeRange = iterator_range< OperandElementTypeIterator >

using	ResultElementTypeRange = iterator_range< ResultElementTypeIterator >

using	AnalysisID = ClassID

using	PassRegistryFunction = std::function< LogicalResult(OpPassManager &, StringRef options)>

using	PassAllocatorFunction = std::function< std::unique_ptr< Pass >()>

using	PassID = ClassID

template<typename KeyT , typename ValueT , typename KeyInfoT = DenseMapInfo<KeyT>, typename BucketT = llvm::detail::DenseMapPair<KeyT, ValueT>>
using	DenseMap = llvm::DenseMap< KeyT, ValueT, KeyInfoT, BucketT >

template<typename ValueT , typename ValueInfoT = DenseMapInfo<ValueT>>
using	DenseSet = llvm::DenseSet< ValueT, ValueInfoT >

template<typename Fn >
using	function_ref = llvm::function_ref< Fn >

template<template< class... > class Op, class... Args>
using	is_detected = typename detail::detector< void, Op, Args... >::value_t

template<typename Callable , typename... Args>
using	is_invocable = is_detected< detail::is_invocable, Callable, Args... >

using	ChunkBufferHandler = function_ref< LogicalResult(std::unique_ptr< llvm::MemoryBuffer > chunkBuffer, raw_ostream &os)>

using	GenFunction = std::function< bool(const llvm::RecordKeeper &recordKeeper, raw_ostream &os)>
	Generator function to invoke. More...

using	Loops = SmallVector< loop::ForOp, 8 >

using	TileLoops = std::pair< Loops, Loops >

using	TranslateSourceMgrToMLIRFunction = std::function< OwningModuleRef(llvm::SourceMgr &sourceMgr, MLIRContext *)>

using	TranslateStringRefToMLIRFunction = std::function< OwningModuleRef(llvm::StringRef, MLIRContext *)>

using	TranslateFromMLIRFunction = std::function< LogicalResult(ModuleOp, llvm::raw_ostream &output)>

using	TranslateFunction = std::function< LogicalResult(llvm::SourceMgr &sourceMgr, llvm::raw_ostream &output, MLIRContext *)>

Enumerations
enum	SDBMExprKind { SDBMExprKind::Add, SDBMExprKind::Stripe, SDBMExprKind::Diff, SDBMExprKind::Constant, SDBMExprKind::DimId, SDBMExprKind::SymbolId, SDBMExprKind::Neg }

enum	CmpFPredicate { CmpFPredicate::FirstValidValue, CmpFPredicate::AlwaysFalse = FirstValidValue, CmpFPredicate::OEQ, CmpFPredicate::OGT, CmpFPredicate::OGE, CmpFPredicate::OLT, CmpFPredicate::OLE, CmpFPredicate::ONE, CmpFPredicate::ORD, CmpFPredicate::UEQ, CmpFPredicate::UGT, CmpFPredicate::UGE, CmpFPredicate::ULT, CmpFPredicate::ULE, CmpFPredicate::UNE, CmpFPredicate::UNO, CmpFPredicate::AlwaysTrue, CmpFPredicate::NumPredicates }

enum	AffineExprKind { AffineExprKind::Add, AffineExprKind::Mul, AffineExprKind::Mod, AffineExprKind::FloorDiv, AffineExprKind::CeilDiv, AffineExprKind::LAST_AFFINE_BINARY_OP = CeilDiv, AffineExprKind::Constant, AffineExprKind::DimId, AffineExprKind::SymbolId }

enum	DiagnosticSeverity { DiagnosticSeverity::Note, DiagnosticSeverity::Warning, DiagnosticSeverity::Error, DiagnosticSeverity::Remark }
	Defines the different supported severity of a diagnostic. More...

enum	OperationProperty { OperationProperty::Commutative = 0x1, OperationProperty::NoSideEffect = 0x2, OperationProperty::Terminator = 0x4, OperationProperty::IsolatedFromAbove = 0x8 }

enum	PassDisplayMode { PassDisplayMode::List, PassDisplayMode::Pipeline }

Functions
void	populateStandardToSPIRVPatterns (MLIRContext *context, SPIRVTypeConverter &typeConverter, OwningRewritePatternList &patterns)

void	getReachableAffineApplyOps (ArrayRef< Value > operands, SmallVectorImpl< Operation *> &affineApplyOps)

LogicalResult	getIndexSet (MutableArrayRef< AffineForOp > forOps, FlatAffineConstraints *domain)

DependenceResult	checkMemrefAccessDependence (const MemRefAccess &srcAccess, const MemRefAccess &dstAccess, unsigned loopDepth, FlatAffineConstraints dependenceConstraints, SmallVector< DependenceComponent, 2 > dependenceComponents, bool allowRAR=false)

bool	hasDependence (DependenceResult result)

void	getDependenceComponents (AffineForOp forOp, unsigned maxLoopDepth, std::vector< SmallVector< DependenceComponent, 2 >> *depCompsVec)

AffineExpr	simplifyAffineExpr (AffineExpr expr, unsigned numDims, unsigned numSymbols)
	Simplify the affine expression by flattening it and reconstructing it. More...

LogicalResult	getFlattenedAffineExpr (AffineExpr expr, unsigned numDims, unsigned numSymbols, SmallVectorImpl< int64_t > flattenedExpr, FlatAffineConstraints cst=nullptr)

LogicalResult	getFlattenedAffineExprs (AffineMap map, std::vector< SmallVector< int64_t, 8 >> flattenedExprs, FlatAffineConstraints cst=nullptr)

LogicalResult	getFlattenedAffineExprs (IntegerSet set, std::vector< SmallVector< int64_t, 8 >> flattenedExprs, FlatAffineConstraints cst=nullptr)

void	buildTripCountMapAndOperands (AffineForOp forOp, AffineMap map, SmallVectorImpl< Value > operands)

Optional< uint64_t >	getConstantTripCount (AffineForOp forOp)

uint64_t	getLargestDivisorOfTripCount (AffineForOp forOp)

DenseSet< Value, DenseMapInfo< Value > >	getInvariantAccesses (Value iv, ArrayRef< Value > indices)

bool	isVectorizableLoopBody (AffineForOp loop, NestedPattern &vectorTransferMatcher)

bool	isVectorizableLoopBody (AffineForOp loop, int *memRefDim, NestedPattern &vectorTransferMatcher)

bool	isInstwiseShiftValid (AffineForOp forOp, ArrayRef< uint64_t > shifts)

bool	defaultFilterFunction (Operation &)

std::unique_ptr< OpPassBase< FuncOp > >	createMemRefBoundCheckPass ()
	Creates a pass to check memref accesses in a Function. More...

std::unique_ptr< OpPassBase< FuncOp > >	createTestMemRefDependenceCheckPass ()
	Creates a pass to check memref access dependences in a Function. More...

std::unique_ptr< OpPassBase< FuncOp > >	createParallelismDetectionTestPass ()
	Creates a pass to test parallelism detection; emits note for parallel loops. More...

void	getForwardSlice (Operation op, llvm::SetVector< Operation > forwardSlice, TransitiveFilter filter=[](Operation ) { return true;})

void	getBackwardSlice (Operation op, llvm::SetVector< Operation > backwardSlice, TransitiveFilter filter=[](Operation ) { return true;})

llvm::SetVector< Operation * >	getSlice (Operation op, TransitiveFilter backwardFilter=[](Operation ) { return true;}, TransitiveFilter forwardFilter=[](Operation *) { return true;})

llvm::SetVector< Operation * >	topologicalSort (const llvm::SetVector< Operation *> &toSort)

void	getLoopIVs (Operation &op, SmallVectorImpl< AffineForOp > *loops)

unsigned	getNestingDepth (Operation &op)

void	getSequentialLoops (AffineForOp forOp, llvm::SmallDenseSet< Value, 8 > *sequentialLoops)

void	getComputationSliceState (Operation depSourceOp, Operation depSinkOp, FlatAffineConstraints dependenceConstraints, unsigned loopDepth, bool isBackwardSlice, ComputationSliceState sliceState)

LogicalResult	computeSliceUnion (ArrayRef< Operation > opsA, ArrayRef< Operation > opsB, unsigned loopDepth, unsigned numCommonLoops, bool isBackwardSlice, ComputationSliceState *sliceUnion)

AffineForOp	insertBackwardComputationSlice (Operation srcOpInst, Operation dstOpInst, unsigned dstLoopDepth, ComputationSliceState *sliceState)

Optional< uint64_t >	getMemRefSizeInBytes (MemRefType memRefType)

template<typename LoadOrStoreOpPointer >
LogicalResult	boundCheckLoadOrStoreOp (LoadOrStoreOpPointer loadOrStoreOp, bool emitError=true)

unsigned	getNumCommonSurroundingLoops (Operation &A, Operation &B)
	Returns the number of surrounding loops common to both A and B. More...

Optional< int64_t >	getMemoryFootprintBytes (AffineForOp forOp, int memorySpace=-1)

bool	isLoopParallel (AffineForOp forOp)
	Returns true if `forOp' is a parallel loop. More...

LogicalResult	verify (Operation *op)

Value	expandAffineExpr (OpBuilder &builder, Location loc, AffineExpr expr, ArrayRef< Value > dimValues, ArrayRef< Value > symbolValues)

void	populateAffineToStdConversionPatterns (OwningRewritePatternList &patterns, MLIRContext *ctx)

Value	lowerAffineLowerBound (AffineForOp op, OpBuilder &builder)

Value	lowerAffineUpperBound (AffineForOp op, OpBuilder &builder)

std::unique_ptr< OpPassBase< ModuleOp > >	createConvertGPUKernelToCubinPass (CubinGenerator cubinGenerator)

std::unique_ptr< OpPassBase< ModuleOp > >	createConvertGpuLaunchFuncToCudaCallsPass ()

void	populateGpuToNVVMConversionPatterns (LLVMTypeConverter &converter, OwningRewritePatternList &patterns)
	Collect a set of patterns to convert from the GPU dialect to NVVM. More...

std::unique_ptr< OpPassBase< ModuleOp > >	createLowerGpuOpsToNVVMOpsPass ()
	Creates a pass that lowers GPU dialect operations to NVVM counterparts. More...

std::unique_ptr< OpPassBase< ModuleOp > >	createLowerGpuOpsToROCDLOpsPass ()
	Creates a pass that lowers GPU dialect operations to ROCDL counterparts. More...

void	populateGPUToSPIRVPatterns (MLIRContext *context, SPIRVTypeConverter &typeConverter, OwningRewritePatternList &patterns, ArrayRef< int64_t > workGroupSize)

std::unique_ptr< OpPassBase< ModuleOp > >	createConvertGPUToSPIRVPass (ArrayRef< int64_t > workGroupSize)

void	populateLinalgToLLVMConversionPatterns (LinalgTypeConverter &converter, OwningRewritePatternList &patterns, MLIRContext *ctx)
	Populate the given list with patterns that convert from Linalg to LLVM. More...

LogicalResult	convertAffineLoopNestToGPULaunch (AffineForOp forOp, unsigned numBlockDims, unsigned numThreadDims)

LogicalResult	convertLoopNestToGPULaunch (loop::ForOp forOp, unsigned numBlockDims, unsigned numThreadDims)

LogicalResult	convertLoopToGPULaunch (loop::ForOp forOp, ArrayRef< Value > numWorkGroups, ArrayRef< Value > workGroupSizes)

std::unique_ptr< OpPassBase< FuncOp > >	createSimpleLoopsToGPUPass (unsigned numBlockDims, unsigned numThreadDims)

std::unique_ptr< OpPassBase< FuncOp > >	createLoopToGPUPass (ArrayRef< int64_t > numWorkGroups, ArrayRef< int64_t > workGroupSize)

void	populateLoopToStdConversionPatterns (OwningRewritePatternList &patterns, MLIRContext *ctx)

std::unique_ptr< Pass >	createLowerToCFGPass ()
	Creates a pass to convert loop.for, loop.if and loop.terminator ops to CFG. More...

void	populateStdToLLVMMemoryConversionPatters (LLVMTypeConverter &converter, OwningRewritePatternList &patterns)

void	populateStdToLLVMNonMemoryConversionPatterns (LLVMTypeConverter &converter, OwningRewritePatternList &patterns)
	Collect a set of patterns to convert from the Standard dialect to LLVM. More...

void	populateStdToLLVMConversionPatterns (LLVMTypeConverter &converter, OwningRewritePatternList &patterns)
	Collect a set of patterns to convert from the Standard dialect to LLVM. More...

std::unique_ptr< OpPassBase< ModuleOp > >	createLowerToLLVMPass (bool useAlloca=false)

std::unique_ptr< OpPassBase< ModuleOp > >	createLowerToLLVMPass (LLVMPatternListFiller patternListFiller, LLVMTypeConverterMaker typeConverterMaker, bool useAlloca=false)

template<typename TypeConverter = LLVMTypeConverter>
std::unique_ptr< OpPassBase< ModuleOp > >	createLowerToLLVMPass (LLVMPatternListFiller patternListFiller, bool useAlloca=false)

void	populateStdLegalizationPatternsForSPIRVLowering (MLIRContext *context, OwningRewritePatternList &patterns)

std::unique_ptr< OpPassBase< ModuleOp > >	createConvertStandardToSPIRVPass ()
	Pass to convert StandardOps to SPIR-V ops. More...

std::unique_ptr< Pass >	createLegalizeStdOpsForSPIRVLoweringPass ()
	Pass to legalize ops that are not directly lowered to SPIR-V. More...

void	populateVectorToLLVMConversionPatterns (LLVMTypeConverter &converter, OwningRewritePatternList &patterns)
	Collect a set of patterns to convert from the Vector dialect to LLVM. More...

OpPassBase< ModuleOp > *	createLowerVectorToLLVMPass ()
	Create a pass to convert vector operations to the LLVMIR dialect. More...

void	populateVectorToAffineLoopsConversionPatterns (MLIRContext *context, OwningRewritePatternList &patterns)
	Collect a set of patterns to convert from the Vector dialect to loops + std. More...

OpPassBase< ModuleOp > *	createLowerVectorToLoopsPass ()
	Create a pass to convert vector operations to affine loops + std dialect. More...

bool	isTopLevelValue (Value value)

bool	isValidDim (Value value)
	Returns true if the given Value can be used as a dimension id. More...

bool	isValidSymbol (Value value)
	Returns true if the given Value can be used as a symbol. More...

void	canonicalizeMapAndOperands (AffineMap map, SmallVectorImpl< Value > operands)

void	canonicalizeSetAndOperands (IntegerSet set, SmallVectorImpl< Value > operands)

AffineApplyOp	makeComposedAffineApply (OpBuilder &b, Location loc, AffineMap map, ArrayRef< Value > operands)

void	fullyComposeAffineMapAndOperands (AffineMap map, SmallVectorImpl< Value > operands)

bool	isForInductionVar (Value val)
	Returns if the provided value is the induction variable of a AffineForOp. More...

AffineForOp	getForInductionVarOwner (Value val)

void	extractForInductionVars (ArrayRef< AffineForOp > forInsts, SmallVectorImpl< Value > *ivs)

template<class AttrElementT , class ElementValueT = typename AttrElementT::ValueType, class CalculationT = function_ref<ElementValueT(ElementValueT, ElementValueT)>>
Attribute	constFoldBinaryOp (ArrayRef< Attribute > operands, const CalculationT &calculate)

void	promoteToWorkgroupMemory (gpu::GPUFuncOp op, unsigned arg)

std::unique_ptr< OpPassBase< ModuleOp > >	createGpuKernelOutliningPass ()

IntInfty	operator+ (IntInfty lhs, IntInfty rhs)

bool	operator< (IntInfty lhs, IntInfty rhs)

bool	operator<= (IntInfty lhs, IntInfty rhs)

bool	operator== (IntInfty lhs, IntInfty rhs)

bool	operator!= (IntInfty lhs, IntInfty rhs)

void	printDimAndSymbolList (Operation::operand_iterator begin, Operation::operand_iterator end, unsigned numDims, OpAsmPrinter &p)
	Prints dimension and symbol list. More...

ParseResult	parseDimAndSymbolList (OpAsmParser &parser, SmallVectorImpl< Value > &operands, unsigned &numDims)
	Parses dimension and symbol list and returns true if parsing failed. More...

raw_ostream &	operator<< (raw_ostream &os, SubViewOp::Range &range)

unsigned	getNumIterators (StringRef name, ArrayAttr iteratorTypes)
	Returns the iterator of a certain type. More...

unsigned	getNumIterators (ArrayAttr iteratorTypes)

Optional< SmallVector< int64_t, 4 > >	shapeRatio (ArrayRef< int64_t > superShape, ArrayRef< int64_t > subShape)

Optional< SmallVector< int64_t, 4 > >	shapeRatio (VectorType superVectorType, VectorType subVectorType)

void	populateVectorToVectorConversionPatterns (MLIRContext *context, OwningRewritePatternList &patterns, ArrayRef< int64_t > coarseVectorShape={}, ArrayRef< int64_t > fineVectorShape={})

void	initializeLLVMPasses ()

inline ::llvm::hash_code	hash_value (AffineExpr arg)
	Make AffineExpr hashable. More...

AffineExpr	operator+ (int64_t val, AffineExpr expr)

AffineExpr	operator* (int64_t val, AffineExpr expr)

AffineExpr	operator- (int64_t val, AffineExpr expr)

AffineExpr	getAffineDimExpr (unsigned position, MLIRContext *context)
	These free functions allow clients of the API to not use classes in detail. More...

AffineExpr	getAffineSymbolExpr (unsigned position, MLIRContext *context)

AffineExpr	getAffineConstantExpr (int64_t constant, MLIRContext *context)

AffineExpr	getAffineBinaryOpExpr (AffineExprKind kind, AffineExpr lhs, AffineExpr rhs)

AffineExpr	toAffineExpr (ArrayRef< int64_t > eq, unsigned numDims, unsigned numSymbols, ArrayRef< AffineExpr > localExprs, MLIRContext *context)

raw_ostream &	operator<< (raw_ostream &os, AffineExpr &expr)

bool	getFlattenedAffineExpr (AffineExpr expr, unsigned numDims, unsigned numSymbols, SmallVectorImpl< int64_t > *flattenedExpr)

bool	getFlattenedAffineExprs (AffineMap map, std::vector< SmallVector< int64_t, 8 >> *flattenedExprs)

bool	getFlattenedAffineExprs (IntegerSet set, std::vector< SmallVector< int64_t, 8 >> *flattenedExprs)

template<typename... AffineExprTy>
void	bindDims (MLIRContext *ctx, AffineExprTy &... exprs)

inline ::llvm::hash_code	hash_value (AffineMap arg)

AffineMap	simplifyAffineMap (AffineMap map)
	Simplify an affine map by simplifying its underlying AffineExpr results. More...

AffineMap	inversePermutation (AffineMap map)

AffineMap	concatAffineMaps (ArrayRef< AffineMap > maps)

raw_ostream &	operator<< (raw_ostream &os, AffineMap map)

raw_ostream &	operator<< (raw_ostream &os, Attribute attr)

inline ::llvm::hash_code	hash_value (Attribute arg)

raw_ostream &	operator<< (raw_ostream &os, const DiagnosticArgument &arg)

raw_ostream &	operator<< (raw_ostream &os, const Diagnostic &diag)

InFlightDiagnostic	emitError (Location loc)
	Utility method to emit an error message using this location. More...

InFlightDiagnostic	emitError (Location loc, const Twine &message)

InFlightDiagnostic	emitWarning (Location loc)
	Utility method to emit a warning message using this location. More...

InFlightDiagnostic	emitWarning (Location loc, const Twine &message)

InFlightDiagnostic	emitRemark (Location loc)
	Utility method to emit a remark message using this location. More...

InFlightDiagnostic	emitRemark (Location loc, const Twine &message)

template<typename... Args>
LogicalResult	emitOptionalError (Optional< Location > loc, Args &&... args)

template<typename... Args>
LogicalResult	emitOptionalWarning (Optional< Location > loc, Args &&... args)

template<typename... Args>
LogicalResult	emitOptionalRemark (Optional< Location > loc, Args &&... args)

void	registerDialectAllocator (const DialectAllocatorFunction &function)

void	registerAllDialects (MLIRContext *context)
	Registers all dialects with the specified MLIRContext. More...

template<typename ConcreteDialect >
void	registerDialect ()

void	registerDialectHooksSetter (const DialectHooksSetter &function)

DialectAsmPrinter &	operator<< (DialectAsmPrinter &p, Attribute attr)

DialectAsmPrinter &	operator<< (DialectAsmPrinter &p, const APFloat &value)

DialectAsmPrinter &	operator<< (DialectAsmPrinter &p, float value)

DialectAsmPrinter &	operator<< (DialectAsmPrinter &p, double value)

DialectAsmPrinter &	operator<< (DialectAsmPrinter &p, Type type)

template<typename T , typename std::enable_if< !std::is_convertible< T &, Attribute &>::value &&!std::is_convertible< T &, Type &>::value &&!std::is_convertible< T &, APFloat &>::value &&!llvm::is_one_of< T, double, float >::value, T >::type * = nullptr>
DialectAsmPrinter &	operator<< (DialectAsmPrinter &p, const T &other)

raw_ostream &	operator<< (raw_ostream &os, Identifier identifier)

bool	operator== (Identifier lhs, Identifier rhs)

bool	operator!= (Identifier lhs, Identifier rhs)

bool	operator== (Identifier lhs, StringRef rhs)

bool	operator!= (Identifier lhs, StringRef rhs)

bool	operator== (StringRef lhs, Identifier rhs)

bool	operator!= (StringRef lhs, Identifier rhs)

llvm::hash_code	hash_value (Identifier arg)

inline ::llvm::hash_code	hash_value (IntegerSet arg)

raw_ostream &	operator<< (raw_ostream &os, const Location &loc)

inline ::llvm::hash_code	hash_value (Location arg)

template<typename AttrT >
detail::constant_op_binder< AttrT >	m_Constant (AttrT *bind_value)

detail::constant_int_value_matcher< 1 >	m_One ()
	Matches a constant scalar / vector splat / tensor splat integer one. More...

template<typename OpClass >
detail::op_matcher< OpClass >	m_Op ()
	Matches the given OpClass. More...

detail::constant_int_value_matcher< 0 >	m_Zero ()
	Matches a constant scalar / vector splat / tensor splat integer zero. More...

detail::constant_int_not_value_matcher< 0 >	m_NonZero ()

template<typename Pattern >
bool	matchPattern (Value value, const Pattern &pattern)
	Entry point for matching a pattern over a Value. More...

template<typename Pattern >
bool	matchPattern (Operation *op, const Pattern &pattern)
	Entry point for matching a pattern over an Operation. More...

detail::constant_int_op_binder	m_ConstantInt (IntegerAttr::ValueType *bind_value)

template<typename OpType , typename... Matchers>
auto	m_Op (Matchers... matchers)

bool	operator== (OpState lhs, OpState rhs)

bool	operator!= (OpState lhs, OpState rhs)

raw_ostream &	operator<< (raw_ostream &os, Operation &op)

raw_ostream &	operator<< (raw_ostream &os, OperationName identifier)

bool	operator== (OperationName lhs, OperationName rhs)

bool	operator!= (OperationName lhs, OperationName rhs)

llvm::hash_code	hash_value (OperationName arg)

OpAsmPrinter &	operator<< (OpAsmPrinter &p, Value value)

template<typename T , typename std::enable_if< std::is_convertible< T &, ValueRange >::value &&!std::is_convertible< T &, Value &>::value, T >::type * = nullptr>
OpAsmPrinter &	operator<< (OpAsmPrinter &p, const T &values)

OpAsmPrinter &	operator<< (OpAsmPrinter &p, Type type)

OpAsmPrinter &	operator<< (OpAsmPrinter &p, Attribute attr)

OpAsmPrinter &	operator<< (OpAsmPrinter &p, bool value)

template<typename IteratorT >
OpAsmPrinter &	operator<< (OpAsmPrinter &p, const iterator_range< ValueTypeIterator< IteratorT >> &types)

bool	applyPatternsGreedily (Operation *op, const OwningRewritePatternList &patterns)

bool	applyPatternsGreedily (MutableArrayRef< Region > regions, const OwningRewritePatternList &patterns)
	Rewrite the given regions, which must be isolated from above. More...

LogicalResult	getStridesAndOffset (MemRefType t, SmallVectorImpl< int64_t > &strides, int64_t &offset)

LogicalResult	getStridesAndOffset (MemRefType t, SmallVectorImpl< AffineExpr > &strides, AffineExpr &offset)

AffineMap	makeStridedLinearLayoutMap (ArrayRef< int64_t > strides, int64_t offset, MLIRContext *context)

MemRefType	canonicalizeStridedLayout (MemRefType t)

bool	isStrided (MemRefType t)
	Return true if the layout for `t` is compatible with strided semantics. More...

raw_ostream &	operator<< (raw_ostream &os, Type type)

inline ::llvm::hash_code	hash_value (Type arg)

Type	getElementTypeOrSelf (Type type)
	Return the element type or return the type itself. More...

Type	getElementTypeOrSelf (Attribute attr)
	Return the element type or return the type itself. More...

Type	getElementTypeOrSelf (Value val)

SmallVector< Type, 10 >	getFlattenedTypes (TupleType t)

bool	isOpaqueTypeWithName (Type type, StringRef dialect, StringRef typeData)

LogicalResult	verifyCompatibleShape (ArrayRef< int64_t > shape1, ArrayRef< int64_t > shape2)

LogicalResult	verifyCompatibleShape (Type type1, Type type2)

raw_ostream &	operator<< (raw_ostream &os, Value value)

inline ::llvm::hash_code	hash_value (Value arg)
	Make Value hashable. More...

OwningModuleRef	parseSourceFile (const llvm::SourceMgr &sourceMgr, MLIRContext *context)

OwningModuleRef	parseSourceFile (llvm::StringRef filename, MLIRContext *context)

OwningModuleRef	parseSourceFile (llvm::StringRef filename, llvm::SourceMgr &sourceMgr, MLIRContext *context)

OwningModuleRef	parseSourceString (llvm::StringRef moduleStr, MLIRContext *context)

Attribute	parseAttribute (llvm::StringRef attrStr, MLIRContext *context)

Attribute	parseAttribute (llvm::StringRef attrStr, Type type)

Attribute	parseAttribute (llvm::StringRef attrStr, MLIRContext *context, size_t &numRead)

Attribute	parseAttribute (llvm::StringRef attrStr, Type type, size_t &numRead)

Type	parseType (llvm::StringRef typeStr, MLIRContext *context)

Type	parseType (llvm::StringRef typeStr, MLIRContext *context, size_t &numRead)

void	registerPassManagerCLOptions ()

void	applyPassManagerCLOptions (PassManager &pm)

void	registerPassPipeline (StringRef arg, StringRef description, const PassRegistryFunction &function)

void	registerPass (StringRef arg, StringRef description, const PassID *passID, const PassAllocatorFunction &function)

LogicalResult	parsePassPipeline (StringRef pipeline, OpPassManager &pm, raw_ostream &errorStream=llvm::errs())

std::unique_ptr< llvm::MemoryBuffer >	openInputFile (llvm::StringRef inputFilename, std::string *errorMessage=nullptr)

std::unique_ptr< llvm::ToolOutputFile >	openOutputFile (llvm::StringRef outputFilename, std::string *errorMessage=nullptr)

int	JitRunnerMain (int argc, char **argv, llvm::function_ref< LogicalResult(mlir::ModuleOp)> mlirTransformer)

LogicalResult	success (bool isSuccess=true)

LogicalResult	failure (bool isFailure=true)

bool	succeeded (LogicalResult result)

bool	failed (LogicalResult result)

int64_t	ceilDiv (int64_t lhs, int64_t rhs)

int64_t	floorDiv (int64_t lhs, int64_t rhs)

int64_t	mod (int64_t lhs, int64_t rhs)

int64_t	lcm (int64_t a, int64_t b)
	Returns the least common multiple of 'a' and 'b'. More...

LogicalResult	MlirOptMain (llvm::raw_ostream &os, std::unique_ptr< llvm::MemoryBuffer > buffer, const PassPipelineCLParser &passPipeline, bool splitInputFile, bool verifyDiagnostics, bool verifyPasses)

template<typename ForwardIterator , typename UnaryFunctor , typename NullaryFunctor , typename = typename std::enable_if< !std::is_constructible<StringRef, UnaryFunctor>::value && !std::is_constructible<StringRef, NullaryFunctor>::value>::type>
void	interleave (ForwardIterator begin, ForwardIterator end, UnaryFunctor each_fn, NullaryFunctor between_fn)

template<typename Container , typename UnaryFunctor , typename NullaryFunctor , typename = typename std::enable_if< !std::is_constructible<StringRef, UnaryFunctor>::value && !std::is_constructible<StringRef, NullaryFunctor>::value>::type>
void	interleave (const Container &c, UnaryFunctor each_fn, NullaryFunctor between_fn)

template<typename Container , typename UnaryFunctor , typename raw_ostream , typename T = detail::ValueOfRange<Container>>
void	interleave (const Container &c, raw_ostream &os, UnaryFunctor each_fn, const StringRef &separator)
	Overload of interleave for the common case of string separator. More...

template<typename Container , typename raw_ostream , typename T = detail::ValueOfRange<Container>>
void	interleave (const Container &c, raw_ostream &os, const StringRef &separator)

template<typename Container , typename UnaryFunctor , typename raw_ostream , typename T = detail::ValueOfRange<Container>>
void	interleaveComma (const Container &c, raw_ostream &os, UnaryFunctor each_fn)

template<typename Container , typename raw_ostream , typename T = detail::ValueOfRange<Container>>
void	interleaveComma (const Container &c, raw_ostream &os)

template<typename ContainerTy >
auto	make_second_range (ContainerTy &&c)
	Given a container of pairs, return a range over the second elements. More...

template<typename ContainerTy >
bool	has_single_element (ContainerTy &&c)
	Returns true of the given range only contains a single element. More...

std::string	convertToSnakeCase (llvm::StringRef input)

std::string	convertToCamelCase (llvm::StringRef input, bool capitalizeFirst=false)

LogicalResult	splitAndProcessBuffer (std::unique_ptr< llvm::MemoryBuffer > originalBuffer, ChunkBufferHandler processChunkBuffer, raw_ostream &os)

std::unique_ptr< llvm::Module >	translateModuleToLLVMIR (ModuleOp m)

OwningModuleRef	translateLLVMIRToModule (std::unique_ptr< llvm::Module > llvmModule, MLIRContext *context)

std::unique_ptr< llvm::Module >	translateModuleToNVVMIR (Operation *m)

std::unique_ptr< llvm::Module >	translateModuleToROCDLIR (Operation *m)

void	populateFuncOpTypeConversionPattern (OwningRewritePatternList &patterns, MLIRContext *ctx, TypeConverter &converter)

LLVM_NODISCARD LogicalResult	applyPartialConversion (ArrayRef< Operation > ops, ConversionTarget &target, const OwningRewritePatternList &patterns, TypeConverter converter=nullptr)

LLVM_NODISCARD LogicalResult	applyPartialConversion (Operation op, ConversionTarget &target, const OwningRewritePatternList &patterns, TypeConverter converter=nullptr)

LLVM_NODISCARD LogicalResult	applyFullConversion (ArrayRef< Operation > ops, ConversionTarget &target, const OwningRewritePatternList &patterns, TypeConverter converter=nullptr)

LLVM_NODISCARD LogicalResult	applyFullConversion (Operation op, ConversionTarget &target, const OwningRewritePatternList &patterns, TypeConverter converter=nullptr)

LLVM_NODISCARD LogicalResult	applyAnalysisConversion (ArrayRef< Operation > ops, ConversionTarget &target, const OwningRewritePatternList &patterns, DenseSet< Operation > &convertedOps, TypeConverter *converter=nullptr)

LLVM_NODISCARD LogicalResult	applyAnalysisConversion (Operation op, ConversionTarget &target, const OwningRewritePatternList &patterns, DenseSet< Operation > &convertedOps, TypeConverter *converter=nullptr)

LogicalResult	inlineRegion (InlinerInterface &interface, Region src, Operation inlinePoint, BlockAndValueMapping &mapper, ArrayRef< Value > resultsToReplace, Optional< Location > inlineLoc=llvm::None, bool shouldCloneInlinedRegion=true)

LogicalResult	inlineRegion (InlinerInterface &interface, Region src, Operation inlinePoint, ArrayRef< Value > inlinedOperands, ArrayRef< Value > resultsToReplace, Optional< Location > inlineLoc=llvm::None, bool shouldCloneInlinedRegion=true)

LogicalResult	inlineCall (InlinerInterface &interface, CallOpInterface call, CallableOpInterface callable, Region *src, bool shouldCloneInlinedRegion=true)

FusionResult	canFuseLoops (AffineForOp srcForOp, AffineForOp dstForOp, unsigned dstLoopDepth, ComputationSliceState *srcSlice)

bool	getLoopNestStats (AffineForOp forOp, LoopNestStats *stats)

int64_t	getComputeCost (AffineForOp forOp, LoopNestStats &stats)

bool	getFusionComputeCost (AffineForOp srcForOp, LoopNestStats &srcStats, AffineForOp dstForOp, LoopNestStats &dstStats, ComputationSliceState slice, int64_t computeCost)

LogicalResult	loopUnrollFull (AffineForOp forOp)
	Unrolls this loop completely. More...

LogicalResult	loopUnrollByFactor (AffineForOp forOp, uint64_t unrollFactor)

LogicalResult	loopUnrollUpToFactor (AffineForOp forOp, uint64_t unrollFactor)

void	getPerfectlyNestedLoops (SmallVectorImpl< AffineForOp > &nestedLoops, AffineForOp root)

void	getPerfectlyNestedLoops (SmallVectorImpl< loop::ForOp > &nestedLoops, loop::ForOp root)

LogicalResult	loopUnrollJamByFactor (AffineForOp forOp, uint64_t unrollJamFactor)
	Unrolls and jams this loop by the specified factor. More...

LogicalResult	loopUnrollJamUpToFactor (AffineForOp forOp, uint64_t unrollJamFactor)

LogicalResult	promoteIfSingleIteration (AffineForOp forOp)

void	promoteSingleIterationLoops (FuncOp f)

void	getCleanupLoopLowerBound (AffineForOp forOp, unsigned unrollFactor, AffineMap map, SmallVectorImpl< Value > operands, OpBuilder &builder)

LLVM_NODISCARD LogicalResult	instBodySkew (AffineForOp forOp, ArrayRef< uint64_t > shifts, bool unrollPrologueEpilogue=false)

LLVM_NODISCARD LogicalResult	tileCodeGen (MutableArrayRef< AffineForOp > band, ArrayRef< unsigned > tileSizes)

void	interchangeLoops (AffineForOp forOpA, AffineForOp forOpB)

bool	isValidLoopInterchangePermutation (ArrayRef< AffineForOp > loops, ArrayRef< unsigned > loopPermMap)

unsigned	interchangeLoops (ArrayRef< AffineForOp > loops, ArrayRef< unsigned > loopPermMap)

AffineForOp	sinkSequentialLoops (AffineForOp forOp)

void	sinkLoop (AffineForOp forOp, unsigned loopDepth)

SmallVector< SmallVector< AffineForOp, 8 >, 8 >	tile (ArrayRef< AffineForOp > forOps, ArrayRef< uint64_t > sizes, ArrayRef< AffineForOp > targets)

SmallVector< Loops, 8 >	tile (ArrayRef< loop::ForOp > forOps, ArrayRef< Value > sizes, ArrayRef< loop::ForOp > targets)

SmallVector< AffineForOp, 8 >	tile (ArrayRef< AffineForOp > forOps, ArrayRef< uint64_t > sizes, AffineForOp target)

Loops	tile (ArrayRef< loop::ForOp > forOps, ArrayRef< Value > sizes, loop::ForOp target)

Loops	tilePerfectlyNested (loop::ForOp rootForOp, ArrayRef< Value > sizes)

uint64_t	affineDataCopyGenerate (Block::iterator begin, Block::iterator end, const AffineCopyOptions &copyOptions, DenseSet< Operation *> &copyNests)

TileLoops	extractFixedOuterLoops (loop::ForOp rootFOrOp, ArrayRef< int64_t > sizes)

void	coalesceLoops (MutableArrayRef< loop::ForOp > loops)

void	mapLoopToProcessorIds (loop::ForOp forOp, ArrayRef< Value > processorId, ArrayRef< Value > numProcessors)

std::unique_ptr< Pass >	createCanonicalizerPass ()
	Creates an instance of the Canonicalizer pass. More...

std::unique_ptr< Pass >	createCSEPass ()
	Creates a pass to perform common sub expression elimination. More...

std::unique_ptr< OpPassBase< FuncOp > >	createVectorizePass (ArrayRef< int64_t > virtualVectorSize)

std::unique_ptr< OpPassBase< FuncOp > >	createVectorizerTestPass ()

std::unique_ptr< OpPassBase< FuncOp > >	createMaterializeVectorsPass (ArrayRef< int64_t > vectorSize)
	Creates a pass to lower super-vectors to target-dependent HW vectors. More...

std::unique_ptr< OpPassBase< FuncOp > >	createLoopUnrollPass (int unrollFactor=-1, int unrollFull=-1, const std::function< unsigned(AffineForOp)> &getUnrollFactor=nullptr)

std::unique_ptr< OpPassBase< FuncOp > >	createLoopUnrollAndJamPass (int unrollJamFactor=-1)

std::unique_ptr< OpPassBase< FuncOp > >	createSimplifyAffineStructuresPass ()

std::unique_ptr< OpPassBase< FuncOp > >	createLoopFusionPass (unsigned fastMemorySpace=0, uint64_t localBufSizeThreshold=0, bool maximalFusion=false)

std::unique_ptr< Pass >	createLoopInvariantCodeMotionPass ()

std::unique_ptr< OpPassBase< FuncOp > >	createAffineLoopInvariantCodeMotionPass ()

std::unique_ptr< OpPassBase< FuncOp > >	createPipelineDataTransferPass ()

std::unique_ptr< OpPassBase< FuncOp > >	createLowerAffinePass ()

std::unique_ptr< OpPassBase< FuncOp > >	createLoopTilingPass (uint64_t cacheSizeBytes)
	Creates a pass to perform tiling on loop nests. More...

std::unique_ptr< OpPassBase< FuncOp > >	createSimpleParametricTilingPass (ArrayRef< int64_t > outerLoopSizes)

std::unique_ptr< OpPassBase< FuncOp > >	createLoopCoalescingPass ()

std::unique_ptr< OpPassBase< FuncOp > >	createAffineDataCopyGenerationPass (unsigned slowMemorySpace, unsigned fastMemorySpace, unsigned tagMemorySpace=0, int minDmaTransferSize=1024, uint64_t fastMemCapacityBytes=std::numeric_limits< uint64_t >::max())

std::unique_ptr< OpPassBase< FuncOp > >	createMemRefDataFlowOptPass ()

std::unique_ptr< OpPassBase< FuncOp > >	createStripDebugInfoPass ()
	Creates a pass to strip debug information from a function. More...

std::unique_ptr< OpPassBase< FuncOp > >	createTestLoopFusionPass ()
	Creates a pass which tests loop fusion utilities. More...

std::unique_ptr< Pass >	createInlinerPass ()

template<typename Range >
bool	areValuesDefinedAbove (Range values, Region &limit)

void	replaceAllUsesInRegionWith (Value orig, Value replacement, Region &region)
	Replace all uses of `orig` within the given region with `replacement`. More...

void	visitUsedValuesDefinedAbove (Region &region, Region &limit, function_ref< void(OpOperand *)> callback)

void	visitUsedValuesDefinedAbove (MutableArrayRef< Region > regions, function_ref< void(OpOperand *)> callback)

void	getUsedValuesDefinedAbove (Region &region, Region &limit, llvm::SetVector< Value > &values)

void	getUsedValuesDefinedAbove (MutableArrayRef< Region > regions, llvm::SetVector< Value > &values)

LogicalResult	simplifyRegions (MutableArrayRef< Region > regions)

LogicalResult	replaceAllMemRefUsesWith (Value oldMemRef, Value newMemRef, ArrayRef< Value > extraIndices={}, AffineMap indexRemap=AffineMap(), ArrayRef< Value > extraOperands={}, ArrayRef< Value > symbolOperands={}, Operation domInstFilter=nullptr, Operation postDomInstFilter=nullptr)

LogicalResult	replaceAllMemRefUsesWith (Value oldMemRef, Value newMemRef, Operation *op, ArrayRef< Value > extraIndices={}, AffineMap indexRemap=AffineMap(), ArrayRef< Value > extraOperands={}, ArrayRef< Value > symbolOperands={})

LogicalResult	normalizeMemRef (AllocOp op)

Operation *	createComposedAffineApplyOp (OpBuilder &builder, Location loc, ArrayRef< Value > operands, ArrayRef< Operation > affineApplyOps, SmallVectorImpl< Value > results)

void	createAffineComputationSlice (Operation opInst, SmallVectorImpl< AffineApplyOp > sliceOps)

void	viewGraph (Block &block, const Twine &name, bool shortNames=false, const Twine &title="", llvm::GraphProgram::Name program=llvm::GraphProgram::DOT)

raw_ostream &	writeGraph (raw_ostream &os, Block &block, bool shortNames=false, const Twine &title="")

std::unique_ptr< OpPassBase< ModuleOp > >	createPrintOpGraphPass (raw_ostream &os=llvm::errs(), bool shortNames=false, const Twine &title="")
	Creates a pass to print op graphs. More...

void	viewGraph (Region &region, const Twine &name, bool shortNames=false, const Twine &title="", llvm::GraphProgram::Name program=llvm::GraphProgram::DOT)

raw_ostream &	writeGraph (raw_ostream &os, Region &region, bool shortNames=false, const Twine &title="")

std::unique_ptr< mlir::OpPassBase< mlir::FuncOp > >	createPrintCFGGraphPass (raw_ostream &os=llvm::errs(), bool shortNames=false, const Twine &title="")
	Creates a pass to print CFG graphs. More...

const llvm::StringMap< TranslateSourceMgrToMLIRFunction > &	getTranslationToMLIRRegistry ()
	Get a read-only reference to the translator registry. More...

const llvm::StringMap< TranslateFromMLIRFunction > &	getTranslationFromMLIRRegistry ()

const llvm::StringMap< TranslateFunction > &	getTranslationRegistry ()

Variables
std::function< llvm::Error(llvm::Module *)>	makeOptimizingTransformer (unsigned optLevel, unsigned sizeLevel, llvm::TargetMachine *targetMachine)

std::function< llvm::Error(llvm::Module *)>	makeLLVMPassesTransformer (llvm::ArrayRef< const llvm::PassInfo > llvmPasses, llvm::Optional< unsigned > mbOptLevel, llvm::TargetMachine targetMachine, unsigned optPassesInsertPos=0)

Detailed Description

This file provides some simple template functional-style sugar to operate on value types. Make sure when using that the stored type is cheap to copy!

TODO(ntv): add some static_assert but we need proper traits for this.

Typedef Documentation

◆ AnalysisID

typedef ClassID mlir::AnalysisID

A special type used by analyses to provide an address that identifies a particular analysis set or a concrete analysis type.

◆ AttributeStorageAllocator

using mlir::AttributeStorageAllocator = typedef StorageUniquer::StorageAllocator

◆ ChunkBufferHandler

using mlir::ChunkBufferHandler = typedef function_ref<LogicalResult( std::unique_ptr<llvm::MemoryBuffer> chunkBuffer, raw_ostream &os)>

◆ CubinGenerator

using mlir::CubinGenerator = typedef std::function<OwnedCubin(const std::string &, Location, StringRef)>

◆ DefaultAttributeStorage

using mlir::DefaultAttributeStorage = typedef AttributeStorage

Default storage type for attributes that require no additional initialization or storage.

◆ DefaultTypeStorage

using mlir::DefaultTypeStorage = typedef TypeStorage

Default storage type for types that require no additional initialization or storage.

◆ DenseMap

template<typename KeyT , typename ValueT , typename KeyInfoT = DenseMapInfo<KeyT>, typename BucketT = llvm::detail::DenseMapPair<KeyT, ValueT>>

using mlir::DenseMap = typedef llvm::DenseMap<KeyT, ValueT, KeyInfoT, BucketT>

◆ DenseSet

template<typename ValueT , typename ValueInfoT = DenseMapInfo<ValueT>>

using mlir::DenseSet = typedef llvm::DenseSet<ValueT, ValueInfoT>

◆ DialectAllocatorFunction

using mlir::DialectAllocatorFunction = typedef std::function<void(MLIRContext *)>

◆ DialectConstantDecodeHook

using mlir::DialectConstantDecodeHook = typedef std::function<bool(const OpaqueElementsAttr, ElementsAttr &)>

◆ DialectConstantFoldHook

using mlir::DialectConstantFoldHook = typedef std::function<LogicalResult( Operation *, ArrayRef<Attribute>, SmallVectorImpl<Attribute> &)>

◆ DialectExtractElementHook

using mlir::DialectExtractElementHook = typedef std::function<Attribute(const OpaqueElementsAttr, ArrayRef<uint64_t>)>

◆ DialectHooksSetter

using mlir::DialectHooksSetter = typedef std::function<void(MLIRContext *)>

◆ DominanceInfoNode

using mlir::DominanceInfoNode = typedef llvm::DomTreeNodeBase<Block>

◆ FilterFunctionType

using mlir::FilterFunctionType = typedef std::function<bool(Operation &)>

A NestedPattern is a nested operation walker that:

recursively matches a substructure in the tree;
uses a filter function to refine matches with extra semantic constraints (passed via a lambda of type FilterFunctionType);
TODO(ntv) optionally applies actions (lambda).

Nested patterns are meant to capture imperfectly nested loops while matching properties over the whole loop nest. For instance, in vectorization we are interested in capturing all the imperfectly nested loops of a certain type and such that all the load and stores have certain access patterns along the loops' induction variables). Such NestedMatches are first captured using the match function and are later processed to analyze properties and apply transformations in a non-greedy way.

The NestedMatches captured in the IR can grow large, especially after aggressive unrolling. As experience has shown, it is generally better to use a plain walk over operations to match flat patterns but the current implementation is competitive nonetheless.

◆ function_ref

template<typename Fn >

using mlir::function_ref = typedef llvm::function_ref<Fn>

◆ GenFunction

using mlir::GenFunction = typedef std::function<bool(const llvm::RecordKeeper &recordKeeper, raw_ostream &os)>

Generator function to invoke.

◆ is_detected

template<template< class... > class Op, class... Args>

using mlir::is_detected = typedef typename detail::detector<void, Op, Args...>::value_t

◆ is_invocable

template<typename Callable , typename... Args>

using mlir::is_invocable = typedef is_detected<detail::is_invocable, Callable, Args...>

◆ LLVMPatternListFiller

using mlir::LLVMPatternListFiller = typedef std::function<void(LLVMTypeConverter &, OwningRewritePatternList &)>

Type for a callback constructing the owning list of patterns for the conversion to the LLVMIR dialect. The callback is expected to append patterns to the owning list provided as the second argument.

◆ LLVMTypeConverterMaker

using mlir::LLVMTypeConverterMaker = typedef std::function<std::unique_ptr<LLVMTypeConverter>(MLIRContext *)>

Type for a callback constructing the type converter for the conversion to the LLVMIR dialect. The callback is expected to return an instance of the converter.

◆ Loops

using mlir::Loops = typedef SmallVector<loop::ForOp, 8>

Performs tiling fo imperfectly nested loops (with interchange) by strip-mining the forOps by sizes and sinking them, in their order of occurrence in forOps, under each of the targets. Returns the new AffineForOps, one per each of (forOps, targets) pair, nested immediately under each of targets.

◆ NamedAttribute

using mlir::NamedAttribute = typedef std::pair<Identifier, Attribute>

NamedAttribute is used for dictionary attributes, it holds an identifier for the name and a value for the attribute. The attribute pointer should always be non-null.

◆ OpAsmSetValueNameFn

using mlir::OpAsmSetValueNameFn = typedef function_ref<void(Value, StringRef)>

A functor used to set the name of the start of a result group of an operation. See 'getAsmResultNames' below for more details.

◆ OperandAdaptor

template<typename OpTy >

using mlir::OperandAdaptor = typedef typename OpTy::OperandAdaptor

This is an adaptor from a list of values to named operands of OpTy. In a generic operation context, e.g., in dialect conversions, an ordered array of Values is treated as operands of OpTy. This adaptor takes a reference to the array and provides accessors with the same names as OpTy for operands. This makes possible to create function templates that operate on either OpTy or OperandAdaptor<OpTy> seamlessly.

◆ OperandElementTypeRange

using mlir::OperandElementTypeRange = typedef iterator_range<OperandElementTypeIterator>

◆ OwnedCubin

using mlir::OwnedCubin = typedef std::unique_ptr<std::vector<char> >

◆ PassAllocatorFunction

using mlir::PassAllocatorFunction = typedef std::function<std::unique_ptr<Pass>()>

◆ PassID

using mlir::PassID = typedef ClassID

A special type used by transformation passes to provide an address that can act as a unique identifier during pass registration.

◆ PassRegistryFunction

using mlir::PassRegistryFunction = typedef std::function<LogicalResult(OpPassManager &, StringRef options)>

A registry function that adds passes to the given pass manager. This should also parse options and return success() if parsing succeeded.

◆ PatternMatchResult

using mlir::PatternMatchResult = typedef Optional<std::unique_ptr<PatternState> >

This is the type returned by a pattern match. A match failure returns a None value. A match success returns a Some value with any state the pattern may need to maintain (but may also be null).

◆ ResultElementTypeRange

using mlir::ResultElementTypeRange = typedef iterator_range<ResultElementTypeIterator>

◆ TileLoops

using mlir::TileLoops = typedef std::pair<Loops, Loops>

◆ TransitiveFilter

using mlir::TransitiveFilter = typedef std::function<bool(Operation *)>

Type of the condition to limit the propagation of transitive use-defs. This can be used in particular to limit the propagation to a given Scope or to avoid passing through certain types of operation in a configurable manner.

◆ TranslateFromMLIRFunction

using mlir::TranslateFromMLIRFunction = typedef std::function<LogicalResult(ModuleOp, llvm::raw_ostream &output)>

Interface of the function that translates MLIR to a different format and outputs the result to a stream. It is allowed to modify the module.

◆ TranslateFunction

using mlir::TranslateFunction = typedef std::function<LogicalResult( llvm::SourceMgr &sourceMgr, llvm::raw_ostream &output, MLIRContext *)>

Interface of the function that performs file-to-file translation involving MLIR. The input file is held in the given MemoryBuffer; the output file should be written to the given raw_ostream. The implementation should create all MLIR constructs needed during the process inside the given context. This can be used for round-tripping external formats through the MLIR system.

◆ TranslateSourceMgrToMLIRFunction

using mlir::TranslateSourceMgrToMLIRFunction = typedef std::function<OwningModuleRef(llvm::SourceMgr &sourceMgr, MLIRContext *)>

Interface of the function that translates the sources managed by sourceMgr to MLIR. The source manager has at least one buffer. The implementation should create a new MLIR ModuleOp in the given context and return a pointer to it, or a nullptr in case of any error.

◆ TranslateStringRefToMLIRFunction

using mlir::TranslateStringRefToMLIRFunction = typedef std::function<OwningModuleRef(llvm::StringRef, MLIRContext *)>

Interface of the function that translates the given string to MLIR. The implementation should create a new MLIR ModuleOp in the given context. If source-related error reporting is required from within the function, use TranslateSourceMgrToMLIRFunction instead.

◆ TypeStorageAllocator

using mlir::TypeStorageAllocator = typedef StorageUniquer::StorageAllocator

◆ VectorizableLoopFun

using mlir::VectorizableLoopFun = typedef std::function<bool(AffineForOp)>

Enumeration Type Documentation

◆ AffineExprKind

enum mlir::AffineExprKind

strong

Enumerator
Add
Mul	RHS of mul is always a constant or a symbolic expression.
Mod	RHS of mod is always a constant or a symbolic expression with a positive value.
FloorDiv	RHS of floordiv is always a constant or a symbolic expression.
CeilDiv	RHS of ceildiv is always a constant or a symbolic expression.
LAST_AFFINE_BINARY_OP	This is a marker for the last affine binary op. The range of binary op's is expected to be this element and earlier.
Constant	Constant integer.
DimId	Dimensional identifier.
SymbolId	Symbolic identifier.

◆ CmpFPredicate

enum mlir::CmpFPredicate

strong

The predicate indicates the type of the comparison to perform: (un)orderedness, (in)equality and less/greater than (or equal to) as well as predicates that are always true or false.

Enumerator
FirstValidValue
AlwaysFalse
OEQ
OGT
OGE
OLT
OLE
ONE
ORD
UEQ
UGT
UGE
ULT
ULE
UNE
UNO
AlwaysTrue
NumPredicates

◆ DiagnosticSeverity

enum mlir::DiagnosticSeverity

strong

Defines the different supported severity of a diagnostic.

Enumerator
Note
Warning
Error
Remark

◆ OperationProperty

enum mlir::OperationProperty

strong

Enumerator
Commutative	This bit is set for an operation if it is a commutative operation: that is a binary operator (two inputs) where "a op b" and "b op a" produce the same results.
NoSideEffect	This bit is set for operations that have no side effects: that means that they do not read or write memory, or access any hidden state.
Terminator	This bit is set for an operation if it is a terminator: that means an operation at the end of a block.
IsolatedFromAbove	This bit is set for operations that are completely isolated from above. This is used for operations whose regions are explicit capture only, i.e. they are never allowed to implicitly reference values defined above the parent operation.

◆ PassDisplayMode

enum mlir::PassDisplayMode

strong

An enum describing the different display modes for the information within the pass manager.

Enumerator
List
Pipeline

◆ SDBMExprKind

enum mlir::SDBMExprKind

strong

Enumerator
Add
Stripe
Diff
Constant
DimId
SymbolId
Neg

Function Documentation

◆ affineDataCopyGenerate()

uint64_t mlir::affineDataCopyGenerate	(	Block::iterator	begin,
		Block::iterator	end,
		const AffineCopyOptions &	copyOptions,
		DenseSet< Operation *> &	copyNests
	)

Performs explicit copying for the contiguous sequence of operations in the block iterator range [`begin', `end'), where `end' can't be past the terminator of the block (since additional operations are potentially inserted right before end. Returns the total size of fast memory space buffers used. copyOptions provides various parameters, and the output argument copyNests is the set of all copy nests inserted, each represented by its root affine.for. Since we generate alloc's and dealloc's for all fast buffers (before and after the range of operations resp. or at a hoisted position), all of the fast memory capacity is assumed to be available for processing this block range.

Generates copies for a contiguous sequence of operations in block in the iterator range [`begin', `end'), where `end' can't be past the terminator of the block (since additional operations are potentially inserted right before `end'. Returns the total size of the fast buffers used.

◆ applyAnalysisConversion() [1/2]

LogicalResult mlir::applyAnalysisConversion	(	ArrayRef< Operation *>	ops,
		ConversionTarget &	target,
		const OwningRewritePatternList &	patterns,
		DenseSet< Operation *> &	convertedOps,
		TypeConverter *	converter = `nullptr`
	)

Apply an analysis conversion on the given operations, and all nested operations. This method analyzes which operations would be successfully converted to the target if a conversion was applied. All operations that were found to be legalizable to the given 'target' are placed within the provided 'convertedOps' set; note that no actual rewrites are applied to the operations on success and only pre-existing operations are added to the set. This method only returns failure if there are unreachable blocks in any of the regions nested within 'ops', or if a type conversion failed. If 'converter' is provided, the signatures of blocks and regions are also considered for conversion.

Apply an analysis conversion on the given operations, and all nested operations. This method analyzes which operations would be successfully converted to the target if a conversion was applied. All operations that were found to be legalizable to the given 'target' are placed within the provided 'convertedOps' set; note that no actual rewrites are applied to the operations on success and only pre-existing operations are added to the set.

◆ applyAnalysisConversion() [2/2]

LogicalResult mlir::applyAnalysisConversion	(	Operation *	op,
		ConversionTarget &	target,
		const OwningRewritePatternList &	patterns,
		DenseSet< Operation *> &	convertedOps,
		TypeConverter *	converter = `nullptr`
	)

◆ applyFullConversion() [1/2]

LogicalResult mlir::applyFullConversion	(	ArrayRef< Operation *>	ops,
		ConversionTarget &	target,
		const OwningRewritePatternList &	patterns,
		TypeConverter *	converter = `nullptr`
	)

Apply a complete conversion on the given operations, and all nested operations. This method returns failure if the conversion of any operation fails, or if there are unreachable blocks in any of the regions nested within 'ops'. If 'converter' is provided, the signatures of blocks and regions are also converted.

Apply a complete conversion on the given operations, and all nested operations. This method will return failure if the conversion of any operation fails.

◆ applyFullConversion() [2/2]

LogicalResult mlir::applyFullConversion	(	Operation *	op,
		ConversionTarget &	target,
		const OwningRewritePatternList &	patterns,
		TypeConverter *	converter = `nullptr`
	)

◆ applyPartialConversion() [1/2]

LogicalResult mlir::applyPartialConversion	(	ArrayRef< Operation *>	ops,
		ConversionTarget &	target,
		const OwningRewritePatternList &	patterns,
		TypeConverter *	converter = `nullptr`
	)

Below we define several entry points for operation conversion. It is important to note that the patterns provided to the conversion framework may have additional constraints. See the PatternRewriter Hooks section of the ConversionPatternRewriter, to see what additional constraints are imposed on the use of the PatternRewriter. Apply a partial conversion on the given operations, and all nested operations. This method converts as many operations to the target as possible, ignoring operations that failed to legalize. This method only returns failure if there are unreachable blocks in any of the regions nested within 'ops'. If 'converter' is provided, the signatures of blocks and regions are also converted.

Apply a partial conversion on the given operations, and all nested operations. This method converts as many operations to the target as possible, ignoring operations that failed to legalize.

◆ applyPartialConversion() [2/2]

LogicalResult mlir::applyPartialConversion	(	Operation *	op,
		ConversionTarget &	target,
		const OwningRewritePatternList &	patterns,
		TypeConverter *	converter = `nullptr`
	)

◆ applyPassManagerCLOptions()

void mlir::applyPassManagerCLOptions ( PassManager & pm )

Apply any values provided to the pass manager options that were registered with 'registerPassManagerOptions'.

◆ applyPatternsGreedily() [1/2]

bool mlir::applyPatternsGreedily	(	Operation *	op,
		const OwningRewritePatternList &	patterns
	)

Rewrite the regions of the specified operation, which must be isolated from above, by repeatedly applying the highest benefit patterns in a greedy work-list driven manner. Return true if no more patterns can be matched in the result operation regions. Note: This does not apply patterns to the top-level operation itself. Note: These methods also perform folding and simple dead-code elimination before attempting to match any of the provided patterns.

Rewrite the regions of the specified operation, which must be isolated from above, by repeatedly applying the highest benefit patterns in a greedy work-list driven manner. Return true if no more patterns can be matched in the result operation regions. Note: This does not apply patterns to the top-level operation itself.

◆ applyPatternsGreedily() [2/2]

bool mlir::applyPatternsGreedily	(	MutableArrayRef< Region >	regions,
		const OwningRewritePatternList &	patterns
	)

Rewrite the given regions, which must be isolated from above.

◆ areValuesDefinedAbove()

template<typename Range >

bool mlir::areValuesDefinedAbove	(	Range	values,
		Region &	limit
	)

Check if all values in the provided range are defined above the limit region. That is, if they are defined in a region that is a proper ancestor of limit.

◆ bindDims()

template<typename... AffineExprTy>

void mlir::bindDims	(	MLIRContext *	ctx,
		AffineExprTy &...	exprs
	)

Bind a list of AffineExpr references to DimExpr at positions: [0 .. sizeof...(exprs)]

◆ boundCheckLoadOrStoreOp()

template<typename LoadOrStoreOpPointer >

LogicalResult mlir::boundCheckLoadOrStoreOp	(	LoadOrStoreOpPointer	loadOrStoreOp,
		bool	emitError = `true`
	)

Checks a load or store op for an out of bound access; returns failure if the access is out of bounds along any of the dimensions, success otherwise. Emits a diagnostic error (with location information) if emitError is true.

◆ buildTripCountMapAndOperands()

void mlir::buildTripCountMapAndOperands	(	AffineForOp	forOp,
		AffineMap *	tripCountMap,
		SmallVectorImpl< Value > *	tripCountOperands
	)

Returns the trip count of the loop as an affine map with its corresponding operands if the latter is expressible as an affine expression, and nullptr otherwise. This method always succeeds as long as the lower bound is not a multi-result map. The trip count expression is simplified before returning. This method only utilizes map composition to construct lower and upper bounds before computing the trip count expressions

Returns the trip count of the loop as an affine expression if the latter is expressible as an affine expression, and nullptr otherwise. The trip count expression is simplified before returning. This method only utilizes map composition to construct lower and upper bounds before computing the trip count expressions.

◆ canFuseLoops()

FusionResult mlir::canFuseLoops	(	AffineForOp	srcForOp,
		AffineForOp	dstForOp,
		unsigned	dstLoopDepth,
		ComputationSliceState *	srcSlice
	)

Checks the feasibility of fusing the loop nest rooted at 'srcForOp' into the loop nest rooted at 'dstForOp' at 'dstLoopDepth'. Returns FusionResult 'Success' if fusion of the src/dst loop nests is feasible (i.e. they are in the same block and dependences would not be violated). Otherwise returns a FusionResult explaining why fusion is not feasible. NOTE: This function is not feature complete and should only be used in testing. TODO(andydavis) Update comments when this function is fully implemented.

◆ canonicalizeMapAndOperands()

void mlir::canonicalizeMapAndOperands	(	AffineMap *	map,
		SmallVectorImpl< Value > *	operands
	)

Modifies both map and operands in-place so as to:

drop duplicate operands
drop unused dims and symbols from map
promote valid symbols to symbolic operands in case they appeared as dimensional operands
propagate constant operands and drop them

◆ canonicalizeSetAndOperands()

void mlir::canonicalizeSetAndOperands	(	IntegerSet *	set,
		SmallVectorImpl< Value > *	operands
	)

Canonicalizes an integer set the same way canonicalizeMapAndOperands does for affine maps.

◆ canonicalizeStridedLayout()

MemRefType mlir::canonicalizeStridedLayout ( MemRefType t )

Return a version of t with identity layout if it can be determined statically that the layout is the canonical contiguous strided layout. Otherwise pass t's layout into simplifyAffineMap and return a copy of t with simplifed layout.

Return a version of t with identity layout if it can be determined statically that the layout is the canonical contiguous strided layout. Otherwise pass t's layout into simplifyAffineMap and return a copy of t with simplifed layout. If t has multiple layout maps or a multi-result layout, just return t.

◆ ceilDiv()

int64_t mlir::ceilDiv	(	int64_t	lhs,
		int64_t	rhs
	)

inline

Returns the result of MLIR's ceildiv operation on constants. The RHS is expected to be positive.

◆ checkMemrefAccessDependence()

DependenceResult mlir::checkMemrefAccessDependence	(	const MemRefAccess &	srcAccess,
		const MemRefAccess &	dstAccess,
		unsigned	loopDepth,
		FlatAffineConstraints *	dependenceConstraints,
		SmallVector< DependenceComponent, 2 > *	dependenceComponents,
		bool	allowRAR = `false`
	)

◆ coalesceLoops()

void mlir::coalesceLoops ( MutableArrayRef< loop::ForOp > loops )

Replace a perfect nest of "for" loops with a single linearized loop. Assumes loops contains a list of perfectly nested loops with bounds and steps independent of any loop induction variable involved in the nest.

◆ computeSliceUnion()

LogicalResult mlir::computeSliceUnion	(	ArrayRef< Operation *>	opsA,
		ArrayRef< Operation *>	opsB,
		unsigned	loopDepth,
		unsigned	numCommonLoops,
		bool	isBackwardSlice,
		ComputationSliceState *	sliceUnion
	)

Computes in 'sliceUnion' the union of all slice bounds computed at 'loopDepth' between all dependent pairs of ops in 'opsA' and 'opsB'. The parameter 'numCommonLoops' is the number of loops common to the operations in 'opsA' and 'opsB'. If 'isBackwardSlice' is true, computes slice bounds for loop nest surrounding ops in 'opsA', as a function of IVs and symbols of loop nest surrounding ops in 'opsB' at 'loopDepth'. If 'isBackwardSlice' is false, computes slice bounds for loop nest surrounding ops in 'opsB', as a function of IVs and symbols of loop nest surrounding ops in 'opsA' at 'loopDepth'. Returns 'success' if union was computed, 'failure' otherwise.

Computes in 'sliceUnion' the union of all slice bounds computed at 'loopDepth' between all dependent pairs of ops in 'opsA' and 'opsB'. Returns 'Success' if union was computed, 'failure' otherwise.

◆ concatAffineMaps()

AffineMap mlir::concatAffineMaps ( ArrayRef< AffineMap > maps )

Concatenates a list of maps into a single AffineMap, stepping over potentially empty maps. Assumes each of the underlying map has 0 symbols. The resulting map has a number of dims equal to the max of maps' dims and the concatenated results as its results. Returns an empty map if all input maps are empty.

Example: When applied to the following list of 3 affine maps,

{
  (i, j, k) -> (i, k),
  (i, j, k) -> (k, j),
  (i, j, k) -> (i, j)
}

Returns the map:

(i, j, k) -> (i, k, k, j, i, j)

◆ constFoldBinaryOp()

template<class AttrElementT , class ElementValueT = typename AttrElementT::ValueType, class CalculationT = function_ref<ElementValueT(ElementValueT, ElementValueT)>>

Attribute mlir::constFoldBinaryOp	(	ArrayRef< Attribute >	operands,
		const CalculationT &	calculate
	)

Performs constant folding calculate with element-wise behavior on the two attributes in operands and returns the result if possible.

◆ convertAffineLoopNestToGPULaunch()

LogicalResult mlir::convertAffineLoopNestToGPULaunch	(	AffineForOp	forOp,
		unsigned	numBlockDims,
		unsigned	numThreadDims
	)

Convert a perfect affine loop nest with the outermost loop identified by forOp into a gpu::Launch operation. Map numBlockDims outer loops to GPU blocks and numThreadDims to GPU threads. The bounds of the loops that are mapped should be independent of the induction variables of the other mapped loops.

No check on the size of the block or grid, or on the validity of parallelization is performed, it is under the responsibility of the caller to strip-mine the loops and to perform the dependence analysis before calling the conversion.

◆ convertLoopNestToGPULaunch()

LogicalResult mlir::convertLoopNestToGPULaunch	(	loop::ForOp	forOp,
		unsigned	numBlockDims,
		unsigned	numThreadDims
	)

Convert a perfect linalg loop nest with the outermost loop identified by forOp into a gpu::Launch operation. Map numBlockDims outer loops to GPU blocks and numThreadDims to GPU threads. The bounds of the loops that are mapped should be independent of the induction variables of the other mapped loops.

No check on the size of the block or grid, or on the validity of parallelization is performed, it is under the responsibility of the caller to strip-mine the loops and to perform the dependence analysis before calling the conversion.

◆ convertLoopToGPULaunch()

LogicalResult mlir::convertLoopToGPULaunch	(	loop::ForOp	forOp,
		ArrayRef< Value >	numWorkGroups,
		ArrayRef< Value >	workGroupSizes
	)

Convert a loop operation into a GPU launch with the values provided in numWorkGroups as the grid size and the values provided in workGroupSizes as the block size. Size of numWorkGroups and workGroupSizes` must be less than or equal to 3. The loop operation can be an imperfectly nested computation with the following restrictions: 1) The loop nest must contain as many perfectly nested loops as the number of values passed in through numWorkGroups. This corresponds to the number of grid dimensions of the launch. All loops within the loop nest must be parallel. 2) The body of the innermost loop of the above perfectly nested loops, must contain statements that satisfy one of the two conditions below: a) A perfect loop nest of depth greater than or equal to the number of values passed in through workGroupSizes, i.e. the number of thread dimensions of the launch. Loops at depth less than or equal to size of workGroupSizes must be parallel. Loops nested deeper can be sequential and are retained as such in the generated GPU launch code. b) Statements that are safe to be executed by all threads within the workgroup. No checks are performed that this is indeed the case. TODO(ravishankarm) : Add checks that verify 2(b) above. The above conditions are assumed to be satisfied by the computation rooted at forOp.

◆ convertToCamelCase()

std::string mlir::convertToCamelCase	(	llvm::StringRef	input,
		bool	capitalizeFirst = `false`
	)

inline

Converts a string from camel-case to snake_case by replacing all occurrences of '_' followed by a lowercase letter with the letter in uppercase. Optionally allow capitalization of the first letter (if it is a lowercase letter)

◆ convertToSnakeCase()

std::string mlir::convertToSnakeCase ( llvm::StringRef input )

inline

Converts a string to snake-case from camel-case by replacing all uppercase letters with '_' followed by the letter in lowercase, except if the uppercase letter is the first character of the string.

◆ createAffineComputationSlice()

void mlir::createAffineComputationSlice	(	Operation *	opInst,
		SmallVectorImpl< AffineApplyOp > *	sliceOps
	)

Given an operation, inserts one or more single result affine apply operations, results of which are exclusively used by this operation. The operands of these newly created affine apply ops are guaranteed to be loop iterators or terminal symbols of a function.

Before

affine.for i = 0 to #map(N) idx = affine.apply (d0) -> (d0 mod 2) (i) send A[idx], ... v = "compute"(idx, ...)

After

affine.for i = 0 to #map(N) idx = affine.apply (d0) -> (d0 mod 2) (i) send A[idx], ... idx_ = affine.apply (d0) -> (d0 mod 2) (i) v = "compute"(idx_, ...) This allows the application of different transformations on send and compute (for eg. different shifts/delays)

Fills sliceOps with the list of affine.apply operations. In the following cases, sliceOps remains empty:

If none of opInst's operands were the result of an affine.apply (i.e., there was no affine computation slice to create).
If all the affine.apply op's supplying operands to this opInst did not have any uses other than those in this opInst.

Given an operation, inserts one or more single result affine apply operations, results of which are exclusively used by this operation operation. The operands of these newly created affine apply ops are guaranteed to be loop iterators or terminal symbols of a function.

Before

affine.for i = 0 to #map(N) idx = affine.apply (d0) -> (d0 mod 2) (i) "send"(idx, A, ...) "compute"(idx)

After

affine.for i = 0 to #map(N) idx = affine.apply (d0) -> (d0 mod 2) (i) "send"(idx, A, ...) idx_ = affine.apply (d0) -> (d0 mod 2) (i) "compute"(idx_)

This allows applying different transformations on send and compute (for eg. different shifts/delays).

Returns nullptr either if none of opInst's operands were the result of an affine.apply and thus there was no affine computation slice to create, or if all the affine.apply op's supplying operands to this opInst did not have any uses besides this opInst; otherwise returns the list of affine.apply operations created in output argument sliceOps.

◆ createAffineDataCopyGenerationPass()

std::unique_ptr< OpPassBase< FuncOp > > mlir::createAffineDataCopyGenerationPass	(	unsigned	slowMemorySpace,
		unsigned	fastMemorySpace,
		unsigned	tagMemorySpace = `0`,
		int	minDmaTransferSize = `1024`,
		uint64_t	fastMemCapacityBytes = `std::numeric_limits<uint64_t>::max()`
	)

Performs packing (or explicit copying) of accessed memref regions into buffers in the specified faster memory space through either pointwise copies or DMA operations.

Generates copies for memref's living in 'slowMemorySpace' into newly created buffers in 'fastMemorySpace', and replaces memory operations to the former by the latter. Only load op's handled for now. TODO(bondhugula): extend this to store op's.

◆ createAffineLoopInvariantCodeMotionPass()

std::unique_ptr< OpPassBase< FuncOp > > mlir::createAffineLoopInvariantCodeMotionPass ( )

Creates a loop invariant code motion pass that hoists loop invariant instructions out of affine loop.

◆ createCanonicalizerPass()

std::unique_ptr< Pass > mlir::createCanonicalizerPass ( )

Creates an instance of the Canonicalizer pass.

Create a Canonicalizer pass.

◆ createComposedAffineApplyOp()

Operation* mlir::createComposedAffineApplyOp	(	OpBuilder &	builder,
		Location	loc,
		ArrayRef< Value >	operands,
		ArrayRef< Operation *>	affineApplyOps,
		SmallVectorImpl< Value > *	results
	)

Creates and inserts into 'builder' a new AffineApplyOp, with the number of its results equal to the number of operands, as a composition of all other AffineApplyOps reachable from input parameter 'operands'. If different operands were drawing results from multiple affine apply ops, these will also be collected into a single (multi-result) affine apply op. The final results of the composed AffineApplyOp are returned in output parameter 'results'. Returns the affine apply op created.

◆ createConvertGPUKernelToCubinPass()

std::unique_ptr< OpPassBase< ModuleOp > > mlir::createConvertGPUKernelToCubinPass ( CubinGenerator cubinGenerator )

Creates a pass to convert kernel functions into CUBIN blobs.

This transformation takes the body of each function that is annotated with the 'nvvm.kernel' attribute, copies it to a new LLVM module, compiles the module with help of the nvptx backend to PTX and then invokes the provided cubinGenerator to produce a binary blob (the cubin). Such blob is then attached as a string attribute named 'nvvm.cubin' to the kernel function. After the transformation, the body of the kernel function is removed (i.e., it is turned into a declaration).

◆ createConvertGpuLaunchFuncToCudaCallsPass()

std::unique_ptr< mlir::OpPassBase< mlir::ModuleOp > > mlir::createConvertGpuLaunchFuncToCudaCallsPass ( )

Creates a pass to convert a gpu.launch_func operation into a sequence of CUDA calls.

This pass does not generate code to call CUDA directly but instead uses a small wrapper library that exports a stable and conveniently typed ABI on top of CUDA.

◆ createConvertGPUToSPIRVPass()

std::unique_ptr< OpPassBase< ModuleOp > > mlir::createConvertGPUToSPIRVPass ( ArrayRef< int64_t > workGroupSize )

Pass to convert GPU Ops to SPIR-V ops. Needs the workgroup size as input since SPIR-V/Vulkan requires the workgroup size to be statically specified.

◆ createConvertStandardToSPIRVPass()

std::unique_ptr< OpPassBase< ModuleOp > > mlir::createConvertStandardToSPIRVPass ( )

Pass to convert StandardOps to SPIR-V ops.

◆ createCSEPass()

std::unique_ptr< Pass > mlir::createCSEPass ( )

Creates a pass to perform common sub expression elimination.

◆ createGpuKernelOutliningPass()

std::unique_ptr< OpPassBase< ModuleOp > > mlir::createGpuKernelOutliningPass ( )

◆ createInlinerPass()

std::unique_ptr< Pass > mlir::createInlinerPass ( )

Creates a pass which inlines calls and callable operations as defined by the CallGraph.

◆ createLegalizeStdOpsForSPIRVLoweringPass()

std::unique_ptr< Pass > mlir::createLegalizeStdOpsForSPIRVLoweringPass ( )

Pass to legalize ops that are not directly lowered to SPIR-V.

◆ createLoopCoalescingPass()

std::unique_ptr< OpPassBase< FuncOp > > mlir::createLoopCoalescingPass ( )

Creates a pass that transforms perfectly nested loops with independent bounds into a single loop.

◆ createLoopFusionPass()

std::unique_ptr< OpPassBase< FuncOp > > mlir::createLoopFusionPass	(	unsigned	fastMemorySpace = `0`,
		uint64_t	localBufSizeThreshold = `0`,
		bool	maximalFusion = `false`
	)

Creates a loop fusion pass which fuses loops. Buffers of size less than or equal to localBufSizeThreshold are promoted to memory space `fastMemorySpace'.

◆ createLoopInvariantCodeMotionPass()

std::unique_ptr< Pass > mlir::createLoopInvariantCodeMotionPass ( )

Creates a loop invariant code motion pass that hoists loop invariant instructions out of the loop.

◆ createLoopTilingPass()

std::unique_ptr< OpPassBase< FuncOp > > mlir::createLoopTilingPass ( uint64_t cacheSizeBytes )

Creates a pass to perform tiling on loop nests.

Creates a pass to perform loop tiling on all suitable loop nests of a Function.

◆ createLoopToGPUPass()

std::unique_ptr< OpPassBase< FuncOp > > mlir::createLoopToGPUPass	(	ArrayRef< int64_t >	numWorkGroups,
		ArrayRef< int64_t >	workGroupSize
	)

Create a pass that converts every loop operation within the body of the FuncOp into a GPU launch. The number of workgroups and workgroup size for the implementation is controlled by SSA values passed into conversion method. For testing, the values are set as constants obtained from a command line flag. See convertLoopToGPULaunch for a description of the required semantics of the converted loop operation.

◆ createLoopUnrollAndJamPass()

std::unique_ptr< OpPassBase< FuncOp > > mlir::createLoopUnrollAndJamPass ( int unrollJamFactor = -1 )

Creates a loop unroll jam pass to unroll jam by the specified factor. A factor of -1 lets the pass use the default factor or the one on the command line if provided.

◆ createLoopUnrollPass()

std::unique_ptr< OpPassBase< FuncOp > > mlir::createLoopUnrollPass	(	int	unrollFactor = `-1`,
		int	unrollFull = `-1`,
		const std::function< unsigned(AffineForOp)> &	getUnrollFactor = `nullptr`
	)

Creates a loop unrolling pass with the provided parameters. 'getUnrollFactor' is a function callback for clients to supply a function that computes an unroll factor - the callback takes precedence over unroll factors supplied through other means. If -1 is passed as the unrollFactor and no callback is provided, anything passed from the command-line (if at all) or the default unroll factor is used (LoopUnroll:kDefaultUnrollFactor).

◆ createLowerAffinePass()

std::unique_ptr< OpPassBase< FuncOp > > mlir::createLowerAffinePass ( )

Lowers affine control flow operations (ForStmt, IfStmt and AffineApplyOp) to equivalent lower-level constructs (flow of basic blocks and arithmetic primitives).

Lowers If and For operations within a function into their lower level CFG equivalent blocks.

◆ createLowerGpuOpsToNVVMOpsPass()

std::unique_ptr< OpPassBase< ModuleOp > > mlir::createLowerGpuOpsToNVVMOpsPass ( )

Creates a pass that lowers GPU dialect operations to NVVM counterparts.

◆ createLowerGpuOpsToROCDLOpsPass()

std::unique_ptr< OpPassBase< ModuleOp > > mlir::createLowerGpuOpsToROCDLOpsPass ( )

Creates a pass that lowers GPU dialect operations to ROCDL counterparts.

◆ createLowerToCFGPass()

std::unique_ptr< Pass > mlir::createLowerToCFGPass ( )

Creates a pass to convert loop.for, loop.if and loop.terminator ops to CFG.

◆ createLowerToLLVMPass() [1/3]

std::unique_ptr< OpPassBase< ModuleOp > > mlir::createLowerToLLVMPass ( bool useAlloca = false )

Creates a pass to convert the Standard dialect into the LLVMIR dialect. By default stdlib malloc/free are used for allocating MemRef payloads. Specifying useAlloca-true emits stack allocations instead. In the future this may become an enum when we have concrete uses for other options.

◆ createLowerToLLVMPass() [2/3]

std::unique_ptr< OpPassBase< ModuleOp > > mlir::createLowerToLLVMPass	(	LLVMPatternListFiller	patternListFiller,
		LLVMTypeConverterMaker	typeConverterMaker,
		bool	useAlloca = `false`
	)

Creates a pass to convert operations to the LLVMIR dialect. The conversion is defined by a list of patterns and a type converter that will be obtained during the pass using the provided callbacks. By default stdlib malloc/free are used for allocating MemRef payloads. Specifying useAlloca-true emits stack allocations instead. In the future this may become an enum when we have concrete uses for other options.

◆ createLowerToLLVMPass() [3/3]

template<typename TypeConverter = LLVMTypeConverter>

std::unique_ptr<OpPassBase<ModuleOp> > mlir::createLowerToLLVMPass	(	LLVMPatternListFiller	patternListFiller,
		bool	useAlloca = `false`
	)

Creates a pass to convert operations to the LLVMIR dialect. The conversion is defined by a list of patterns obtained during the pass using the provided callback and an optional type conversion class, an instance is created during the pass. By default stdlib malloc/free are used for allocating MemRef payloads. Specifying useAlloca-true emits stack allocations instead. In the future this may become an enum when we have concrete uses for other options.

◆ createLowerVectorToLLVMPass()

OpPassBase< ModuleOp > * mlir::createLowerVectorToLLVMPass ( )

Create a pass to convert vector operations to the LLVMIR dialect.

◆ createLowerVectorToLoopsPass()

OpPassBase<ModuleOp>* mlir::createLowerVectorToLoopsPass ( )

Create a pass to convert vector operations to affine loops + std dialect.

◆ createMaterializeVectorsPass()

std::unique_ptr<OpPassBase<FuncOp> > mlir::createMaterializeVectorsPass ( ArrayRef< int64_t > vectorSize )

Creates a pass to lower super-vectors to target-dependent HW vectors.

◆ createMemRefBoundCheckPass()

std::unique_ptr< OpPassBase< FuncOp > > mlir::createMemRefBoundCheckPass ( )

Creates a pass to check memref accesses in a Function.

◆ createMemRefDataFlowOptPass()

std::unique_ptr< OpPassBase< FuncOp > > mlir::createMemRefDataFlowOptPass ( )

Creates a pass to perform optimizations relying on memref dataflow such as store to load forwarding, elimination of dead stores, and dead allocs.

◆ createParallelismDetectionTestPass()

std::unique_ptr< OpPassBase< FuncOp > > mlir::createParallelismDetectionTestPass ( )

Creates a pass to test parallelism detection; emits note for parallel loops.

◆ createPipelineDataTransferPass()

std::unique_ptr< OpPassBase< FuncOp > > mlir::createPipelineDataTransferPass ( )

Creates a pass to pipeline explicit movement of data across levels of the memory hierarchy.

◆ createPrintCFGGraphPass()

std::unique_ptr< mlir::OpPassBase< mlir::FuncOp > > mlir::createPrintCFGGraphPass	(	raw_ostream &	os = `llvm::errs()`,
		bool	shortNames = `false`,
		const Twine &	title = `""`
	)

Creates a pass to print CFG graphs.

◆ createPrintOpGraphPass()

std::unique_ptr< OpPassBase< ModuleOp > > mlir::createPrintOpGraphPass	(	raw_ostream &	os = `llvm::errs()`,
		bool	shortNames = `false`,
		const Twine &	title = `""`
	)

Creates a pass to print op graphs.

◆ createSimpleLoopsToGPUPass()

std::unique_ptr< OpPassBase< FuncOp > > mlir::createSimpleLoopsToGPUPass	(	unsigned	numBlockDims,
		unsigned	numThreadDims
	)

Create a pass that converts loop nests into GPU kernels. It considers top-level affine.for and linalg.for operations as roots of loop nests and converts them to the gpu.launch operations if possible.

No check on the size of the block or grid, or on the validity of parallelization is performed, it is under the responsibility of the caller to strip-mine the loops and to perform the dependence analysis before calling the conversion.

◆ createSimpleParametricTilingPass()

std::unique_ptr<OpPassBase<FuncOp> > mlir::createSimpleParametricTilingPass ( ArrayRef< int64_t > outerLoopSizes )

Creates a pass that performs parametric tiling so that the outermost loops have the given fixed number of iterations. Assumes outermost loop nests are permutable.

◆ createSimplifyAffineStructuresPass()

std::unique_ptr< OpPassBase< FuncOp > > mlir::createSimplifyAffineStructuresPass ( )

Creates a simplification pass for affine structures (maps and sets). In addition, this pass also normalizes memrefs to have the trivial (identity) layout map.

◆ createStripDebugInfoPass()

std::unique_ptr< OpPassBase< FuncOp > > mlir::createStripDebugInfoPass ( )

Creates a pass to strip debug information from a function.

◆ createTestLoopFusionPass()

std::unique_ptr<OpPassBase<FuncOp> > mlir::createTestLoopFusionPass ( )

Creates a pass which tests loop fusion utilities.

◆ createTestMemRefDependenceCheckPass()

std::unique_ptr< OpPassBase< FuncOp > > mlir::createTestMemRefDependenceCheckPass ( )

Creates a pass to check memref access dependences in a Function.

◆ createVectorizePass()

std::unique_ptr< OpPassBase< FuncOp > > mlir::createVectorizePass ( ArrayRef< int64_t > virtualVectorSize )

Creates a pass to vectorize loops, operations and data types using a target-independent, n-D super-vector abstraction.

◆ createVectorizerTestPass()

std::unique_ptr<OpPassBase<FuncOp> > mlir::createVectorizerTestPass ( )

Creates a pass to allow independent testing of vectorizer functionality with FileCheck.

◆ defaultFilterFunction()

bool mlir::defaultFilterFunction ( Operation & )

inline

◆ emitError() [1/2]

InFlightDiagnostic mlir::emitError ( Location loc )

Utility method to emit an error message using this location.

Emit an error message using this location.

◆ emitError() [2/2]

InFlightDiagnostic mlir::emitError	(	Location	loc,
		const Twine &	message
	)

◆ emitOptionalError()

template<typename... Args>

LogicalResult mlir::emitOptionalError	(	Optional< Location >	loc,
		Args &&...	args
	)

Overloads of the above emission functions that take an optionally null location. If the location is null, no diagnostic is emitted and a failure is returned. Given that the provided location may be null, these methods take the diagnostic arguments directly instead of relying on the returned InFlightDiagnostic.

◆ emitOptionalRemark()

template<typename... Args>

LogicalResult mlir::emitOptionalRemark	(	Optional< Location >	loc,
		Args &&...	args
	)

◆ emitOptionalWarning()

template<typename... Args>

LogicalResult mlir::emitOptionalWarning	(	Optional< Location >	loc,
		Args &&...	args
	)

◆ emitRemark() [1/2]

InFlightDiagnostic mlir::emitRemark ( Location loc )

Utility method to emit a remark message using this location.

Emit a remark message using this location.

◆ emitRemark() [2/2]

InFlightDiagnostic mlir::emitRemark	(	Location	loc,
		const Twine &	message
	)

◆ emitWarning() [1/2]

InFlightDiagnostic mlir::emitWarning ( Location loc )

Utility method to emit a warning message using this location.

Emit a warning message using this location.

◆ emitWarning() [2/2]

InFlightDiagnostic mlir::emitWarning	(	Location	loc,
		const Twine &	message
	)

◆ expandAffineExpr()

mlir::Value mlir::expandAffineExpr	(	OpBuilder &	builder,
		Location	loc,
		AffineExpr	expr,
		ArrayRef< Value >	dimValues,
		ArrayRef< Value >	symbolValues
	)

Emit code that computes the given affine expression using standard arithmetic operations applied to the provided dimension and symbol values.

◆ extractFixedOuterLoops()

TileLoops mlir::extractFixedOuterLoops	(	loop::ForOp	rootFOrOp,
		ArrayRef< int64_t >	sizes
	)

Tile a nest of standard for loops rooted at rootForOp by finding such parametric tile sizes that the outer loops have a fixed number of iterations as defined in sizes.

◆ extractForInductionVars()

void mlir::extractForInductionVars	(	ArrayRef< AffineForOp >	forInsts,
		SmallVectorImpl< Value > *	ivs
	)

Extracts the induction variables from a list of AffineForOps and places them in the output argument ivs.

Extracts the induction variables from a list of AffineForOps and returns them.

◆ failed()

bool mlir::failed ( LogicalResult result )

inline

Utility function that returns true if the provided LogicalResult corresponds to a failure value.

◆ failure()

LogicalResult mlir::failure ( bool isFailure = true )

inline

Utility function to generate a LogicalResult. If isFailure is true a failure result is generated, otherwise a 'success' result is generated.

◆ floorDiv()

int64_t mlir::floorDiv	(	int64_t	lhs,
		int64_t	rhs
	)

inline

Returns the result of MLIR's floordiv operation on constants. The RHS is expected to be positive.

◆ fullyComposeAffineMapAndOperands()

void mlir::fullyComposeAffineMapAndOperands	(	AffineMap *	map,
		SmallVectorImpl< Value > *	operands
	)

Given an affine map map and its input operands, this method composes into map, maps of AffineApplyOps whose results are the values in operands, iteratively until no more of operands are the result of an AffineApplyOp. When this function returns, map becomes the composed affine map, and each Value in operands is guaranteed to be either a loop IV or a terminal symbol, i.e., a symbol defined at the top level or a block/function argument.

◆ getAffineBinaryOpExpr()

AffineExpr mlir::getAffineBinaryOpExpr	(	AffineExprKind	kind,
		AffineExpr	lhs,
		AffineExpr	rhs
	)

◆ getAffineConstantExpr()

AffineExpr mlir::getAffineConstantExpr	(	int64_t	constant,
		MLIRContext *	context
	)

◆ getAffineDimExpr()

AffineExpr mlir::getAffineDimExpr	(	unsigned	position,
		MLIRContext *	context
	)

These free functions allow clients of the API to not use classes in detail.

◆ getAffineSymbolExpr()

AffineExpr mlir::getAffineSymbolExpr	(	unsigned	position,
		MLIRContext *	context
	)

◆ getBackwardSlice()

void mlir::getBackwardSlice	(	Operation *	op,
		llvm::SetVector< Operation >	backwardSlice,
		TransitiveFilter	filter = `[](Operation *) { return true;}`
	)

Fills backwardSlice with the computed backward slice (i.e. all the transitive defs of op), without including that operation.

This additionally takes a TransitiveFilter which acts as a frontier: when looking at defs transitively, a operation that does not pass the filter is never propagated through. This allows in particular to carve out the scope within a ForInst or the scope within an IfInst.

The implementation traverses the def chains in postorder traversal for efficiency reasons: if a operation is already in backwardSlice, no need to traverse its definitions again. Since useuse-def chains form a DAG, this terminates.

Upon return to the root call, backwardSlice is filled with a postorder list of defs. This happens to be a topological order, from the point of view of the use-def chains.

Example starting from node 8

1 2 3 4 |_______| |______| | | | | 5 6 |___|_____________| | | 7 8 |_______________| | 9

Assuming all local orders match the numbering order: {1, 2, 5, 3, 4, 6}

◆ getCleanupLoopLowerBound()

void mlir::getCleanupLoopLowerBound	(	AffineForOp	forOp,
		unsigned	unrollFactor,
		AffineMap *	map,
		SmallVectorImpl< Value > *	operands,
		OpBuilder &	b
	)

Computes the cleanup loop lower bound of the loop being unrolled with the specified unroll factor; this bound will also be upper bound of the main part of the unrolled loop. Computes the bound as an AffineMap with its operands or a null map when the trip count can't be expressed as an affine expression.

◆ getComputationSliceState()

void mlir::getComputationSliceState	(	Operation *	depSourceOp,
		Operation *	depSinkOp,
		FlatAffineConstraints *	dependenceConstraints,
		unsigned	loopDepth,
		bool	isBackwardSlice,
		ComputationSliceState *	sliceState
	)

Computes the computation slice loop bounds for one loop nest as affine maps of the other loop nest's IVs and symbols, using 'dependenceConstraints' computed between 'depSourceAccess' and 'depSinkAccess'. If 'isBackwardSlice' is true, a backwards slice is computed in which the slice bounds of loop nest surrounding 'depSourceAccess' are computed in terms of loop IVs and symbols of the loop nest surrounding 'depSinkAccess' at 'loopDepth'. If 'isBackwardSlice' is false, a forward slice is computed in which the slice bounds of loop nest surrounding 'depSinkAccess' are computed in terms of loop IVs and symbols of the loop nest surrounding 'depSourceAccess' at 'loopDepth'. The slice loop bounds and associated operands are returned in 'sliceState'.

◆ getComputeCost()

int64_t mlir::getComputeCost	(	AffineForOp	forOp,
		LoopNestStats &	stats
	)

Computes the total cost of the loop nest rooted at 'forOp' using 'stats'. Currently, the total cost is computed by counting the total operation instance count (i.e. total number of operations in the loop body * loop trip count) for the entire loop nest.

◆ getConstantTripCount()

Optional< uint64_t > mlir::getConstantTripCount ( AffineForOp forOp )

Returns the trip count of the loop if it's a constant, None otherwise. This uses affine expression analysis and is able to determine constant trip count in non-trivial cases.

Returns the trip count of the loop if it's a constant, None otherwise. This method uses affine expression analysis (in turn using getTripCount) and is able to determine constant trip count in non-trivial cases.

◆ getDependenceComponents()

void mlir::getDependenceComponents	(	AffineForOp	forOp,
		unsigned	maxLoopDepth,
		std::vector< SmallVector< DependenceComponent, 2 >> *	depCompsVec
	)

Returns in 'depCompsVec', dependence components for dependences between all load and store ops in loop nest rooted at 'forOp', at loop depths in range [1, maxLoopDepth].

Gathers dependence components for dependences between all ops in loop nest rooted at 'forOp' at loop depths in range [1, maxLoopDepth].

◆ getElementTypeOrSelf() [1/3]

Type mlir::getElementTypeOrSelf ( Type type )

Return the element type or return the type itself.

◆ getElementTypeOrSelf() [2/3]

Type mlir::getElementTypeOrSelf ( Attribute attr )

Return the element type or return the type itself.

◆ getElementTypeOrSelf() [3/3]

Type mlir::getElementTypeOrSelf ( Value val )

◆ getFlattenedAffineExpr() [1/2]

bool mlir::getFlattenedAffineExpr	(	AffineExpr	expr,
		unsigned	numDims,
		unsigned	numSymbols,
		SmallVectorImpl< int64_t > *	flattenedExpr
	)

Flattens 'expr' into 'flattenedExpr'. Returns true on success or false if 'expr' could not be flattened (i.e., semi-affine is not yet handled). See documentation for AffineExprFlattener on how mod's and div's are flattened.

◆ getFlattenedAffineExpr() [2/2]

LogicalResult mlir::getFlattenedAffineExpr	(	AffineExpr	expr,
		unsigned	numDims,
		unsigned	numSymbols,
		SmallVectorImpl< int64_t > *	flattenedExpr,
		FlatAffineConstraints *	cst = `nullptr`
	)

Flattens 'expr' into 'flattenedExpr', which contains the coefficients of the dimensions, symbols, and additional variables that represent floor divisions of dimensions, symbols, and in turn other floor divisions. Returns failure if 'expr' could not be flattened (i.e., semi-affine is not yet handled). 'cst' contains constraints that connect newly introduced local identifiers to existing dimensional and symbolic identifiers. See documentation for AffineExprFlattener on how mod's and div's are flattened.

◆ getFlattenedAffineExprs() [1/4]

bool mlir::getFlattenedAffineExprs	(	AffineMap	map,
		std::vector< SmallVector< int64_t, 8 >> *	flattenedExprs
	)

Flattens the result expressions of the map to their corresponding flattened forms and set in 'flattenedExprs'. Returns true on success or false if any expression in the map could not be flattened (i.e., semi-affine is not yet handled). For all affine expressions that share the same operands (like those of an affine map), this method should be used instead of repeatedly calling getFlattenedAffineExpr since local variables added to deal with div's and mod's will be reused across expressions.

Flattens the expressions in map. Returns true on success or false if 'expr' was unable to be flattened (i.e., semi-affine expressions not handled yet).

◆ getFlattenedAffineExprs() [2/4]

bool mlir::getFlattenedAffineExprs	(	IntegerSet	set,
		std::vector< SmallVector< int64_t, 8 >> *	flattenedExprs
	)

◆ getFlattenedAffineExprs() [3/4]

LogicalResult mlir::getFlattenedAffineExprs	(	AffineMap	map,
		std::vector< SmallVector< int64_t, 8 >> *	flattenedExprs,
		FlatAffineConstraints *	localVarCst = `nullptr`
	)

Flattens the result expressions of the map to their corresponding flattened forms and set in 'flattenedExprs'. Returns failure if any expression in the map could not be flattened (i.e., semi-affine is not yet handled). 'cst' contains constraints that connect newly introduced local identifiers to existing dimensional and / symbolic identifiers. See documentation for AffineExprFlattener on how mod's and div's are flattened. For all affine expressions that share the same operands (like those of an affine map), this method should be used instead of repeatedly calling getFlattenedAffineExpr since local variables added to deal with div's and mod's will be reused across expressions.

Flattens the expressions in map. Returns failure if 'expr' was unable to be flattened (i.e., semi-affine expressions not handled yet).

◆ getFlattenedAffineExprs() [4/4]

LogicalResult mlir::getFlattenedAffineExprs	(	IntegerSet	set,
		std::vector< SmallVector< int64_t, 8 >> *	flattenedExprs,
		FlatAffineConstraints *	cst = `nullptr`
	)

◆ getFlattenedTypes()

SmallVector< Type, 10 > mlir::getFlattenedTypes ( TupleType t )

Get the types within a nested Tuple. A helper for the class method that handles storage concerns, which is tricky to do in tablegen.

◆ getForInductionVarOwner()

ForOp mlir::loop::getForInductionVarOwner ( Value val )

Returns the loop parent of an induction variable. If the provided value is not an induction variable, then return nullptr.

◆ getForwardSlice()

void mlir::getForwardSlice	(	Operation *	op,
		llvm::SetVector< Operation >	forwardSlice,
		TransitiveFilter	filter = `[](Operation *) { return true;}`
	)

Fills forwardSlice with the computed forward slice (i.e. all the transitive uses of op), without including that operation.

This additionally takes a TransitiveFilter which acts as a frontier: when looking at uses transitively, a operation that does not pass the filter is never propagated through. This allows in particular to carve out the scope within a ForInst or the scope within an IfInst.

The implementation traverses the use chains in postorder traversal for efficiency reasons: if a operation is already in forwardSlice, no need to traverse its uses again. Since use-def chains form a DAG, this terminates.

Upon return to the root call, forwardSlice is filled with a postorder list of uses (i.e. a reverse topological order). To get a proper topological order, we just just reverse the order in forwardSlice before returning.

Example starting from node 0

___________|___________ 1 2 3 4 |_______| |______| | | | | 5 6 |___|_____________| | | 7 8 |_______________| | 9

Assuming all local orders match the numbering order:

after getting back to the root getForwardSlice, forwardSlice may contain: {9, 7, 8, 5, 1, 2, 6, 3, 4}
reversing the result of 1. gives: {4, 3, 6, 2, 1, 5, 8, 7, 9}

◆ getFusionComputeCost()

bool mlir::getFusionComputeCost	(	AffineForOp	srcForOp,
		LoopNestStats &	srcStats,
		AffineForOp	dstForOp,
		LoopNestStats &	dstStats,
		ComputationSliceState *	slice,
		int64_t *	computeCost
	)

Computes and returns in 'computeCost', the total compute cost of fusing the 'slice' of the loop nest rooted at 'srcForOp' into 'dstForOp'. Currently, the total cost is computed by counting the total operation instance count (i.e. total number of operations in the loop body * loop trip count) for the entire loop nest. Returns true on success, failure otherwise (e.g. non-constant trip counts).

Computes and returns in 'computeCost', the total compute cost of fusing the 'slice' of the loop nest rooted at 'srcForOp' into 'dstForOp'. Currently, the total cost is computed by counting the total operation instance count (i.e. total number of operations in the loop body * loop trip count) for the entire loop nest.

◆ getIndexSet()

LogicalResult mlir::getIndexSet	(	MutableArrayRef< AffineForOp >	forOps,
		FlatAffineConstraints *	domain
	)

Builds a system of constraints with dimensional identifiers corresponding to the loop IVs of the forOps appearing in that order. Bounds of the loop are used to add appropriate inequalities. Any symbols founds in the bound operands are added as symbols in the system. Returns failure for the yet unimplemented cases.

◆ getInvariantAccesses()

DenseSet< Value > mlir::getInvariantAccesses	(	Value	iv,
		ArrayRef< Value >	indices
	)

Given an induction variable iv of type AffineForOp and indices of type IndexType, returns the set of indices that are independent of iv.

Prerequisites (inherited from isAccessInvariant above):

iv and indices of the proper type;
at most one affine.apply is reachable from each index in indices;

Emits a note if it encounters a chain of affine.apply and conservatively those cases.

◆ getLargestDivisorOfTripCount()

uint64_t mlir::getLargestDivisorOfTripCount ( AffineForOp forOp )

Returns the greatest known integral divisor of the trip count. Affine expression analysis is used (indirectly through getTripCount), and this method is thus able to determine non-trivial divisors.

◆ getLoopIVs()

void mlir::getLoopIVs	(	Operation &	op,
		SmallVectorImpl< AffineForOp > *	loops
	)

Populates 'loops' with IVs of the loops surrounding 'op' ordered from the outermost 'affine.for' operation to the innermost one.

◆ getLoopNestStats()

bool mlir::getLoopNestStats	(	AffineForOp	forOpRoot,
		LoopNestStats *	stats
	)

Collect loop nest statistics (eg. loop trip count and operation count) in 'stats' for loop nest rooted at 'forOp'. Returns true on success, returns false otherwise.

◆ getMemoryFootprintBytes()

Optional< int64_t > mlir::getMemoryFootprintBytes	(	AffineForOp	forOp,
		int	memorySpace = `-1`
	)

Gets the memory footprint of all data touched in the specified memory space in bytes; if the memory space is unspecified, considers all memory spaces.

◆ getMemRefSizeInBytes()

Optional< uint64_t > mlir::getMemRefSizeInBytes ( MemRefType memRefType )

Returns the size of memref data in bytes if it's statically shaped, None otherwise.

Returns the size of memref data in bytes if it's statically shaped, None otherwise. If the element of the memref has vector type, takes into account size of the vector as well.

◆ getNestingDepth()

unsigned mlir::getNestingDepth ( Operation & op )

Returns the nesting depth of this operation, i.e., the number of loops surrounding this operation.

Returns the nesting depth of this statement, i.e., the number of loops surrounding this statement.

◆ getNumCommonSurroundingLoops()

unsigned mlir::getNumCommonSurroundingLoops	(	Operation &	A,
		Operation &	B
	)

Returns the number of surrounding loops common to both A and B.

Returns the number of surrounding loops common to 'loopsA' and 'loopsB', where each lists loops from outer-most to inner-most in loop nest.

◆ getNumIterators() [1/2]

unsigned mlir::getNumIterators	(	StringRef	name,
		ArrayAttr	iteratorTypes
	)

inline

Returns the iterator of a certain type.

◆ getNumIterators() [2/2]

unsigned mlir::getNumIterators ( ArrayAttr iteratorTypes )

inline

◆ getPerfectlyNestedLoops() [1/2]

void mlir::getPerfectlyNestedLoops	(	SmallVectorImpl< AffineForOp > &	nestedLoops,
		AffineForOp	root
	)

Get perfectly nested sequence of loops starting at root of loop nest (the first op being another AffineFor, and the second op - a terminator). A loop is perfectly nested iff: the first op in the loop's body is another AffineForOp, and the second op is a terminator).

◆ getPerfectlyNestedLoops() [2/2]

void mlir::getPerfectlyNestedLoops	(	SmallVectorImpl< loop::ForOp > &	nestedLoops,
		loop::ForOp	root
	)

◆ getReachableAffineApplyOps()

void mlir::getReachableAffineApplyOps	(	ArrayRef< Value >	operands,
		SmallVectorImpl< Operation *> &	affineApplyOps
	)

Returns in affineApplyOps, the sequence of those AffineApplyOp Operations that are reachable via a search starting from operands and ending at those operands that are not the result of an AffineApplyOp.

Returns the sequence of AffineApplyOp Operations operation in 'affineApplyOps', which are reachable via a search starting from 'operands', and ending at operands which are not defined by AffineApplyOps.

◆ getSequentialLoops()

void mlir::getSequentialLoops	(	AffineForOp	forOp,
		llvm::SmallDenseSet< Value, 8 > *	sequentialLoops
	)

Returns in 'sequentialLoops' all sequential loops in loop nest rooted at 'forOp'.

◆ getSlice()

SetVector< Operation * > mlir::getSlice	(	Operation *	op,
		TransitiveFilter	backwardFilter = `[](Operation *) { return true; }`,
		TransitiveFilter	forwardFilter = `[](Operation *) { return true; }`
	)

Iteratively computes backward slices and forward slices until a fixed point is reached. Returns an llvm::SetVector<Operation *> which includes the original operation.

This allows building a slice (i.e. multi-root DAG where everything that is reachable from an Value in forward and backward direction is contained in the slice). This is the abstraction we need to materialize all the operations for supervectorization without worrying about orderings and Value replacements.

Example starting from any node

1 2 3 4 |_______| |______| | | | | | 5 6___| |___|_____________| | | | | 7 8 | |_______________| | | | 9 10

Return the whole DAG in some topological order.

The implementation works by just filling up a worklist with iterative alternate calls to getBackwardSlice and getForwardSlice.

The following section describes some additional implementation considerations for a potentially more efficient implementation but they are just an intuition without proof, we still use a worklist for now.

Additional implementation considerations

Consider the defs-op-uses hourglass.

\ / defs (in some topological order) \/ op /\ / \ uses (in some topological order) /____\

We want to iteratively apply getSlice to construct the whole list of Operation that are reachable by (use|def)+ from op. We want the resulting slice in topological order. Ideally we would like the ordering to be maintained in-place to avoid copying Operation at each step. Keeping this ordering by construction seems very unclear, so we list invariants in the hope of seeing whether useful properties pop up.

In the following: we use |= for set inclusion; we use << for set topological ordering (i.e. each pair is ordered).

Assumption:

We wish to maintain the following property by a recursive argument: """ defs << {op} <<uses are in topological order. """ The property clearly holds for 0 and 1-sized uses and defs;

Invariants:

defs and uses are in topological order internally, by construction;
for any {x} |= defs, defs(x) |= defs; because all go through op
for any {x} |= uses, defs |= defs(x); because all go through op
for any {x} |= defs, uses |= uses(x); because all go through op
for any {x} |= uses, uses(x) |= uses; because all go through op

Intuitively, we should be able to recurse like: preorder(defs) - op - postorder(uses) and keep things ordered but this is still hand-wavy and not worth the trouble for now: punt to a simple worklist-based solution.

◆ getStridesAndOffset() [1/2]

LogicalResult mlir::getStridesAndOffset	(	MemRefType	t,
		SmallVectorImpl< int64_t > &	strides,
		int64_t &	offset
	)

Returns the strides of the MemRef if the layout map is in strided form. MemRefs with layout maps in strided form include:

empty or identity layout map, in which case the stride information is the canonical form computed from sizes;
single affine map layout of the form K + k0 * d0 + ... kn * dn, where K and ki's are constants or symbols.

A stride specification is a list of integer values that are either static or dynamic (encoded with getDynamicStrideOrOffset()). Strides encode the distance in the number of elements between successive entries along a particular dimension. For example, memref<42x16xf32, (64 * d0 + d1)> specifies a view into a non-contiguous memory region of 42 by 16 f32 elements in which the distance between two consecutive elements along the outer dimension is 1 and the distance between two consecutive elements along the inner dimension is 64.

If a simple strided form cannot be extracted from the composition of the layout map, returns llvm::None.

The convention is that the strides for dimensions d0, .. dn appear in order to make indexing intuitive into the result.

◆ getStridesAndOffset() [2/2]

LogicalResult mlir::getStridesAndOffset	(	MemRefType	t,
		SmallVectorImpl< AffineExpr > &	strides,
		AffineExpr &	offset
	)

In practice, a strided memref must be internally non-aliasing. Test against 0 as a proxy. TODO(ntv) static cases can have more advanced checks. TODO(ntv) dynamic cases would require a way to compare symbolic expressions and would probably need an affine set context propagated everywhere.

◆ getTranslationFromMLIRRegistry()

const llvm::StringMap< TranslateFromMLIRFunction > & mlir::getTranslationFromMLIRRegistry ( )

◆ getTranslationRegistry()

const llvm::StringMap< TranslateFunction > & mlir::getTranslationRegistry ( )

◆ getTranslationToMLIRRegistry()

const llvm::StringMap< TranslateSourceMgrToMLIRFunction > & mlir::getTranslationToMLIRRegistry ( )

Get a read-only reference to the translator registry.

◆ getUsedValuesDefinedAbove() [1/2]

void mlir::getUsedValuesDefinedAbove	(	Region &	region,
		Region &	limit,
		llvm::SetVector< Value > &	values
	)

Fill values with a list of values defined at the ancestors of the limit region and used within region or its descendants.

◆ getUsedValuesDefinedAbove() [2/2]

void mlir::getUsedValuesDefinedAbove	(	MutableArrayRef< Region >	regions,
		llvm::SetVector< Value > &	values
	)

Fill values with a list of values used within any of the regions provided but defined in one of the ancestors.

◆ has_single_element()

template<typename ContainerTy >

bool mlir::has_single_element ( ContainerTy && c )

Returns true of the given range only contains a single element.

◆ hasDependence()

bool mlir::hasDependence ( DependenceResult result )

inline

Utility function that returns true if the provided DependenceResult corresponds to a dependence result.

◆ hash_value() [1/9]

llvm::hash_code mlir::hash_value ( Identifier arg )

inline

◆ hash_value() [2/9]

inline ::llvm::hash_code mlir::hash_value ( IntegerSet arg )

◆ hash_value() [3/9]

inline ::llvm::hash_code mlir::hash_value ( AffineMap arg )

◆ hash_value() [4/9]

inline ::llvm::hash_code mlir::hash_value ( AffineExpr arg )

Make AffineExpr hashable.

◆ hash_value() [5/9]

llvm::hash_code mlir::hash_value ( OperationName arg )

inline

◆ hash_value() [6/9]

inline ::llvm::hash_code mlir::hash_value ( Type arg )

◆ hash_value() [7/9]

inline ::llvm::hash_code mlir::hash_value ( Location arg )

◆ hash_value() [8/9]

inline ::llvm::hash_code mlir::hash_value ( Value arg )

Make Value hashable.

◆ hash_value() [9/9]

inline ::llvm::hash_code mlir::hash_value ( Attribute arg )

◆ initializeLLVMPasses()

void mlir::initializeLLVMPasses ( )

Initialize LLVM passes that can be when running MLIR code using ExecutionEngine.

◆ inlineCall()

LogicalResult mlir::inlineCall	(	InlinerInterface &	interface,
		CallOpInterface	call,
		CallableOpInterface	callable,
		Region *	src,
		bool	shouldCloneInlinedRegion = `true`
	)

This function inlines a given region, 'src', of a callable operation, 'callable', into the location defined by the given call operation. This function returns failure if inlining is not possible, success otherwise. On failure, no changes are made to the module. 'shouldCloneInlinedRegion' corresponds to whether the source region should be cloned into the 'call' or spliced directly.

◆ inlineRegion() [1/2]

LogicalResult mlir::inlineRegion	(	InlinerInterface &	interface,
		Region *	src,
		Operation *	inlinePoint,
		BlockAndValueMapping &	mapper,
		ArrayRef< Value >	resultsToReplace,
		Optional< Location >	inlineLoc = `llvm::None`,
		bool	shouldCloneInlinedRegion = `true`
	)

This function inlines a region, 'src', into another. This function returns failure if it is not possible to inline this function. If the function returned failure, then no changes to the module have been made.

The provided 'inlinePoint' must be within a region, and corresponds to the location where the 'src' region should be inlined. 'mapping' contains any remapped operands that are used within the region, and must include remappings for the entry arguments to the region. 'resultsToReplace' corresponds to any results that should be replaced by terminators within the inlined region. 'inlineLoc' is an optional Location that, if provided, will be used to update the inlined operations' location information. 'shouldCloneInlinedRegion' corresponds to whether the source region should be cloned into the 'inlinePoint' or spliced directly.

Handle the terminators for each of the new blocks.

◆ inlineRegion() [2/2]

LogicalResult mlir::inlineRegion	(	InlinerInterface &	interface,
		Region *	src,
		Operation *	inlinePoint,
		ArrayRef< Value >	inlinedOperands,
		ArrayRef< Value >	resultsToReplace,
		Optional< Location >	inlineLoc = `llvm::None`,
		bool	shouldCloneInlinedRegion = `true`
	)

This function is an overload of the above 'inlineRegion' that allows for providing the set of operands ('inlinedOperands') that should be used in-favor of the region arguments when inlining.

◆ insertBackwardComputationSlice()

AffineForOp mlir::insertBackwardComputationSlice	(	Operation *	srcOpInst,
		Operation *	dstOpInst,
		unsigned	dstLoopDepth,
		ComputationSliceState *	sliceState
	)

Creates a clone of the computation contained in the loop nest surrounding 'srcOpInst', slices the iteration space of src loop based on slice bounds in 'sliceState', and inserts the computation slice at the beginning of the operation block of the loop at 'dstLoopDepth' in the loop nest surrounding 'dstOpInst'. Returns the top-level loop of the computation slice on success, returns nullptr otherwise.

Creates a computation slice of the loop nest surrounding 'srcOpInst', updates the slice loop bounds with any non-null bound maps specified in 'sliceState', and inserts this slice into the loop nest surrounding 'dstOpInst' at loop depth 'dstLoopDepth'.

◆ instBodySkew()

LogicalResult mlir::instBodySkew	(	AffineForOp	forOp,
		ArrayRef< uint64_t >	shifts,
		bool	unrollPrologueEpilogue = `false`
	)

Skew the operations in the body of a 'affine.for' operation with the specified operation-wise shifts. The shifts are with respect to the original execution order, and are multiplied by the loop 'step' before being applied.

Skew the operations in the body of a 'affine.for' operation with the specified operation-wise shifts. The shifts are with respect to the original execution order, and are multiplied by the loop 'step' before being applied. A shift of zero for each operation will lead to no change.

◆ interchangeLoops() [1/2]

void mlir::interchangeLoops	(	AffineForOp	forOpA,
		AffineForOp	forOpB
	)

Performs loop interchange on 'forOpA' and 'forOpB'. Requires that 'forOpA' and 'forOpB' are part of a perfectly nested sequence of loops.

Performs loop interchange on 'forOpA' and 'forOpB', where 'forOpB' is nested within 'forOpA' as the only non-terminator operation in its block.

◆ interchangeLoops() [2/2]

unsigned mlir::interchangeLoops	(	ArrayRef< AffineForOp >	loops,
		ArrayRef< unsigned >	loopPermMap
	)

Performs a sequence of loop interchanges on perfectly nested 'loops', as specified by permutation 'loopPermMap' (loop 'i' in 'loops' is mapped to location 'j = 'loopPermMap[i]' after the loop interchange).

Performs a sequence of loop interchanges of loops in perfectly nested sequence of loops in 'loops', as specified by permutation in 'loopPermMap'.

◆ interleave() [1/4]

template<typename ForwardIterator , typename UnaryFunctor , typename NullaryFunctor , typename = typename std::enable_if< !std::is_constructible<StringRef, UnaryFunctor>::value && !std::is_constructible<StringRef, NullaryFunctor>::value>::type>

void mlir::interleave	(	ForwardIterator	begin,
		ForwardIterator	end,
		UnaryFunctor	each_fn,
		NullaryFunctor	between_fn
	)

inline

An STL-style algorithm similar to std::for_each that applies a second functor between every pair of elements.

This provides the control flow logic to, for example, print a comma-separated list:

interleave(names.begin(), names.end(),
           [&](StringRef name) { os << name; },
           [&] { os << ", "; });

◆ interleave() [2/4]

template<typename Container , typename UnaryFunctor , typename NullaryFunctor , typename = typename std::enable_if< !std::is_constructible<StringRef, UnaryFunctor>::value && !std::is_constructible<StringRef, NullaryFunctor>::value>::type>

void mlir::interleave	(	const Container &	c,
		UnaryFunctor	each_fn,
		NullaryFunctor	between_fn
	)

inline

◆ interleave() [3/4]

template<typename Container , typename UnaryFunctor , typename raw_ostream , typename T = detail::ValueOfRange<Container>>

void mlir::interleave	(	const Container &	c,
		raw_ostream &	os,
		UnaryFunctor	each_fn,
		const StringRef &	separator
	)

inline

Overload of interleave for the common case of string separator.

◆ interleave() [4/4]

template<typename Container , typename raw_ostream , typename T = detail::ValueOfRange<Container>>

void mlir::interleave	(	const Container &	c,
		raw_ostream &	os,
		const StringRef &	separator
	)

inline

◆ interleaveComma() [1/2]

template<typename Container , typename UnaryFunctor , typename raw_ostream , typename T = detail::ValueOfRange<Container>>

void mlir::interleaveComma	(	const Container &	c,
		raw_ostream &	os,
		UnaryFunctor	each_fn
	)

inline

◆ interleaveComma() [2/2]

template<typename Container , typename raw_ostream , typename T = detail::ValueOfRange<Container>>

void mlir::interleaveComma	(	const Container &	c,
		raw_ostream &	os
	)

inline

◆ inversePermutation()

AffineMap mlir::inversePermutation ( AffineMap map )

Returns a map of codomain to domain dimensions such that the first codomain dimension for a particular domain dimension is selected. Returns an empty map if the input map is empty or if map is not invertible (i.e. map does not contain a subset that is a permutation of full domain rank).

Prerequisites:

map has no symbols.

Example 1:

(d0, d1, d2) -> (d1, d1, d0, d2, d1, d2, d1, d0)

0 2 3

returns:

(d0, d1, d2, d3, d4, d5, d6, d7) -> (d2, d0, d3)

Example 2:

(d0, d1, d2) -> (d1, d0 + d1, d0, d2, d1, d2, d1, d0)

0 2 3

returns:

(d0, d1, d2, d3, d4, d5, d6, d7) -> (d2, d0, d3)

◆ isForInductionVar()

bool mlir::isForInductionVar ( Value val )

Returns if the provided value is the induction variable of a AffineForOp.

◆ isInstwiseShiftValid()

bool mlir::isInstwiseShiftValid	(	AffineForOp	forOp,
		ArrayRef< uint64_t >	shifts
	)

Checks where SSA dominance would be violated if a for op's body operations are shifted by the specified shifts. This method checks if a 'def' and all its uses have the same shift factor.

Checks whether SSA dominance would be violated if a for op's body operations are shifted by the specified shifts. This method checks if a 'def' and all its uses have the same shift factor.

◆ isLoopParallel()

bool mlir::isLoopParallel ( AffineForOp forOp )

Returns true if `forOp' is a parallel loop.

Returns true if 'forOp' is parallel.

◆ isOpaqueTypeWithName()

bool mlir::isOpaqueTypeWithName	(	Type	type,
		StringRef	dialect,
		StringRef	typeData
	)

Return true if the specified type is an opaque type with the specified dialect and typeData.

◆ isStrided()

bool mlir::isStrided ( MemRefType t )

Return true if the layout for t is compatible with strided semantics.

◆ isTopLevelValue()

bool mlir::isTopLevelValue ( Value value )

A utility function to check if a value is defined at the top level of a function. A value of index type defined at the top level is always a valid symbol.

◆ isValidDim()

bool mlir::isValidDim ( Value value )

Returns true if the given Value can be used as a dimension id.

◆ isValidLoopInterchangePermutation()

bool mlir::isValidLoopInterchangePermutation	(	ArrayRef< AffineForOp >	loops,
		ArrayRef< unsigned >	loopPermMap
	)

Checks if the loop interchange permutation 'loopPermMap', of the perfectly nested sequence of loops in 'loops', would violate dependences (loop 'i' in 'loops' is mapped to location 'j = 'loopPermMap[i]' in the interchange).

Checks if the loop interchange permutation 'loopPermMap' of the perfectly nested sequence of loops in 'loops' would violate dependences.

◆ isValidSymbol()

bool mlir::isValidSymbol ( Value value )

Returns true if the given Value can be used as a symbol.

◆ isVectorizableLoopBody() [1/2]

bool mlir::isVectorizableLoopBody	(	AffineForOp	loop,
		NestedPattern &	vectorTransferMatcher
	)

Checks whether the loop is structurally vectorizable; i.e.:

no conditionals are nested under the loop;
all nested load/stores are to scalar MemRefs. TODO(ntv): relax the no-conditionals restriction

◆ isVectorizableLoopBody() [2/2]

bool mlir::isVectorizableLoopBody	(	AffineForOp	loop,
		int *	memRefDim,
		NestedPattern &	vectorTransferMatcher
	)

Checks whether the loop is structurally vectorizable and that all the LoadOp and StoreOp matched have access indexing functions that are are either:

invariant along the loop induction variable created by 'loop';
varying along at most one memory dimension. If such a unique dimension is found, it is written into memRefDim.

◆ JitRunnerMain()

int mlir::JitRunnerMain	(	int	argc,
		char **	argv,
		llvm::function_ref< LogicalResult(mlir::ModuleOp)>	mlirTransformer
	)

◆ lcm()

int64_t mlir::lcm	(	int64_t	a,
		int64_t	b
	)

inline

Returns the least common multiple of 'a' and 'b'.

◆ loopUnrollByFactor()

LogicalResult mlir::loopUnrollByFactor	(	AffineForOp	forOp,
		uint64_t	unrollFactor
	)

Unrolls this for operation by the specified unroll factor. Returns failure if the loop cannot be unrolled either due to restrictions or due to invalid unroll factors.

Unrolls this loop by the specified factor. Returns success if the loop is successfully unrolled.

◆ loopUnrollFull()

LogicalResult mlir::loopUnrollFull ( AffineForOp forOp )

Unrolls this loop completely.

Unrolls this for operation completely if the trip count is known to be constant. Returns failure otherwise.

◆ loopUnrollJamByFactor()

LogicalResult mlir::loopUnrollJamByFactor	(	AffineForOp	forOp,
		uint64_t	unrollJamFactor
	)

Unrolls and jams this loop by the specified factor.

Unrolls and jams this loop by the specified factor. Returns success if the loop is successfully unroll-jammed.

◆ loopUnrollJamUpToFactor()

LogicalResult mlir::loopUnrollJamUpToFactor	(	AffineForOp	forOp,
		uint64_t	unrollJamFactor
	)

Unrolls and jams this loop by the specified factor or by the trip count (if constant), whichever is lower.

◆ loopUnrollUpToFactor()

LogicalResult mlir::loopUnrollUpToFactor	(	AffineForOp	forOp,
		uint64_t	unrollFactor
	)

Unrolls this loop by the specified unroll factor or its trip count, whichever is lower.

Unrolls and jams this loop by the specified factor or by the trip count (if constant) whichever is lower.

◆ lowerAffineLowerBound()

Value mlir::lowerAffineLowerBound	(	AffineForOp	op,
		OpBuilder &	builder
	)

Emit code that computes the lower bound of the given affine loop using standard arithmetic operations.

◆ lowerAffineUpperBound()

Value mlir::lowerAffineUpperBound	(	AffineForOp	op,
		OpBuilder &	builder
	)

Emit code that computes the upper bound of the given affine loop using standard arithmetic operations.

◆ m_Constant()

template<typename AttrT >

detail::constant_op_binder<AttrT> mlir::m_Constant ( AttrT * bind_value )

inline

Matches a value from a constant foldable operation and writes the value to bind_value.

◆ m_ConstantInt()

detail::constant_int_op_binder mlir::m_ConstantInt ( IntegerAttr::ValueType * bind_value )

inline

Matches a constant holding a scalar/vector/tensor integer (splat) and writes the integer value to bind_value.

◆ m_NonZero()

detail::constant_int_not_value_matcher<0> mlir::m_NonZero ( )

inline

Matches a constant scalar / vector splat / tensor splat integer that is any non-zero value.

◆ m_One()

detail::constant_int_value_matcher<1> mlir::m_One ( )

inline

Matches a constant scalar / vector splat / tensor splat integer one.

◆ m_Op() [1/2]

template<typename OpClass >

detail::op_matcher<OpClass> mlir::m_Op ( )

inline

Matches the given OpClass.

◆ m_Op() [2/2]

template<typename OpType , typename... Matchers>

auto mlir::m_Op ( Matchers... matchers )

◆ m_Zero()

detail::constant_int_value_matcher<0> mlir::m_Zero ( )

inline

Matches a constant scalar / vector splat / tensor splat integer zero.

◆ make_second_range()

template<typename ContainerTy >

auto mlir::make_second_range ( ContainerTy && c )

Given a container of pairs, return a range over the second elements.

◆ makeComposedAffineApply()

AffineApplyOp mlir::makeComposedAffineApply	(	OpBuilder &	b,
		Location	loc,
		AffineMap	map,
		ArrayRef< Value >	operands
	)

Returns a composed AffineApplyOp by composing map and operands with other AffineApplyOps supplying those operands. The operands of the resulting AffineApplyOp do not change the length of AffineApplyOp chains.

◆ makeStridedLinearLayoutMap()

AffineMap mlir::makeStridedLinearLayoutMap	(	ArrayRef< int64_t >	strides,
		int64_t	offset,
		MLIRContext *	context
	)

Given a list of strides (in which MemRefType::getDynamicStrideOrOffset() represents a dynamic value), return the single result AffineMap which represents the linearized strided layout map. Dimensions correspond to the offset followed by the strides in order. Symbols are inserted for each dynamic dimension in order. A stride cannot take value 0.

Examples:

For offset: 0 strides: ?, ?, 1 return (i, j, k)[M, N]->(M * i + N * j + k)
For offset: 3 strides: 32, ?, 16 return (i, j, k)[M]->(3 + 32 * i + M * j + 16 * k)
For offset: ? strides: ?, ?, ? return (i, j, k)[off, M, N, P]->(off + M * i + N * j + P * k)

◆ mapLoopToProcessorIds()

void mlir::mapLoopToProcessorIds	(	loop::ForOp	forOp,
		ArrayRef< Value >	processorId,
		ArrayRef< Value >	numProcessors
	)

Maps forOp for execution on a parallel grid of virtual processorIds of size given by numProcessors. This is achieved by embedding the SSA values corresponding to processorIds and numProcessors into the bounds and step of the forOp. No check is performed on the legality of the rewrite, it is the caller's responsibility to ensure legality.

Requires that processorIds and numProcessors have the same size and that for each idx, processorIds[idx] takes, at runtime, all values between 0 and numProcessors[idx] - 1. This corresponds to traditional use cases for:

GPU (threadIdx, get_local_id(), ...)
MPI (MPI_Comm_rank)
OpenMP (omp_get_thread_num)

Example: Assuming a 2-d grid with processorIds = [blockIdx.x, threadIdx.x] and numProcessors = [gridDim.x, blockDim.x], the loop:

loop.for %i = %lb to %ub step %step {
  ...
}

is rewritten into a version resembling the following pseudo-IR:

loop.for %i = %lb + %step * (threadIdx.x + blockIdx.x * blockDim.x)
   to %ub step %gridDim.x * blockDim.x * %step {
  ...
}

◆ matchPattern() [1/2]

template<typename Pattern >

bool mlir::matchPattern	(	Value	value,
		const Pattern &	pattern
	)

inline

Entry point for matching a pattern over a Value.

◆ matchPattern() [2/2]

template<typename Pattern >

bool mlir::matchPattern	(	Operation *	op,
		const Pattern &	pattern
	)

inline

Entry point for matching a pattern over an Operation.

◆ MlirOptMain()

LogicalResult mlir::MlirOptMain	(	llvm::raw_ostream &	os,
		std::unique_ptr< llvm::MemoryBuffer >	buffer,
		const PassPipelineCLParser &	passPipeline,
		bool	splitInputFile,
		bool	verifyDiagnostics,
		bool	verifyPasses
	)

◆ mod()

int64_t mlir::mod	(	int64_t	lhs,
		int64_t	rhs
	)

inline

Returns MLIR's mod operation on constants. MLIR's mod operation yields the remainder of the Euclidean division of 'lhs' by 'rhs', and is therefore not C's % operator. The RHS is always expected to be positive, and the result is always non-negative.

◆ normalizeMemRef()

LogicalResult mlir::normalizeMemRef ( AllocOp op )

Rewrites the memref defined by this alloc op to have an identity layout map and updates all its indexing uses. Returns failure if any of its uses escape (while leaving the IR in a valid state).

◆ openInputFile()

std::unique_ptr<llvm::MemoryBuffer> mlir::openInputFile	(	llvm::StringRef	inputFilename,
		std::string *	errorMessage = `nullptr`
	)

Open the file specified by its name for reading. Write the error message to errorMessage if errors occur and errorMessage is not nullptr.

◆ openOutputFile()

std::unique_ptr<llvm::ToolOutputFile> mlir::openOutputFile	(	llvm::StringRef	outputFilename,
		std::string *	errorMessage = `nullptr`
	)

Open the file specified by its name for writing. Write the error message to errorMessage if errors occur and errorMessage is not nullptr.

◆ operator!=() [1/6]

bool mlir::operator!=	(	Identifier	lhs,
		Identifier	rhs
	)

inline

◆ operator!=() [2/6]

bool mlir::operator!=	(	Identifier	lhs,
		StringRef	rhs
	)

inline

◆ operator!=() [3/6]

bool mlir::operator!=	(	StringRef	lhs,
		Identifier	rhs
	)

inline

◆ operator!=() [4/6]

bool mlir::operator!=	(	IntInfty	lhs,
		IntInfty	rhs
	)

inline

◆ operator!=() [5/6]

bool mlir::operator!=	(	OperationName	lhs,
		OperationName	rhs
	)

inline

◆ operator!=() [6/6]

bool mlir::operator!=	(	OpState	lhs,
		OpState	rhs
	)

inline

◆ operator*()

AffineExpr mlir::operator*	(	int64_t	val,
		AffineExpr	expr
	)

inline

◆ operator+() [1/2]

IntInfty mlir::operator+	(	IntInfty	lhs,
		IntInfty	rhs
	)

inline

◆ operator+() [2/2]

AffineExpr mlir::operator+	(	int64_t	val,
		AffineExpr	expr
	)

inline

◆ operator-()

AffineExpr mlir::operator-	(	int64_t	val,
		AffineExpr	expr
	)

inline

◆ operator<()

bool mlir::operator<	(	IntInfty	lhs,
		IntInfty	rhs
	)

inline

◆ operator<<() [1/24]

DialectAsmPrinter& mlir::operator<<	(	DialectAsmPrinter &	p,
		Attribute	attr
	)

inline

◆ operator<<() [2/24]

DialectAsmPrinter& mlir::operator<<	(	DialectAsmPrinter &	p,
		const APFloat &	value
	)

inline

◆ operator<<() [3/24]

DialectAsmPrinter& mlir::operator<<	(	DialectAsmPrinter &	p,
		float	value
	)

inline

◆ operator<<() [4/24]

DialectAsmPrinter& mlir::operator<<	(	DialectAsmPrinter &	p,
		double	value
	)

inline

◆ operator<<() [5/24]

raw_ostream& mlir::operator<<	(	raw_ostream &	os,
		Identifier	identifier
	)

inline

◆ operator<<() [6/24]

DialectAsmPrinter& mlir::operator<<	(	DialectAsmPrinter &	p,
		Type	type
	)

inline

◆ operator<<() [7/24]

template<typename T , typename std::enable_if< !std::is_convertible< T &, Attribute &>::value &&!std::is_convertible< T &, Type &>::value &&!std::is_convertible< T &, APFloat &>::value &&!llvm::is_one_of< T, double, float >::value, T >::type * = nullptr>

DialectAsmPrinter& mlir::operator<<	(	DialectAsmPrinter &	p,
		const T &	other
	)

inline

◆ operator<<() [8/24]

raw_ostream& mlir::operator<<	(	raw_ostream &	os,
		const Location &	loc
	)

inline

◆ operator<<() [9/24]

raw_ostream& mlir::operator<<	(	raw_ostream &	os,
		Attribute	attr
	)

inline

◆ operator<<() [10/24]

OpAsmPrinter& mlir::operator<<	(	OpAsmPrinter &	p,
		Value	value
	)

inline

◆ operator<<() [11/24]

template<typename T , typename std::enable_if< std::is_convertible< T &, ValueRange >::value &&!std::is_convertible< T &, Value &>::value, T >::type * = nullptr>

OpAsmPrinter & mlir::operator<<	(	OpAsmPrinter &	p,
		const T &	values
	)

inline

◆ operator<<() [12/24]

OpAsmPrinter& mlir::operator<<	(	OpAsmPrinter &	p,
		Type	type
	)

inline

◆ operator<<() [13/24]

raw_ostream& mlir::operator<<	(	raw_ostream &	os,
		const DiagnosticArgument &	arg
	)

inline

◆ operator<<() [14/24]

OpAsmPrinter& mlir::operator<<	(	OpAsmPrinter &	p,
		Attribute	attr
	)

inline

◆ operator<<() [15/24]

OpAsmPrinter& mlir::operator<<	(	OpAsmPrinter &	p,
		bool	value
	)

inline

◆ operator<<() [16/24]

raw_ostream& mlir::operator<<	(	raw_ostream &	os,
		Type	type
	)

inline

◆ operator<<() [17/24]

template<typename IteratorT >

OpAsmPrinter& mlir::operator<<	(	OpAsmPrinter &	p,
		const iterator_range< ValueTypeIterator< IteratorT >> &	types
	)

inline

◆ operator<<() [18/24]

raw_ostream& mlir::operator<<	(	raw_ostream &	os,
		Value	value
	)

inline

◆ operator<<() [19/24]

raw_ostream& mlir::operator<<	(	raw_ostream &	os,
		AffineMap	map
	)

inline

◆ operator<<() [20/24]

raw_ostream & mlir::operator<<	(	raw_ostream &	os,
		AffineExpr &	expr
	)

◆ operator<<() [21/24]

raw_ostream& mlir::operator<<	(	raw_ostream &	os,
		OperationName	identifier
	)

inline

◆ operator<<() [22/24]

raw_ostream& mlir::operator<<	(	raw_ostream &	os,
		const Diagnostic &	diag
	)

inline

◆ operator<<() [23/24]

raw_ostream & mlir::operator<<	(	raw_ostream &	os,
		SubViewOp::Range &	range
	)

◆ operator<<() [24/24]

raw_ostream& mlir::operator<<	(	raw_ostream &	os,
		Operation &	op
	)

inline

◆ operator<=()

bool mlir::operator<=	(	IntInfty	lhs,
		IntInfty	rhs
	)

inline

◆ operator==() [1/6]

bool mlir::operator==	(	Identifier	lhs,
		Identifier	rhs
	)

inline

◆ operator==() [2/6]

bool mlir::operator==	(	Identifier	lhs,
		StringRef	rhs
	)

inline

◆ operator==() [3/6]

bool mlir::operator==	(	StringRef	lhs,
		Identifier	rhs
	)

inline

◆ operator==() [4/6]

bool mlir::operator==	(	IntInfty	lhs,
		IntInfty	rhs
	)

inline

◆ operator==() [5/6]

bool mlir::operator==	(	OperationName	lhs,
		OperationName	rhs
	)

inline

◆ operator==() [6/6]

bool mlir::operator==	(	OpState	lhs,
		OpState	rhs
	)

inline

◆ parseAttribute() [1/4]

Attribute mlir::parseAttribute	(	llvm::StringRef	attrStr,
		MLIRContext *	context
	)

This parses a single MLIR attribute to an MLIR context if it was valid. If not, an error message is emitted through a new SourceMgrDiagnosticHandler constructed from a new SourceMgr with a single a MemoryBuffer wrapping attrStr. If the passed attrStr has additional tokens that were not part of the type, an error is emitted.

◆ parseAttribute() [2/4]

Attribute mlir::parseAttribute	(	llvm::StringRef	attrStr,
		Type	type
	)

◆ parseAttribute() [3/4]

Attribute mlir::parseAttribute	(	llvm::StringRef	attrStr,
		MLIRContext *	context,
		size_t &	numRead
	)

This parses a single MLIR attribute to an MLIR context if it was valid. If not, an error message is emitted through a new SourceMgrDiagnosticHandler constructed from a new SourceMgr with a single a MemoryBuffer wrapping attrStr. The number of characters of attrStr parsed in the process is returned in numRead.

◆ parseAttribute() [4/4]

Attribute mlir::parseAttribute	(	llvm::StringRef	attrStr,
		Type	type,
		size_t &	numRead
	)

◆ parseDimAndSymbolList()

ParseResult mlir::parseDimAndSymbolList	(	OpAsmParser &	parser,
		SmallVectorImpl< Value > &	operands,
		unsigned &	numDims
	)

Parses dimension and symbol list and returns true if parsing failed.

◆ parsePassPipeline()

LogicalResult mlir::parsePassPipeline	(	StringRef	pipeline,
		OpPassManager &	pm,
		raw_ostream &	errorStream = `llvm::errs()`
	)

This function parses the textual representation of a pass pipeline, and adds the result to 'pm' on success. This function returns failure if the given pipeline was invalid. 'errorStream' is the output stream used to emit errors found during parsing.

This function parses the textual representation of a pass pipeline, and adds the result to 'pm' on success. This function returns failure if the given pipeline was invalid. 'errorStream' is an optional parameter that, if non-null, will be used to emit errors found during parsing.

◆ parseSourceFile() [1/3]

OwningModuleRef mlir::parseSourceFile	(	const llvm::SourceMgr &	sourceMgr,
		MLIRContext *	context
	)

This parses the file specified by the indicated SourceMgr and returns an MLIR module if it was valid. If not, the error message is emitted through the error handler registered in the context, and a null pointer is returned.

This parses the file specified by the indicated SourceMgr and returns an MLIR module if it was valid. If not, it emits diagnostics and returns null.

◆ parseSourceFile() [2/3]

OwningModuleRef mlir::parseSourceFile	(	llvm::StringRef	filename,
		MLIRContext *	context
	)

This parses the file specified by the indicated filename and returns an MLIR module if it was valid. If not, the error message is emitted through the error handler registered in the context, and a null pointer is returned.

◆ parseSourceFile() [3/3]

OwningModuleRef mlir::parseSourceFile	(	llvm::StringRef	filename,
		llvm::SourceMgr &	sourceMgr,
		MLIRContext *	context
	)

This parses the file specified by the indicated filename using the provided SourceMgr and returns an MLIR module if it was valid. If not, the error message is emitted through the error handler registered in the context, and a null pointer is returned.

◆ parseSourceString()

OwningModuleRef mlir::parseSourceString	(	llvm::StringRef	moduleStr,
		MLIRContext *	context
	)

This parses the module string to a MLIR module if it was valid. If not, the error message is emitted through the error handler registered in the context, and a null pointer is returned.

◆ parseType() [1/2]

Type mlir::parseType	(	llvm::StringRef	typeStr,
		MLIRContext *	context
	)

This parses a single MLIR type to an MLIR context if it was valid. If not, an error message is emitted through a new SourceMgrDiagnosticHandler constructed from a new SourceMgr with a single a MemoryBuffer wrapping typeStr. If the passed typeStr has additional tokens that were not part of the type, an error is emitted.

◆ parseType() [2/2]

Type mlir::parseType	(	llvm::StringRef	typeStr,
		MLIRContext *	context,
		size_t &	numRead
	)

This parses a single MLIR type to an MLIR context if it was valid. If not, an error message is emitted through a new SourceMgrDiagnosticHandler constructed from a new SourceMgr with a single a MemoryBuffer wrapping typeStr. The number of characters of typeStr parsed in the process is returned in numRead.

◆ populateAffineToStdConversionPatterns()

void mlir::populateAffineToStdConversionPatterns	(	OwningRewritePatternList &	patterns,
		MLIRContext *	ctx
	)

Collect a set of patterns to convert from the Affine dialect to the Standard dialect, in particular convert structured affine control flow into CFG branch-based control flow.

◆ populateFuncOpTypeConversionPattern()

void mlir::populateFuncOpTypeConversionPattern	(	OwningRewritePatternList &	patterns,
		MLIRContext *	ctx,
		TypeConverter &	converter
	)

Add a pattern to the given pattern list to convert the signature of a FuncOp with the given type converter.

◆ populateGpuToNVVMConversionPatterns()

void mlir::populateGpuToNVVMConversionPatterns	(	LLVMTypeConverter &	converter,
		OwningRewritePatternList &	patterns
	)

Collect a set of patterns to convert from the GPU dialect to NVVM.

◆ populateGPUToSPIRVPatterns()

void mlir::populateGPUToSPIRVPatterns	(	MLIRContext *	context,
		SPIRVTypeConverter &	typeConverter,
		OwningRewritePatternList &	patterns,
		ArrayRef< int64_t >	workGroupSize
	)

Appends to a pattern list additional patterns for translating GPU Ops to SPIR-V ops. Needs the workgroup size as input since SPIR-V/Vulkan requires the workgroup size to be statically specified.

◆ populateLinalgToLLVMConversionPatterns()

void mlir::populateLinalgToLLVMConversionPatterns	(	LinalgTypeConverter &	converter,
		OwningRewritePatternList &	patterns,
		MLIRContext *	ctx
	)

Populate the given list with patterns that convert from Linalg to LLVM.

◆ populateLoopToStdConversionPatterns()

void mlir::populateLoopToStdConversionPatterns	(	OwningRewritePatternList &	patterns,
		MLIRContext *	ctx
	)

Collect a set of patterns to lower from loop.for, loop.if, and loop.terminator to CFG operations within the Standard dialect, in particular convert structured control flow into CFG branch-based control flow.

◆ populateStandardToSPIRVPatterns()

void mlir::populateStandardToSPIRVPatterns	(	MLIRContext *	context,
		SPIRVTypeConverter &	typeConverter,
		OwningRewritePatternList &	patterns
	)

Appends to a pattern list additional patterns for translating StandardOps to SPIR-V ops. Also adds the patterns legalize ops not directly translated to SPIR-V dialect.

◆ populateStdLegalizationPatternsForSPIRVLowering()

void mlir::populateStdLegalizationPatternsForSPIRVLowering	(	MLIRContext *	context,
		OwningRewritePatternList &	patterns
	)

Appends to a pattern list patterns to legalize ops that are not directly lowered to SPIR-V.

◆ populateStdToLLVMConversionPatterns()

void mlir::populateStdToLLVMConversionPatterns	(	LLVMTypeConverter &	converter,
		OwningRewritePatternList &	patterns
	)

Collect a set of patterns to convert from the Standard dialect to LLVM.

◆ populateStdToLLVMMemoryConversionPatters()

void mlir::populateStdToLLVMMemoryConversionPatters	(	LLVMTypeConverter &	converter,
		OwningRewritePatternList &	patterns
	)

Collect a set of patterns to convert memory-related operations from the Standard dialect to the LLVM dialect, excluding the memory-related operations.

◆ populateStdToLLVMNonMemoryConversionPatterns()

void mlir::populateStdToLLVMNonMemoryConversionPatterns	(	LLVMTypeConverter &	converter,
		OwningRewritePatternList &	patterns
	)

Collect a set of patterns to convert from the Standard dialect to LLVM.

Collect a set of patterns to convert from the Standard dialect to the LLVM dialect, excluding the memory-related operations.

◆ populateVectorToAffineLoopsConversionPatterns()

void mlir::populateVectorToAffineLoopsConversionPatterns	(	MLIRContext *	context,
		OwningRewritePatternList &	patterns
	)

Collect a set of patterns to convert from the Vector dialect to loops + std.

◆ populateVectorToLLVMConversionPatterns()

void mlir::populateVectorToLLVMConversionPatterns	(	LLVMTypeConverter &	converter,
		OwningRewritePatternList &	patterns
	)

Collect a set of patterns to convert from the Vector dialect to LLVM.

Populate the given list with patterns that convert from Vector to LLVM.

◆ populateVectorToVectorConversionPatterns()

void mlir::populateVectorToVectorConversionPatterns	(	MLIRContext *	context,
		OwningRewritePatternList &	patterns,
		ArrayRef< int64_t >	coarseVectorShape = `{}`,
		ArrayRef< int64_t >	fineVectorShape = `{}`
	)

Collect a set of patterns to convert from the Vector dialect to itself. Should be merged with populateVectorToAffineLoopsConversionPatterns.

◆ printDimAndSymbolList()

void mlir::printDimAndSymbolList	(	Operation::operand_iterator	begin,
		Operation::operand_iterator	end,
		unsigned	numDims,
		OpAsmPrinter &	p
	)

Prints dimension and symbol list.

◆ promoteIfSingleIteration()

LogicalResult mlir::promoteIfSingleIteration ( AffineForOp forOp )

Promotes the loop body of a AffineForOp to its containing block if the AffineForOp was known to have a single iteration.

Promotes the loop body of a forOp to its containing block if the forOp was known to have a single iteration.

◆ promoteSingleIterationLoops()

void mlir::promoteSingleIterationLoops ( FuncOp f )

Promotes all single iteration AffineForOp's in the Function, i.e., moves their body into the containing Block.

Promotes all single iteration for op's in the FuncOp, i.e., moves their body into the containing Block.

◆ promoteToWorkgroupMemory()

void mlir::promoteToWorkgroupMemory	(	gpu::GPUFuncOp	op,
		unsigned	arg
	)

Promotes a function argument to workgroup memory in the given function. The copies will be inserted in the beginning and in the end of the function.

◆ registerAllDialects()

void mlir::registerAllDialects ( MLIRContext * context )

Registers all dialects with the specified MLIRContext.

Registers all dialects and their const folding hooks with the specified MLIRContext.

◆ registerDialect()

template<typename ConcreteDialect >

void mlir::registerDialect ( )

Utility to register a dialect. Client can register their dialect with the global registry by calling registerDialect<MyDialect>();

◆ registerDialectAllocator()

void mlir::registerDialectAllocator ( const DialectAllocatorFunction & function )

Registers a specific dialect creation function with the system, typically used through the DialectRegistration template.

◆ registerDialectHooksSetter()

void mlir::registerDialectHooksSetter ( const DialectHooksSetter & function )

Registers a function that will set hooks in the registered dialects based on information coming from DialectHooksRegistration.

Registers a function to set specific hooks for a specific dialect, typically used through the DialectHooksRegistration template.

◆ registerPass()

void mlir::registerPass	(	StringRef	arg,
		StringRef	description,
		const PassID *	passID,
		const PassAllocatorFunction &	function
	)

Register a specific dialect pass allocator function with the system, typically used through the PassRegistration template.

◆ registerPassManagerCLOptions()

void mlir::registerPassManagerCLOptions ( )

Register a set of useful command-line options that can be used to configure a pass manager. The values of these options can be applied via the 'applyPassManagerCLOptions' method below.

◆ registerPassPipeline()

void mlir::registerPassPipeline	(	StringRef	arg,
		StringRef	description,
		const PassRegistryFunction &	function
	)

Register a specific dialect pipeline registry function with the system, typically used through the PassPipelineRegistration template.

◆ replaceAllMemRefUsesWith() [1/2]

LogicalResult mlir::replaceAllMemRefUsesWith	(	Value	oldMemRef,
		Value	newMemRef,
		ArrayRef< Value >	extraIndices = `{}`,
		AffineMap	indexRemap = `AffineMap()`,
		ArrayRef< Value >	extraOperands = `{}`,
		ArrayRef< Value >	symbolOperands = `{}`,
		Operation *	domInstFilter = `nullptr`,
		Operation *	postDomInstFilter = `nullptr`
	)

Replaces all "dereferencing" uses of oldMemRef with newMemRef while optionally remapping the old memref's indices using the supplied affine map, indexRemap. The new memref could be of a different shape or rank. extraIndices provides any additional access indices to be added to the start.

indexRemap remaps indices of the old memref access to a new set of indices that are used to index the memref. Additional input operands to indexRemap can be optionally provided in extraOperands, and they occupy the start of its input list. indexRemap's dimensional inputs are expected to correspond to memref's indices, and its symbolic inputs if any should be provided in symbolOperands.

domInstFilter, if non-null, restricts the replacement to only those operations that are dominated by the former; similarly, postDomInstFilter restricts replacement to only those operations that are postdominated by it.

Returns true on success and false if the replacement is not possible, whenever a memref is used as an operand in a non-dereferencing context, except for dealloc's on the memref which are left untouched. See comments at function definition for an example.

◆ replaceAllMemRefUsesWith() [2/2]

LogicalResult mlir::replaceAllMemRefUsesWith	(	Value	oldMemRef,
		Value	newMemRef,
		Operation *	op,
		ArrayRef< Value >	extraIndices = `{}`,
		AffineMap	indexRemap = `AffineMap()`,
		ArrayRef< Value >	extraOperands = `{}`,
		ArrayRef< Value >	symbolOperands = `{}`
	)

Performs the same replacement as the other version above but only for the dereferencing uses of oldMemRef in op.

◆ replaceAllUsesInRegionWith()

void mlir::replaceAllUsesInRegionWith	(	Value	orig,
		Value	replacement,
		Region &	region
	)

Replace all uses of orig within the given region with replacement.

◆ shapeRatio() [1/2]

Optional< SmallVector< int64_t, 4 > > mlir::shapeRatio	(	ArrayRef< int64_t >	superShape,
		ArrayRef< int64_t >	subShape
	)

Computes and returns the multi-dimensional ratio of superShape to subShape. This is calculated by performing a traversal from minor to major dimensions (i.e. in reverse shape order). If integral division is not possible, returns None. The ArrayRefs are assumed (and enforced) to only contain > 1 values. This constraint comes from the fact that they are meant to be used with VectorTypes, for which the property holds by construction.

Examples:

shapeRatio({3, 4, 5, 8}, {2, 5, 2}) returns {3, 2, 1, 4}
shapeRatio({3, 4, 4, 8}, {2, 5, 2}) returns None
shapeRatio({1, 2, 10, 32}, {2, 5, 2}) returns {1, 1, 2, 16}

◆ shapeRatio() [2/2]

Optional< SmallVector< int64_t, 4 > > mlir::shapeRatio	(	VectorType	superVectorType,
		VectorType	subVectorType
	)

Computes and returns the multi-dimensional ratio of the shapes of superVector to subVector. If integral division is not possible, returns None. Assumes and enforces that the VectorTypes have the same elemental type.

◆ simplifyAffineExpr()

AffineExpr mlir::simplifyAffineExpr	(	AffineExpr	expr,
		unsigned	numDims,
		unsigned	numSymbols
	)

Simplify the affine expression by flattening it and reconstructing it.

Simplify an affine expression by flattening and some amount of simple analysis. This has complexity linear in the number of nodes in 'expr'. Returns the simplified expression, which is the same as the input expression if it can't be simplified.

◆ simplifyAffineMap()

AffineMap mlir::simplifyAffineMap ( AffineMap map )

Simplify an affine map by simplifying its underlying AffineExpr results.

◆ simplifyRegions()

LogicalResult mlir::simplifyRegions ( MutableArrayRef< Region > regions )

Run a set of structural simplifications over the given regions. This includes transformations like unreachable block elimination, dead argument elimination, as well as some other DCE. This function returns success if any of the regions were simplified, failure otherwise.

◆ sinkLoop()

void mlir::sinkLoop	(	AffineForOp	forOp,
		unsigned	loopDepth
	)

Sinks 'forOp' by 'loopDepth' levels by performing a series of loop interchanges. Requires that 'forOp' is part of a perfect nest with 'loopDepth' AffineForOps consecutively nested under it.

Performs a series of loop interchanges to sink 'forOp' 'loopDepth' levels deeper in the loop nest.

◆ sinkSequentialLoops()

AffineForOp mlir::sinkSequentialLoops ( AffineForOp forOp )

◆ splitAndProcessBuffer()

LogicalResult mlir::splitAndProcessBuffer	(	std::unique_ptr< llvm::MemoryBuffer >	originalBuffer,
		ChunkBufferHandler	processChunkBuffer,
		raw_ostream &	os
	)

Splits the specified buffer on a marker (// -----), processes each chunk independently according to the normal processChunkBuffer logic, and writes all results to os.

This is used to allow a large number of small independent tests to be put into a single file.

◆ succeeded()

bool mlir::succeeded ( LogicalResult result )

inline

Utility function that returns true if the provided LogicalResult corresponds to a success value.

◆ success()

LogicalResult mlir::success ( bool isSuccess = true )

inline

Utility function to generate a LogicalResult. If isSuccess is true a success result is generated, otherwise a 'failure' result is generated.

◆ tile() [1/4]

SmallVector< SmallVector< AffineForOp, 8 >, 8 > mlir::tile	(	ArrayRef< AffineForOp >	forOps,
		ArrayRef< uint64_t >	sizes,
		ArrayRef< AffineForOp >	targets
	)

◆ tile() [2/4]

SmallVector< Loops, 8 > mlir::tile	(	ArrayRef< loop::ForOp >	forOps,
		ArrayRef< Value >	sizes,
		ArrayRef< loop::ForOp >	targets
	)

◆ tile() [3/4]

SmallVector< AffineForOp, 8 > mlir::tile	(	ArrayRef< AffineForOp >	forOps,
		ArrayRef< uint64_t >	sizes,
		AffineForOp	target
	)

Performs tiling (with interchange) by strip-mining the forOps by sizes and sinking them, in their order of occurrence in forOps, under target. Returns the new AffineForOps, one per forOps, nested immediately under target.

◆ tile() [4/4]

Loops mlir::tile	(	ArrayRef< loop::ForOp >	forOps,
		ArrayRef< Value >	sizes,
		loop::ForOp	target
	)

◆ tileCodeGen()

LogicalResult mlir::tileCodeGen	(	MutableArrayRef< AffineForOp >	band,
		ArrayRef< unsigned >	tileSizes
	)

Tiles the specified band of perfectly nested loops creating tile-space loops and intra-tile loops. A band is a contiguous set of loops.

◆ tilePerfectlyNested()

Loops mlir::tilePerfectlyNested	(	loop::ForOp	rootForOp,
		ArrayRef< Value >	sizes
	)

Tile a nest of loop::ForOp loops rooted at rootForOp with the given (parametric) sizes. Sizes are expected to be strictly positive values at runtime. If more sizes than loops are provided, discard the trailing values in sizes. Assumes the loop nest is permutable. Returns the newly created intra-tile loops.

◆ toAffineExpr()

AffineExpr mlir::toAffineExpr	(	ArrayRef< int64_t >	eq,
		unsigned	numDims,
		unsigned	numSymbols,
		ArrayRef< AffineExpr >	localExprs,
		MLIRContext *	context
	)

Constructs an affine expression from a flat ArrayRef. If there are local identifiers (neither dimensional nor symbolic) that appear in the sum of products expression, 'localExprs' is expected to have the AffineExpr for it, and is substituted into. The ArrayRef 'eq' is expected to be in the format [dims, symbols, locals, constant term].

◆ topologicalSort()

llvm::SetVector<Operation *> mlir::topologicalSort ( const llvm::SetVector< Operation *> & toSort )

Multi-root DAG topological sort. Performs a topological sort of the Operation in the toSort SetVector. Returns a topologically sorted SetVector.

◆ translateLLVMIRToModule()

OwningModuleRef mlir::translateLLVMIRToModule	(	std::unique_ptr< llvm::Module >	llvmModule,
		MLIRContext *	context
	)

Convert the given LLVM module into MLIR's LLVM dialect. The LLVM context is extracted from the registered LLVM IR dialect. In case of error, report it to the error handler registered with the MLIR context, if any (obtained from the MLIR module), and return {}.

◆ translateModuleToLLVMIR()

std::unique_ptr< llvm::Module > mlir::translateModuleToLLVMIR ( ModuleOp m )

Convert the given MLIR module into LLVM IR. The LLVM context is extracted from the registered LLVM IR dialect. In case of error, report it to the error handler registered with the MLIR context, if any (obtained from the MLIR module), and return nullptr.

◆ translateModuleToNVVMIR()

std::unique_ptr< llvm::Module > mlir::translateModuleToNVVMIR ( Operation * m )

Convert the given LLVM-module-like operation into NVVM IR. This conversion requires the registration of the LLVM IR dialect and will extract the LLVM context from the registered LLVM IR dialect. In case of error, report it to the error handler registered with the MLIR context, if any (obtained from the MLIR module), and return nullptr.

◆ translateModuleToROCDLIR()

std::unique_ptr< llvm::Module > mlir::translateModuleToROCDLIR ( Operation * m )

Convert the given LLVM-module-like operation into ROCDL IR. This conversion requires the registration of the LLVM IR dialect and will extract the LLVM context from the registered LLVM IR dialect. In case of error, report it to the error handler registered with the MLIR context, if any (obtained from the MLIR module), and return nullptr.

◆ verify()

LogicalResult mlir::verify ( Operation * op )

Perform (potentially expensive) checks of invariants, used to detect compiler bugs, on this operation and any nested operations. On error, this reports the error through the MLIRContext and returns failure.

Perform (potentially expensive) checks of invariants, used to detect compiler bugs. On error, this reports the error through the MLIRContext and returns failure.

◆ verifyCompatibleShape() [1/2]

LogicalResult mlir::verifyCompatibleShape	(	ArrayRef< int64_t >	shape1,
		ArrayRef< int64_t >	shape2
	)

Returns success if the given two shapes are compatible. That is, they have the same size and each pair of the elements are equal or one of them is dynamic.

◆ verifyCompatibleShape() [2/2]

LogicalResult mlir::verifyCompatibleShape	(	Type	type1,
		Type	type2
	)

Returns success if the given two types have compatible shape. That is, they are both scalars (not shaped), or they are both shaped types and at least one is unranked or they have compatible dimensions. Dimensions are compatible if at least one is dynamic or both are equal. The element type does not matter.

◆ viewGraph() [1/2]

void mlir::viewGraph	(	Block &	block,
		const Twine &	name,
		bool	shortNames = `false`,
		const Twine &	title = `""`,
		llvm::GraphProgram::Name	program = `llvm::GraphProgram::DOT`
	)

Displays the graph in a window. This is for use from the debugger and depends on Graphviz to generate the graph.

◆ viewGraph() [2/2]

void mlir::viewGraph	(	Region &	region,
		const Twine &	name,
		bool	shortNames = `false`,
		const Twine &	title = `""`,
		llvm::GraphProgram::Name	program = `llvm::GraphProgram::DOT`
	)

Displays the CFG in a window. This is for use from the debugger and depends on Graphviz to generate the graph.

◆ visitUsedValuesDefinedAbove() [1/2]

void mlir::visitUsedValuesDefinedAbove	(	Region &	region,
		Region &	limit,
		function_ref< void(OpOperand *)>	callback
	)

Calls callback for each use of a value within region or its descendants that was defined at the ancestors of the limit.

◆ visitUsedValuesDefinedAbove() [2/2]

void mlir::visitUsedValuesDefinedAbove	(	MutableArrayRef< Region >	regions,
		function_ref< void(OpOperand *)>	callback
	)

Calls callback for each use of a value within any of the regions provided that was defined in one of the ancestors.

◆ writeGraph() [1/2]

raw_ostream & mlir::writeGraph	(	raw_ostream &	os,
		Block &	block,
		bool	shortNames = `false`,
		const Twine &	title = `""`
	)

◆ writeGraph() [2/2]

raw_ostream & mlir::writeGraph	(	raw_ostream &	os,
		Region &	region,
		bool	shortNames = `false`,
		const Twine &	title = `""`
	)

Variable Documentation

◆ makeLLVMPassesTransformer

std::function< llvm::Error(llvm::Module *)> mlir::makeLLVMPassesTransformer

Create a module transformer function for MLIR ExecutionEngine that runs LLVM IR passes explicitly specified, plus an optional optimization level, Any optimization passes, if present, will be inserted before the pass at position optPassesInsertPos. If not null, targetMachine is used to initialize passes that provide target-specific information to the LLVM optimizer. targetMachine must outlive the returned std::function.

◆ makeOptimizingTransformer

std::function< llvm::Error(llvm::Module *)> mlir::makeOptimizingTransformer

Create a module transformer function for MLIR ExecutionEngine that runs LLVM IR passes corresponding to the given speed and size optimization levels (e.g. -O2 or -Os). If not null, targetMachine is used to initialize passes that provide target-specific information to the LLVM optimizer. targetMachine must outlive the returned std::function.

Namespaces

Classes

Typedefs

Enumerations

Functions

Variables

Detailed Description

Typedef Documentation

◆ AnalysisID

◆ AttributeStorageAllocator

◆ ChunkBufferHandler

◆ CubinGenerator

◆ DefaultAttributeStorage

◆ DefaultTypeStorage

◆ DenseMap

◆ DenseSet

◆ DialectAllocatorFunction

◆ DialectConstantDecodeHook

◆ DialectConstantFoldHook

◆ DialectExtractElementHook

◆ DialectHooksSetter

◆ DominanceInfoNode

◆ FilterFunctionType

◆ function_ref

◆ GenFunction

◆ is_detected

◆ is_invocable

◆ LLVMPatternListFiller

◆ LLVMTypeConverterMaker

◆ Loops

◆ NamedAttribute

◆ OpAsmSetValueNameFn

◆ OperandAdaptor

◆ OperandElementTypeRange

◆ OwnedCubin

◆ PassAllocatorFunction

◆ PassID

◆ PassRegistryFunction

◆ PatternMatchResult

◆ ResultElementTypeRange

◆ TileLoops

◆ TransitiveFilter

◆ TranslateFromMLIRFunction

◆ TranslateFunction

◆ TranslateSourceMgrToMLIRFunction

◆ TranslateStringRefToMLIRFunction

◆ TypeStorageAllocator

◆ VectorizableLoopFun

Enumeration Type Documentation

◆ AffineExprKind

◆ CmpFPredicate

◆ DiagnosticSeverity

◆ OperationProperty

◆ PassDisplayMode

◆ SDBMExprKind

Function Documentation

◆ affineDataCopyGenerate()

◆ applyAnalysisConversion() [1/2]

◆ applyAnalysisConversion() [2/2]

◆ applyFullConversion() [1/2]

◆ applyFullConversion() [2/2]

◆ applyPartialConversion() [1/2]

◆ applyPartialConversion() [2/2]

◆ applyPassManagerCLOptions()

◆ applyPatternsGreedily() [1/2]

◆ applyPatternsGreedily() [2/2]

◆ areValuesDefinedAbove()

◆ bindDims()

◆ boundCheckLoadOrStoreOp()

◆ buildTripCountMapAndOperands()

◆ canFuseLoops()

◆ canonicalizeMapAndOperands()

◆ canonicalizeSetAndOperands()

◆ canonicalizeStridedLayout()

◆ ceilDiv()

◆ checkMemrefAccessDependence()

◆ coalesceLoops()

◆ computeSliceUnion()

◆ concatAffineMaps()

◆ constFoldBinaryOp()