44 Commits

Author SHA1 Message Date
Tianqi Chen 4e5b869c27 [REFACTOR][RUNTIME] Phase out profiling.h heavy types, rename to timer.h (#19455) 2026-04-27 20:27:07 -04:00
Tianqi Chen 9edd5bd958 [REFACTOR] Remove tvm.runtime.packed_func and container shims; route via tvm_ffi (#19442)
## Summary

- Delete the three Python shim modules that re-exported tvm-ffi types
under `tvm.runtime` / `tvm.ir`:
`python/tvm/runtime/packed_func.py`, `python/tvm/runtime/container.py`,
`python/tvm/ir/container.py`.
- Drop the matching re-exports from `tvm.runtime`, `tvm.ir`, and `tvm`
package init files, so
`tvm.runtime.PackedFunc`, `tvm.runtime.ShapeTuple`,
`tvm.runtime.String`, `tvm.ir.Array`,
  `tvm.ir.Map`, and `tvm.container.Array` no longer exist.
- Migrate every productive caller, test, and tutorial to the canonical
names: `tvm_ffi.Function`,
`tvm_ffi.Shape`, `tvm_ffi.core.String`, `tvm_ffi.Array`, and
`tvm_ffi.Map`.

## Test plan

- [x] `pytest tests/python/all-platform-minimal-test` (75 passed, 77
skipped)
- [x] `pytest tests/python/runtime/test_runtime_container.py
tests/python/all-platform-minimal-test/test_runtime_packed_func.py` (20
passed)
- [x] `pytest tests/python/ir/test_node_reflection.py
tests/python/ir/test_container_structural_equal.py` (32 passed)
- [x] `pytest tests/python/relax/test_vm_build.py
tests/python/relax/test_vm_execbuilder.py
tests/python/relax/test_vm_codegen_only.py` (125 passed, 2 xfailed)
- [x] `pytest tests/python/relax/test_runtime_builtin.py
tests/python/relax/test_op_misc.py` (19 passed)
- [x] `pytest tests/python/target/test_target_target.py` (37 passed, 3
skipped)
- [x] `pre-commit run` clean on touched files
2026-04-25 11:02:08 -04:00
Tianqi Chen 9a8320acbd [LINT][PYTHON] Modernize annotations with ruff UP rules (#18830)
This PR enables ruff pyupgrade (UP) rules with py310 target, auto-fixing
~5600 annotation modernizations (PEP 585 generics, PEP 604 unions,
deprecated typing imports).

Also removes from __future__ import annotations from ir/module.py and
rmsnorm.py, bumps requires-python to >=3.10, and removes absolute_import
aliases from topi/contrib files.
2026-02-27 21:29:47 -05:00
Tianqi Chen 33dcea1686 [REFACTOR][LINT] Modernize ruff config (#18810)
This PR removes the extra lint violations from the codebase so lint
aligns with the latest style
2026-02-23 07:29:21 -05:00
Tianqi Chen aa2e609136 [LINT] Modernize lint to use pre-commit hooks (#18807)
This PR migrates existing lint to use pre-commit hooks
2026-02-22 11:03:21 -05:00
Tianqi Chen 543e64dbb1 [FFI][REFACTOR] Cleanup tvm_ffi python API and types (#18277)
This PR cleans up the python API to make things more consistent
with existing python array api and torch.

Device update
- device_id => index, to be consistent with torch
- device_type => dlpack_device_type() returns int
- added type property same as torch.device

API updates:

- Move the convenient method like cpu() out into tvm runtime to keep device minimal
- tvm_ffi._init_api => tvm_ffi.init_ffi_api
- tvm_ffi.register_func => tvm_ffi.register_global_func
2025-09-07 10:38:50 -04:00
Tianqi Chen 3c36ce2ec6 [FFI][REFACTOR][ABI] Rename NDArray to Tensor (#18275)
This PR Updates the NDArray => Tensor.

Both tensor and ndarray are commonly used terms.

Because the term Tensor is getting more common in the context of ML,
we do the rename to stay more aligned with torch.Tensor and DLTensor.
2025-09-06 14:33:59 -07:00
Tianqi Chen a7a0168be5 [FFI][REFACTOR] Establish tvm_ffi python module (#18226)
* [FFI][REFACTOR] Establish tvm_ffi as a standalone python module

This PR establishes tvm_ffi as a standalone python module.
The ffi is structured as a minimal pip module that can be
directly install by path or url.

examples/get_started provided a minimal example.
This is a major change as we are decoupling tvm_ffi as a
separate package, users need to install tvm_ffi separately.

Thanks to its minimal dependency, tvm_ffi can be easily installed
even just from the source by pip install ./ffi

This change would enable future improvement for library plugins
to have lightweight dependencies by just working on top of
the tvm_ffi, while the main compiler toolchain and runtime
can be layered on top.

* [FFI] Improve traceback setups

This PR improves traceback related setups
2025-08-24 15:46:20 -07:00
Ruihang Lai 61e7c8fb63 [Refactor] Rename relax_vm to vm (#18049)
This PR renames the filenames/namespaces of `relax_vm`
to `vm`.

Previously, both VMs of relay and relax exist, and to avoid the
name conflicts, we added the prefix `relax_` to relax VM.
With the Relay runtime being phased out, we can now rename
`relax_vm` to `vm` for conciseness.
2025-06-06 18:54:28 -04:00
Tianqi Chen ccaa534b2c [REFACTOR] Phase out relay python components (#17656)
This PR starts the step 0 to phase out relay from the current
development main branch.  This PR focuses on the python
components of relay, autotvm, auto_scheduler. To make the change
manageable, we will also do followup steps on te.Schedule and
c++ components in followup PRs.

To continue support community members who depends on
legacy flows, the [v0.19.0](https://github.com/apache/tvm/tree/v0.19.0)
branch will continue contain these components.


As noted in [discussion on phasing out legacy components](https://discuss.tvm.apache.org/t/phasing-out-legacy-components/17703/30),
this would help us to do two purposes:

- By removing outdated or redundant elements, we can significantly
reduce complexity and improve maintainability.
- Unify our focus: Concentrating our efforts on the new unity flow
will allow for more efficient development and innovation.

It is also a good opportunity for us to revisit and reduce CI time.
The past relay legacy flow contains a lot of end to end tests that
requires hardware resources to run and causing long CI time.
Moving onwards, we can focus more on unit-tests that focuses
on structural equality and runs within seconds, while be mindful
about tests that requires hardware resources (by restricting them
to specific folders and CI nightly in some cases).

---

Co-authored-by: Siyuan Feng <hzfengsy@sjtu.edu.cn>
2025-02-15 13:48:28 -05:00
Krzysztof Parzyszek 4f041c96e7 [Runtime,RPC] Use f-strings for string formatting, NFC (#14967)
* [Runtime,RPC] Use f-strings for string formatting, NFC

Replace uses of % and .format() with f-strings.

Reformat modified files.

* Fix linter
2023-05-27 16:14:15 +09:00
Ehsan M. Kermani f44ef53a6b [Relay][ONNX] Add Optional, OptionalHasElement and OptionalGetElement ops based on Sequence op (#13189)
[Relay][ONNX] Add Optional, OptionalHasElement and OptionalGetElement based on Sequence op

Co-authored-by: Ehsan M. Kermani <ehsanmok@users.noreply.github.com>
2022-10-25 19:42:48 -07:00
Valery Chernov 3e3d900c66 [Virtual Machine] Implementation of 'set_output_zero_copy' (#11358)
There is python API function 'set_output' which save external outputs in VM outputs_ field (map) for specified func name. It looks like 'set_input' method.
During 'invoke' outputs_ are saved in register_file. For this the register indices of output tensors are found from code_ field. I observed in tests for different models that AllocTensor and AllocADT ops are used for result tensors. Let's consider these two cases: result index is destination for AllocTensor op or AllocADT op. At the first case instead of construction new NDArray the outside output tensor is used. At the second one the fields of AllocADT are analyzed and register indices are extracted. During tests I observed that ReshapeTensor operation is rarely used as final one (SqueezeNet-v1.0 and DUC). Mechanism for replacement by external output tensors was also implemented for this op.
2022-09-29 09:22:35 -07:00
Robert Kimball effcd2251b Add methods to get and set late-bound constants. (#12664)
* Add methods to read and restore late-bound constants on Executable.

* Add bindings for new functions

* Cleanup

* Fix function name

* Add tests for python API to access new load/save functions

* Add another tests for python API to access new load/save functions where there are no constants
2022-09-01 08:57:40 -07:00
Egor Churaev c3c7c4ccc3 [Profiler] Fix graph_executor_debug hang (#12382)
For some operations such as `__nop` or `__copy` the measured inference
time is equal to 0. In this case we are in infinite loop and we won't
exit from it. Added new parameter `limit_zero_time_iterations ` which specify the
maximum number of repeats then the inference time is equal to 0. When
we exceed this value then we will exit from a loop.
2022-08-12 09:26:38 -07:00
Alexey Voronov 6c8a353099 Add cooldown interval logic for the profiling functional (#11465)
* Add cooldown interval logic for the profiling functional.

* Remove string serialize hack from RunIndividual functions

* Update src/runtime/graph_executor/debug/graph_executor_debug.cc

Co-authored-by: Tristan Konolige <tristan.konolige@gmail.com>
2022-06-28 09:10:09 -07:00
Valery Chernov fafabc96c1 [VirtualMachine] Zero copy in set_input when input is DLTensor (#11003)
* method of creating of NDArray from external DLTensor was implemented

* set input without copying for DLTensor source

* code clean up

* update description and comments after review

Co-authored-by: Valery Chernov <valery.chernov@deelvin.com>
2022-04-15 15:31:01 -07:00
mawnja 2cc0451c28 added surpport for arg type of numeric float16 and testcase, fixed the (#10797)
cierror
2022-03-28 06:14:09 +09:00
Valery Chernov d62a364ba7 [VirtualMachine] new method allowing to set one input tensor by its index or name (#10293)
* set_input_with_index was implemented for VM

* clean code

* add getInputIndexFromName. add function descriptions. lint fix

* fix lint

* transfer comparison of parameter names number and assigned devices number to VMFunction constructor

* add GetVMFunctionWithName to Executable API

* clean code

* add SetInputWithName (set_input_with_name) to VM API

* join SetInputWithIndex and SetInputWithName to SetOneInputTensor (set_one_input) to VM API, the joined methods were removed

* fix lint

* some fixes after review

* add set_one_input method to python API of VirtualMachine

* pytests for set_input and set_one_input methods of VirtualMachine were implemented and checked

* CI restart

* construct simple model for pytests by relay instead of onnx tools (need for correct CI)

Co-authored-by: Valery Chernov <valery.chernov@deelvin.com>
2022-02-26 05:46:01 +09:00
Michal Piszczek 3b85f7c768 [VM] Remove undesired arg to load_late_bound_consts (#9870)
* Remove undesired arg to vm exec load_late_bound_consts

* No-op for ci
2022-01-10 11:06:17 -08:00
Mark Shields e7b5d2dde8 [Relay] Support large constants saved/loaded outside of VM executable (#9734)
* [Relay] Support large constants.

This allows constant tensors at or above a given byte limit to be marked as
'late bound' and saved/reloaded to a file independently of the overall
executable. Since the executable is often embedded in the data segment of
generated runtime Modules this avoids problems with external tools which can't
handle multi-gigabyte data segments.

[ACE-466 in OctoML JIRA]

* [checkpoint] fix latent bytecode/code bug
2021-12-14 22:38:41 -08:00
Mark Shields 3047709f2a [Relay] Use LowerTEPass in VM (#9483)
We replace use of the TECompiler::{Lower,LowerShapeFunc} methods from the VM's
compiler.cc with LowerTEPass. This clears the way for performing post-lowering
IRModule->IRModule transformations which combine Relay and TIR analysis. In particular,
it will allow us to use the PlanDevices pass to propagate memory scope constraints
across PrimFuncs.

We run LowerTEPass fairly early in the pipeline, which required quite a few passes
to become 'post-lowering friendly'. In particular, ManifestAlloc is now run after
rather than before lowering, and so must now work in a mixed Function/PrimFunc world.

The "vm.shape_func" operator has been removed since a) lowering has already generated
the necessary dynamic shape function, and b) the call to that function can be
represented by an 'ordinary' vm.invoke_tvm_op call.

We worked our way through the following glitches:
 - Dynamic shape functions are now given their true type (rather than the type of
   the primitive function they are paired with).
 - Lowering was choosing definitional GlobalVars which were not pointer-equal to the
   referential GlobalVars left behind in the rewritten Calls. We fixed that in
   te_compiler.cc, though better would be to push GlobalVars deeper into the
   lowering machinery.
 - device_copy was rewritten to a call to @__copy without any definition. Though we
   tried adding it as a global this (obviously in retrospect...) won't typecheck if
   there are multiple device_copies in the program. Instead leave device_copy unchanged
   during lowering and update each executor codegen to look for them specially.
 - Calls to already-compiled BYOC functions were indistinguishable from calls
   to (non-primitive) Relay functions. We move them into the call_lowered calling
   convention, and leave behind a Function tagged with "ExternalSymbol". Better would
   be a first-class representatn for externals in the IRModule but one step at a time.
 - Functions with dynamic shapes tagged for BYOC compilation were not tracking their
   connection to their dynamic shape function. We now use exactly the same attributes
   as for non-BYOC primitives.
 - VerilatorRuntime can legitimately be deleted before initialized.
 - IRModule attributes must be preserved. In particular, since LowerTEPass can
   be invoked more than once we need to be careful to preserve any existing external
   modules and other attributes gatherd from an earlier LowerTEPass.
 - GetUniqueName accounts for existing definitions in the module, but is not used
   for external functions since their intended names are communicated to the codegen
   toolchain via the already fixed "global_symbol" attribute.
2021-11-30 23:15:07 -08:00
Mark Shields 289bd90cfc Prepare for switching VM to LowerTEPass. (#9550)
This is a grab bag of fallout changes from switching the VM to use LoweTEPass
which can be easily split out of the main #9483 PR.

- AnnotateSpans can be used from C++ (though, unfortunately, it didn't help
  me with debugging since spans are universally dropped in most passes).
- Can get a human readable dump of the VM's PackedFunc names and indexes for
  debugging.
- If TVM_LOG_DEBUG defined then include types and ids of GlobalVars. I had
  a lot of difficulty tracking down where duplicate GlobalVars for the same
  name_hint were getting created and propagated.
- GetCallLoweredProps follows same API as GetDeviceCopy and GetOnDevice
  where will return 'null' properties if call/expr is not of call_lowered
  form. Mildly more convenient, though switching all the above to ICHECK
  and push 'if (op == the relevant op)' into all use sites would also be just
  fine.
- Misc VLOG improvements made while tracking down issues in #9483.
2021-11-24 17:09:53 +00:00
Mark Shields be03d62e5b Switch PlanDevices pass to be w.r.t. SEScopes instead of DLDeviceTypes. (#9326)
* Switch PlanDevices pass to be w.r.t. SEScopes instead of DLDeviceTypes.

CAUTION: Breaking VM executable serialization change. I needed a new 'virtual devices' array in the executable so that instructions can continue to refer to devices by a simple index yet the VM can respect both the device type and id for runtime devices.

Continuing from #9313, and as part of apache/tvm-rfcs#38, we switch PlanDevices to plan with respect to SEScopes instead of just DLDeviceTypes. Our ultimate goal is to be able to flow memory scopes between PrimFuncs by re-running PlanDevices after the LowerTE pass. This PR at least gets us to being able to flow the memory scopes, but the actual changes to PlanDevices to look inside PrimFuncs is still two PR's in the future.

However, we get two nice side effects right away:
 - Since SEScopes contain Targets we can isolate all the device-to-target resolution machinery within PlanDevices (with the help of CompilationConfig). After PlanDevices has run we can retrieve the Target for any sub-expression directly from that sub-expression's SEScope. For now we retain the one-Target-per-DLDeviceType constraint since it baked into the public 'TargetMap' API, but the path to breaking that constraint is clearer.
 - Device ids are now respected all the way from annotation to executor. Previously though we had a bit of plumbing using Devices the device_id therein was ignored or defaulted to zero.

 The Python "on_device" annotation helpers still work w.r.t. devices. Thus though they now respect device ids, they do not allow the user to specify a Target or memory scope as supported by the underlying SEScope.

* [checkpoint] Revert emitter.py, must have run 'black .' by mistake.

* [checkpoint] Address PR comments

Also add back SplitArgs pass in build_module.cc which somehow got lost in the shuffle.

(try again -- flaky test_crt.py test_autotune?)

* [checkpoint] Fix after rebase on CallLowered.
2021-11-12 09:11:12 -08:00
Mark Shields aeff3ea711 BUG: alloc_tensor offset and reshape shape should be on the CPU (#9421)
* BUG: alloc_tensor offset and reshape shape should be on the CPU

The VM ManifestAlloc pass was allocating constants in a few places I
forgot to tag with on_device for the host/cpu. As a result the runtime
would (silently) do the x-device copy, which destroys perf.

To make this easier to spot in the future added a 'constants' property
to the VM Executable to dump the shape & device for all VM constants.

This is CORE-102 in OctoML JIRA.

* [checkpoint] Older compilers can't handle << overload

* [checkpoint] Woops, forgot requires_cuda
2021-11-03 06:22:32 +09:00
Tristan Konolige 37cd9837ff Fix end to end benchmark with rpc devices (#9175)
* Ensure that device used in end to end rpc is a local device

* fix vm; add actually failing tests

* bump roi_align test tolerances
2021-10-05 11:57:35 -07:00
Tristan Konolige c650f9ac15 [PROFILING] Profiling over RPC (#8885)
* [PROFILING] Profiling over RPC

Allow for profiling over RPC by serializing the returned report before
sending it. Also remove collectors argument when profiling over rpc
because it cannot be serialized.

* lint

* fixes

* add comments
2021-09-09 14:53:31 +09:00
Tristan Konolige 55bb8b60b7 [Graph Executor, VM] Add end to end benchmarking of models (#8858)
Add benchmarking that includes ovearhead of transfering inputs and
outputs to and from the device. This should give an accurate measurement
of the runtime a user would see when using the model. This is
accomplished by adding functions that run from inputs to return values
into the graph executor and the VM.
2021-08-27 14:39:03 -07:00
Tristan Konolige f1ca91d4e4 [GRAPH EXECUTOR,VM] Add benchmarking function to graph executor and vm (#8807)
* [GRAPH EXECUTOR,VM] Add benchmarking function to graph executor and vm

This new benchmarking function is just a convenience function for
calling time_evaluator on the underlying module. Hopefully this should
make it easier for users to get good benchmarks of their code.

* formatting

* import order

* more test, more comments, more precision

* fix tests

* add seconds descriptions to doc
2021-08-25 20:25:29 -05:00
Hua Jiang e1bb7ac880 [VM] Add get_input_index support. (#8661) 2021-08-06 17:24:09 +09:00
Chenfan 88dd31b600 [VM] Bug fix for numpy scalar input in vm (#8553)
* Bug fix for numpy scalar input in vm

* Bug fix

* Re-triggle CI

* Update

* Update UT

* Re-triggle CI
2021-07-30 14:11:51 +09:00
Haichen Shen 720e7b1ebd [Refactor] Rename asnumpy -> numpy in NDArray (#8083) 2021-05-21 08:32:27 -04:00
Tristan Konolige f57830bd8b [FIX] Fix RPC for the VM (#7810)
* [FIX] Fix RPC for the VM
2021-04-15 11:03:27 -04:00
Jared Roesch fd18751e68 Add support for using the VM across the RPC boundary. (#7746)
* Get basic verison of VM RPC working

* Test case passes

* Clean up PR

* Lint

* Format

* Address Andrew R and TK feedback

* Add comment for Andrew

* Address Zhi's comment

* Format

* Fix broken test
2021-03-30 02:01:40 -07:00
Haichen Shen fbfeee4ccb [Refactor] Rename TVMContext to Device (#7721) 2021-03-26 07:56:57 -04:00
Yanming Wang 485dfd6a3c Fix typo in relay.vm.Executable (#7543)
Co-authored-by: Yanming Wang <yanmwang@amazon.com>
2021-02-28 05:31:29 +09:00
Tianqi Chen c8064b3ca6 [REFACTOR] Remainings of util => utils (#6778) 2020-10-29 13:46:31 -04:00
Jared Roesch f13fed55cf [Format] Convert all Python code w/o CI (#6448)
* Add black setup

* Tweak pyproject.toml

* Fix syntax issues

* Fix

* Tweak

* Black all Python code
2020-09-11 22:17:24 +09:00
Zhi 1224d56ca9 [RELAY][VM] Enable heterogeneous execution for Relay VM (#6337)
* vm heterogeneous execution

* context analysis on module

* fix profiler

* fix memory plan

* add more unification

* add serialization

* add gpu tests for test_adt

* cache visited functions

* path compression

* C++ context analysis

* remove python context analysis

* add tests

* clean

* lint

* fix

* enable gpu test for dynamic namespace

* remove GetParamsContext

* fix comments and add doc for context analysis

* cache context

* cache allocator

* rebase and fix comments
2020-09-03 09:47:03 -07:00
Haichen Shen 922e0a05c8 [Relay][VM] Allow to config allocator type and refactor vm code structure (#6105)
* [Relay][VM] Allow to config allocator type and refactor vm code structure

* fix doc

* fix

* update

* trigger ci

* trigger ci

* trigger ci

* trigger ci

* fix doc warning
2020-07-24 15:49:45 -07:00
Leon Wang 5d445ca4c6 Fix some typo errors in license header (#5956)
Signed-off-by: leonwanghui <wanghui71leon@gmail.com>
2020-06-29 20:28:30 -07:00
Zhi eacfe89066 [RUNTIME] Introduce MetadataModule to separate code compilation/interpretation and weight initialization (#5770) 2020-06-18 15:18:29 -07:00
Tianqi Chen 9816efc2df [REFACTOR][PY][API-CHANGE] Remove legacy python files. (#4943)
* [REFACTOR][PY][API-CHANGE] Remove legacy python files.

Remove legacy python files.
Use the te namespace for most of the tensor expression primitives.

- tvm.create_schedule -> tvm.te.create_schedule
- tvm.placeholder -> tvm.te.placeholder
- tvm.compute -> tvm.te.compute

* Remove top-level exposures.
2020-02-26 21:10:47 -08:00
Zhi 502cf264d5 [Refactor] move vm.py under runtime and adt to runtime.container.py (#4855) 2020-02-11 12:04:04 -08:00