[Bug] Cannot find intrinsic declaration, possible type mismatch: llvm.exp

[Cookiee235]

The tvm.tir.exp intrinsic function does not support integer types (e.g., int32) as input, resulting in an InternalError during code generation for LLVM targets. This is inconsistent with libraries like PyTorch and NumPy, which support integer inputs for their exponential functions by implicitly converting them to floating-point types.

Actual behavior

Traceback (most recent call last):
  File "/share_container/LLMFuzz/TirFuzz/bug_tp/tir_exp.py", line 23, in <module>
    tvm.build(mod, target='llvm')
  File "/software/tvm-latest/python/tvm/driver/build_module.py", line 59, in build
    return tvm.tir.build(mod, target, pipeline)
           ^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^
  File "/software/tvm-latest/python/tvm/tir/build.py", line 239, in build
    return tir_to_runtime(host_mod, device_mod_dict, target_host)
           ^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^
  File "/software/tvm-latest/python/tvm/tir/build.py", line 149, in tir_to_runtime
    mhost = codegen_build(mhost_all, target_host)
            ^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^
  File "/software/tvm-latest/python/tvm/tir/build.py", line 131, in codegen_build
    return bf(mod, target)
           ^^^^^^^^^^^^^^^
  File "python/tvm_ffi/cython/function.pxi", line 758, in core.Function.__call__
  File "<unknown>", line 0, in tvm::codegen::LLVMModuleNode::Init(tvm::IRModule const&, tvm::Target const&)
  File "<unknown>", line 0, in tvm::codegen::CodeGenCPU::AddFunction(tvm::GlobalVar const&, tvm::tir::PrimFunc const&)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::AddFunctionInternal(tvm::GlobalVar const&, tvm::tir::PrimFunc const&)
  File "<unknown>", line 0, in tvm::codegen::CodeGenCPU::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenCPU::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::LetStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenCPU::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::LetStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenCPU::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::LetStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::LetStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenCPU::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenCPU::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::LetStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::LetStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::LetStmtNode const*)
  [Previous line repeated 1 more time]
  File "<unknown>", line 0, in tvm::codegen::CodeGenCPU::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenCPU::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::LetStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::LetStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::LetStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenCPU::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenCPU::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::SeqStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenCPU::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenCPU::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenCPU::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenCPU::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenCPU::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::SeqStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenCPU::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenCPU::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenCPU::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenCPU::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::AssertStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::SeqStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenCPU::CreateComputeScope(tvm::tir::AttrStmtNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::CreateSerialFor(llvm::Value*, llvm::Value*, llvm::Value*, tvm::tir::Var const&, tvm::tir::Stmt const&)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitStmt_(tvm::tir::BufferStoreNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::VisitExpr_(tvm::tir::CallNode const*)
  File "<unknown>", line 0, in tvm::codegen::CodeGenLLVM::CreateIntrinsic(tvm::tir::CallNode const*)
  File "<unknown>", line 0, in tvm::runtime::detail::LogFatal::~LogFatal() [clone .constprop.0]
  File "<unknown>", line 0, in tvm::runtime::detail::LogFatal::Entry::Finalize()
tvm.error.InternalError: Check failed: (f) is false: Cannot find intrinsic declaration, possible type mismatch: llvm.exp

Environment

tvm-0.22.dev0

Steps to reproduce

import tvm

tir_str = """# from tvm.script import ir as I
# from tvm.script import tir as T

@I.ir_module
class Module:
    @T.prim_func
    def main(x: T.Buffer((128,), "int32"), compute: T.Buffer((128,), "int32")):
        T.func_attr({"target": T.target({"keys": ["cpu"], "kind": "llvm", "mtriple": "x86_64-unknown-linux-gnu", "tag": ""}), "tir.noalias": True})
        # with T.block("root"):
        for i0 in range(128):
            with T.block("compute"):
                v_i0 = T.axis.spatial(128, i0)
                T.reads(x[v_i0])
                T.writes(compute[v_i0])
                compute[v_i0] = T.exp(x[v_i0])
"""

mod = tvm.script.from_source(tir_str)
mod.show()
with tvm.transform.PassContext(0):
    tvm.build(mod, target='llvm')

Triage

• needs-triage
• tir