HLS Backend Example

HLS Backend Example#

TVM supports Xilinx FPGA board with SDAccel. Here is a tutorial for how to deploy TVM to AWS F1 FPGA instance.

备注

This feature is still experimental. We cannot use SDAccel to deploy an end to end neural networks for now.

We use two python scripts for this tutorial.

build.py - a script to synthesize FPGA bitstream.

import tvm
from tvm import te

tgt= tvm.target.Target("sdaccel", host="llvm")

n = te.var("n")
A = te.placeholder((n,), name='A')
B = te.placeholder((n,), name='B')
C = te.compute(A.shape, lambda i: A[i] + B[i], name="C")

s = te.create_schedule(C.op)
px, x = s[C].split(C.op.axis[0], nparts=1)

s[C].bind(px, tvm.te.thread_axis("pipeline"))

fadd = tvm.build(s, [A, B, C], tgt, name="myadd")

fadd.save("myadd.o")
fadd.imported_modules[0].save("myadd.xclbin")

tvm.contrib.cc.create_shared("myadd.so", ["myadd.o"])

run.py - a script to use FPGA as an accelerator.

import tvm
import numpy as np
import os

tgt="sdaccel"

fadd = tvm.runtime.load_module("myadd.so")
if os.environ.get("XCL_EMULATION_MODE"):
    fadd_dev = tvm.runtime.load_module("myadd.xclbin")
else:
    fadd_dev = tvm.runtime.load_module("myadd.awsxclbin")
fadd.import_module(fadd_dev)

dev = tvm.device(tgt, 0)

n = 1024
a = tvm.nd.array(np.random.uniform(size=n).astype("float32"), dev)
b = tvm.nd.array(np.random.uniform(size=n).astype("float32"), dev)
c = tvm.nd.array(np.zeros(n, dtype="float32"), dev)

fadd(a, b, c)
tvm.testing.assert_allclose(c.numpy(), a.numpy() + b.numpy())

Setup#

Launch an instance using the FPGA Developer AMI. We don't need an F1 instance for emulation and synthesis, so it is recommended to use a lower cost instance for them.

Setup AWS FPGA development kit.

git clone https://github.com/aws/aws-fpga.git
cd aws-fpga
source sdaccel_setup.sh
source ${XILINX_SDX}/settings64.sh

Setup TVM with OpenCL enabled.

Emulation#

Create emconfig.json for emulation.

emconfigutil --platform ${AWS_PLATFORM} --nd 1

Copy emconfig.json to the python binary directory. It is because the current Xilinx toolkit assumes that both host binary and the emconfig.json file are in the same path.
```
cp emconfig.json $(dirname $(which python))
```

Run software emulation

export XCL_EMULATION_MODE=1
export XCL_TARGET=sw_emu

python build.py
python run.py

Run hardware emulation

export XCL_EMULATION_MODE=1
export XCL_TARGET=hw_emu

python build.py
python run.py

Synthesis#

Run synthesis with the following script.

unset XCL_EMULATION_MODE
export XCL_TARGET=hw

python build.py

Create AWS FPGA image and upload it to AWS S3.

${SDACCEL_DIR}/tools/create_sdaccel_afi.sh \
    -xclbin=myadd.xclbin -o=myadd \
    -s3_bucket=<bucket-name> -s3_dcp_key=<dcp-folder-name> \
    -s3_logs_key=<logs-folder-name>

This also generates an awsxclbin file, which is necessary to use the AWS FPGA image on F1 instances.

Run#

Launch Amazon EC2 F1 instance.
Copy myadd.so, myadd.awsxclbin, and run.py to the F1 instance.