1.TVM安裝

這部分之前就寫過，為了方便，這里再復(fù)制一遍。

首先下載代碼

git clone --recursive https://github.com/dmlc/tvm

這個(gè)地方最好使用--recursive選項(xiàng)，不然會(huì)缺dlpack這些庫，原因是

子模組 'HalideIR' (https://github.com/dmlc/HalideIR) 未對路徑 '3rdparty/HalideIR' 注冊
子模組 'dlpack' (https://github.com/dmlc/dlpack) 未對路徑 '3rdparty/dlpack' 注冊
子模組 'dmlc-core' (https://github.com/dmlc/dmlc-core) 未對路徑 '3rdparty/dmlc-core' 注冊
子模組 '3rdparty/rang' (https://github.com/agauniyal/rang) 未對路徑 '3rdparty/rang' 注冊

sudoapt-get update

sudoapt-get install-y python python-dev python-setuptools gcclibtinfo-dev zlib1g-dev

創(chuàng)建要編譯生成so的文件夾（該文件夾位于tvm源碼目錄下，與src同級(jí)）

mkdir build

拷貝一份官方的cmake文件進(jìn)行（測試時(shí)先使用官方的，之后這個(gè)config.camke文件我們要進(jìn)行修改以支持更多的設(shè)備）

cp cmake/config.cmake build

修改該文件，這里我們的服務(wù)器上支持CUDA和LLVM的環(huán)境，因此將這兩個(gè)配置打開

set(USE_CUDA OFF)

set(USE_LLVM OFF)

修改為：

set(USE_CUDA ON)

set(USE_LLVM ON)

修改好配置文件后，進(jìn)行編譯。因?yàn)樾薷牧藘蓚€(gè)編譯選項(xiàng)，因此首先需要cmake重新生成Makefile，以后每次新添加了文件和文件夾，一定要重新cmake，否則文件很可能沒有編譯。

cd build

好像最新版本編譯出來的默認(rèn)不是debug版本，為了保險(xiǎn)，手動(dòng)選擇Debug選項(xiàng)

cmake-DCMAKE_BUILD_TYPE=Debug ..

make -j4

上邊的三個(gè)步驟，非常關(guān)鍵，建議不要隨便改變

在無錯(cuò)誤編譯完成后，build目錄下形成了libtvm*.so之類的文件，我們因?yàn)橐薷膖vm，所以不建議移動(dòng)這些so文件到python目錄下，建議添加響應(yīng)的配置。具體配置如下：

在.bashrc文件中，添加

export TVM_PATH=path/to/tvm

export PYTHONPATH=$TVM_PATH/python:$TVM_PATH/topi/python:$TVM_PATH/nnvm/python:${PYTHONPATH}

2.TVM測試代碼

在安裝完成后，進(jìn)入python的命令行，使用

>>>import tvm

測試tvm使用可以使用

報(bào)錯(cuò)找不到module的原因可能是，配置不對，或者配置沒有生效。簡單的查看配置路徑是否正確，可以按照以下命令進(jìn)行

>>>import sys

>>>sys.path

查看python尋找module的文件夾，以排查錯(cuò)誤。

下邊列出生成CUDA代碼的的一段python測試程序，程序來源于tvm.ai，略有改動(dòng)

生成的結(jié)果如下：

import tvm
import numpy as np
import timeit
import pdb
import os

raw_input(os.getpid())
# The size of the matrix
# (M, K) x (K, N)
# You are free to try out different shapes, sometimes TVM optimization outperforms numpy with MKL.
M = 1024
K = 1024
N = 1024

# The default tensor type in tvm
dtype = "float32"

# using Intel AVX2(Advanced Vector Extensions) ISA for SIMD
# To get the best performance, please change the following line
# to llvm -mcpu=core-avx2, or specific type of CPU you use
tgt = "cuda"
tgt_host="llvm"

n = tvm.var("n")
A = tvm.placeholder((n,), name='A')
B = tvm.placeholder((n,), name='B')
C = tvm.compute(A.shape, lambda i: A[i] + B[i], name="C")
s = tvm.create_schedule(C.op)
#pdb.set_trace()
bx, tx = s[C].split(C.op.axis[0], factor=64)
if tgt == "cuda":
    s[C].bind(bx, tvm.thread_axis("blockIdx.x"))
    s[C].bind(tx, tvm.thread_axis("threadIdx.x"))

fadd = tvm.build(s, [A, B, C], tgt, target_host=tgt_host, name="myadd")
ctx = tvm.context(tgt, 0)

n = 1024
a = tvm.nd.array(np.random.uniform(size=n).astype(A.dtype), ctx)
b = tvm.nd.array(np.random.uniform(size=n).astype(B.dtype), ctx)
c = tvm.nd.array(np.zeros(n, dtype=C.dtype), ctx)
# fadd(a, b, c)
# tvm.testing.assert_allclose(c.asnumpy(), a.asnumpy() + b.asnumpy())

if tgt == "cuda":
    dev_module = fadd.imported_modules[0]
    print("-----GPU code-----")
    print(dev_module.get_source())
else:
    print(fadd.get_source())

3.gdb與pdb配置

Pdb一般是隨著python安裝包進(jìn)行安裝的，如果使用pdb命令失敗，可以重新安裝python。

上邊的測試程序有可能會(huì)失敗，報(bào)缺失decorator的錯(cuò)誤，這是一個(gè)python的wheel，需要手動(dòng)安裝。

先說pdb的調(diào)試，pdb調(diào)試與gdb使用方式類似，都是使用pdb xxx.py進(jìn)行。這個(gè)時(shí)候，程序會(huì)自動(dòng)運(yùn)行到程序的第一行。之后使用命令進(jìn)行，網(wǎng)上相關(guān)的文檔非常多，不再進(jìn)行贅述。

這里介紹下python pdb特有的一種調(diào)試方法，在源碼中可以使用

import pdb

pdb.set_trace()

然后使用python xxx.py運(yùn)行程序，程序會(huì)自動(dòng)斷在pdb.set_trace()那一行，從該行起開始調(diào)試，這里僅作為介紹。

因?yàn)閠vm是一個(gè)使用python接口，但是大部分實(shí)現(xiàn)是使用的C++的開源包。存在很多python和C++的交互，因此調(diào)試tvm的過程需要python和C++的聯(lián)合調(diào)試工具。就是python代碼需要pdb，C++代碼需要gdb。因此后邊介紹gdb對C++代碼的調(diào)試。

由于進(jìn)入的時(shí)候是python代碼，因此想要使用gdb下斷點(diǎn)非常困難。我們需要使用gdb附加進(jìn)程的方式進(jìn)行。這個(gè)過程需要root的支持。但是目前使用的Ubuntu系統(tǒng)中沒有root用戶。使用以下命令來添加：

echo "0" | sudo tee /proc/sys/kernel/yama/ptrace_scope

為了方便調(diào)試，我在剛才的測試代碼中加入了一個(gè)

import os

raw_input(os.getpid())

來方便進(jìn)行調(diào)試。

4.調(diào)試過程

使用命令運(yùn)行程序

jourluohua@jour:~/work/python/tvm$ python gen_cuda.py

19143

19143是運(yùn)行該python程序時(shí)的pid

在另一個(gè)窗口中使用gdb attach該pid劫持python程序的線程

jourluohua@jour:~$ gdb attach 19143

然后對想要進(jìn)行下斷點(diǎn)的函數(shù)或者行進(jìn)行下斷點(diǎn)

(gdb) b tvm::codegen::CodeGenCUDA::PrintType

Breakpoint 1 at 0x7f22ae4e9df0 (2 locations)

在gdb所在的窗口使用c命令使程序執(zhí)行起來

(gdb) c

Continuing.

然后再python對應(yīng)的窗口輸入回車，繼續(xù)執(zhí)行，就會(huì)斷到斷點(diǎn)所在的位置。

總結(jié)

以上是生活随笔為你收集整理的TVM调试指南的全部內(nèi)容，希望文章能夠幫你解決所遇到的問題。

如果覺得生活随笔網(wǎng)站內(nèi)容還不錯(cuò)，歡迎將生活随笔推薦給好友。