tilelang.transform package#
Submodules#
Module contents#
Wrapping transformations.
- tilelang.transform.AnnotateDeviceRegions()#
- Returns:
fpass – The result pass
- Return type:
tvm.transform.Pass
- tilelang.transform.ClusterPlanning()#
- Returns:
fpass – The result pass
- Return type:
tvm.transform.Pass
- tilelang.transform.FrontendLegalize()#
- Returns:
fpass – The result pass
- Return type:
tvm.transform.Pass
- tilelang.transform.IfStmtBinding()#
- Returns:
fpass – The result pass
- Return type:
tvm.transform.Pass
- tilelang.transform.InjectFenceProxy()#
- Returns:
fpass – The result pass
- Return type:
tvm.transform.Pass
- tilelang.transform.InjectPTXAsyncCopy()#
Rewrite global to shared memory copy on CUDA with asynchronous copy.
- Returns:
fpass – The result pass
- Return type:
tvm.transform.Pass
- tilelang.transform.InjectSoftwarePipeline()#
- Returns:
fpass – The result pass
- Return type:
tvm.transform.Pass
- tilelang.transform.LayoutInference()#
- Returns:
fpass – The result pass
- Return type:
tvm.transform.Pass
- tilelang.transform.LegalizeSafeMemoryAccess()#
LegalizeLoopVectorize
- Returns:
fpass – The result pass
- Return type:
tvm.transform.Pass
- tilelang.transform.LegalizeVectorizedLoop()#
LegalizeLoopVectorize
- Returns:
fpass – The result pass
- Return type:
tvm.transform.Pass
- tilelang.transform.LoopVectorizeDynamic()#
Try to vectorize loop with dynamic shape.
- Returns:
fpass (tvm.transform.Pass) – The result pass
—-
- tilelang.transform.LowerDeviceStorageAccessInfo()#
Lower attached storage access information on device.
- Returns:
fpass – The result pass
- Return type:
tvm.transform.Pass
Note
Run this pass after all storage access analysis finish.
- tilelang.transform.LowerHopperIntrin()#
- Returns:
fpass – The result pass
- Return type:
tvm.transform.Pass
- tilelang.transform.LowerTileOp()#
- Returns:
fpass – The result pass
- Return type:
tvm.transform.Pass
- tilelang.transform.MakePackedAPI()#
- Returns:
fpass – The result pass
- Return type:
tvm.transform.Pass
- tilelang.transform.MergeIfStmt()#
- Returns:
fpass – The result pass
- Return type:
tvm.transform.Pass
- tilelang.transform.MultiVersionBuffer()#
WarpSpecializedPipeline
- Returns:
fpass – The result pass
- Return type:
tvm.transform.Pass
- tilelang.transform.PipelinePlanning()#
infer the fragment/shared memory layout
- Returns:
fpass – The result pass
- Return type:
tvm.transform.Pass
- tilelang.transform.RewriteWgmmaSync()#
- Returns:
fpass – The result pass
- Return type:
tvm.transform.Pass
- tilelang.transform.ThreadPartialSync(storage_scope: str)#
Insert partial sync.
- Parameters:
storage_scope (str) – The target storage scope.
- Returns:
fpass – The result pass
- Return type:
tvm.transform.Pass
- tilelang.transform.ThreadSync(storage_scope: str)#
Insert sync between parallel read/write of shared buffers.
- Parameters:
storage_scope (str) – The target storage scope.
- Returns:
fpass – The result pass
- Return type:
tvm.transform.Pass
- tilelang.transform.VectorizeLoop(enable_vectorize: bool = True)#
- Returns:
fpass – The result pass
- Return type:
tvm.transform.Pass
- tilelang.transform.WarpSpecialized()#
WarpSpecializedPipeline
- Returns:
fpass – The result pass
- Return type:
tvm.transform.Pass
- tilelang.transform.WarpSpecializedPipeline()#
- Returns:
fpass – The result pass
- Return type:
tvm.transform.Pass