Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

[Bugfix] Add nullptr checking for AttrStmt with coproc_uop_scope attr key #9123

Merged
merged 3 commits into from
Sep 28, 2021

Conversation

syang-ng
Copy link
Contributor

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:

# (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)
(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. 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.

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);
    }
    // ...
}

@syang-ng syang-ng changed the title [Bugfix] Add nullptr checking for AttrStmt with coproc_uop_scope attr key. [Bugfix] Add nullptr checking for AttrStmt with coproc_uop_scope attr key Sep 26, 2021
@syang-ng
Copy link
Contributor Author

Hi all, I also discuss this problem in this post. In short, TVM seems to has an implicit requirement on the value of AttrStmt, with some specific keys. And what the document said is the value can be PrimExpr, so if we pass a tir.const(1) as the value, it will crash during building.

@syang-ng
Copy link
Contributor Author

Could you help review this PR :-) @tqchen @junrushao1994 @kparzysz-quic

@tqchen tqchen merged commit 4251103 into apache:main Sep 28, 2021
AndrewZhaoLuo added a commit to AndrewZhaoLuo/tvm that referenced this pull request Sep 28, 2021
* main: (37 commits)
  [ONNX] [Relay] Dynamic squeeze (apache#9095)
  [Meta Schedule][M3b] Database (apache#9061)
  [Bugfix] Add nullptr checking for `AttrStmt` with `coproc_uop_scope` attr key (apache#9123)
  [Codegen] Swap out analyzer when outlining (apache#9117)
  [CI] bash.sh, build.sh: add option to set the container name and hostname (apache#9110)
  Ensure google-mock is installed and setup (apache#9107)
  Arm(R) Ethos(TM)-U NPU TIR to CS for Conv2D (apache#8811)
  Frontend: add onnx GlobalLpPool op (apache#8845)
  [LLVM] Refactor MakeCallPacked, NFC (apache#9118)
  prevent casting handle to other types (apache#9114)
  fix annotation of tir generic (apache#9119)
  [Relay] Register layout conversion function to more reduce ops (apache#9048)
  Fix the missing `dtype` attribute of `tir.Shuffle` in Python level (apache#9131)
  add `multiply` and remove `subtract` for dnnl json runtime (apache#9120)
  relu of dnnl json runtime only support 4-dims input (apache#9122)
  [Meta Schedule][M3a] SpaceGenerator  (apache#9079)
  [TensorIR][Bugfix] Disallow fusing loops with dependency (apache#9112)
  adding Jorn to reviewers list (apache#9105)
  [BYOC] Fix incorrect conv2d padding handling of `dnnl with c source runtime` (apache#9097)
  [Frontend][TFLite] fix apache#9078 (apache#9099)
  ...
ylc pushed a commit to ylc/tvm that referenced this pull request Sep 29, 2021
@syang-ng syang-ng deleted the check-attr-string-value branch October 26, 2021 06:03
ylc pushed a commit to ylc/tvm that referenced this pull request Jan 13, 2022
Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment
Labels
None yet
Projects
None yet
Development

Successfully merging this pull request may close these issues.

2 participants