[sparse] Fix allocator initialization (taichi-dev#2010)

taichi/program/program.cpp

@@ -337,7 +337,7 @@ void Program::initialize_runtime_system(StructCompiler *scomp) {
                snodes[i]->id, node_size);
       auto rt = llvm_runtime;
       runtime->call<void *, int, std::size_t>(
-          "runtime_NodeAllocator_initialize", rt, i, node_size);
+          "runtime_NodeAllocator_initialize", rt, snodes[i]->id, node_size);
       TI_TRACE("Allocating ambient element for snode {} (node size {})",
                snodes[i]->id, node_size);
       runtime->call<void *, int>("runtime_allocate_ambient", rt, i, node_size);

tests/python/test_sparse_activate.py

@@ -27,3 +27,30 @@ def func():
     activate()
     func()
     assert s[None] == 32
+
+
+@ti.test(require=ti.extension.sparse)
+def test_non_dfs_snode_order():
+    x = ti.field(dtype=ti.i32)
+    y = ti.field(dtype=ti.i32)
+
+    grid1 = ti.root.dense(ti.i, 1)
+    grid2 = ti.root.dense(ti.i, 1)
+    ptr = grid1.pointer(ti.i, 1)
+    ptr.place(x)
+    grid2.place(y)
+    '''
+    This SNode tree has node ids that do not follow DFS order:
+    S0root
+      S1dense
+        S3pointer
+          S4place<i32>
+      S2dense
+        S5place<i32>
+    '''
+    @ti.kernel
+    def foo():
+        ti.activate(ptr, [0])
+
+    foo()  # Just make sure it doesn't crash
+    ti.sync()