RuntimeError: [offload.cpp:visit@731] struct_for cannot be nested inside a kernel, stmt=$14

[Cleaf-y]

Environment

macos 14.4 aarch64, [Taichi] version 1.6.0, llvm 15.0.7, commit f1c6fbb, osx, python 3.11.8

Issue

When dealing with SNode, I encounterd a strange problem with the following simplified code:

import taichi as ti

ti.init(arch=ti.cpu)

data = ti.field(dtype=ti.u8)
block = ti.root.pointer(ti.ij, (50, 50))
block.place(data)
data[0, 1] = 1


@ti.kernel
def test():
    for _ in range(100000):
        for i, j in data:
            print(i, j)


test()

Then it gots the error:

[E 03/17/24 20:16:46.217 526947] [offload.cpp:visit@731] struct_for cannot be nested inside a kernel, stmt=$14
Traceback (most recent call last):
File "/Users/shuyechen/LabWork/iFIM/cross_validation.py", line 18, in
test()
File "/Users/shu/miniconda3/envs/fmm4/lib/python3.11/site-packages/taichi/lang/kernel_impl.py", line 974, in wrapped
return primal(*args, **kwargs)
^^^^^^^^^^^^^^^^^^^^^^^
File "/Users/shu/miniconda3/envs/fmm4/lib/python3.11/site-packages/taichi/lang/kernel_impl.py", line 906, in call
return self.runtime.compiled_functionskey
^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^
File "/Users/shuminiconda3/envs/fmm4/lib/python3.11/site-packages/taichi/lang/kernel_impl.py", line 817, in func__
raise e from None
File "/Users/shu/miniconda3/envs/fmm4/lib/python3.11/site-packages/taichi/lang/kernel_impl.py", line 814, in func__
t_kernel(launch_ctx)
RuntimeError: [offload.cpp:visit@731] struct_for cannot be nested inside a kernel, stmt=$14

Could you please help me check this? I initially planning to use the SNode pointer to respresent a list of active points which will be calculated in one iteration. I tried to implement a count_available points function(which goes through all points, checks wheter it is acitve and update the counter and finally return it, looks reaaly dumb) to be used as the while loop condition, but got the same error, and I tried this for loop, still in vain. Maybe I use this in the wrong way? Thanks for your time!

[Ives0721]

In the the-struct-for-statement doc of TaiChi v1.6, it had mention that :

The struct for statement must be at the outermost scope of the kernel, and it cannot be terminated by a break statement even when it is run serially.

So the loop style written in your example code (i.e. A struct-for nested in a range-for) is not allowed in TaiChi v1.6.