apache / apache/tvm

[Bug] inconsistent results for the CPU and CUDA targets.

Open
#17,965 3 comments 0 reactions 0 assignees View on GitHub
needs-triage type: bug
Dominant language
Python
Stars
13.7k
Forks
4k
Avg merge
2d 19h
Merged PRs (30d)
111

Description

### Expected behavior

TVM should output consistent results for the CPU and GPU targets.

### Actual behavior
For the following model:

![Image](https://github.com/user-attachments/assets/a95460b1-d289-4885-9d82-8ef27f993927)

when compile the model for the CPU target, the output is:
```c
cpu: [[[[nan nan nan]
[nan nan nan]
[nan nan nan]]

[[nan nan nan]
[nan nan nan]
[nan nan nan]]

[[nan nan nan]
[nan nan nan]
[nan nan nan]]]]
```

However, when the target is CUDA, the output is:
```c
gpu: [[[[ 9.5653236e-01 8.9820576e-01 8.9820576e-01]
[ 9.5653236e-01 -3.4028231e+38 -3.4028231e+38]
[-3.4028231e+38 -3.4028231e+38 -3.4028231e+38]]

[[ 9.5653236e-01 8.9820576e-01 8.9820576e-01]
[ 9.5653236e-01 -3.4028231e+38 -3.4028231e+38]
[-3.4028231e+38 -3.4028231e+38 -3.4028231e+38]]

[[ 9.5653236e-01 8.9820576e-01 8.9820576e-01]
[ 9.5653236e-01 -3.4028231e+38 -3.4028231e+38]
[-3.4028231e+38 -3.4028231e+38 -3.4028231e+38]]]]
```

### Environment

OS: Ubuntu 20.04
TVM: 0.21.dev0(bcb68b130)
CUDA: 11.8
GPU: NVIDIA GeForce RTX 3080

### Steps to reproduce
This bug can be reproduced by the following code with the model in the attachment.
```python
import sys

import numpy as np
import onnx
import onnxruntime

import tvm
import tvm.testing
from tvm import relax
from tvm.relax.frontend.onnx import from_onnx

import pickle


def main():
onnx_model = onnx.load("a249.onnx")

with open("inputs.pkl", "rb") as fp:
inputs = pickle.load(fp)

# Convert the onnx model into relax through the onnx importer.
tvm_model = from_onnx(onnx_model, keep_params_in_input=True)
# Convert operators for inference mode.
tvm_model = relax.transform.DecomposeOpsForInference()(tvm_model)
# Legalize any relax ops into tensorir.
tvm_model = relax.transform.LegalizeOps()(tvm_model)

# Separate model from parameters.
tvm_model, params = relax.frontend.detach_params(tvm_model)

# Prepare inputs.
input_list = [
inputs[key.name_hint] for key in tvm_model["main"].params if key.name_hint in inputs
]
if params:
input_list += params["main"]

# Compile the relax graph into a VM then run.
#----------------------cpu-----------------------
with tvm.transform.PassContext(opt_level=0):
ex = relax.build(tvm_model, target="llvm")
vm = relax.VirtualMachine(ex, tvm.cpu())

# Run model and check outputs.
vm.set_input("main", *input_list)
vm.invoke_stateful("main")
tvm_cpu_output = vm.get_outputs("main")

print("cpu: ", tvm_cpu_output)
#----------------------cpu-----------------------

#----------------------cuda-----------------------
with tvm.target.Target("cuda"):
tvm_model = tvm.tir.transform.DefaultGPUSchedule()(tvm_model)

with tvm.transform.PassContext(opt_level=3):
ex = tvm.compile(tvm_model, target="cuda")
vm1 = relax.VirtualMachine(ex, tvm.cuda())

vm1.set_input("main", *input_list)
vm1.invoke_stateful("main")
tvm_gpu_output = vm1.get_outputs("main")

print("gpu: ", tvm_gpu_output)
#----------------------cuda-----------------------

if __name__ == "__main__":

main()

```

[testcase.zip](https://github.com/user-attachments/files/20180366/testcase.zip)

### Triage

* needs-triage

Contributor guide

No contributing guide indexed for this repository

Research direction

Start by running the provided Python reproduction with testcase.zip on the stated Ubuntu, TVM, CUDA, and GPU environment. Compare the outputs after from_onnx, DecomposeOpsForInference, LegalizeOps, and DefaultGPUSchedule for the LLVM and CUDA targets. Done means the same model produces consistent CPU and GPU results, with a focused regression test if the failing path can be isolated.

Written by the indexing model from the issue text.

Assessment

Tech stack
python
Domain
compilers, machine-learning
Issue type
Bug
Difficulty
4/5
Estimated time
3-5 days
Activity status
Quiet
Clarity
Mostly clear
Newbie friendliness
35/100

Get new issues in your inbox

A short digest of beginner-friendly GitHub issues.