I have a custom relay operator that computes a few identically shaped te.Tensor
s using te_list = [te.compute(...) for par in input_params]
and then I concatenate using topi.concatenate(te_list, axis=1)
.
This breaks with an llvm verification error when doing kernel tuning and I can’t figure out why:
Found 1 tasks
[Task 1/ 1] Current/Best: 0.00/ 0.00 GFLOPS | Progress: (0/2) | 0.00 sDEBUG:asyncio:Using selector: EpollSelector
INFO:autotvm:Get devices for measurement successfully!
DEBUG:autotvm:No: 1 GFLOPS: 0.00/0.00
result: MeasureResult(costs=(TVMError('
Traceback (most recent call last):
[bt] (5) /home/ubuntu/tvm/build/libtvm.so(TVMFuncCall+0x65) [0x7f72ce9a5495]
[bt] (4) /home/ubuntu/tvm/build/libtvm.so(tvm::runtime::TypedPackedFunc<tvm::runtime::Module (tvm::IRModule, tvm::Target const&)>::AssignTypedLambda<tvm::runtime::Module (*)(tvm::IRModule, tvm::Target const&)>(tvm::runtime::Module (*)(tvm::IRModule, tvm::Target const&))::{lambda(tvm::runtime::TVMArgs const&, tvm::runtime::TVMRetValue*)#1}::operator()(tvm::runtime::TVMArgs const&, tvm::runtime::TVMRetValue*) const+0xef2) [0x7f72ce3d32e2]
[bt] (3) /home/ubuntu/tvm/build/libtvm.so(tvm::codegen::Build(tvm::IRModule, tvm::Target const&)+0x787) [0x7f72ce3d1227]
[bt] (2) /home/ubuntu/tvm/build/libtvm.so(+0xe13b23) [0x7f72ce984b23]
[bt] (1) /home/ubuntu/tvm/build/libtvm.so(tvm::codegen::LLVMModuleNode::Init(tvm::IRModule const&, std::__cxx11::basic_string<char, std::char_traits<char>, std::allocator<char> >)+0x1085) [0x7f72ce988e45]
[bt] (0) /home/ubuntu/tvm/build/libtvm.so(dmlc::LogMessageFatal::~LogMessageFatal()+0x43) [0x7f72cdf74c03]
File "/home/ubuntu/tvm/src/target/llvm/llvm_module.cc", line 230
TVMError: LLVM module verification failed with the following errors:
Invalid operand types for FCmp instruction
%166 = fcmp oeq i8* %32, %165
',),), error_no=2, all_cost=0.10512018203735352, timestamp=1594717891.4763963) [('dummy_knob', 0)],None,0
If I just return te_list[0]
(without concatenating) this step works well, which is why it might be a compatibility issue between topi.concatenate
and te.Tensor
s, or maybe I don’t understand how to use them together.
Any ideas on how to debug this? Thanks!
Edit: same error occurs when using a cascading te.if_then_else
structure instead of topi.concatenate
, which is consistent since topi.concatenate
lowers to if_then_else
.