[CRT] [AOT] '__tvm_module_ctx' undeclared when building Module using the CRT runtime

Troubleshooting
#1

As explained in this post I am currently trying to build a TFLite model which uses the topi.vision.non_max_supression operator. As explained in the post, I found a workaround to be able to build the model, but now there are problems in the generated C code.

TVM is generating the following code as part of the non_max_suppression operator:

if (TVMBackendGetFuncFromEnv(__tvm_module_ctx, "tvm.contrib.sort.argsort_nms", &tvm_contrib_sort_argsort_nms_packed) != 0) {
      return -1;
    }

When I try to compile this code, I get that __tvm_module_ctx is not declared. As far as I can tell, this is because there is a specific case where this variable is generated (see file src/target/source/codegen_c_host.cc). According to this, this __tvm_module_ctx is only generated when the AOT executor is defined (it is) and when the runtime name is Cpp (it is not).

So this is the first question: can the CRT runtime also be supported here?

I also noticed that the TVMBackendGetFuncFromEnv is not in the generated code. As far as i understand, this line of code that is generating trouble is generated because of the te.extern and the tvm.tir.call_packed in the argsort declared in tvm/topi/sort.py.

So this is the second question: can this argsort be generated without calling call_packed and te.extern?

#2

I’m facing the same issue. It’s been a while. Has anyone found a solution or has any insights on how to resolve this? :thinking:

According to https://github.com/apache/tvm/blob/d7e0af2d88f75e2ab21c6dbde43813a033c0fb35/src/runtime/c_runtime_api.cc#L522 All we need is to link a runtime module, which contains the function we need. (In my case, the module is an OperatorModule, and the functions are __tvm_set_device and my opencl kernel default_function_kernel_packed.

But I have no idea about how.

Using ‘c’ as the target often encounters some obstacles. (see AttributeError: Module has no function ‘tvm_main’ - Questions - Apache TVM Discuss).

Is there any challenge preventing us from treating ‘c’ the same as ‘llvm’?

#3

Particularly in my case, I solved it by implementing my own argsort version, to avoid the call of TVMBackendGetFuncFromEnv.

There is probably a correct way of solving it, my version was extremely hacky.

1 Like