diff --git a/.github/workflows/conda-gpu-build.yaml b/.github/workflows/conda-gpu-build.yaml new file mode 100644 index 0000000000..1f5b9f7fc2 --- /dev/null +++ b/.github/workflows/conda-gpu-build.yaml @@ -0,0 +1,51 @@ +# +# Licensed to the Apache Software Foundation (ASF) under one or more +# contributor license agreements. See the NOTICE file distributed with +# this work for additional information regarding copyright ownership. +# The ASF licenses this file to You under the Apache License, Version 2.0 +# (the "License"); you may not use this file except in compliance with +# the License. You may obtain a copy of the License at +# +# http://www.apache.org/licenses/LICENSE-2.0 +# +# Unless required by applicable law or agreed to in writing, software +# distributed under the License is distributed on an "AS IS" BASIS, +# WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +# See the License for the specific language governing permissions and +# limitations under the License. +# + +name: Conda GPU Build + +on: + push: + pull_request: + workflow_dispatch: + +jobs: + build-gpu-conda-package: + runs-on: ubuntu-latest + + env: + CUDA: "12.2" + DIST: "OFF" + TEST_COMMAND: python -c "import singa; import singa.singa_wrap; print(singa.__version__)" + + steps: + - uses: actions/checkout@v7 + + - name: setup-miniconda + uses: conda-incubator/setup-miniconda@v4.0.1 + with: + auto-update-conda: true + channels: nvidia,conda-forge,nusdbsystem,defaults + channel-priority: flexible + + - name: install-conda-build + shell: bash -el {0} + run: conda install -y conda-build + + - name: build-gpu-package + shell: bash -el {0} + run: | + conda-build tool/conda/singa -c nvidia -c conda-forge -c nusdbsystem -c defaults --python 3.8 --no-anaconda-upload --output-folder conda-bld diff --git a/CMakeLists.txt b/CMakeLists.txt index cc025fe165..be07ae478f 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -119,7 +119,7 @@ IF (USE_CUDA) #LIST(APPEND SINGA_LINKER_LIBS cnmem) SET(global_cuda_objs "") # add support cuda fp16 - SET(CUDA_NVCC_FLAGS "${CUDA_NVCC_FLAGS} --gpu-architecture=compute_75") + SET(CUDA_NVCC_FLAGS "${CUDA_NVCC_FLAGS} --gpu-architecture=compute_75 --std=c++11") ENDIF() # TODO(wangwei) detect the ev lib diff --git a/src/core/tensor/math_kernel.h b/src/core/tensor/math_kernel.h index 668df4d545..16a4ca7177 100644 --- a/src/core/tensor/math_kernel.h +++ b/src/core/tensor/math_kernel.h @@ -24,6 +24,7 @@ #include "singa/singa_config.h" #ifdef USE_CUDA +#include #include #include #include diff --git a/src/model/layer/cudnn_convolution.cc b/src/model/layer/cudnn_convolution.cc index 44e1fef8ba..54324ee473 100644 --- a/src/model/layer/cudnn_convolution.cc +++ b/src/model/layer/cudnn_convolution.cc @@ -19,6 +19,7 @@ #ifdef USE_CUDNN #include #include +#include #include "./cudnn_utils.h" #include "singa/utils/logging.h" @@ -88,6 +89,62 @@ void CudnnConvolution::InitCudnn(const Tensor &input) { channels_, kernel_h_, kernel_w_)); if (prefer_ == "fastest" || prefer_ == "limited_workspace" || prefer_ == "no_workspace") { +#if CUDNN_MAJOR >= 8 + const int max_algos = 16; + int num_fp_alg = 0, num_bp_filt_alg = 0, num_bp_data_alg = 0; + cudnnConvolutionFwdAlgoPerf_t fp_alg_perf[max_algos]; + cudnnConvolutionBwdFilterAlgoPerf_t bp_filt_perf[max_algos]; + cudnnConvolutionBwdDataAlgoPerf_t bp_data_perf[max_algos]; + const size_t workspace_limit = + prefer_ == "fastest" ? std::numeric_limits::max() + : (prefer_ == "limited_workspace" ? + workspace_byte_limit_ : 0); + + CUDNN_CHECK(cudnnGetConvolutionForwardAlgorithm_v7( + ctx->cudnn_handle, x_desc_, filter_desc_, conv_desc_, y_desc_, + max_algos, &num_fp_alg, fp_alg_perf)); + bool found_alg = false; + for (int i = 0; i < num_fp_alg; ++i) { + if (fp_alg_perf[i].status == CUDNN_STATUS_SUCCESS && + fp_alg_perf[i].memory <= workspace_limit) { + fp_alg_ = fp_alg_perf[i].algo; + found_alg = true; + break; + } + } + CHECK(found_alg) << "No cuDNN forward convolution algorithm matched " + << prefer_ << " workspace preference"; + + CUDNN_CHECK(cudnnGetConvolutionBackwardFilterAlgorithm_v7( + ctx->cudnn_handle, x_desc_, y_desc_, conv_desc_, filter_desc_, + max_algos, &num_bp_filt_alg, bp_filt_perf)); + found_alg = false; + for (int i = 0; i < num_bp_filt_alg; ++i) { + if (bp_filt_perf[i].status == CUDNN_STATUS_SUCCESS && + bp_filt_perf[i].memory <= workspace_limit) { + bp_filter_alg_ = bp_filt_perf[i].algo; + found_alg = true; + break; + } + } + CHECK(found_alg) << "No cuDNN backward-filter convolution algorithm matched " + << prefer_ << " workspace preference"; + + CUDNN_CHECK(cudnnGetConvolutionBackwardDataAlgorithm_v7( + ctx->cudnn_handle, filter_desc_, y_desc_, conv_desc_, x_desc_, + max_algos, &num_bp_data_alg, bp_data_perf)); + found_alg = false; + for (int i = 0; i < num_bp_data_alg; ++i) { + if (bp_data_perf[i].status == CUDNN_STATUS_SUCCESS && + bp_data_perf[i].memory <= workspace_limit) { + bp_data_alg_ = bp_data_perf[i].algo; + found_alg = true; + break; + } + } + CHECK(found_alg) << "No cuDNN backward-data convolution algorithm matched " + << prefer_ << " workspace preference"; +#else cudnnConvolutionFwdPreference_t fwd_pref; cudnnConvolutionBwdFilterPreference_t bwd_filt_pref; cudnnConvolutionBwdDataPreference_t bwd_data_pref; @@ -110,10 +167,10 @@ void CudnnConvolution::InitCudnn(const Tensor &input) { CUDNN_CHECK(cudnnGetConvolutionBackwardFilterAlgorithm( ctx->cudnn_handle, x_desc_, y_desc_, conv_desc_, filter_desc_, bwd_filt_pref, workspace_byte_limit_, &bp_filter_alg_)); - // deprecated in cudnn v7 CUDNN_CHECK(cudnnGetConvolutionBackwardDataAlgorithm( ctx->cudnn_handle, filter_desc_, y_desc_, conv_desc_, x_desc_, bwd_data_pref, workspace_byte_limit_, &bp_data_alg_)); +#endif // CUDNN_MAJOR >= 8 } else if (prefer_ == "autotune") { const int topk = 1; int num_fp_alg, num_bp_filt_alg, num_bp_data_alg; diff --git a/src/model/layer/cudnn_rnn.cc b/src/model/layer/cudnn_rnn.cc index eb2bfd3516..2654880b88 100755 --- a/src/model/layer/cudnn_rnn.cc +++ b/src/model/layer/cudnn_rnn.cc @@ -144,7 +144,14 @@ void CudnnRNN::SetRNNDescriptor(shared_ptr dev) { rnn_mode = CUDNN_RNN_TANH; else if (rnn_mode_ == "gru") rnn_mode = CUDNN_GRU; -#if CUDNN_MAJOR <= 5 +#if CUDNN_MAJOR >= 8 + CUDNN_CHECK(cudnnSetRNNDescriptor_v8( + rnn_desc_, CUDNN_RNN_ALGO_STANDARD, rnn_mode, CUDNN_RNN_DOUBLE_BIAS, + direction, input_mode, dtype_, dtype_, CUDNN_DEFAULT_MATH, + static_cast(input_size_), static_cast(hidden_size_), + static_cast(hidden_size_), static_cast(num_stacks_), + dropout_desc_, 0)); +#elif CUDNN_MAJOR <= 5 CUDNN_CHECK(cudnnSetRNNDescriptor(rnn_desc_, hidden_size_, num_stacks_, dropout_desc_, input_mode, direction, rnn_mode, dtype_)); diff --git a/src/model/operation/convolution.cc b/src/model/operation/convolution.cc index 96313d20b1..5096be23e9 100644 --- a/src/model/operation/convolution.cc +++ b/src/model/operation/convolution.cc @@ -23,6 +23,7 @@ #include "convolution.h" #include +#include namespace singa { @@ -502,6 +503,62 @@ CudnnConvHandle::CudnnConvHandle( bp_data_alg = CUDNN_CONVOLUTION_BWD_DATA_ALGO_1; } else if (prefer == "fastest" || prefer == "limited_workspace" || prefer == "no_workspace") { +#if CUDNN_MAJOR >= 8 + const int max_algos = 16; + int num_fp_alg = 0, num_bp_filt_alg = 0, num_bp_data_alg = 0; + cudnnConvolutionFwdAlgoPerf_t fp_alg_perf[max_algos]; + cudnnConvolutionBwdFilterAlgoPerf_t bp_filt_perf[max_algos]; + cudnnConvolutionBwdDataAlgoPerf_t bp_data_perf[max_algos]; + const size_t workspace_limit = + prefer == "fastest" ? std::numeric_limits::max() + : (prefer == "limited_workspace" ? + workspace_byte_limit : 0); + + CUDNN_CHECK(cudnnGetConvolutionForwardAlgorithm_v7( + ctx->cudnn_handle, x_desc, filter_desc, conv_desc, y_desc, max_algos, + &num_fp_alg, fp_alg_perf)); + bool found_alg = false; + for (int i = 0; i < num_fp_alg; ++i) { + if (fp_alg_perf[i].status == CUDNN_STATUS_SUCCESS && + fp_alg_perf[i].memory <= workspace_limit) { + fp_alg = fp_alg_perf[i].algo; + found_alg = true; + break; + } + } + CHECK(found_alg) << "No cuDNN forward convolution algorithm matched " + << prefer << " workspace preference"; + + CUDNN_CHECK(cudnnGetConvolutionBackwardFilterAlgorithm_v7( + ctx->cudnn_handle, x_desc, y_desc, conv_desc, filter_desc, max_algos, + &num_bp_filt_alg, bp_filt_perf)); + found_alg = false; + for (int i = 0; i < num_bp_filt_alg; ++i) { + if (bp_filt_perf[i].status == CUDNN_STATUS_SUCCESS && + bp_filt_perf[i].memory <= workspace_limit) { + bp_filter_alg = bp_filt_perf[i].algo; + found_alg = true; + break; + } + } + CHECK(found_alg) << "No cuDNN backward-filter convolution algorithm matched " + << prefer << " workspace preference"; + + CUDNN_CHECK(cudnnGetConvolutionBackwardDataAlgorithm_v7( + ctx->cudnn_handle, filter_desc, y_desc, conv_desc, x_desc, max_algos, + &num_bp_data_alg, bp_data_perf)); + found_alg = false; + for (int i = 0; i < num_bp_data_alg; ++i) { + if (bp_data_perf[i].status == CUDNN_STATUS_SUCCESS && + bp_data_perf[i].memory <= workspace_limit) { + bp_data_alg = bp_data_perf[i].algo; + found_alg = true; + break; + } + } + CHECK(found_alg) << "No cuDNN backward-data convolution algorithm matched " + << prefer << " workspace preference"; +#else cudnnConvolutionFwdPreference_t fwd_pref; cudnnConvolutionBwdFilterPreference_t bwd_filt_pref; cudnnConvolutionBwdDataPreference_t bwd_data_pref; @@ -524,10 +581,10 @@ CudnnConvHandle::CudnnConvHandle( CUDNN_CHECK(cudnnGetConvolutionBackwardFilterAlgorithm( ctx->cudnn_handle, x_desc, y_desc, conv_desc, filter_desc, bwd_filt_pref, workspace_byte_limit, &bp_filter_alg)); - // deprecated in cudnn v7 CUDNN_CHECK(cudnnGetConvolutionBackwardDataAlgorithm( ctx->cudnn_handle, filter_desc, y_desc, conv_desc, x_desc, bwd_data_pref, workspace_byte_limit, &bp_data_alg)); +#endif // CUDNN_MAJOR >= 8 } else if (prefer == "autotune") { const int topk = 1; int num_fp_alg, num_bp_filt_alg, num_bp_data_alg; diff --git a/src/model/operation/rnn.cc b/src/model/operation/rnn.cc index bc8edfd013..a09ba6d515 100644 --- a/src/model/operation/rnn.cc +++ b/src/model/operation/rnn.cc @@ -111,12 +111,23 @@ void CudnnRNNHandle::init_rnn_desc() { RNNMode = CUDNN_LSTM; else if (mode == 3) RNNMode = CUDNN_GRU; +#if CUDNN_MAJOR >= 8 + CUDNN_CHECK(cudnnSetRNNDescriptor_v8( + rnnDesc, cudnnRNNAlgo, RNNMode, + bias ? CUDNN_RNN_DOUBLE_BIAS : CUDNN_RNN_NO_BIAS, + bidirectional ? CUDNN_BIDIRECTIONAL : CUDNN_UNIDIRECTIONAL, + CUDNN_LINEAR_INPUT, cudnnDataType, cudnnDataType, CUDNN_DEFAULT_MATH, + static_cast(feature_size), static_cast(hidden_size), + static_cast(hidden_size), static_cast(num_layers), + dropoutDesc, 0)); +#else CUDNN_CHECK(cudnnSetRNNDescriptor( ctx->cudnn_handle, rnnDesc, hidden_size, num_layers, dropoutDesc, CUDNN_LINEAR_INPUT, bidirectional ? CUDNN_BIDIRECTIONAL : CUDNN_UNIDIRECTIONAL, RNNMode, cudnnRNNAlgo, // CUDNN_RNN_ALGO_STANDARD, cudnnDataType)); +#endif // CUDNN_MAJOR >= 8 } void CudnnRNNHandle::init_dropout_desc() { /* drop out */ diff --git a/tool/conda/singa/README.md b/tool/conda/singa/README.md index c35a8b4f26..eacb0ff014 100644 --- a/tool/conda/singa/README.md +++ b/tool/conda/singa/README.md @@ -23,10 +23,10 @@ ## Environment variables -We export the CUDA version if SINGA is compiled with CUDA enabled. The cuDNN version is fixed by SINGA and cuDNN is installed from [anaconda cloud](https://anaconda.org/anaconda/cudnn). +We export the CUDA version if SINGA is compiled with CUDA enabled. The value must be in `x.y` form and must match a CUDA variant in `conda_build_config.yaml`. The CUDA and cuDNN packages are installed into the conda-build environment from the configured conda channels. - # for SINGA with GPU, e.g. cuda9.0-cudnn7.3.1 - export CUDA=9.0 + # for SINGA with GPU, e.g. cuda12.2-cudnn8.9.2 + export CUDA=12.2 Then, we export a flag DIST to indicate if SINGA is compiled with distributed training enabled. @@ -39,6 +39,7 @@ We need to export both CUDA and DIST for GPU version. For CPU-only version, we d After exporting the environment variables, we need to add the necessary conda channels + conda config --add channels nvidia conda config --add channels conda-forge conda config --add channels nusdbsystem @@ -46,7 +47,7 @@ Then, we can execute the following commands to compile SINGA and package it conda-build . --python 3.6 -You will see the package path from the screen output, e.g., `xx/yy/singa-1.2.0-cpu.tar.bz2` or `xx/yy/singa-1.2.0-cudnn7.3.1_cuda9.0.tar.bz2`. +You will see the package path from the screen output, e.g., `xx/yy/singa-1.2.0-cpu.tar.bz2` or `xx/yy/singa-1.2.0-cudnn8.9.2_cuda12.2.tar.bz2`. To clean the cache diff --git a/tool/conda/singa/conda_build_config.yaml b/tool/conda/singa/conda_build_config.yaml index 1226855ffe..742827e645 100644 --- a/tool/conda/singa/conda_build_config.yaml +++ b/tool/conda/singa/conda_build_config.yaml @@ -18,16 +18,25 @@ # c_compiler_version: # [linux] - - 5.4 # [linux] + - 11 # [linux] cxx_compiler_version: # [linux] - - 5.4 # [linux] + - 11 # [linux] # https://docs.conda.io/projects/conda-build/en/latest/resources/compiler-tools.html#macos-sdk CONDA_BUILD_SYSROOT: - "/Applications/Xcode_11.7.app/Contents/Developer/Platforms/MacOSX.platform/Developer/SDKs/MacOSX.sdk" # [osx] cudnn: # [linux] + - "8.9.2.26 cuda12_0" # [environ.get("CUDA")=="12.2"] - "7.6.5 cuda10.2_0" # [environ.get("CUDA")=="10.2"] # - "7.6.5 cuda10.0_0" # [environ.get("CUDA")=="10.0"] # - "7.6.5 cuda9.0_0" # [environ.get("CUDA")=="9.0"] +cuda_nvcc: # [linux] + - 12.2.140 # [environ.get("CUDA")=="12.2"] +cuda_libraries: # [linux] + - 12.2.2 # [environ.get("CUDA")=="12.2"] +cuda_cudart: # [linux] + - 12.2.140 # [environ.get("CUDA")=="12.2"] +cuda_cccl: # [linux] + - 12.2.140 # [environ.get("CUDA")=="12.2"] dnnl: - 1.1 python: @@ -40,10 +49,11 @@ nccl: mpich: - 3.3.2 build_str: - - "cudnn7.6.5_cuda10.2" # [environ.get("CUDA")=="10.2"] && [environ.get("DIST")=="OFF"] - - "cudnn7.6.5_cuda10.0" # [environ.get("CUDA")=="10.0"] && [environ.get("DIST")=="OFF"] - - "cudnn7.6.5_cuda9.0" # [environ.get("CUDA")=="9.0"] && [environ.get("DIST")=="OFF"] + - "cudnn8.9.2_cuda12.2" # [environ.get("CUDA")=="12.2" and environ.get("DIST")=="OFF"] + - "cudnn7.6.5_cuda10.2" # [environ.get("CUDA")=="10.2" and environ.get("DIST")=="OFF"] + - "cudnn7.6.5_cuda10.0" # [environ.get("CUDA")=="10.0" and environ.get("DIST")=="OFF"] + - "cudnn7.6.5_cuda9.0" # [environ.get("CUDA")=="9.0" and environ.get("DIST")=="OFF"] - "cpu" # [environ.get("CUDA", "")== ""] - - "cudnn7.6.5_cuda10.2_nccl2.6.4.1_mpich3.3.2" # [environ.get("CUDA")=="10.2"] && [environ.get("DIST")=="ON"] - - "cudnn7.6.5_cuda10.0_nccl2.4.8.1_mpich3.3.2" # [environ.get("CUDA")=="10.0"] && [environ.get("DIST")=="ON"] - - "cudnn7.6.5_cuda9.0_nccl2.4.8.1_mpich3.3.2" # [environ.get("CUDA")=="9.0"] && [environ.get("DIST")=="ON"] + - "cudnn7.6.5_cuda10.2_nccl2.6.4.1_mpich3.3.2" # [environ.get("CUDA")=="10.2" and environ.get("DIST")=="ON"] + - "cudnn7.6.5_cuda10.0_nccl2.4.8.1_mpich3.3.2" # [environ.get("CUDA")=="10.0" and environ.get("DIST")=="ON"] + - "cudnn7.6.5_cuda9.0_nccl2.4.8.1_mpich3.3.2" # [environ.get("CUDA")=="9.0" and environ.get("DIST")=="ON"] diff --git a/tool/conda/singa/meta.yaml b/tool/conda/singa/meta.yaml index e95396d73d..868fb07561 100644 --- a/tool/conda/singa/meta.yaml +++ b/tool/conda/singa/meta.yaml @@ -41,7 +41,7 @@ requirements: build: - {{ compiler('cxx') }} - {{ compiler('c') }} - - cmake >=3.12.2 + - cmake >=3.12.2,<4 - make # [unix] host: @@ -53,6 +53,11 @@ requirements: - numpy >=1.16,<2.0 - pytest - deprecated 1.2.7 + - cuda-nvcc {{ cuda_nvcc }} # ['cuda' in str(build_str)] + - cuda-cccl {{ cuda_cccl }} # ['cuda' in str(build_str)] + - cuda-cudart {{ cuda_cudart }} # ['cuda' in str(build_str)] + - cuda-cudart-dev {{ cuda_cudart }} # ['cuda' in str(build_str)] + - cuda-libraries-dev {{ cuda_libraries }} # ['cuda' in str(build_str)] - cudnn {{ cudnn }} # ['cudnn' in str(build_str)] - dnnl {{ dnnl }} - python {{ python }} @@ -63,6 +68,8 @@ requirements: - {{ pin_compatible('glog', max_pin='x.x') }} - {{ pin_compatible('numpy', max_pin='x.x') }} - {{ pin_compatible('dnnl', max_pin='x.x') }} + - cuda-cudart {{ cuda_cudart }} # ['cuda' in str(build_str)] + - cuda-libraries {{ cuda_libraries }} # ['cuda' in str(build_str)] - cudnn {{ cudnn }} # ['cudnn' in str(build_str)] - python {{ python }} - nccl {{ nccl }} # ['nccl' in str(build_str)]