Create test statements for various TIR constructs.
()
| 619 | |
| 620 | |
| 621 | def create_test_statements(): |
| 622 | """Create test statements for various TIR constructs.""" |
| 623 | x = tir.Var("x", "int32") |
| 624 | tir.Var("y", "int32") |
| 625 | |
| 626 | # IntImm |
| 627 | int_imm = tir.IntImm("int32", 10) |
| 628 | |
| 629 | # Simple expression |
| 630 | add_expr = tir.Add(x, int_imm) |
| 631 | |
| 632 | # Evaluate |
| 633 | evaluate_stmt = tir.Evaluate(add_expr) |
| 634 | |
| 635 | # Bind + SeqStmt (was LetStmt) |
| 636 | let_stmt = tir.SeqStmt([tir.Bind(x, int_imm), evaluate_stmt]) |
| 637 | |
| 638 | # For loop |
| 639 | for_loop = tir.For(x, 0, 10, tir.ForKind.SERIAL, evaluate_stmt) |
| 640 | |
| 641 | # While loop |
| 642 | while_loop = tir.While(tir.LT(x, int_imm), evaluate_stmt) |
| 643 | |
| 644 | # Buffer operations |
| 645 | buffer_var = tir.Var("buf", "handle") |
| 646 | buffer = tir.decl_buffer((10,), "int32", buffer_var.name) |
| 647 | buffer_store = tir.BufferStore(buffer, add_expr, [int_imm]) |
| 648 | |
| 649 | # Sequence of statements |
| 650 | seq_stmt = tir.SeqStmt([evaluate_stmt, for_loop]) |
| 651 | |
| 652 | # Block with iteration variables |
| 653 | iter_var = tir.IterVar(Range(0, 10), x, 0) |
| 654 | block = tir.SBlock([iter_var], [], [], "block", evaluate_stmt) |
| 655 | block_realize = tir.SBlockRealize([int_imm], tir.IntImm("bool", 1), block) |
| 656 | |
| 657 | # IfThenElse statement |
| 658 | if_then_else = tir.IfThenElse(tir.LT(x, int_imm), evaluate_stmt, evaluate_stmt) |
| 659 | |
| 660 | # Break and continue statements inside a for loop |
| 661 | @T.prim_func |
| 662 | def func(A: T.Buffer((10,), "int32")): |
| 663 | for x in range(10): |
| 664 | A[x] = x + 1 |
| 665 | if x == 5: |
| 666 | break |
| 667 | continue |
| 668 | |
| 669 | # DeclBuffer |
| 670 | buffer_decl = tir.DeclBuffer(T.buffer((10,), "int32"), evaluate_stmt) |
| 671 | |
| 672 | # TilePrimitiveCall — extract the TilePrimitiveCall from the kernel body, then wrap in an SBlock |
| 673 | @T.prim_func |
| 674 | def op_call(A: T.Buffer((10,), "int32"), B: T.Buffer((10,), "int32")): |
| 675 | T.device_entry() |
| 676 | Tx.add(A, B, 1.0) |
| 677 | |
| 678 | # op_call.body is the tirx.device_entry AttrStmt, op_call.body.body is TilePrimitiveCall |
no test coverage detected
searching dependent graphs…