Failed to export to C code. Any method to export sparse data structure to C?

[xuhao1]

Describe the bug
I am trying to export the taichi kernels to C code because my own taichi project TaichiSLAM requires massive real-time IO.
I followed the instructions and tried to record my kernel, it raise

[I 05/20/21 14:57:28.509] [action_recorder.cpp:start_recording@27] ActionRecorder: start recording to [./export/TaichiSLAM.yml]
[Taichi] Starting on arch=cc
[Tina] Hint: MMB to orbit, Shift+MMB to pan, wheel to zoom
[Taichi] materializing...
[E 05/20/21 14:57:30.417] [struct_cc.cpp:generate_types@36] SNodeType=pointer not supported on C backend

Sparse data structure is improtant to my SLAM project (that's why I trying to write them in Taichi), and the IO speed of python too slow for a real-world aerial robot.

Does Taichi has the plan to support SNodeType=pointer on C backend or I have another way to call Taichi kernels in C/C++?

To Reproduce

git clone https://github.com/xuhao1/TaichiSLAM
python examples/TaichiSLAM_demo.py --record

# My data structure
  B = ti.root.pointer(ti.ijk, (block_num_xy, block_num_xy, block_num_z))
  B = B.dense(ti.ijk, (block_size, block_size, block_size))
  B.place(self.occupy, self.W_TSDF,self.TSDF, self.ESDF)
....
# Record to yml
if args.record:
    ti.core.start_recording('./export/TaichiSLAM.yml')
    ti.init(arch=ti.cc)
else:
    if args.cuda:
        ti.init(arch=ti.cuda)
    else:
        ti.init(arch=ti.cpu)

Log/Screenshots

(base) ➜  TaichiSLAM git:(main) ✗ python examples/TaichiSLAM_demo.py --record          
[Taichi] mode=release
[Taichi] preparing sandbox at /tmp/taichi-8bhk1fyu
[Taichi] version 0.7.19, llvm 10.0.0, commit 19356808, linux, python 3.8.5
Failed to load Python extension for LZ4 support. LZ4 compression will not be available.
[Tina] version 0.1.1
[Tina] Taichi properties hacked

Res [1024x768] GPU False RVIZ False scale of map [100, 10] grid 0.05 
[I 05/20/21 15:05:34.745] [action_recorder.cpp:start_recording@27] ActionRecorder: start recording to [./export/TaichiSLAM.yml]
[Taichi] Starting on arch=cc
The map voxel is:[2048x2048x256] all 1.02e+03M grid scale [0.049x0.049x0.039] map scale:[100mx100mx10m] tree depth [11, 8]
[Tina] Hint: MMB to orbit, Shift+MMB to pan, wheel to zoom
No data input, using random generate maps
[Taichi] materializing...
[E 05/20/21 15:05:36.469] [struct_cc.cpp:generate_types@36] SNodeType=pointer not supported on C backend


***********************************
* Taichi Compiler Stack Traceback *
***********************************
/home/xuhao/anaconda3/lib/python3.8/site-packages/taichi/core/../lib/taichi_core.so: taichi::Logger::error(std::__cxx11::basic_string<char, std::char_traits<char>, std::allocator<char> > const&, bool)
/home/xuhao/anaconda3/lib/python3.8/site-packages/taichi/core/../lib/taichi_core.so: taichi::lang::cccp::CCLayoutGen::generate_types(taichi::lang::SNode*)
/home/xuhao/anaconda3/lib/python3.8/site-packages/taichi/core/../lib/taichi_core.so: taichi::lang::cccp::CCLayoutGen::generate_children(taichi::lang::SNode*)
/home/xuhao/anaconda3/lib/python3.8/site-packages/taichi/core/../lib/taichi_core.so: taichi::lang::cccp::CCLayoutGen::generate_types(taichi::lang::SNode*)
/home/xuhao/anaconda3/lib/python3.8/site-packages/taichi/core/../lib/taichi_core.so: taichi::lang::cccp::CCLayoutGen::compile()
/home/xuhao/anaconda3/lib/python3.8/site-packages/taichi/core/../lib/taichi_core.so: taichi::lang::cccp::CCProgram::compile_layout(taichi::lang::SNode*)
/home/xuhao/anaconda3/lib/python3.8/site-packages/taichi/core/../lib/taichi_core.so: taichi::lang::Program::materialize_layout()
/home/xuhao/anaconda3/lib/python3.8/site-packages/taichi/core/../lib/taichi_core.so: taichi::lang::layout(std::function<void ()> const&)
/home/xuhao/anaconda3/lib/python3.8/site-packages/taichi/core/../lib/taichi_core.so(+0x40a1e9) [0x7fe4818431e9]
/home/xuhao/anaconda3/lib/python3.8/site-packages/taichi/core/../lib/taichi_core.so(+0x35fda4) [0x7fe481798da4]
python(PyCFunction_Call+0x56) [0x56471736cf76]
python(_PyObject_MakeTpCall+0x22f) [0x56471732a85f]
python(_PyEval_EvalFrameDefault+0x4596) [0x5647173b1f56]
python(_PyEval_EvalCodeWithName+0x7df) [0x564717377f9f]
python(_PyFunction_Vectorcall+0x1e3) [0x564717378943]
python(+0x10075e) [0x5647172ed75e]
python(_PyEval_EvalCodeWithName+0x7df) [0x564717377f9f]
python(+0x18bd20) [0x564717378d20]
python(+0x10011a) [0x5647172ed11a]
python(_PyEval_EvalCodeWithName+0x2d2) [0x564717377a92]
python(_PyFunction_Vectorcall+0x1e3) [0x564717378943]
python(_PyObject_FastCallDict+0x24b) [0x5647173794cb]
python(_PyObject_Call_Prepend+0x63) [0x564717379733]
python(+0x18c83a) [0x56471737983a]
python(PyObject_Call+0x70) [0x56471732a200]
python(_PyEval_EvalFrameDefault+0x1fdb) [0x5647173af99b]
python(_PyEval_EvalCodeWithName+0x2d2) [0x564717377a92]
python(_PyObject_FastCallDict+0x20c) [0x56471737948c]
python(_PyObject_Call_Prepend+0x63) [0x564717379733]
python(+0x18c83a) [0x56471737983a]
python(_PyObject_MakeTpCall+0x22f) [0x56471732a85f]
python(_PyEval_EvalFrameDefault+0x4596) [0x5647173b1f56]
python(_PyEval_EvalCodeWithName+0x2d2) [0x564717377a92]
python(PyEval_EvalCodeEx+0x44) [0x564717378754]
python(PyEval_EvalCode+0x1c) [0x564717406edc]
python(+0x219f84) [0x564717406f84]
python(+0x24c1f4) [0x5647174391f4]
python(PyRun_FileExFlags+0xa1) [0x5647173016e1]
python(PyRun_SimpleFileExFlags+0x3b4) [0x564717301ac6]
python(+0x11598b) [0x56471730298b]
python(Py_BytesMain+0x39) [0x56471743bd19]
/lib/x86_64-linux-gnu/libc.so.6: __libc_start_main
python(+0x1dee93) [0x5647173cbe93]

Internal error occurred. Check out this page for possible solutions:
https://taichi.readthedocs.io/en/stable/install.html#troubleshooting

[k-ye]

Thanks! Supporting pointer on the CC backend is not on our near roadmap yet. We can try to figure sth out regarding how users can better utilize the LLVM backend's compiled results.

[xuhao1]

@k-ye Thanks for your reply! Directly using LLVM's compiled result is more reasonable than use C as a intermediate representation . Do you have any roadmap on this?

[k-ye]

Hey @xuhao1 !

Basically, you can dump the .ll file using print_kernel_llvm_ir and print_kernel_llvm_ir_optimized (see also https://taichi.readthedocs.io/en/latest/contributor_guide.html#tips-on-the-taichi-compiler-development), then convert the ll to an actual library via LLVM CLI tools.

Inside the .ll file, you should be able to find functions like cpu_parallel_range_for and cpu_parallel_range_for_task (if you choose the ti.cpu backend), among various other arcane stuff... . Basically, cpu_parallel_range_for is the entry point of the CPU-backend's thread loop, and the body parameter is the LLVM IR generated from your @ti.kernel.

taichi/taichi/runtime/llvm/runtime.cpp

Lines 1278 to 1287 in a07b5a5

	void cpu_parallel_range_for(Context *context,
	int num_threads,
	int begin,
	int end,
	int step,
	int block_dim,
	range_for_xlogue prologue,
	RangeForTaskFunc *body,
	range_for_xlogue epilogue,
	std::size_t tls_size) {

Sorry I know this is not a great answer and lacks lots of details. But if you start playing around with a simple taichi kernel, it might become intuitive to figure out what's going on... We are trying to implement something that can dump your taichi kernels into a file and load it separately, please stay tuned! :-)

[xuhao1]

@k-ye Well I am taking a look of this method and trying to understand the LLVM layer of taichi. Also hoping we can see your new easy-to-use dumping tool soon~
Thanks for your excellent work on taichi again.