|
tvm
|
TIR specific transformation passes. More...
#include <tvm/ir/transform.h>#include <tvm/tir/expr.h>#include <tvm/tir/function.h>#include <string>Go to the source code of this file.
Namespaces | |
| tvm | |
| Performance counters for profiling via the PAPI library. | |
| tvm::tir | |
| tvm::tir::transform | |
Functions | |
| Pass | tvm::tir::transform::CreatePrimFuncPass (const runtime::TypedPackedFunc< PrimFunc(PrimFunc, IRModule, PassContext)> &pass_func, int opt_level, String name, tvm::Array< String > required) |
| Pass | tvm::tir::transform::InjectPrefetch () |
| Inject prefetch instructions into stmt. More... | |
| Pass | tvm::tir::transform::StorageFlatten (int cache_line_size, bool create_bound_attribute=false) |
| Flatten the multi-dimensional read/write to single dimensional Load/Store. More... | |
| Pass | tvm::tir::transform::InjectCopyIntrin (String pragma_key, runtime::PackedFunc fintrin) |
| Inject copy intrinsics with optional pad. More... | |
| Pass | tvm::tir::transform::CoProcSync () |
| Detect and insert sync points to co-processor. More... | |
| Pass | tvm::tir::transform::LiftAttrScope (String attr_key) |
| Lift common attrs with attr_key to outer scope. More... | |
| Pass | tvm::tir::transform::LoopPartition () |
| partition loops in the stmt. More... | |
| Pass | tvm::tir::transform::VectorizeLoop (bool enable_vectorize=true) |
| Lower vectorization loops. More... | |
| Pass | tvm::tir::transform::InjectVirtualThread () |
| Inject virtual thread loops. More... | |
| Pass | tvm::tir::transform::InjectDoubleBuffer () |
| Inject double buffer statements. More... | |
| Pass | tvm::tir::transform::StorageRewrite () |
| Rewrite storage allocation pattern. Moves the allocation to outer most possible scope. Trying to share space between allocations to make a static allocation plan when possible. More... | |
| Pass | tvm::tir::transform::UnrollLoop () |
| unroll the constant loop marked by unroll. This pass also automatically attach pragma unroll tag to loops which meets the standard. More... | |
| Pass | tvm::tir::transform::RemoveNoOp () |
| Remove No Op from the Stmt. More... | |
| Pass | tvm::tir::transform::RewriteUnsafeSelect () |
| Detect and rewrite unsafe select that contains memory access. More... | |
| Pass | tvm::tir::transform::Simplify () |
| Run arithmetic simplifications on the statements and expressions. More... | |
| Pass | tvm::tir::transform::InstrumentBoundCheckers () |
| Instruments bound checkers. More... | |
| Pass | tvm::tir::transform::MakePackedAPI (int num_unpacked_args) |
| Transform the high-level PrimFunc to a low-level version that can be used as an API function. More... | |
| Pass | tvm::tir::transform::MakeUnpackedAPI () |
| Transform the high-level PrimFunc to a C signature that can be used to call the operator directly. More... | |
| Pass | tvm::tir::transform::RemapThreadAxis (Map< String, IterVar > axis_map) |
| Remap the thread axis. More... | |
| Pass | tvm::tir::transform::LowerCustomDatatypes () |
| Lower custom datatypes. More... | |
| Pass | tvm::tir::transform::DecorateDeviceScope () |
| Decorate all the function's body as device function. More... | |
| Pass | tvm::tir::transform::SplitHostDevice () |
| Split the function into a host function and device functions. More... | |
| Pass | tvm::tir::transform::SkipAssert () |
| skip assert stmt. More... | |
| Pass | tvm::tir::transform::ThreadSync (String storage_scope) |
| Insert sync between parallel read/write of shared buffers. More... | |
| Pass | tvm::tir::transform::LowerThreadAllreduce () |
| Lower cross thread alleduce. More... | |
| Pass | tvm::tir::transform::InferFragment () |
| Infer the TensorCore fragment infomation using tensor intrinsics. More... | |
| Pass | tvm::tir::transform::LowerTVMBuiltin () |
| Lower builtin intrinsics. More... | |
| Pass | tvm::tir::transform::LowerIntrin () |
| Lower the target specific function intrinsics in each of the function. More... | |
| Pass | tvm::tir::transform::LowerWarpMemory () |
| Lower warp memory access to low-level device related function calls. More... | |
| Pass | tvm::tir::transform::LowerDeviceStorageAccessInfo () |
| Lower attached storage access information on device. More... | |
| Pass | tvm::tir::transform::CombineContextCall () |
| Combine context calls in the host function. More... | |
| Pass | tvm::tir::transform::NarrowDataType (int target_bits) |
| Narrow down PrimExpr datatype in stmt to target_bits. More... | |
| Pass | tvm::tir::transform::BF16Legalize () |
| Legalize bf16 typed Ops. Add a cast to fp32 before Ops, then add a cast back to bf16. More... | |
| Pass | tvm::tir::transform::PointerValueTypeRewrite () |
| Rewrite the pointer content type of arguments, as well as Alloc internal to the function to use the most frequently accessed type for load/store to avoid pointer casting in backend when possible. More... | |
| Pass | tvm::tir::transform::HoistIfThenElse () |
| Hoist loop-invariant IfThenElse nodes to outside the elligible loops. More... | |
| Pass | tvm::tir::transform::LowerInitBlock () |
| Lower block init stmt into IfThenElse stmts. More... | |
| Pass | tvm::tir::transform::PlanAndUpdateBufferAllocationLocation () |
| Locate the buffer allocation to the exact position (usually is the lca of buffer access). This pass will inject opaque block with alloc_buffers at the allocation site. More... | |
| Pass | tvm::tir::transform::ConvertBlocksToOpaque () |
| Substitute all the block vars with the PrimExprs they are bound to, indicated by the corresponding iter_values in BlockRealize, for opaque blocks by removing all . the iter_values in BlockRealize and iter_vars in Block. More... | |
| Pass | tvm::tir::transform::CompactBufferAllocation () |
| Compact the buffer access region by removing the buffer regions that are not accessed, i.e. narrowing the buffer shape and adjust the access region if necessary. More... | |
| Pass | tvm::tir::transform::LegalizePackedCalls () |
| Pass | tvm::tir::transform::LowerMatchBuffer () |
| Remove match buffers inside the block. Also, it will validate the binding. More... | |
| Pass | tvm::tir::transform::FlattenBuffer () |
| Flatten the multi-dimensional BufferLoad and BufferStore to single dimensional Load/Store. Also remove Block to ensure that the flattened TIR can not be scheduled again. More... | |
| Pass | tvm::tir::transform::TextureFlatten () |
| Pass | tvm::tir::transform::UnifyThreadBinding () |
| Unify all the thread bindings for "blockIdx.x/y/z", "threadIdx.x/y/z", and "vthread.x/y/z". Before the unification, two vars that are bound to a thread axis (e.g., "threadIdx.x") use different IterVars and variables in their AttrStmts. After the unification, we use a consolidated IterVar and a variable for them. More... | |
| Pass | tvm::tir::transform::MergeDynamicSharedMemoryAllocations () |
| Pass | tvm::tir::transform::ConvertForLoopsToSerial () |
| This pass is post-scheduling pass to convert all Parallel For loops to Serial ones. This is run to attain lesser memory and/or executor/backend does not support parallel launch of For loops. More... | |
TIR specific transformation passes.
1.8.13