From 595122ed7a00bc9462e3f648def38772327cc3ce Mon Sep 17 00:00:00 2001 From: Tianqi Chen Date: Fri, 30 Mar 2018 17:46:10 -0700 Subject: [PATCH] [DOCS] Try upgrade build (#1066) --- Jenkinsfile | 39 +--------- python/tvm/_ffi/function.py | 6 +- python/tvm/api.py | 1 + tests/ci_build/Dockerfile.emscripten | 6 +- tests/ci_build/Dockerfile.gpu | 9 ++- tests/ci_build/install/ubuntu_install_core.sh | 2 +- tests/ci_build/install/ubuntu_install_llvm.sh | 5 ++ .../ci_build/install/ubuntu_install_opencl.sh | 6 +- .../ci_build/install/ubuntu_install_sphinx.sh | 2 +- .../python/topi/cuda/conv2d_transpose_nchw.py | 1 + topi/python/topi/cuda/dense.py | 1 + topi/python/topi/cuda/depthwise_conv2d.py | 2 + topi/python/topi/cuda/pooling.py | 2 + topi/python/topi/cuda/reduction.py | 2 + topi/python/topi/mali/dense.py | 1 + topi/python/topi/mali/depthwise_conv2d.py | 1 + topi/python/topi/nn/bnn.py | 1 + topi/python/topi/opengl/conv2d_nchw.py | 1 + topi/python/topi/opengl/dense.py | 1 + topi/python/topi/opengl/pooling.py | 2 + topi/python/topi/rasp/depthwise_conv2d.py | 1 + topi/python/topi/x86/binary_dense.py | 1 + .../deployment/cross_compilation_and_rpc.py | 71 ++++++++++--------- 23 files changed, 78 insertions(+), 86 deletions(-) diff --git a/Jenkinsfile b/Jenkinsfile index b667359f0f2b1..7fb8e6b8b5d1f 100644 --- a/Jenkinsfile +++ b/Jenkinsfile @@ -87,7 +87,6 @@ stage('Build') { cp make/config.mk . echo USE_CUDNN=1 >> config.mk echo USE_CUDA=1 >> config.mk - echo USE_OPENCL=1 >> config.mk echo USE_OPENGL=1 >> config.mk echo LLVM_CONFIG=llvm-config-4.0 >> config.mk echo USE_RPC=1 >> config.mk @@ -105,6 +104,7 @@ stage('Build') { sh "mv lib/libtvm.so lib/libtvm_llvm60.so" pack_lib('gpu', tvm_multilib) sh """ + echo USE_OPENCL=1 >> config.mk echo USE_ROCM=1 >> config.mk echo ROCM_PATH=/opt/rocm >> config.mk echo USE_VULKAN=1 >> config.mk @@ -152,31 +152,6 @@ stage('Build') { pack_lib('i386', tvm_multilib) } } - }, - 'web': { - node('emcc') { - ws('workspace/tvm/build-weblib') { - init_git() - sh """ - cp make/config.mk . - echo USE_CUDA=0 >> config.mk - echo USE_OPENCL=0 >> config.mk - echo LLVM_CONFIG=llvm-config >> config.mk - echo USE_RPC=0 >> config.mk - """ - sh "${docker_run} emscripten echo testing javascript..." - timeout(time: max_time, unit: 'MINUTES') { - try { - sh "${docker_run} emscripten ./tests/scripts/task_web_build.sh" - } catch (exc) { - echo 'Incremental compilation failed. Fall back to build from scratch' - sh "${docker_run} emscripten make clean" - sh "${docker_run} emscripten ./tests/scripts/task_web_build.sh" - } - } - pack_lib('weblib', tvm_lib) - } - } } } @@ -256,18 +231,6 @@ stage('Integration Test') { } } }, - 'web': { - node('emcc') { - ws('workspace/tvm/it-weblib') { - init_git() - unpack_lib('weblib', tvm_lib) - sh "${docker_run} emscripten echo testing javascript..." - timeout(time: max_time, unit: 'MINUTES') { - sh "${docker_run} emscripten ./tests/scripts/task_web_test.sh" - } - } - } - }, 'docs': { node('GPU' && 'linux') { ws('workspace/tvm/docs-python-gpu') { diff --git a/python/tvm/_ffi/function.py b/python/tvm/_ffi/function.py index e0f85be6f1a92..cfda2a35f9b98 100644 --- a/python/tvm/_ffi/function.py +++ b/python/tvm/_ffi/function.py @@ -181,10 +181,10 @@ def register(myf): myf = convert_to_tvm_func(myf) check_call(_LIB.TVMFuncRegisterGlobal( c_str(func_name), myf.handle, ioverride)) + return myf if f: - register(f) - else: - return register + return register(f) + return register def get_global_func(name, allow_missing=False): diff --git a/python/tvm/api.py b/python/tvm/api.py index 66c154bc9f001..b827cce72896b 100644 --- a/python/tvm/api.py +++ b/python/tvm/api.py @@ -652,6 +652,7 @@ def _make_reduce(expr, axis, where=None): for i in range(size)) return outputs[0] if size == 1 else outputs + # pylint: disable=keyword-arg-before-vararg def reducer(expr, axis, where=None, *args): if isinstance(axis, (_schedule.IterVar, list, tuple)): assert not args diff --git a/tests/ci_build/Dockerfile.emscripten b/tests/ci_build/Dockerfile.emscripten index 59bf02ea7d2cc..b4d5a63c52eff 100644 --- a/tests/ci_build/Dockerfile.emscripten +++ b/tests/ci_build/Dockerfile.emscripten @@ -15,4 +15,8 @@ RUN bash /install/ubuntu_install_emscripten.sh COPY install/ubuntu_install_python_package.sh /install/ubuntu_install_python_package.sh RUN bash /install/ubuntu_install_python_package.sh -RUN cp /root/.emscripten /emsdk-portable/ \ No newline at end of file +RUN chmod a+rwx -R /emsdk-portable +RUN cp -r /emsdk-portable /emsdk-portable-backup +RUN mv /emsdk-portable /emsdk-portable-x +RUN mv /emsdk-portable-backup /emsdk-portable +RUN cp /root/.emscripten /emsdk-portable/ diff --git a/tests/ci_build/Dockerfile.gpu b/tests/ci_build/Dockerfile.gpu index e49e498b8d406..4b461ebf19c68 100644 --- a/tests/ci_build/Dockerfile.gpu +++ b/tests/ci_build/Dockerfile.gpu @@ -1,7 +1,6 @@ FROM nvidia/cuda:8.0-cudnn7-devel # Base scripts -RUN apt-get update --fix-missing COPY install/ubuntu_install_core.sh /install/ubuntu_install_core.sh RUN bash /install/ubuntu_install_core.sh @@ -12,9 +11,6 @@ RUN bash /install/ubuntu_install_python.sh COPY install/ubuntu_install_llvm.sh /install/ubuntu_install_llvm.sh RUN bash /install/ubuntu_install_llvm.sh -COPY install/ubuntu_install_opencl.sh /install/ubuntu_install_opencl.sh -RUN bash /install/ubuntu_install_opencl.sh - COPY install/ubuntu_install_iverilog.sh /install/ubuntu_install_iverilog.sh RUN bash /install/ubuntu_install_iverilog.sh @@ -40,8 +36,11 @@ RUN bash /install/ubuntu_install_rocm.sh COPY install/ubuntu_install_opengl.sh /install/ubuntu_install_opengl.sh RUN bash /install/ubuntu_install_opengl.sh +COPY install/ubuntu_install_opencl.sh /install/ubuntu_install_opencl.sh +RUN bash /install/ubuntu_install_opencl.sh + # Enable doxygen for c++ doc build -RUN apt-get install -y doxygen graphviz +RUN apt-get update && apt-get install -y doxygen graphviz # Install vulkan COPY install/ubuntu_install_vulkan.sh /install/ubuntu_install_vulkan.sh diff --git a/tests/ci_build/install/ubuntu_install_core.sh b/tests/ci_build/install/ubuntu_install_core.sh index 9823ae0788ac6..efc69c946b978 100644 --- a/tests/ci_build/install/ubuntu_install_core.sh +++ b/tests/ci_build/install/ubuntu_install_core.sh @@ -1,5 +1,5 @@ # install libraries for building c++ core on ubuntu -apt-get install -y --no-install-recommends --force-yes \ +apt-get update && apt-get install -y --no-install-recommends --force-yes \ git make libgtest-dev cmake wget unzip libtinfo-dev libz-dev\ libcurl4-openssl-dev libopenblas-dev g++ sudo diff --git a/tests/ci_build/install/ubuntu_install_llvm.sh b/tests/ci_build/install/ubuntu_install_llvm.sh index e5b28b911f61e..ba0afcd18cc9d 100644 --- a/tests/ci_build/install/ubuntu_install_llvm.sh +++ b/tests/ci_build/install/ubuntu_install_llvm.sh @@ -8,6 +8,11 @@ echo deb http://apt.llvm.org/xenial/ llvm-toolchain-xenial-5.0 main\ echo deb-src http://apt.llvm.org/xenial/ llvm-toolchain-xenial-5.0 main\ >> /etc/apt/sources.list.d/llvm.list +echo deb http://apt.llvm.org/xenial/ llvm-toolchain-xenial-6.0 main\ + >> /etc/apt/sources.list.d/llvm.list +echo deb-src http://apt.llvm.org/xenial/ llvm-toolchain-xenial-6.0 main\ + >> /etc/apt/sources.list.d/llvm.list + echo deb http://apt.llvm.org/xenial/ llvm-toolchain-xenial main\ >> /etc/apt/sources.list.d/llvm.list echo deb-src http://apt.llvm.org/xenial/ llvm-toolchain-xenial main\ diff --git a/tests/ci_build/install/ubuntu_install_opencl.sh b/tests/ci_build/install/ubuntu_install_opencl.sh index 636236539a984..ca4d1d04fd5c2 100644 --- a/tests/ci_build/install/ubuntu_install_opencl.sh +++ b/tests/ci_build/install/ubuntu_install_opencl.sh @@ -1,8 +1,8 @@ # Install OpenCL runtime in nvidia docker. -apt-get install -y --no-install-recommends --force-yes \ - ocl-icd-libopencl1 \ +apt-get update && apt-get install -y --no-install-recommends --force-yes \ + ocl-icd-opencl-dev \ clinfo && \ - rm -rf /var/lib/apt/lists/* + rm -rf /var/lib/apt/lists/* mkdir -p /etc/OpenCL/vendors && \ echo "libnvidia-opencl.so.1" > /etc/OpenCL/vendors/nvidia.icd diff --git a/tests/ci_build/install/ubuntu_install_sphinx.sh b/tests/ci_build/install/ubuntu_install_sphinx.sh index 767643f104886..30d4596aabf29 100644 --- a/tests/ci_build/install/ubuntu_install_sphinx.sh +++ b/tests/ci_build/install/ubuntu_install_sphinx.sh @@ -1 +1 @@ -pip install sphinx==1.6.2 sphinx-gallery sphinx_rtd_theme matplotlib Image commonmark>=0.7.3 docutils>=0.11 +pip install sphinx sphinx-gallery sphinx_rtd_theme matplotlib Image commonmark>=0.7.3 docutils>=0.11 diff --git a/topi/python/topi/cuda/conv2d_transpose_nchw.py b/topi/python/topi/cuda/conv2d_transpose_nchw.py index edd255a8fac03..0d439bfdfdeaa 100644 --- a/topi/python/topi/cuda/conv2d_transpose_nchw.py +++ b/topi/python/topi/cuda/conv2d_transpose_nchw.py @@ -74,6 +74,7 @@ def schedule(temp, Filter, Output): conv2d_56_64_64(s, Filter, temp_S, Filter_S, Out, Out_L) def traverse(OP): + """Internal travserse function""" # inline all one-to-one-mapping operators except the last stage (output) if tag.is_injective(OP.tag): if OP not in s.outputs: diff --git a/topi/python/topi/cuda/dense.py b/topi/python/topi/cuda/dense.py index 6207c14220d05..7c62fab743f5b 100644 --- a/topi/python/topi/cuda/dense.py +++ b/topi/python/topi/cuda/dense.py @@ -87,6 +87,7 @@ def _schedule(Dense): s[Out].set_store_predicate(thread_x.var.equal(0)) def traverse(OP): + """Internal travserse function""" # inline all one-to-one-mapping operators except the last stage (output) if tag.is_broadcast(OP.tag): if OP not in s.outputs: diff --git a/topi/python/topi/cuda/depthwise_conv2d.py b/topi/python/topi/cuda/depthwise_conv2d.py index 851a00db0a481..3045774296d02 100644 --- a/topi/python/topi/cuda/depthwise_conv2d.py +++ b/topi/python/topi/cuda/depthwise_conv2d.py @@ -102,6 +102,7 @@ def _schedule(PaddedInput, Filter, DepthwiseConv2d): s[FS].bind(tx, thread_x) def traverse(OP): + """Internal travserse function""" # inline all one-to-one-mapping operators except the last stage (output) if tag.is_broadcast(OP.tag): if OP not in s.outputs: @@ -178,6 +179,7 @@ def _schedule(temp, Filter, DepthwiseConv2d): s[FS].bind(fused, thread_x) def traverse(OP): + """Internal travserse function""" # inline all one-to-one-mapping operators except the last stage (output) if tag.is_broadcast(OP.tag): if OP not in s.outputs: diff --git a/topi/python/topi/cuda/pooling.py b/topi/python/topi/cuda/pooling.py index 4ed5ae66c19b9..95ef3ce67c5cf 100644 --- a/topi/python/topi/cuda/pooling.py +++ b/topi/python/topi/cuda/pooling.py @@ -47,6 +47,7 @@ def _schedule(Pool): s[Pool].compute_at(s[Out], tx) def traverse(OP): + """Internal travserse function""" # inline all one-to-one-mapping operators except the last stage (output) if tag.is_broadcast(OP.tag): if OP not in s.outputs: @@ -101,6 +102,7 @@ def _schedule(PaddedInput, Pool): s[Pool].compute_at(s[Out], tx) def traverse(OP): + """Internal travserse function""" # inline all one-to-one-mapping operators except the last stage (output) if tag.is_broadcast(OP.tag): if OP not in s.outputs: diff --git a/topi/python/topi/cuda/reduction.py b/topi/python/topi/cuda/reduction.py index 932f2aae30988..601d11ab9a2ff 100644 --- a/topi/python/topi/cuda/reduction.py +++ b/topi/python/topi/cuda/reduction.py @@ -87,6 +87,7 @@ def schedule_reduce(outs): sch = tvm.create_schedule([x.op for x in outs]) def traverse_before_reduce(operator): + """Internal travserse function""" if isinstance(operator, tvm.tensor.PlaceholderOp): return elif tag.is_injective(operator.tag): @@ -97,6 +98,7 @@ def traverse_before_reduce(operator): raise RuntimeError("Unsupported operator: %s" % operator.tag) def traverse_after_reduce(operator): + """Internal travserse function""" if tag.is_broadcast(operator.tag): raise RuntimeError("Not yet support ewise after reduce") elif operator.tag == 'comm_reduce': diff --git a/topi/python/topi/mali/dense.py b/topi/python/topi/mali/dense.py index d3edeafed3b3f..ff88ce51866fa 100644 --- a/topi/python/topi/mali/dense.py +++ b/topi/python/topi/mali/dense.py @@ -82,6 +82,7 @@ def fuse_and_bind(s, tensor, axis=None, num_thread=None): # print(tvm.lower(s, [data, weight, bias, outs[0]], simple_mode=True)) def traverse(OP): + """Internal travserse function""" # inline all one-to-one-mapping operators except the last stage (output) if tag.is_broadcast(OP.tag): if OP not in s.outputs: diff --git a/topi/python/topi/mali/depthwise_conv2d.py b/topi/python/topi/mali/depthwise_conv2d.py index 46ce7f747def9..4281405505287 100644 --- a/topi/python/topi/mali/depthwise_conv2d.py +++ b/topi/python/topi/mali/depthwise_conv2d.py @@ -87,6 +87,7 @@ def tile_and_bind3d(tensor, z, y, x, z_factor=2, y_factor=None, x_factor=None): s[conv].compute_at(s[output], ji) def traverse(op): + """Internal travserse function""" # inline all one-to-one-mapping operators except the last stage (output) if tag.is_broadcast(op.tag): if op not in s.outputs: diff --git a/topi/python/topi/nn/bnn.py b/topi/python/topi/nn/bnn.py index 39b9d2a15a1b7..591a082e7d306 100644 --- a/topi/python/topi/nn/bnn.py +++ b/topi/python/topi/nn/bnn.py @@ -43,6 +43,7 @@ def _binarize_pack(*indices): if j == 31: return packed packed = packed << 1 + raise RuntimeError("not resach") return tvm.compute(oshape, _binarize_pack, name=name, tag='binarize_pack') diff --git a/topi/python/topi/opengl/conv2d_nchw.py b/topi/python/topi/opengl/conv2d_nchw.py index c633d8a21e6ea..7e8b7275f75db 100644 --- a/topi/python/topi/opengl/conv2d_nchw.py +++ b/topi/python/topi/opengl/conv2d_nchw.py @@ -31,6 +31,7 @@ def _schedule(conv2d, data): s[data].opengl() def traverse(OP): + """Internal travserse function""" # inline all one-to-one-mapping operators except the last stage (output) if tag.is_broadcast(OP.tag): if OP not in s.outputs: diff --git a/topi/python/topi/opengl/dense.py b/topi/python/topi/opengl/dense.py index e7cf008ae2400..e4d327afa4d62 100644 --- a/topi/python/topi/opengl/dense.py +++ b/topi/python/topi/opengl/dense.py @@ -31,6 +31,7 @@ def _schedule(Dense): s[Out].opengl() def traverse(OP): + """Internal travserse function""" # inline all one-to-one-mapping operators except the last stage (output) if tag.is_broadcast(OP.tag): if OP not in s.outputs: diff --git a/topi/python/topi/opengl/pooling.py b/topi/python/topi/opengl/pooling.py index 5c26c56bb1ac2..dc4f47609f923 100644 --- a/topi/python/topi/opengl/pooling.py +++ b/topi/python/topi/opengl/pooling.py @@ -30,6 +30,7 @@ def _schedule(Pool): s[Out].opengl() def traverse(OP): + """Internal travserse function""" # inline all one-to-one-mapping operators except the last stage (output) if tag.is_broadcast(OP.tag): if OP not in s.outputs: @@ -75,6 +76,7 @@ def _schedule(PaddedInput, Pool): s[Out].opengl() def traverse(OP): + """Internal travserse function""" # inline all one-to-one-mapping operators except the last stage (output) if tag.is_broadcast(OP.tag): if OP not in s.outputs: diff --git a/topi/python/topi/rasp/depthwise_conv2d.py b/topi/python/topi/rasp/depthwise_conv2d.py index a6fd691f843f9..b7cb6570987a7 100644 --- a/topi/python/topi/rasp/depthwise_conv2d.py +++ b/topi/python/topi/rasp/depthwise_conv2d.py @@ -164,6 +164,7 @@ def schedule_depthwise_conv2d(outs): s = tvm.create_schedule([x.op for x in outs]) def traverse(op): + """Internal travserse function""" # inline all one-to-one-mapping operators except the last stage (output) if tag.is_broadcast(op.tag): if op not in s.outputs: diff --git a/topi/python/topi/x86/binary_dense.py b/topi/python/topi/x86/binary_dense.py index 11fccba7b6446..8b28dd7288428 100644 --- a/topi/python/topi/x86/binary_dense.py +++ b/topi/python/topi/x86/binary_dense.py @@ -35,6 +35,7 @@ def _schedule(A, B, C): s[Out].vectorize(xi) def traverse(OP): + """Internal travserse function""" # inline all one-to-one-mapping operators except the last stage (output) if tag.is_broadcast(OP.tag): if OP not in s.outputs: diff --git a/tutorials/deployment/cross_compilation_and_rpc.py b/tutorials/deployment/cross_compilation_and_rpc.py index ccaf9e79e3de3..f06bbfca64077 100644 --- a/tutorials/deployment/cross_compilation_and_rpc.py +++ b/tutorials/deployment/cross_compilation_and_rpc.py @@ -108,8 +108,6 @@ import numpy as np from tvm.contrib import rpc, util -server = rpc.Server(host='0.0.0.0', port=9090, use_popen=True) - ###################################################################### # Declare and Cross Compile Kernel on Local Machine # ------------------------------------------------- @@ -241,47 +239,52 @@ # But here we set 'llvm' to enable this tutorial to run locally. # # Also we need to build the runtime with the flag `USE_OPENCL=1`. - # build kernel (different from cpu, we need bind axis for OpenCL) -s = tvm.create_schedule(B.op) -xo, xi = s[B].split(B.op.axis[0], factor=32) -s[B].bind(xo, tvm.thread_axis("blockIdx.x")) -s[B].bind(xi, tvm.thread_axis("threadIdx.x")) -f = tvm.build(s, [A, B], "opencl", target_host="llvm", name="myadd") +# +# The following functions shows how we can deploy CL +def deploy_cl(): + s = tvm.create_schedule(B.op) + xo, xi = s[B].split(B.op.axis[0], factor=32) + s[B].bind(xo, tvm.thread_axis("blockIdx.x")) + s[B].bind(xi, tvm.thread_axis("threadIdx.x")) + f = tvm.build(s, [A, B], "opencl", target_host="llvm", name="myadd") -# save files -path_o = temp.relpath("myadd.o") -path_cl = temp.relpath("myadd.cl") -path_json = temp.relpath("myadd.tvm_meta.json") -f.save(path_o) -f.imported_modules[0].save(path_cl) + # save files + path_o = temp.relpath("myadd.o") + path_cl = temp.relpath("myadd.cl") + path_json = temp.relpath("myadd.tvm_meta.json") + f.save(path_o) + f.imported_modules[0].save(path_cl) -# upload files -remote.upload(path_o) -remote.upload(path_cl) -remote.upload(path_json) + # upload files + remote.upload(path_o) + remote.upload(path_cl) + remote.upload(path_json) -# load files on remote device -fhost = remote.load_module("myadd.o") -fdev = remote.load_module("myadd.cl") -fhost.import_module(fdev) + # load files on remote device + fhost = remote.load_module("myadd.o") + fdev = remote.load_module("myadd.cl") + fhost.import_module(fdev) + + # run + ctx = remote.cl(0) + a = tvm.nd.array(np.random.uniform(size=1024).astype(A.dtype), ctx) + b = tvm.nd.array(np.zeros(1024, dtype=A.dtype), ctx) + fhost(a, b) + np.testing.assert_equal(b.asnumpy(), a.asnumpy() + 1) -# run -ctx = remote.cl(0) -a = tvm.nd.array(np.random.uniform(size=1024).astype(A.dtype), ctx) -b = tvm.nd.array(np.zeros(1024, dtype=A.dtype), ctx) -fhost(a, b) -np.testing.assert_equal(b.asnumpy(), a.asnumpy() + 1) ##################################################################### # Instead of uploading files separately, there is a more convinient way. # You can export libraray as a tar ball. -path_tar = temp.relpath("myadd.tar") -f.export_library(path_tar) -remote.upload(path_tar) -fhost = remote.load_module("myadd.tar") -fhost(a, b) -np.testing.assert_equal(b.asnumpy(), a.asnumpy() + 1) +# The following functions shows how we can deploy by tar ball +def deploy_cl_by_tar(): + path_tar = temp.relpath("myadd.tar") + f.export_library(path_tar) + remote.upload(path_tar) + fhost = remote.load_module("myadd.tar") + fhost(a, b) + np.testing.assert_equal(b.asnumpy(), a.asnumpy() + 1) # terminate the server after experiment server.terminate()