TVM 手册#
- TVM 准备
- Apache TVM 文档
- 概述
- Installing TVM
- Development Guides
- Design and Architecture
- TensorIR
- Relax
- Python API
- tvm.error
- tvm.ir
- tvm.instrument
- tvm.transform
ModulePassPassPassContextPassContext.current()PassContext.get_current_trace()PassContext.get_trace_stack()PassContext.get_trace_stack_size()PassContext.get_tuning_api_database()PassContext.inc_num_evals()PassContext.list_configs()PassContext.override_instruments()PassContext.pop_trace()PassContext.push_trace()PassContext.set_num_evals()
PassInfoSequentialApplyPassToFunction()PrintIR()_wrap_class_module_pass()module_pass()
- tvm.target
- tvm.driver
- tvm.runtime
- tvm.runtime.ndarray
- tvm.runtime.relax_vm
VMInstrumentReturnKindVirtualMachineVirtualMachine.__init__()VirtualMachine._convert()VirtualMachine._convert_func_named_args()VirtualMachine._setup_device()VirtualMachine.get_outputs()VirtualMachine.invoke_closure()VirtualMachine.invoke_stateful()VirtualMachine.profile()VirtualMachine.save_function()VirtualMachine.set_input()VirtualMachine.set_instrument()VirtualMachine.time_evaluator()
- tvm.runtime.disco
DModuleDPackedFuncDRefProcessSessionSessionSession._sync_worker()Session.allgather()Session.allreduce()Session.broadcast()Session.broadcast_from_worker0()Session.call_packed()Session.copy_from_worker_0()Session.copy_to_worker_0()Session.empty()Session.gather_to_worker0()Session.get_global_func()Session.import_python_module()Session.init_ccl()Session.load_vm_module()Session.scatter()Session.scatter_from_worker0()Session.shutdown()Session.sync_worker_0()Session.num_workers
SocketSessionThreadedSession
- tvm.runtime.profiling
- tvm.relax
BindingBindingBlockCallConstantDataTypeImmDataflowBlockDataflowBlockRewriteDataflowVarDynTensorTypeExecBuilderExecutableExprExprFunctorExternFuncFuncStructInfoFunctionIdIfMatchCastObjectStructInfoObjectTypePackedFuncTypePrimStructInfoPrimValuePyExprMutatorPyExprMutator.__init__()PyExprMutator.get_var_remap()PyExprMutator.lookup_binding()PyExprMutator.set_var_remap()PyExprMutator.visit_binding()PyExprMutator.visit_binding_block()PyExprMutator.visit_binding_block_()PyExprMutator.visit_call_()PyExprMutator.visit_constant_()PyExprMutator.visit_data_type_imm_()PyExprMutator.visit_dataflow_block_()PyExprMutator.visit_dataflow_var_()PyExprMutator.visit_dataflow_var_def_()PyExprMutator.visit_expr()PyExprMutator.visit_expr_post_order()PyExprMutator.visit_extern_func_()PyExprMutator.visit_function_()PyExprMutator.visit_global_var_()PyExprMutator.visit_if_()PyExprMutator.visit_match_cast_()PyExprMutator.visit_op_()PyExprMutator.visit_prim_value_()PyExprMutator.visit_seq_expr_()PyExprMutator.visit_shape_expr_()PyExprMutator.visit_span()PyExprMutator.visit_string_imm_()PyExprMutator.visit_tuple_()PyExprMutator.visit_tuple_getitem_()PyExprMutator.visit_var_()PyExprMutator.visit_var_binding_()PyExprMutator.visit_var_def()PyExprMutator.visit_var_def_()PyExprMutator.visit_with_new_scope()PyExprMutator.with_struct_info()
PyExprVisitorPyExprVisitor.visit_binding()PyExprVisitor.visit_binding_block()PyExprVisitor.visit_binding_block_()PyExprVisitor.visit_call_()PyExprVisitor.visit_constant_()PyExprVisitor.visit_data_type_imm_()PyExprVisitor.visit_dataflow_block_()PyExprVisitor.visit_dataflow_var_()PyExprVisitor.visit_dataflow_var_def_()PyExprVisitor.visit_expr()PyExprVisitor.visit_extern_func_()PyExprVisitor.visit_function_()PyExprVisitor.visit_global_var_()PyExprVisitor.visit_if_()PyExprVisitor.visit_match_cast_()PyExprVisitor.visit_op_()PyExprVisitor.visit_prim_value_()PyExprVisitor.visit_seq_expr_()PyExprVisitor.visit_shape_expr_()PyExprVisitor.visit_span()PyExprVisitor.visit_string_imm_()PyExprVisitor.visit_tuple_()PyExprVisitor.visit_tuple_getitem_()PyExprVisitor.visit_var_()PyExprVisitor.visit_var_binding_()PyExprVisitor.visit_var_def()PyExprVisitor.visit_var_def_()
SeqExprShapeExprShapeStructInfoShapeTypeStringImmStructInfoTensorStructInfoTupleTupleGetItemTupleStructInfoVMInstrumentReturnKindVarVarBindingVirtualMachineVirtualMachine.__init__()VirtualMachine._convert()VirtualMachine._convert_func_named_args()VirtualMachine._setup_device()VirtualMachine.get_outputs()VirtualMachine.invoke_closure()VirtualMachine.invoke_stateful()VirtualMachine.profile()VirtualMachine.save_function()VirtualMachine.set_input()VirtualMachine.set_instrument()VirtualMachine.time_evaluator()
build()call_dps_packed()call_pure_packed()call_tir()call_tir_inplace()call_tir_with_grad()const()convert_to_expr()extern()get_pipeline()get_shape_of()register_pipeline()
- tvm.relax.analysis
BaseCheckResultall_global_vars()all_vars()bound_vars()collect_non_negative_expressions()computable_at_compile_time()contains_impure_call()definable_tir_vars_in_struct_info()defined_symbolic_vars()derive_call_ret_struct_info()detect_recursion()erase_to_well_defined()estimate_memory_usage()free_symbolic_vars()free_vars()get_static_type()get_var2val()has_reshape_pattern()name_to_binding()post_order_visit()remove_all_unused()struct_info_base_check()struct_info_lca()suggest_layout_transforms()tir_vars_in_struct_info()udchain()well_formed()
- tvm.relax.block_builder
BlockBuilderBlockBuilder._normalize_python_tuple()BlockBuilder.add_func()BlockBuilder.begin_scope()BlockBuilder.call_te()BlockBuilder.call_te_with_grad()BlockBuilder.current()BlockBuilder.current_block_is_dataflow()BlockBuilder.dataflow()BlockBuilder.emit()BlockBuilder.emit_func_output()BlockBuilder.emit_normalized()BlockBuilder.emit_output()BlockBuilder.emit_te()BlockBuilder.end_scope()BlockBuilder.finalize()BlockBuilder.function()BlockBuilder.get()BlockBuilder.get_unique_name()BlockBuilder.lookup_binding()BlockBuilder.match_cast()BlockBuilder.normalize()BlockBuilder.testing_scope()BlockBuilder.update_func()
DataflowScopeFunctionScopeTestingScope
- tvm.relax.frontend
- tvm.relax.op
- tvm.relax.op
abs()acos()acosh()add()arange()argmax()argmin()argsort()asin()asinh()assert_op()astype()atan()atanh()bitwise_and()bitwise_not()bitwise_or()bitwise_xor()broadcast_to()call_builtin_with_ctx()call_dps_packed()call_inplace_packed()call_pure_packed()call_tir()call_tir_inplace()call_tir_with_grad()ceil()clip()collapse_sum_like()collapse_sum_to()concat()cos()cosh()cumprod()cumsum()dequantize()divide()dynamic_strided_slice()einsum()equal()erf()ewise_fma()exp()expand_dims()eye()eye_like()flatten()flip()floor()floor_divide()floor_mod()full()full_like()greater()greater_equal()hint_on_device()invoke_closure()invoke_pure_closure()isfinite()isinf()isnan()layout_transform()left_shift()less()less_equal()linear()log()logical_and()logical_not()logical_or()logical_xor()make_closure()masked_fill()matmul()max()maximum()mean()min()minimum()mod()multinomial_from_uniform()multiply()negative()nonzero()not_equal()null_value()one_hot()ones()ones_like()permute_dims()power()print()prod()quantize()register_gradient()repeat()reshape()right_shift()round()rsqrt()scatter_elements()scatter_nd()shape_of()shape_to_tensor()sigmoid()sign()sin()sinh()sort()split()sqrt()square()squeeze()std()strided_slice()subtract()sum()take()tan()tanh()tensor_to_shape()tile()to_vdevice()topk()tril()triu()unique()variance()where()wrap_param()zeros()zeros_like()
- tvm.relax.op.nn
adaptive_avg_pool1d()adaptive_avg_pool2d()adaptive_avg_pool3d()attention()attention_var_len()avg_pool1d()avg_pool2d()avg_pool3d()batch_norm()conv1d()conv1d_transpose()conv2d()conv2d_transpose()conv3d()cross_entropy_with_logits()dropout()gelu()gelu_tanh()group_norm()layer_norm()leakyrelu()log_softmax()max_pool1d()max_pool2d()max_pool3d()nll_loss()pad()relu()rms_norm()silu()softmax()
- tvm.relax.op.builtin
- tvm.relax.op.ccl
- tvm.relax.op.distributed
- tvm.relax.op.grad
- tvm.relax.op.image
- tvm.relax.op.memory
- tvm.relax.op.op_attrs
AdaptivePool2DAttrsArgmaxArgminAttrsArgsortAttrsAstypeAttrsBatchNormAttrsCallTIRWithGradAttrsConcatAttrsConv2DAttrsConv2DTransposeAttrsConv3DAttrsDropoutAttrsEinsumAttrsExpandDimsAttrsFlipAttrsInitAttrsLayerNormAttrsLayoutTransformAttrsMatmulAttrsPermuteDimsAttrsPool2DAttrsRepeatAttrsResize2DAttrsScanopAttrsSoftmaxAttrsSortAttrsSplitAttrsSqueezeAttrsStatisticalAttrsStridedSliceAttrsTakeAttrsTileAttrsTopKAttrsTriluAttrs
- tvm.relax.op
- tvm.relax.transform
AttachExternModulesDataflowBlockPassFastMathTransformFunctionPassFuseTransposeMatmulFusionPatternIPCAllReduceRewriteLazyTransformParamsLowerGPUIPCAllocStorageOptimizeLayoutTransformPatternCheckContextRemoveRedundantReshapeAdjustMatmulOrder()AllocateWorkspace()AlterOpImpl()AnnotateTIROpPattern()AttachAttrLayoutFreeBuffers()AttachGlobalSymbol()BindParams()BindSymbolicVars()BundleModelParams()CallTIRRewrite()CanonicalizeBindings()CombineParallelMatmul()ComputePrimValue()ConvertLayout()ConvertToDataflow()DataflowUseInplaceCalls()DeadCodeElimination()DecomposeOpsForInference()DecomposeOpsForTraining()EliminateCommonSubexpr()ExpandMatmulOfSum()ExpandTupleArguments()FewShotTuning()FoldConstant()FuseOps()FuseOpsByPattern()FuseTIR()Gradient()InlinePrivateFunctions()KillAfterLastUse()LambdaLift()LazyGetInput()LazySetOutput()LegalizeOps()LiftTransformParams()LowerAllocTensor()LowerRuntimeBuiltin()MergeCompositeFunctions()MetaScheduleApplyDatabase()MetaScheduleTuneIRMod()MetaScheduleTuneTIR()Normalize()NormalizeGlobalVar()RealizeVDevice()RemovePurityChecking()RemoveUnusedOutputs()RemoveUnusedParameters()ReorderPermuteDimsAfterConcat()ReorderTakeAfterMatmul()RewriteCUDAGraph()RewriteDataflowReshape()RunCodegen()SplitCallTIRByPattern()SplitLayoutRewritePreproc()StaticPlanBlockMemory()ToMixedPrecision()ToNonDataflow()TopologicalSort()UpdateParamStructInfo()UpdateVDevice()VMBuiltinLower()VMShapeLower()dataflowblock_pass()function_pass()
- tvm.tir
AddAllocateAllocateConstAndAnyAssertStmtAttrStmtBijectiveLayoutBlockBlockDependenceInfoBlockRealizeBroadcastBufferBufferLoadBufferRealizeBufferRegionBufferStoreCallCallEffectKindCastCommReducerDataProducerDeclBufferDivEQEvaluateFloatImmFloorDivFloorModForForKindGEGTIfThenElseIndexMapIntImmIterVarLELTLayoutLetLetStmtMatchBufferRegionMaxMinModMulNENotOrPrefetchPrimFuncProducerLoadProducerRealizeProducerStoreRampReduceSelectSeqStmtShuffleSizeVarStmtStringImmSubTensorIntrinVarWhileTVMBackendAllocWorkspace()TVMBackendFreeWorkspace()abs()acos()acosh()add()address_of()all()any()asin()asinh()assume()atan()atan2()atanh()bijective_layout()bitwise_and()bitwise_not()bitwise_or()bitwise_xor()call_cpacked()call_cpacked_lowered()call_extern()call_intrin()call_llvm_intrin()call_llvm_pure_intrin()call_packed()call_packed_lowered()call_pure_extern()call_tir()ceil()ceildiv()clz()comm_reducer()copysign()cos()cosh()create_barriers()decl_buffer()div()dp4a()end_profile_intrinsic()erf()exp()exp10()exp2()floor()floordiv()floormod()fmod()get_active_lane_mask()get_vscale_expr()hypot()if_then_else()indexdiv()indexmod()infinity()isfinite()isinf()isnan()isnullptr()layout()ldexp()likely()log()log10()log1p()log2()lookup_param()make_filled_simdgroup_matrix()max()max_value()min()min_value()mma_fill()mma_store()multiply()nearbyint()nextafter()popcount()pow()power()ptx_arrive_barrier()ptx_arrive_barrier_expect_tx()ptx_commit_group()ptx_cp_async()ptx_cp_async_barrier()ptx_cp_async_bulk()ptx_init_barrier_thread_count()ptx_ldmatrix()ptx_mma()ptx_mma_sp()ptx_wait_barrier()ptx_wait_group()q_multiply_shift()q_multiply_shift_per_axis()reinterpret()ret()round()rsqrt()shift_left()shift_right()sigmoid()simdgroup_load()simdgroup_multiply_accumulate()simdgroup_store()sin()sinh()sqrt()start_profile_intrinsic()stmt_list()stmt_seq()subtract()sum()tan()tanh()trace()trunc()truncdiv()truncmod()tvm_access_ptr()tvm_bmma_sync()tvm_check_return()tvm_fill_fragment()tvm_load_matrix_sync()tvm_mma_sync()tvm_stack_alloca()tvm_stack_make_array()tvm_stack_make_shape()tvm_store_matrix_sync()tvm_struct_get()tvm_struct_set()tvm_thread_allreduce()tvm_throw_last_error()tvm_tuple()type_annotation()undef()vectorcombine()vectorhigh()vectorlow()vscale()
- tvm.tir.analysis
OOBChecker()apply_prim_func_arg_and_result_memory_constraints()assert_pure_function()calculate_allocated_bytes()calculate_constant_bytes()calculate_workspace_bytes()detect_buffer_access_lca()estimate_tir_flops()expr_deep_equal()find_anchor_block()get_block_access_region()get_block_read_write_region()get_prim_func_arg_and_result_memory_constraints()get_vtcm_compaction_passes()is_pure_function()undefined_vars()verify_gpu_code()verify_memory()verify_ssa()verify_well_formed()
- tvm.tir.schedule
ScheduleErrorBlockRVBlockScopeDepKindDependencyExprRVInstructionInstructionKindLoopRVScheduleSchedule.__init__()Schedule._create_non_traced()Schedule.add_unit_loop()Schedule.annotate()Schedule.annotate_buffer_access()Schedule.bind()Schedule.blockize()Schedule.cache_index()Schedule.cache_inplace()Schedule.cache_read()Schedule.cache_write()Schedule.can_decompose_padding()Schedule.compute_at()Schedule.compute_inline()Schedule.copy()Schedule.decompose_padding()Schedule.decompose_reduction()Schedule.enter_postproc()Schedule.fork_seed()Schedule.fuse()Schedule.get()Schedule.get_block()Schedule.get_child_blocks()Schedule.get_consumers()Schedule.get_loops()Schedule.get_output_blocks()Schedule.get_producers()Schedule.get_sref()Schedule.loop_partition()Schedule.merge()Schedule.pad_einsum()Schedule.parallel()Schedule.reindex()Schedule.reindex_cache_read()Schedule.reindex_cache_write()Schedule.remove_rv()Schedule.reorder()Schedule.reorder_block_iter_var()Schedule.reverse_compute_at()Schedule.reverse_compute_inline()Schedule.rfactor()Schedule.rolling_buffer()Schedule.sample_categorical()Schedule.sample_compute_location()Schedule.sample_partitioned_tile()Schedule.sample_perfect_tile()Schedule.seed()Schedule.set_axis_separator()Schedule.set_scope()Schedule.show()Schedule.split()Schedule.storage_align()Schedule.tensorize()Schedule.transform_block_layout()Schedule.transform_layout()Schedule.unannotate()Schedule.unroll()Schedule.unsafe_hide_buffer_access()Schedule.unsafe_set_dtype()Schedule.vectorize()Schedule.work_on()Schedule.func_working_onSchedule.modSchedule.stateSchedule.trace
ScheduleDebugMaskScheduleStateStmtSRefTrace
- tvm.tir.stmt_functor
- tvm.tir.transform
HoistedConditionalsHoistedLetBindingsPrimFuncPassAnnotateDeviceRegions()AnnotateEntryFunc()Apply()ApplyLayoutTransforms()BF16ComputeLegalize()BF16StorageLegalize()BindTarget()CoProcSync()CombineContextCall()CommonSubexprElimTIR()CompactBufferAllocation()ConvertBlocksToOpaque()ConvertForLoopsToSerial()ConvertSSA()DecorateDeviceScope()DefaultGPUSchedule()ExtractPrimFuncConstants()FP8ComputeLegalize()FP8StorageLegalize()Filter()FlattenBuffer()ForceNarrowIndexToInt32()HoistExpression()HoistIfThenElse()InferFragment()InjectCopyIntrin()InjectDoubleBuffer()InjectPTXAsyncCopy()InjectPermutedLayout()InjectPrefetch()InjectRollingBuffer()InjectSoftwarePipeline()InjectVirtualThread()InlinePrivateFunctions()InstallDebugSpans()InstrumentBoundCheckers()InstrumentProfileIntrinsics()LegalizePackedCalls()LiftAttrScope()LiftThreadBinding()LoopPartition()LowerAutoCopy()LowerCrossThreadReduction()LowerCustomDatatypes()LowerDeviceKernelLaunch()LowerDeviceStorageAccessInfo()LowerInitBlock()LowerIntrin()LowerMatchBuffer()LowerOpaqueBlock()LowerTVMBuiltin()LowerThreadAllreduce()LowerWarpMemory()MakePackedAPI()MakeUnpackedAPI()ManifestSharedMemoryLocalStage()MergeSharedMemoryAllocations()NarrowDataType()PlanAndUpdateBufferAllocationLocation()PointerValueTypeRewrite()ReduceBranchingThroughOvercompute()RemoveAssume()RemoveNoOp()RemoveStoreUndef()RemoveWeightLayoutRewriteBlock()RenormalizeSplitPattern()RewriteUnsafeSelect()Simplify()SkipAssert()SplitHostDevice()StorageFlatten()StorageRewrite()TextureFlatten()ThreadSync()TransformMmaBufferLayout()UnifyThreadBinding()UnrollLoop()UseAssumeToReduceBranches()VectorizeLoop()VerifyMemory()VerifyVTCMLimit()prim_func_pass()
- tvm.te
- tvm.te.hybrid
- tvm.topi
- tvm.meta_schedule
- tvm.dlight
- tvm.rpc
- tvm.contrib
- tvm.contrib.cblas
- tvm.contrib.clang
- tvm.contrib.cc
- tvm.contrib.cublas
- tvm.contrib.dlpack
- tvm.contrib.emcc
- tvm.contrib.miopen
- tvm.contrib.mxnet
- tvm.contrib.ndk
- tvm.contrib.nnpack
- tvm.contrib.nvcc
- tvm.contrib.pickle_memoize
- tvm.contrib.random
- tvm.contrib.relay_viz
- tvm.contrib.rocblas
- tvm.contrib.rocm
- tvm.contrib.sparse
- tvm.contrib.spirv
- tvm.contrib.tar
- tvm.contrib.utils
- tvm.contrib.xcode
- tvm.relay
- tvm.relay.frontend
- tvm.relay.nn
- tvm.relay.vision
- tvm.relay.image
- tvm.relay.transform
- tvm.relay.analysis
- tvm.relay.backend
- tvm.relay.dataflow_pattern
- tvm.relay.testing
get_net()get_workload()get_net()get_workload()residual_unit()resnet()deconv2d()deconv2d_bn_relu()get_net()get_workload()conv_block()get_workload()mobile_net()separable_conv_block()get_net()get_workload()lstm_cell()get_net()get_workload()get_net()get_workload()get_classifier()get_feature()get_net()get_workload()_make_dense_block()_make_dense_layer()_make_dense_net()_make_transition()get_workload()
- tvm.autotvm
apply_history_best()- tvm.autotvm.measure
- tvm.autotvm.tuner
- tvm.autotvm.task
FlopCalculationErrorMissingTaskTaskTaskTemplate_register_customized_task()_register_task_compute()_register_task_schedule()args_to_workload()compute_flop()create()deserialize_args()get_config()serialize_args()template()InstantiationErrorAnnotateEntityAnnotateSpaceAxisConfigEntityConfigSpaceConfigSpace.__getitem__()ConfigSpace.__len__()ConfigSpace._add_new_transform()ConfigSpace.add_flop()ConfigSpace.axis()ConfigSpace.clear_cache()ConfigSpace.define_annotate()ConfigSpace.define_knob()ConfigSpace.define_reorder()ConfigSpace.define_split()ConfigSpace.get()ConfigSpace.get_next_index()ConfigSpace.get_rand_index()ConfigSpace.is_index_valid()ConfigSpace.knob2point()ConfigSpace.multi_filter()ConfigSpace.point2knob()ConfigSpace.raise_error()ConfigSpace.random_walk()ConfigSpace.reduce_axis()ConfigSpace.sample_ints()ConfigSpace.subrange_length()ConfigSpace.valid()ConfigSpace.dimsConfigSpace.range_length
FallbackConfigEntityOtherOptionEntityOtherOptionSpaceReorderEntityReorderSpaceSplitEntitySplitSpaceTransformSpaceVirtualAxisget_factors()get_pow2s()ApplyConfigApplyFixedConfigApplyGraphBestApplyHistoryBestDispatchContextFallbackContextclear_fallback_cache()TaskExtractEnvget_workload()register_topi_compute()register_topi_schedule()
- tvm.autotvm.record
- tvm.auto_scheduler
- tvm.micro
- tvm.contrib.graph_executor
GraphModuleGraphModule.__getitem__()GraphModule.benchmark()GraphModule.debug_get_output()GraphModule.get_input()GraphModule.get_input_index()GraphModule.get_input_info()GraphModule.get_num_inputs()GraphModule.get_num_outputs()GraphModule.get_output()GraphModule.load_params()GraphModule.run()GraphModule.set_input()GraphModule.set_input_zero_copy()GraphModule.set_output_zero_copy()GraphModule.share_params()
create()get_device()
- Other APIs
- How To Guides
- Deploy Models and Integrate TVM
- Build the TVM runtime library
- Cross compile the TVM runtime for other architectures
- Optimize and tune models for target devices
- Deploy optimized model on target devices
- Deploy TVM Module using C++ API
- Deploy to Android
- Deploy to Adreno™ GPU
- Integrate TVM into Your Project
- HLS Backend Example
- Relay Arm® Compute Library Integration
- Relay TensorRT Integration
- Vitis AI Integration
- Relay BNNS Integration
- Marvell Machine Learning Integration
- 1. Introduction
- 2. Building TVM with mrvl support
- 2.1 Clone TVM repo
- 2.2 Build and start the TVM - mrvl docker container
- 3. Compiling a model using TVMC command line
- 3.1 TVMC Compilation Flow for a model
- 3.2. TVMC - Command line option(s): Syntax for mrvl target
- 3.3. TVMC Compiler: mrvl specific Command Line Options
- 4. Compile ONNX model using the TVMC flow
- 4.1 Compile and Run ONNX model for Simulator + LLVM / x86_64 target
- 4.2 Compile and Run ONNX model for Octeon target
- 5. Compiling a model using Python APIs
- Additional Deployment How-Tos
- Profile Models
- Frequently Asked Questions
- Deploy Models and Integrate TVM
- Developer Tutorial
- Developer How-To Guide
- Debugging TVM
- Adding an Operator to Relay
- 1. Defining an Attribute Node
- 2. Writing a Type Relation
- 3. Relating the Arity and Attributes to an Operation
- 4. Defining the Compute of the Operation
- 1. Hooking up Compute and Strategy with Relay
- 6. Creating a Relay Call Node and Exposing a Python Hook
- 7. Including a Cleaner Python API Hook
- 8. Writing Unit Tests!
- Other Topics
- Adding a Compiler Pass to Relay
- Bring Your Own Codegen To TVM
- Language Reference
- microTVM: TVM on bare-metal
- VTA: Versatile Tensor Accelerator
- Contributor Guide
- TVM Community Guidelines
- Submit a Pull Request
- Code Reviews
- Committer Guide
- Documentation
- Code Guide and Tips
- Git Usage Tips
- Using TVM's CI
- Release Process
- Error Handling Guide
- Publications
- Security Guide
- Index
- API 参考
- vta
- Submodules
- vta.autotvm
- vta.bitstream
- vta.build_module
- vta.environment
- Classes
- Functions
- Module Contents
DevContextDevContext.get_task_qid()DevContext.ALU_OPCODE_ADDDevContext.ALU_OPCODE_MAXDevContext.ALU_OPCODE_MINDevContext.ALU_OPCODE_MULDevContext.ALU_OPCODE_SHRDevContext.DEBUG_NO_SYNCDevContext.MEM_ID_ACCDevContext.MEM_ID_ACC_8BITDevContext.MEM_ID_INPDevContext.MEM_ID_OUTDevContext.MEM_ID_UOPDevContext.MEM_ID_WGTDevContext.QID_COMPUTEDevContext.QID_LOAD_INPDevContext.QID_LOAD_OUTDevContext.QID_LOAD_WGTDevContext.QID_STORE_OUTDevContext.command_handleDevContext.gemmDevContext.vta_axisDevContext.vta_push_uop
EnvironmentEnvironment.__enter__()Environment.__exit__()Environment.ACC_BUFF_SIZEEnvironment.ACC_ELEM_BITSEnvironment.ACC_ELEM_BYTESEnvironment.ACC_WIDTHEnvironment.BATCHEnvironment.BITSTREAMEnvironment.BLOCK_INEnvironment.BLOCK_OUTEnvironment.DEBUG_DUMP_INSNEnvironment.DEBUG_DUMP_UOPEnvironment.DEBUG_SKIP_READ_BARRIEREnvironment.DEBUG_SKIP_WRITE_BARRIEREnvironment.INP_BUFF_SIZEEnvironment.INP_ELEM_BITSEnvironment.INP_ELEM_BYTESEnvironment.INP_WIDTHEnvironment.MAX_XFEREnvironment.MODELEnvironment.OUT_BUFF_SIZEEnvironment.OUT_ELEM_BITSEnvironment.OUT_ELEM_BYTESEnvironment.OUT_WIDTHEnvironment.UOP_BUFF_SIZEEnvironment.WGT_BUFF_SIZEEnvironment.WGT_ELEM_BITSEnvironment.WGT_ELEM_BYTESEnvironment.WGT_WIDTHEnvironment._dev_ctxEnvironment._last_envEnvironment._mock_envEnvironment.acc_dtypeEnvironment.acc_scopeEnvironment.aluEnvironment.cfg_dictEnvironment.currentEnvironment.devEnvironment.dma_copyEnvironment.gemmEnvironment.inp_dtypeEnvironment.inp_scopeEnvironment.mockEnvironment.mock_modeEnvironment.out_dtypeEnvironment.pkgEnvironment.targetEnvironment.target_hostEnvironment.target_vta_cpuEnvironment.wgt_dtypeEnvironment.wgt_scope
_init_env()get_env()get_vta_hw_path()pkg_config()
- vta.exec
- vta.intrin
- vta.libinfo
- vta.program_bitstream
- vta.rpc_client
- vta.testing
- vta.top
- vta.transform
- Attributes
- Classes
- Functions
- Package Contents
EnvironmentEnvironment.__enter__()Environment.__exit__()Environment.ACC_BUFF_SIZEEnvironment.ACC_ELEM_BITSEnvironment.ACC_ELEM_BYTESEnvironment.ACC_WIDTHEnvironment.BATCHEnvironment.BITSTREAMEnvironment.BLOCK_INEnvironment.BLOCK_OUTEnvironment.DEBUG_DUMP_INSNEnvironment.DEBUG_DUMP_UOPEnvironment.DEBUG_SKIP_READ_BARRIEREnvironment.DEBUG_SKIP_WRITE_BARRIEREnvironment.INP_BUFF_SIZEEnvironment.INP_ELEM_BITSEnvironment.INP_ELEM_BYTESEnvironment.INP_WIDTHEnvironment.MAX_XFEREnvironment.MODELEnvironment.OUT_BUFF_SIZEEnvironment.OUT_ELEM_BITSEnvironment.OUT_ELEM_BYTESEnvironment.OUT_WIDTHEnvironment.UOP_BUFF_SIZEEnvironment.WGT_BUFF_SIZEEnvironment.WGT_ELEM_BITSEnvironment.WGT_ELEM_BYTESEnvironment.WGT_WIDTHEnvironment._dev_ctxEnvironment._last_envEnvironment._mock_envEnvironment.acc_dtypeEnvironment.acc_scopeEnvironment.aluEnvironment.cfg_dictEnvironment.currentEnvironment.devEnvironment.dma_copyEnvironment.gemmEnvironment.inp_dtypeEnvironment.inp_scopeEnvironment.mockEnvironment.mock_modeEnvironment.out_dtypeEnvironment.pkgEnvironment.targetEnvironment.target_hostEnvironment.target_vta_cpuEnvironment.wgt_dtypeEnvironment.wgt_scope
download_bitstream()get_bitstream_path()get_env()module_loader()program_fpga()reconfig_runtime()__version__
- Submodules
- vta