tvm.driver¶
Namespace for driver APIs
-
tvm.
lower
(sch, args, name='main', binds=None, simple_mode=False)¶ Lowering step before build into target.
- Parameters
sch (tvm.te.schedule.Schedule) – The schedule to be built
args (list of Buffer or Tensor or Var) – The argument lists to the function.
name (str, optional) – The name of result function.
binds (dict of
Tensor
toBuffer
, optional) – Dictionary that maps the Tensor to Buffer which specified the data layout requirement of the function. By default, a new compact buffer is created for each tensor in the argument.simple_mode (bool, optional) – Whether only output simple and compact statement, this will skip LoopPartition, api wrapper generation and Unrolling.
- Returns
m – The result IRModule, if simple_mode=False Then the Stmt before make api is returned.
- Return type
-
tvm.
build
(inputs, args=None, target=None, target_host=None, name='default_function', binds=None)¶ Build a function with arguments as signature. Code will be generated for devices coupled with target information.
- Parameters
inputs (tvm.te.Schedule, IRModule, or dict of target to IRModule) – The schedule to be built
args (list of Buffer or Tensor or Var, optional) – The argument lists to the function.
target (str or
tvm.target.Target
, optional) – The target and option of the compilation.target_host (str or
tvm.target.Target
optional) – Host compilation target, if target is device. When TVM compiles device specific program such as CUDA, we also need host(CPU) side code to interact with the driver setup the dimensions and parameters correctly. target_host is used to specify the host side codegen target. By default, llvm is used if it is enabled, otherwise a stackvm intepreter is used.name (str, optional) – The name of result function.
binds (dict, optional) – Dictionary that maps the binding of symbolic buffer to Tensor. By default, a new buffer is created for each tensor in the argument.
- Returns
ret – A module that combines both host and device code.
- Return type
tvm.module
Examples
There are two typical example uses of this function depending on the type of the argument inputs: 1. it is an IRModule.
n = 2 A = te.placeholder((n,), name='A') B = te.placeholder((n,), name='B') C = te.compute(A.shape, lambda *i: A(*i) + B(*i), name='C') s = tvm.te.create_schedule(C.op) m = tvm.lower(s, [A, B, C], name="test_add") rt_mod = tvm.build(m, target="llvm")
it is a dict of compilation target to IRModule.
n = 2 A = te.placeholder((n,), name='A') B = te.placeholder((n,), name='B') C = te.compute(A.shape, lambda *i: A(*i) + B(*i), name='C') s1 = tvm.te.create_schedule(C.op) with tvm.target.cuda() as cuda_tgt: s2 = topi.cuda.schedule_injective(cuda_tgt, [C]) m1 = tvm.lower(s1, [A, B, C], name="test_add1") m2 = tvm.lower(s2, [A, B, C], name="test_add2") rt_mod = tvm.build({"llvm": m1, "cuda": m2}, target_host="llvm")
Note
See the note on
tvm.target
on target string format.