From 50f08d3d38fa8d857ce048d9cca34083cd2aabaa Mon Sep 17 00:00:00 2001 From: SFMDI <36741818+SFMDI@users.noreply.github.com> Date: Tue, 16 Mar 2021 19:42:38 +0900 Subject: [PATCH 1/6] replace floor/ceil to floorf/ceilf to eliminate cuda compilation errors in the latest version of Visual Studio 16.9 --- mmcv/ops/csrc/deform_conv_cuda_kernel.cuh | 12 ++++++------ mmcv/ops/csrc/deform_roi_pool_cuda_kernel.cuh | 8 ++++---- mmcv/ops/csrc/modulated_deform_conv_cuda_kernel.cuh | 12 ++++++------ mmcv/ops/csrc/roi_align_cuda_kernel.cuh | 8 ++++---- mmcv/ops/csrc/roi_pool_cuda_kernel.cuh | 8 ++++---- 5 files changed, 24 insertions(+), 24 deletions(-) diff --git a/mmcv/ops/csrc/deform_conv_cuda_kernel.cuh b/mmcv/ops/csrc/deform_conv_cuda_kernel.cuh index 1c4d684202..6b4d1bbd85 100644 --- a/mmcv/ops/csrc/deform_conv_cuda_kernel.cuh +++ b/mmcv/ops/csrc/deform_conv_cuda_kernel.cuh @@ -85,8 +85,8 @@ __device__ T deformable_im2col_bilinear(const T *input, const int data_width, return 0; } - int h_low = floor(h); - int w_low = floor(w); + int h_low = floorf(h); + int w_low = floorf(w); int h_high = h_low + 1; int w_high = w_low + 1; @@ -122,8 +122,8 @@ __device__ T get_gradient_weight(T argmax_h, T argmax_w, const int h, return 0; } - int argmax_h_low = floor(argmax_h); - int argmax_w_low = floor(argmax_w); + int argmax_h_low = floorf(argmax_h); + int argmax_w_low = floorf(argmax_w); int argmax_h_high = argmax_h_low + 1; int argmax_w_high = argmax_w_low + 1; @@ -149,8 +149,8 @@ __device__ T get_coordinate_weight(T argmax_h, T argmax_w, const int height, return 0; } - int argmax_h_low = floor(argmax_h); - int argmax_w_low = floor(argmax_w); + int argmax_h_low = floorf(argmax_h); + int argmax_w_low = floorf(argmax_w); int argmax_h_high = argmax_h_low + 1; int argmax_w_high = argmax_w_low + 1; diff --git a/mmcv/ops/csrc/deform_roi_pool_cuda_kernel.cuh b/mmcv/ops/csrc/deform_roi_pool_cuda_kernel.cuh index e6b59b3999..a3385cff16 100644 --- a/mmcv/ops/csrc/deform_roi_pool_cuda_kernel.cuh +++ b/mmcv/ops/csrc/deform_roi_pool_cuda_kernel.cuh @@ -42,10 +42,10 @@ __global__ void deform_roi_pool_forward_cuda_kernel( int roi_bin_grid_h = (sampling_ratio > 0) ? sampling_ratio - : static_cast(ceil(roi_height / pooled_height)); + : static_cast(ceilf(roi_height / pooled_height)); int roi_bin_grid_w = (sampling_ratio > 0) ? sampling_ratio - : static_cast(ceil(roi_width / pooled_width)); + : static_cast(ceilf(roi_width / pooled_width)); // Compute roi offset if (offset != NULL) { @@ -113,10 +113,10 @@ __global__ void deform_roi_pool_backward_cuda_kernel( int roi_bin_grid_h = (sampling_ratio > 0) ? sampling_ratio - : static_cast(ceil(roi_height / pooled_height)); + : static_cast(ceilf(roi_height / pooled_height)); int roi_bin_grid_w = (sampling_ratio > 0) ? sampling_ratio - : static_cast(ceil(roi_width / pooled_width)); + : static_cast(ceilf(roi_width / pooled_width)); // Compute roi offset if (offset != NULL) { diff --git a/mmcv/ops/csrc/modulated_deform_conv_cuda_kernel.cuh b/mmcv/ops/csrc/modulated_deform_conv_cuda_kernel.cuh index 9953a09457..04bf5c308d 100644 --- a/mmcv/ops/csrc/modulated_deform_conv_cuda_kernel.cuh +++ b/mmcv/ops/csrc/modulated_deform_conv_cuda_kernel.cuh @@ -75,8 +75,8 @@ template __device__ T dmcn_im2col_bilinear(const T *input, const int data_width, const int height, const int width, T h, T w) { - int h_low = floor(h); - int w_low = floor(w); + int h_low = floorf(h); + int w_low = floorf(w); int h_high = h_low + 1; int w_high = w_low + 1; @@ -112,8 +112,8 @@ __device__ T dmcn_get_gradient_weight(T argmax_h, T argmax_w, const int h, return 0; } - int argmax_h_low = floor(argmax_h); - int argmax_w_low = floor(argmax_w); + int argmax_h_low = floorf(argmax_h); + int argmax_w_low = floorf(argmax_w); int argmax_h_high = argmax_h_low + 1; int argmax_w_high = argmax_w_low + 1; @@ -140,8 +140,8 @@ __device__ T dmcn_get_coordinate_weight(T argmax_h, T argmax_w, return 0; } - int argmax_h_low = floor(argmax_h); - int argmax_w_low = floor(argmax_w); + int argmax_h_low = floorf(argmax_h); + int argmax_w_low = floorf(argmax_w); int argmax_h_high = argmax_h_low + 1; int argmax_w_high = argmax_w_low + 1; diff --git a/mmcv/ops/csrc/roi_align_cuda_kernel.cuh b/mmcv/ops/csrc/roi_align_cuda_kernel.cuh index c6f350b340..46e8872b64 100644 --- a/mmcv/ops/csrc/roi_align_cuda_kernel.cuh +++ b/mmcv/ops/csrc/roi_align_cuda_kernel.cuh @@ -54,10 +54,10 @@ __global__ void roi_align_forward_cuda_kernel( int roi_bin_grid_h = (sampling_ratio > 0) ? sampling_ratio - : static_cast(ceil(roi_height / pooled_height)); + : static_cast(ceilf(roi_height / pooled_height)); int roi_bin_grid_w = (sampling_ratio > 0) ? sampling_ratio - : static_cast(ceil(roi_width / pooled_width)); + : static_cast(ceilf(roi_width / pooled_width)); if (pool_mode == 0) { // We do max pooling inside a bin @@ -168,11 +168,11 @@ __global__ void roi_align_backward_cuda_kernel( int roi_bin_grid_h = (sampling_ratio > 0) ? sampling_ratio - : static_cast(ceil(roi_height / pooled_height)); + : static_cast(ceilf(roi_height / pooled_height)); int roi_bin_grid_w = (sampling_ratio > 0) ? sampling_ratio - : static_cast(ceil(roi_width / pooled_width)); + : static_cast(ceilf(roi_width / pooled_width)); // We do average (integral) pooling inside a bin const T count = roi_bin_grid_h * roi_bin_grid_w; // e.g. = 4 diff --git a/mmcv/ops/csrc/roi_pool_cuda_kernel.cuh b/mmcv/ops/csrc/roi_pool_cuda_kernel.cuh index d499ade76e..72a2750ef8 100644 --- a/mmcv/ops/csrc/roi_pool_cuda_kernel.cuh +++ b/mmcv/ops/csrc/roi_pool_cuda_kernel.cuh @@ -36,10 +36,10 @@ __global__ void roi_pool_forward_cuda_kernel( T bin_size_h = roi_h / static_cast(pooled_height); // the corresponding bin region - int bin_x1 = floor(static_cast(pw) * bin_size_w + roi_x1); - int bin_y1 = floor(static_cast(ph) * bin_size_h + roi_y1); - int bin_x2 = ceil(static_cast(pw + 1) * bin_size_w + roi_x1); - int bin_y2 = ceil(static_cast(ph + 1) * bin_size_h + roi_y1); + int bin_x1 = floorf(static_cast(pw) * bin_size_w + roi_x1); + int bin_y1 = floorf(static_cast(ph) * bin_size_h + roi_y1); + int bin_x2 = ceilf(static_cast(pw + 1) * bin_size_w + roi_x1); + int bin_y2 = ceilf(static_cast(ph + 1) * bin_size_h + roi_y1); // add roi offsets and clip to input boundaries bin_x1 = min(max(bin_x1, 0), width); From 1e0dc219f09f8e5fa82bfc8519a22081bfadc76e Mon Sep 17 00:00:00 2001 From: wangruohui <12756472+wangruohui@users.noreply.github.com> Date: Wed, 31 Mar 2021 13:18:00 +0800 Subject: [PATCH 2/6] fix lint error --- mmcv/ops/csrc/deform_roi_pool_cuda_kernel.cuh | 14 ++++++++------ mmcv/ops/csrc/roi_align_cuda_kernel.cuh | 7 ++++--- 2 files changed, 12 insertions(+), 9 deletions(-) diff --git a/mmcv/ops/csrc/deform_roi_pool_cuda_kernel.cuh b/mmcv/ops/csrc/deform_roi_pool_cuda_kernel.cuh index a3385cff16..51dba10618 100644 --- a/mmcv/ops/csrc/deform_roi_pool_cuda_kernel.cuh +++ b/mmcv/ops/csrc/deform_roi_pool_cuda_kernel.cuh @@ -43,9 +43,10 @@ __global__ void deform_roi_pool_forward_cuda_kernel( (sampling_ratio > 0) ? sampling_ratio : static_cast(ceilf(roi_height / pooled_height)); - int roi_bin_grid_w = (sampling_ratio > 0) - ? sampling_ratio - : static_cast(ceilf(roi_width / pooled_width)); + int roi_bin_grid_w = + (sampling_ratio > 0) + ? sampling_ratio + : static_cast(ceilf(roi_width / pooled_width)); // Compute roi offset if (offset != NULL) { @@ -114,9 +115,10 @@ __global__ void deform_roi_pool_backward_cuda_kernel( (sampling_ratio > 0) ? sampling_ratio : static_cast(ceilf(roi_height / pooled_height)); - int roi_bin_grid_w = (sampling_ratio > 0) - ? sampling_ratio - : static_cast(ceilf(roi_width / pooled_width)); + int roi_bin_grid_w = + (sampling_ratio > 0) + ? sampling_ratio + : static_cast(ceilf(roi_width / pooled_width)); // Compute roi offset if (offset != NULL) { diff --git a/mmcv/ops/csrc/roi_align_cuda_kernel.cuh b/mmcv/ops/csrc/roi_align_cuda_kernel.cuh index 46e8872b64..d350843bec 100644 --- a/mmcv/ops/csrc/roi_align_cuda_kernel.cuh +++ b/mmcv/ops/csrc/roi_align_cuda_kernel.cuh @@ -55,9 +55,10 @@ __global__ void roi_align_forward_cuda_kernel( (sampling_ratio > 0) ? sampling_ratio : static_cast(ceilf(roi_height / pooled_height)); - int roi_bin_grid_w = (sampling_ratio > 0) - ? sampling_ratio - : static_cast(ceilf(roi_width / pooled_width)); + int roi_bin_grid_w = + (sampling_ratio > 0) + ? sampling_ratio + : static_cast(ceilf(roi_width / pooled_width)); if (pool_mode == 0) { // We do max pooling inside a bin From c0a4cd1f2162deafe24ab7d52bdbd27c49a79a2e Mon Sep 17 00:00:00 2001 From: wangruohui <12756472+wangruohui@users.noreply.github.com> Date: Wed, 31 Mar 2021 13:55:01 +0800 Subject: [PATCH 3/6] fix lint --- mmcv/ops/csrc/deform_roi_pool_cuda_kernel.cuh | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/mmcv/ops/csrc/deform_roi_pool_cuda_kernel.cuh b/mmcv/ops/csrc/deform_roi_pool_cuda_kernel.cuh index 51dba10618..4e03fd3692 100644 --- a/mmcv/ops/csrc/deform_roi_pool_cuda_kernel.cuh +++ b/mmcv/ops/csrc/deform_roi_pool_cuda_kernel.cuh @@ -115,7 +115,7 @@ __global__ void deform_roi_pool_backward_cuda_kernel( (sampling_ratio > 0) ? sampling_ratio : static_cast(ceilf(roi_height / pooled_height)); - int roi_bin_grid_w = + int roi_bin_grid_w = (sampling_ratio > 0) ? sampling_ratio : static_cast(ceilf(roi_width / pooled_width)); From fb027beca92c024933270fe627fcb4b40cc4c755 Mon Sep 17 00:00:00 2001 From: wangruohui <12756472+wangruohui@users.noreply.github.com> Date: Thu, 1 Apr 2021 02:19:13 +0800 Subject: [PATCH 4/6] fix lint --- .pre-commit-config.yaml | 52 +++++++++---------- mmcv/ops/csrc/deform_roi_pool_cuda_kernel.cuh | 2 +- 2 files changed, 27 insertions(+), 27 deletions(-) diff --git a/.pre-commit-config.yaml b/.pre-commit-config.yaml index 056c046592..975194f8c1 100644 --- a/.pre-commit-config.yaml +++ b/.pre-commit-config.yaml @@ -1,21 +1,21 @@ exclude: ^tests/data/ repos: - - repo: https://gitlab.com/pycqa/flake8.git - rev: 3.8.3 - hooks: - - id: flake8 - - repo: https://github.com/asottile/seed-isort-config - rev: v2.2.0 - hooks: - - id: seed-isort-config - - repo: https://github.com/timothycrosley/isort - rev: 4.3.21 - hooks: - - id: isort - - repo: https://github.com/pre-commit/mirrors-yapf - rev: v0.30.0 - hooks: - - id: yapf + # - repo: https://gitlab.com/pycqa/flake8.git + # rev: 3.8.3 + # hooks: + # - id: flake8 + # - repo: https://github.com/asottile/seed-isort-config + # rev: v2.2.0 + # hooks: + # - id: seed-isort-config + # - repo: https://github.com/timothycrosley/isort + # rev: 4.3.21 + # hooks: + # - id: isort + # - repo: https://github.com/pre-commit/mirrors-yapf + # rev: v0.30.0 + # hooks: + # - id: yapf - repo: https://github.com/pre-commit/pre-commit-hooks rev: v3.1.0 hooks: @@ -29,16 +29,16 @@ repos: args: ["--remove"] - id: mixed-line-ending args: ["--fix=lf"] - - repo: https://github.com/jumanjihouse/pre-commit-hooks - rev: 2.1.4 - hooks: - - id: markdownlint - args: ["-r", "~MD002,~MD013,~MD029,~MD033,~MD034"] - - repo: https://github.com/myint/docformatter - rev: v1.3.1 - hooks: - - id: docformatter - args: ["--in-place", "--wrap-descriptions", "79"] + # - repo: https://github.com/jumanjihouse/pre-commit-hooks + # rev: 2.1.4 + # hooks: + # - id: markdownlint + # args: ["-r", "~MD002,~MD013,~MD029,~MD033,~MD034"] + # - repo: https://github.com/myint/docformatter + # rev: v1.3.1 + # hooks: + # - id: docformatter + # args: ["--in-place", "--wrap-descriptions", "79"] # - repo: local # hooks: # - id: clang-format diff --git a/mmcv/ops/csrc/deform_roi_pool_cuda_kernel.cuh b/mmcv/ops/csrc/deform_roi_pool_cuda_kernel.cuh index 4e03fd3692..cddb8d5e9e 100644 --- a/mmcv/ops/csrc/deform_roi_pool_cuda_kernel.cuh +++ b/mmcv/ops/csrc/deform_roi_pool_cuda_kernel.cuh @@ -43,7 +43,7 @@ __global__ void deform_roi_pool_forward_cuda_kernel( (sampling_ratio > 0) ? sampling_ratio : static_cast(ceilf(roi_height / pooled_height)); - int roi_bin_grid_w = + int roi_bin_grid_w = (sampling_ratio > 0) ? sampling_ratio : static_cast(ceilf(roi_width / pooled_width)); From 5975ebc8491aafef4dcbbf3abd7e0e5f81f46c19 Mon Sep 17 00:00:00 2001 From: WRH <12756472+wangruohui@users.noreply.github.com> Date: Sat, 10 Apr 2021 22:28:17 +0800 Subject: [PATCH 5/6] Update .pre-commit-config.yaml --- .pre-commit-config.yaml | 68 ++++++++++++++++++++--------------------- 1 file changed, 34 insertions(+), 34 deletions(-) diff --git a/.pre-commit-config.yaml b/.pre-commit-config.yaml index 975194f8c1..714e03eb43 100644 --- a/.pre-commit-config.yaml +++ b/.pre-commit-config.yaml @@ -1,21 +1,21 @@ exclude: ^tests/data/ repos: - # - repo: https://gitlab.com/pycqa/flake8.git - # rev: 3.8.3 - # hooks: - # - id: flake8 - # - repo: https://github.com/asottile/seed-isort-config - # rev: v2.2.0 - # hooks: - # - id: seed-isort-config - # - repo: https://github.com/timothycrosley/isort - # rev: 4.3.21 - # hooks: - # - id: isort - # - repo: https://github.com/pre-commit/mirrors-yapf - # rev: v0.30.0 - # hooks: - # - id: yapf + - repo: https://gitlab.com/pycqa/flake8.git + rev: 3.8.3 + hooks: + - id: flake8 + - repo: https://github.com/asottile/seed-isort-config + rev: v2.2.0 + hooks: + - id: seed-isort-config + - repo: https://github.com/timothycrosley/isort + rev: 4.3.21 + hooks: + - id: isort + - repo: https://github.com/pre-commit/mirrors-yapf + rev: v0.30.0 + hooks: + - id: yapf - repo: https://github.com/pre-commit/pre-commit-hooks rev: v3.1.0 hooks: @@ -29,21 +29,21 @@ repos: args: ["--remove"] - id: mixed-line-ending args: ["--fix=lf"] - # - repo: https://github.com/jumanjihouse/pre-commit-hooks - # rev: 2.1.4 - # hooks: - # - id: markdownlint - # args: ["-r", "~MD002,~MD013,~MD029,~MD033,~MD034"] - # - repo: https://github.com/myint/docformatter - # rev: v1.3.1 - # hooks: - # - id: docformatter - # args: ["--in-place", "--wrap-descriptions", "79"] - # - repo: local - # hooks: - # - id: clang-format - # name: clang-format - # description: Format files with ClangFormat - # entry: clang-format -style=google -i - # language: system - # files: \.(c|cc|cxx|cpp|cu|h|hpp|hxx|cuh|proto)$ + - repo: https://github.com/jumanjihouse/pre-commit-hooks + rev: 2.1.4 + hooks: + - id: markdownlint + args: ["-r", "~MD002,~MD013,~MD029,~MD033,~MD034"] + - repo: https://github.com/myint/docformatter + rev: v1.3.1 + hooks: + - id: docformatter + args: ["--in-place", "--wrap-descriptions", "79"] + - repo: local + hooks: + - id: clang-format + name: clang-format + description: Format files with ClangFormat + entry: clang-format -style=google -i + language: system + files: \.(c|cc|cxx|cpp|cu|h|hpp|hxx|cuh|proto)$ From 2c496b67cdc1802b9fd60ffa1a21187745c22350 Mon Sep 17 00:00:00 2001 From: WRH <12756472+wangruohui@users.noreply.github.com> Date: Sat, 10 Apr 2021 22:29:26 +0800 Subject: [PATCH 6/6] Update .pre-commit-config.yaml --- .pre-commit-config.yaml | 16 ++++++++-------- 1 file changed, 8 insertions(+), 8 deletions(-) diff --git a/.pre-commit-config.yaml b/.pre-commit-config.yaml index 714e03eb43..056c046592 100644 --- a/.pre-commit-config.yaml +++ b/.pre-commit-config.yaml @@ -39,11 +39,11 @@ repos: hooks: - id: docformatter args: ["--in-place", "--wrap-descriptions", "79"] - - repo: local - hooks: - - id: clang-format - name: clang-format - description: Format files with ClangFormat - entry: clang-format -style=google -i - language: system - files: \.(c|cc|cxx|cpp|cu|h|hpp|hxx|cuh|proto)$ + # - repo: local + # hooks: + # - id: clang-format + # name: clang-format + # description: Format files with ClangFormat + # entry: clang-format -style=google -i + # language: system + # files: \.(c|cc|cxx|cpp|cu|h|hpp|hxx|cuh|proto)$