syang-ng opened a new pull request #9123:
URL: https://github.com/apache/tvm/pull/9123
## Bug Description
I found there will be a segfault if we try to pass other types of
parameters(except String) to `tir.AttrStmt` when `attr_key` is equal to
`coproc_uop_scope`.
Here is the code to trigger the segmentation fault error:
```python
# (tvm-build) ➜ cat test_coproc_uop_scope.py
import tvm
from tvm import ir, tir
a = tir.Var("a", "int32")
b = tir.Var("b", "handle")
iter_var = tir.IterVar(ir.Range(0,1 ), a, 1)
buffer = tir.buffer.decl_buffer((1,))
buffer_map = {b: buffer}
store = tir.Store(buffer.data, tir.const(1), tir.const(1))
attr_stmt = tir.AttrStmt(iter_var, "coproc_uop_scope", tir.const(1), store)
f = tir.PrimFunc({a, b}, body=attr_stmt, buffer_map=buffer_map)
mod = tvm.lower(f)
tvm.build(mod)
```
```shell
(tvm-build) ➜ python3 test_coproc_uop_scope.py
[1] 28685 segmentation fault (core dumped) python3
test_coproc_uop_scope.py
```
## Bug Analysis
The bug located in
[src/target/llvm/codegen_cpu.cc](https://github.com/apache/tvm/blob/main/src/target/llvm/codegen_cpu.cc#L944).
It assumes the value is a valid pointer after casting, but actually, it can be
invalid. And we can find this function will check whether the value is a valid
pointer for other `attr_key`, such as `tir::attr::pragma_import_llvm`. So to
prevent this crash happen, I add a check after casting the `value` to
`StringImmNode` to verify its validity.
```c++
void CodeGenCPU::VisitStmt_(const AttrStmtNode* op) {
if (op->attr_key == tir::attr::coproc_uop_scope) {
this->CreateStaticInit(op->value.as<StringImmNode>()->value, op->body);
}
// ...
} else if (op->attr_key == tir::attr::pragma_import_llvm) {
const StringImmNode* value = op->value.as<StringImmNode>();
ICHECK(value != nullptr);
this->HandleImport(value->value);
this->VisitStmt(op->body);
}
// ...
}
```
--
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
To unsubscribe, e-mail: [email protected]
For queries about this service, please contact Infrastructure at:
[email protected]
