From a0d6c63cca28c5bb6602784ddf6edddc1781f03f Mon Sep 17 00:00:00 2001 From: HAOCHENYE <21724054@zju.edu.cn> Date: Sat, 2 Nov 2024 00:48:24 +0800 Subject: [PATCH 1/4] [Fix] Fix pre-commit-hook for python3.12 --- .github/workflows/lint.yml | 6 +++--- .pre-commit-config-zh-cn.yaml | 15 ++++++++++----- .pre-commit-config.yaml | 13 +++++++++---- mmcv/cnn/bricks/drop.py | 3 ++- mmcv/cnn/rfsearch/search.py | 8 ++++---- mmcv/cnn/rfsearch/utils.py | 4 ++-- mmcv/ops/active_rotated_filter.py | 6 +++--- mmcv/ops/cc_attention.py | 2 +- mmcv/ops/chamfer_distance.py | 3 ++- mmcv/ops/correlation.py | 2 +- mmcv/ops/deform_conv.py | 4 +++- mmcv/ops/iou3d.py | 2 +- mmcv/ops/rotated_feature_align.py | 6 +++--- mmcv/ops/scatter_points.py | 2 +- mmcv/ops/sparse_modules.py | 13 ++++++------- mmcv/ops/sparse_structure.py | 2 +- mmcv/ops/tin_shift.py | 8 +++++--- mmcv/ops/upfirdn2d.py | 6 +++--- mmcv/transforms/formatting.py | 1 + mmcv/transforms/processing.py | 2 +- mmcv/transforms/wrappers.py | 2 +- mmcv/video/optflow.py | 4 ++-- mmcv/visualization/image.py | 2 +- 23 files changed, 66 insertions(+), 50 deletions(-) diff --git a/.github/workflows/lint.yml b/.github/workflows/lint.yml index ac5afd597a..db6b570b17 100644 --- a/.github/workflows/lint.yml +++ b/.github/workflows/lint.yml @@ -11,10 +11,10 @@ jobs: runs-on: ubuntu-22.04 steps: - uses: actions/checkout@v2 - - name: Set up Python 3.7 + - name: Set up Python 3.10.15 uses: actions/setup-python@v2 with: - python-version: 3.7 + python-version: '3.10.15' - name: Install pre-commit hook run: | pip install pre-commit @@ -22,7 +22,7 @@ jobs: - name: Linting run: pre-commit run --all-files - name: Format c/cuda codes with clang-format - uses: DoozyX/clang-format-lint-action@v0.11 + uses: DoozyX/clang-format-lint-action@v0.13 with: source: mmcv/ops/csrc extensions: h,c,cpp,hpp,cu,cuh diff --git a/.pre-commit-config-zh-cn.yaml b/.pre-commit-config-zh-cn.yaml index 313c83c84b..f705ec904a 100644 --- a/.pre-commit-config-zh-cn.yaml +++ b/.pre-commit-config-zh-cn.yaml @@ -1,7 +1,11 @@ exclude: ^tests/data/ repos: - - repo: https://gitee.com/openmmlab/mirrors-flake8 - rev: 5.0.4 + - repo: https://github.com/pre-commit/pre-commit + rev: v4.0.0 + hooks: + - id: validate_manifest + - repo: https://github.com/PyCQA/flake8 + rev: 7.1.1 hooks: - id: flake8 - repo: https://gitee.com/openmmlab/mirrors-isort @@ -13,7 +17,7 @@ repos: hooks: - id: yapf - repo: https://gitee.com/openmmlab/mirrors-pre-commit-hooks - rev: v4.3.0 + rev: v5.0.0 hooks: - id: trailing-whitespace - id: check-yaml @@ -53,8 +57,8 @@ repos: hooks: - id: check-copyright args: ["mmcv", "tests", "--excludes", "mmcv/ops"] - - repo: https://gitee.com/openmmlab/mirrors-mypy - rev: v0.812 + - repo: https://github.com/pre-commit/mirrors-mypy + rev: v1.2.0 hooks: - id: mypy exclude: |- @@ -62,6 +66,7 @@ repos: ^test | ^docs ) + additional_dependencies: ["types-setuptools", "types-requests"] # - repo: local # hooks: # - id: clang-format diff --git a/.pre-commit-config.yaml b/.pre-commit-config.yaml index 76ded556f9..e3028c5bbe 100644 --- a/.pre-commit-config.yaml +++ b/.pre-commit-config.yaml @@ -1,7 +1,11 @@ exclude: ^tests/data/ repos: + - repo: https://github.com/pre-commit/pre-commit + rev: v4.0.0 + hooks: + - id: validate_manifest - repo: https://github.com/PyCQA/flake8 - rev: 5.0.4 + rev: 7.1.1 hooks: - id: flake8 - repo: https://github.com/PyCQA/isort @@ -13,7 +17,7 @@ repos: hooks: - id: yapf - repo: https://github.com/pre-commit/pre-commit-hooks - rev: v4.3.0 + rev: v5.0.0 hooks: - id: trailing-whitespace - id: check-yaml @@ -39,7 +43,7 @@ repos: - mdformat_frontmatter - linkify-it-py - repo: https://github.com/myint/docformatter - rev: v1.3.1 + rev: 06907d0 hooks: - id: docformatter args: ["--in-place", "--wrap-descriptions", "79"] @@ -54,7 +58,7 @@ repos: - id: check-copyright args: ["mmcv", "tests", "--excludes", "mmcv/ops"] - repo: https://github.com/pre-commit/mirrors-mypy - rev: v0.812 + rev: v1.2.0 hooks: - id: mypy exclude: |- @@ -62,6 +66,7 @@ repos: ^test | ^docs ) + additional_dependencies: ["types-setuptools", "types-requests"] # - repo: local # hooks: # - id: clang-format diff --git a/mmcv/cnn/bricks/drop.py b/mmcv/cnn/bricks/drop.py index fe82a25605..c3daaff8d6 100644 --- a/mmcv/cnn/bricks/drop.py +++ b/mmcv/cnn/bricks/drop.py @@ -13,7 +13,8 @@ def drop_path(x: torch.Tensor, residual blocks). We follow the implementation - https://github.com/rwightman/pytorch-image-models/blob/a2727c1bf78ba0d7b5727f5f95e37fb7f8866b1f/timm/models/layers/drop.py # noqa: E501 + https://github.com/rwightman/pytorch-image-models/blob/a2727c1bf78ba0d7b5727f5f95e37fb7f8866b1f/timm/models/layers/drop.py + # noqa: E501 """ if drop_prob == 0. or not training: return x diff --git a/mmcv/cnn/rfsearch/search.py b/mmcv/cnn/rfsearch/search.py index f4add4b23a..9ba5b9dc01 100644 --- a/mmcv/cnn/rfsearch/search.py +++ b/mmcv/cnn/rfsearch/search.py @@ -71,7 +71,7 @@ def __init__(self, self.by_epoch = by_epoch def init_model(self, model: nn.Module): - """init model with search ability. + """Init model with search ability. Args: model (nn.Module): pytorch model @@ -132,7 +132,7 @@ def step(self, model: nn.Module, work_dir: str) -> None: ) def estimate_and_expand(self, model: nn.Module) -> None: - """estimate and search for RFConvOp. + """Estimate and search for RFConvOp. Args: model (nn.Module): pytorch model @@ -146,7 +146,7 @@ def wrap_model(self, model: nn.Module, search_op: str = 'Conv2d', prefix: str = '') -> None: - """wrap model to support searchable conv op. + """Wrap model to support searchable conv op. Args: model (nn.Module): pytorch model @@ -187,7 +187,7 @@ def set_model(self, search_op: str = 'Conv2d', init_rates: Optional[int] = None, prefix: str = '') -> None: - """set model based on config. + """Set model based on config. Args: model (nn.Module): pytorch model diff --git a/mmcv/cnn/rfsearch/utils.py b/mmcv/cnn/rfsearch/utils.py index 4c8168e343..9ced5d040a 100644 --- a/mmcv/cnn/rfsearch/utils.py +++ b/mmcv/cnn/rfsearch/utils.py @@ -4,7 +4,7 @@ def write_to_json(config: dict, filename: str): - """save config to json file. + """Save config to json file. Args: config (dict): Config to be saved. @@ -16,7 +16,7 @@ def write_to_json(config: dict, filename: str): def expand_rates(dilation: tuple, config: dict) -> list: - """expand dilation rate according to config. + """Expand dilation rate according to config. Args: dilation (int): _description_ diff --git a/mmcv/ops/active_rotated_filter.py b/mmcv/ops/active_rotated_filter.py index b8ba43dd41..1ac324126a 100644 --- a/mmcv/ops/active_rotated_filter.py +++ b/mmcv/ops/active_rotated_filter.py @@ -16,9 +16,9 @@ class ActiveRotatedFilterFunction(Function): """Encoding the orientation information and generating orientation- sensitive features. - The details are described in the paper `Align Deep Features for Oriented - Object Detection _`. - """ + The details are described in the paper + `Align Deep Features for Oriented Object Detection _`. + """ # noqa: E501 @staticmethod def forward(ctx, input: torch.Tensor, diff --git a/mmcv/ops/cc_attention.py b/mmcv/ops/cc_attention.py index efde7b703c..4e61357025 100644 --- a/mmcv/ops/cc_attention.py +++ b/mmcv/ops/cc_attention.py @@ -51,7 +51,7 @@ def __init__(self, in_channels: int) -> None: self.in_channels = in_channels def forward(self, x: torch.Tensor) -> torch.Tensor: - """forward function of Criss-Cross Attention. + """Forward function of Criss-Cross Attention. Args: x (torch.Tensor): Input feature with the shape of diff --git a/mmcv/ops/chamfer_distance.py b/mmcv/ops/chamfer_distance.py index d95bd47747..759ef98e90 100644 --- a/mmcv/ops/chamfer_distance.py +++ b/mmcv/ops/chamfer_distance.py @@ -16,7 +16,8 @@ class ChamferDistanceFunction(Function): """This is an implementation of the 2D Chamfer Distance. It has been used in the paper `Oriented RepPoints for Aerial Object - Detection (CVPR 2022) _`. + Detection (CVPR 2022) + _`. """ @staticmethod diff --git a/mmcv/ops/correlation.py b/mmcv/ops/correlation.py index 319b764678..9524610d4f 100644 --- a/mmcv/ops/correlation.py +++ b/mmcv/ops/correlation.py @@ -116,7 +116,7 @@ def _output_size(ctx, input1): class Correlation(nn.Module): - r"""Correlation operator + r"""Correlation operator. This correlation operator works for optical flow correlation computation. diff --git a/mmcv/ops/deform_conv.py b/mmcv/ops/deform_conv.py index c6cbba6779..1bde04d735 100644 --- a/mmcv/ops/deform_conv.py +++ b/mmcv/ops/deform_conv.py @@ -449,7 +449,9 @@ def _load_from_state_dict(self, state_dict, prefix, local_metadata, strict, @MODELS.register_module('DCN', force=True) class DeformConv2dPack_MLU(DeformConv2d): - """This class is the DCN implementation of the MLU device. The MLU + """This class is the DCN implementation of the MLU device. + + The MLU backend support of the operator has been implemented in torchvision. The mmcv registration mechanism is used for multiplexing here. The torchvision implementation of DCN is called. diff --git a/mmcv/ops/iou3d.py b/mmcv/ops/iou3d.py index 94e2057ad2..dd47993917 100755 --- a/mmcv/ops/iou3d.py +++ b/mmcv/ops/iou3d.py @@ -42,7 +42,7 @@ def boxes_iou3d(boxes_a: Tensor, boxes_b: Tensor) -> Tensor: Returns: torch.Tensor: 3D IoU result with shape (M, N). """ - assert boxes_a.shape[1] == boxes_b.shape[1] == 7,\ + assert boxes_a.shape[1] == boxes_b.shape[1] == 7, \ 'Input boxes shape should be (N, 7)' boxes_a_height_max = (boxes_a[:, 2] + boxes_a[:, 5] / 2).view(-1, 1) diff --git a/mmcv/ops/rotated_feature_align.py b/mmcv/ops/rotated_feature_align.py index 0132c04862..893b029443 100644 --- a/mmcv/ops/rotated_feature_align.py +++ b/mmcv/ops/rotated_feature_align.py @@ -17,9 +17,9 @@ class RotatedFeatureAlignFunction(Function): correspond to the refined rotate anchors and reconstruct the feature maps in pixel-wise manner to achieve feature alignment. - The details are described in the paper - `R3Det: Refined Single-Stage Detector with Feature Refinement for Rotating - Object `_. + The details are described in the paper `R3Det: Refined Single-Stage + Detector with Feature Refinement for Rotating Object + `_. """ @staticmethod diff --git a/mmcv/ops/scatter_points.py b/mmcv/ops/scatter_points.py index 5d881bfe63..7b03e8f1da 100644 --- a/mmcv/ops/scatter_points.py +++ b/mmcv/ops/scatter_points.py @@ -20,7 +20,7 @@ def forward(ctx: Any, feats: torch.Tensor, coors: torch.Tensor, reduce_type: str = 'max') -> Tuple[torch.Tensor, torch.Tensor]: - """convert kitti points(N, >=3) to voxels. + """Convert kitti points(N, >=3) to voxels. Args: feats (torch.Tensor): [N, C]. Points features to be reduced diff --git a/mmcv/ops/sparse_modules.py b/mmcv/ops/sparse_modules.py index 20a92aa279..ed2935343b 100644 --- a/mmcv/ops/sparse_modules.py +++ b/mmcv/ops/sparse_modules.py @@ -47,16 +47,15 @@ def _mean_update(vals: Union[int, List], m_vals: Union[int, List], class SparseModule(nn.Module): - """place holder, All module subclass from this will take sptensor in + """Place holder, All module subclass from this will take sptensor in SparseSequential.""" pass class SparseSequential(SparseModule): - r"""A sequential container. - Modules will be added to it in the order they are passed in the - constructor. - Alternatively, an ordered dict of modules can also be passed in. + r"""A sequential container. Modules will be added to it in the order they + are passed in the constructor. Alternatively, an ordered dict of modules + can also be passed in. To make it easier to understand, given is a small example:: @@ -189,14 +188,14 @@ def fused(self): class ToDense(SparseModule): - """convert SparseConvTensor to NCHW dense tensor.""" + """Convert SparseConvTensor to NCHW dense tensor.""" def forward(self, x: SparseConvTensor): return x.dense() class RemoveGrid(SparseModule): - """remove pre-allocated grid buffer.""" + """Remove pre-allocated grid buffer.""" def forward(self, x: SparseConvTensor): x.grid = None diff --git a/mmcv/ops/sparse_structure.py b/mmcv/ops/sparse_structure.py index 83907ab556..9d7cb6e1ff 100644 --- a/mmcv/ops/sparse_structure.py +++ b/mmcv/ops/sparse_structure.py @@ -6,7 +6,7 @@ def scatter_nd(indices: torch.Tensor, updates: torch.Tensor, shape: torch.Tensor) -> torch.Tensor: - """pytorch edition of tensorflow scatter_nd. + """Pytorch edition of tensorflow scatter_nd. this function don't contain except handle code. so use this carefully when indice repeats, don't support repeat add which is supported in tensorflow. diff --git a/mmcv/ops/tin_shift.py b/mmcv/ops/tin_shift.py index 473231cc0d..7d2df2e77a 100755 --- a/mmcv/ops/tin_shift.py +++ b/mmcv/ops/tin_shift.py @@ -55,10 +55,12 @@ class TINShift(nn.Module): Temporal Interlace shift is a differentiable temporal-wise frame shifting which is proposed in "Temporal Interlacing Network" - Please refer to `Temporal Interlacing Network - `_ for more details. + Please refer to + `Temporal Interlacing Network `_ + for more details. - Code is modified from https://github.com/mit-han-lab/temporal-shift-module + Code is modified from + https://github.com/mit-han-lab/temporal-shift-module """ def forward(self, input, shift): diff --git a/mmcv/ops/upfirdn2d.py b/mmcv/ops/upfirdn2d.py index 857e840c1b..e015095033 100644 --- a/mmcv/ops/upfirdn2d.py +++ b/mmcv/ops/upfirdn2d.py @@ -19,7 +19,7 @@ def _parse_scaling(scaling): - """parse scaling into list [x, y]""" + """Parse scaling into list [x, y]""" if isinstance(scaling, int): scaling = [scaling, scaling] assert isinstance(scaling, (list, tuple)) @@ -30,7 +30,7 @@ def _parse_scaling(scaling): def _parse_padding(padding): - """parse padding into list [padx0, padx1, pady0, pady1]""" + """Parse padding into list [padx0, padx1, pady0, pady1]""" if isinstance(padding, int): padding = [padding, padding] assert isinstance(padding, (list, tuple)) @@ -43,7 +43,7 @@ def _parse_padding(padding): def _get_filter_size(filter): - """get width and height of filter kernel.""" + """Get width and height of filter kernel.""" if filter is None: return 1, 1 assert isinstance(filter, torch.Tensor) and filter.ndim in [1, 2] diff --git a/mmcv/transforms/formatting.py b/mmcv/transforms/formatting.py index 02089215e1..89d9d096bb 100644 --- a/mmcv/transforms/formatting.py +++ b/mmcv/transforms/formatting.py @@ -110,6 +110,7 @@ def __init__(self, keys: dict) -> None: def transform(self, results: dict) -> dict: """Transform function to convert image in results to :obj:`torch.Tensor` and transpose the channel order. + Args: results (dict): Result dict contains the image data to convert. Returns: diff --git a/mmcv/transforms/processing.py b/mmcv/transforms/processing.py index 8b65b112b6..d490745a65 100644 --- a/mmcv/transforms/processing.py +++ b/mmcv/transforms/processing.py @@ -1457,7 +1457,7 @@ class RandomResize(BaseTransform): def __init__( self, scale: Union[Tuple[int, int], Sequence[Tuple[int, int]]], - ratio_range: Tuple[float, float] = None, + ratio_range: Optional[Tuple[float, float]] = None, resize_type: str = 'Resize', **resize_kwargs, ) -> None: diff --git a/mmcv/transforms/wrappers.py b/mmcv/transforms/wrappers.py index 27e07d6335..ed20449d1e 100644 --- a/mmcv/transforms/wrappers.py +++ b/mmcv/transforms/wrappers.py @@ -177,7 +177,7 @@ class KeyMapper(BaseTransform): """ def __init__(self, - transforms: Union[Transform, List[Transform]] = None, + transforms: Union[Transform, List[Transform], None] = None, mapping: Optional[Dict] = None, remapping: Optional[Dict] = None, auto_remap: Optional[bool] = None, diff --git a/mmcv/video/optflow.py b/mmcv/video/optflow.py index edd3e42069..4101b6c8f5 100644 --- a/mmcv/video/optflow.py +++ b/mmcv/video/optflow.py @@ -149,8 +149,8 @@ def dequantize_flow(dx: np.ndarray, dx, dy = (dequantize(d, -max_val, max_val, 255) for d in [dx, dy]) if denorm: - dx *= dx.shape[1] - dy *= dx.shape[0] + dx *= dx.shape[1] # type: ignore + dy *= dx.shape[0] # type: ignore flow = np.dstack((dx, dy)) return flow diff --git a/mmcv/visualization/image.py b/mmcv/visualization/image.py index e7ac4c1817..0c4624aabd 100644 --- a/mmcv/visualization/image.py +++ b/mmcv/visualization/image.py @@ -91,7 +91,7 @@ def imshow_bboxes(img: Union[str, np.ndarray], def imshow_det_bboxes(img: Union[str, np.ndarray], bboxes: np.ndarray, labels: np.ndarray, - class_names: List[str] = None, + class_names: Optional[List[str]] = None, score_thr: float = 0, bbox_color: ColorType = 'green', text_color: ColorType = 'green', From eb571966f8f2cedb7a4917d018e8b0ff8929e113 Mon Sep 17 00:00:00 2001 From: HAOCHENYE <21724054@zju.edu.cn> Date: Sat, 2 Nov 2024 02:31:53 +0800 Subject: [PATCH 2/4] [Fix] clang lint --- .github/workflows/lint.yml | 2 +- .../common/cuda/bezier_align_cuda_kernel.cuh | 6 +- .../cuda/riroi_align_rotated_cuda_kernel.cuh | 6 +- .../cuda/roi_align_rotated_cuda_kernel.cuh | 6 +- .../ops/csrc/common/mlu/common_mlu_helper.hpp | 2 +- mmcv/ops/csrc/common/pytorch_mlu_helper.hpp | 4 +- mmcv/ops/csrc/common/pytorch_npu_util.hpp | 4 +- .../ops/csrc/common/utils/spconv/paramsgrid.h | 2 +- .../utils/spconv/tensorview/tensorview.h | 4 +- mmcv/ops/csrc/pytorch/cpu/bezier_align.cpp | 12 ++-- mmcv/ops/csrc/pytorch/cpu/roi_align.cpp | 18 +++--- .../csrc/pytorch/cpu/roi_align_rotated.cpp | 12 ++-- mmcv/ops/csrc/pytorch/cuda/filtered_lrelu.cu | 52 +++++++--------- .../ops/csrc/pytorch/cuda/upfirdn2d_kernel.cu | 6 +- .../csrc/pytorch/npu/bbox_overlaps_npu.cpp | 2 +- .../csrc/pytorch/npu/box_iou_quadri_npu.cpp | 5 +- .../csrc/pytorch/npu/box_iou_rotated_npu.cpp | 3 +- .../pytorch/npu/boxes_overlap_bev_npu.cpp | 21 ++++--- .../pytorch/npu/points_in_box_npu_all.cpp | 15 ++--- .../pytorch/npu/roi_align_rotated_v2_npu.cpp | 62 ++++++++----------- .../pytorch/npu/roipoint_pool3d_forward.cpp | 2 +- mmcv/ops/csrc/pytorch/pybind.cpp | 22 ++++--- .../ops/csrc/pytorch/roi_align_rotated_v2.cpp | 49 ++++++++------- 23 files changed, 155 insertions(+), 162 deletions(-) diff --git a/.github/workflows/lint.yml b/.github/workflows/lint.yml index db6b570b17..f670cbd918 100644 --- a/.github/workflows/lint.yml +++ b/.github/workflows/lint.yml @@ -22,7 +22,7 @@ jobs: - name: Linting run: pre-commit run --all-files - name: Format c/cuda codes with clang-format - uses: DoozyX/clang-format-lint-action@v0.13 + uses: DoozyX/clang-format-lint-action@v0.18 with: source: mmcv/ops/csrc extensions: h,c,cpp,hpp,cu,cuh diff --git a/mmcv/ops/csrc/common/cuda/bezier_align_cuda_kernel.cuh b/mmcv/ops/csrc/common/cuda/bezier_align_cuda_kernel.cuh index 537610416e..44fbc4f338 100644 --- a/mmcv/ops/csrc/common/cuda/bezier_align_cuda_kernel.cuh +++ b/mmcv/ops/csrc/common/cuda/bezier_align_cuda_kernel.cuh @@ -222,9 +222,9 @@ __global__ void bezier_align_backward_cuda_kernel( atomicAdd(offset_bottom_diff + y_high * width + x_high, static_cast(g4)); } // if - } // ix - } // iy - } // CUDA_1D_KERNEL_LOOP + } // ix + } // iy + } // CUDA_1D_KERNEL_LOOP } // BezierAlignBackward #endif // BEZIER_ALIGN_CUDA_KERNEL_CUH diff --git a/mmcv/ops/csrc/common/cuda/riroi_align_rotated_cuda_kernel.cuh b/mmcv/ops/csrc/common/cuda/riroi_align_rotated_cuda_kernel.cuh index 4383d9e82c..4c0ff1f7c5 100644 --- a/mmcv/ops/csrc/common/cuda/riroi_align_rotated_cuda_kernel.cuh +++ b/mmcv/ops/csrc/common/cuda/riroi_align_rotated_cuda_kernel.cuh @@ -234,9 +234,9 @@ __global__ void riroi_align_rotated_backward_cuda_kernel( g4 * l_var); } // if - } // ix - } // iy - } // CUDA_1D_KERNEL_LOOP + } // ix + } // iy + } // CUDA_1D_KERNEL_LOOP } // RiRoIAlignBackward #endif // RIROI_ALIGN_ROTATED_CUDA_KERNEL_CUH diff --git a/mmcv/ops/csrc/common/cuda/roi_align_rotated_cuda_kernel.cuh b/mmcv/ops/csrc/common/cuda/roi_align_rotated_cuda_kernel.cuh index 8274dc50c7..2fd465eaa2 100644 --- a/mmcv/ops/csrc/common/cuda/roi_align_rotated_cuda_kernel.cuh +++ b/mmcv/ops/csrc/common/cuda/roi_align_rotated_cuda_kernel.cuh @@ -194,9 +194,9 @@ __global__ void roi_align_rotated_backward_cuda_kernel( atomicAdd(offset_bottom_diff + y_high * width + x_low, g3); atomicAdd(offset_bottom_diff + y_high * width + x_high, g4); } // if - } // ix - } // iy - } // CUDA_1D_KERNEL_LOOP + } // ix + } // iy + } // CUDA_1D_KERNEL_LOOP } // RoIAlignBackward #endif // ROI_ALIGN_ROTATED_CUDA_KERNEL_CUH diff --git a/mmcv/ops/csrc/common/mlu/common_mlu_helper.hpp b/mmcv/ops/csrc/common/mlu/common_mlu_helper.hpp index 8527372241..0518806208 100644 --- a/mmcv/ops/csrc/common/mlu/common_mlu_helper.hpp +++ b/mmcv/ops/csrc/common/mlu/common_mlu_helper.hpp @@ -33,7 +33,7 @@ #define PAD_DOWN(x, y) (((x) / (y)) * (y)) #endif -#define CEIL_ALIGN(x, y) (((x) + (y)-1) / (y) * (y)) +#define CEIL_ALIGN(x, y) (((x) + (y) - 1) / (y) * (y)) template __mlu_func__ inline scalar_t min(scalar_t a, scalar_t b) { diff --git a/mmcv/ops/csrc/common/pytorch_mlu_helper.hpp b/mmcv/ops/csrc/common/pytorch_mlu_helper.hpp index e49572ca84..1b264a9a1e 100644 --- a/mmcv/ops/csrc/common/pytorch_mlu_helper.hpp +++ b/mmcv/ops/csrc/common/pytorch_mlu_helper.hpp @@ -21,9 +21,9 @@ #define PAD_DOWN(x, y) (((x) / (y)) * (y)) -#define CEIL_DIV(x, y) (((x) + (y)-1) / (y)) +#define CEIL_DIV(x, y) (((x) + (y) - 1) / (y)) -#define CEIL_ALIGN(x, y) (((x) + (y)-1) / (y) * (y)) +#define CEIL_ALIGN(x, y) (((x) + (y) - 1) / (y) * (y)) inline int32_t getJobLimitCapability() { CNcontext drv_ctx; diff --git a/mmcv/ops/csrc/common/pytorch_npu_util.hpp b/mmcv/ops/csrc/common/pytorch_npu_util.hpp index 3c3712a933..aed56ca1e0 100644 --- a/mmcv/ops/csrc/common/pytorch_npu_util.hpp +++ b/mmcv/ops/csrc/common/pytorch_npu_util.hpp @@ -468,7 +468,7 @@ void ReleaseConvertTypes(Tuple &t) { } template -constexpr auto ConvertTypes(Ts &... args) { +constexpr auto ConvertTypes(Ts &...args) { return std::make_tuple(ConvertType(args)...); } @@ -506,7 +506,7 @@ void AddParamToBuf(const string &); void AddParamToBuf(); template -void AddParamToBuf(const T &arg, Args &... args) { +void AddParamToBuf(const T &arg, Args &...args) { AddParamToBuf(arg); AddParamToBuf(args...); } diff --git a/mmcv/ops/csrc/common/utils/spconv/paramsgrid.h b/mmcv/ops/csrc/common/utils/spconv/paramsgrid.h index f23ff44823..a4b9024715 100644 --- a/mmcv/ops/csrc/common/utils/spconv/paramsgrid.h +++ b/mmcv/ops/csrc/common/utils/spconv/paramsgrid.h @@ -40,7 +40,7 @@ void assigner(TT &src, std::vector counter, std::vector &arg) { template void assigner(TT &src, std::vector counter, std::vector &arg, - std::vector &... args) { + std::vector &...args) { std::get(src) = arg[counter[Idx]]; assigner(src, counter, args...); } diff --git a/mmcv/ops/csrc/common/utils/spconv/tensorview/tensorview.h b/mmcv/ops/csrc/common/utils/spconv/tensorview/tensorview.h index 66e01a8ed1..88eca8084c 100644 --- a/mmcv/ops/csrc/common/utils/spconv/tensorview/tensorview.h +++ b/mmcv/ops/csrc/common/utils/spconv/tensorview/tensorview.h @@ -108,7 +108,7 @@ struct CPU {}; template struct SimpleVector { public: - TV_HOST_DEVICE_INLINE SimpleVector(){}; + TV_HOST_DEVICE_INLINE SimpleVector() {}; TV_HOST_DEVICE_INLINE SimpleVector(std::initializer_list q) { TV_ASSERT(q.size() <= MaxDim); mSize = 0; @@ -315,7 +315,7 @@ struct Slice { template struct ShapeBase : public SimpleVector { - TV_HOST_DEVICE_INLINE ShapeBase() : SimpleVector(){}; + TV_HOST_DEVICE_INLINE ShapeBase() : SimpleVector() {}; TV_HOST_DEVICE_INLINE ShapeBase(std::initializer_list shape) : SimpleVector(shape) {} diff --git a/mmcv/ops/csrc/pytorch/cpu/bezier_align.cpp b/mmcv/ops/csrc/pytorch/cpu/bezier_align.cpp index 7eb0e5b940..21e2c5fc87 100644 --- a/mmcv/ops/csrc/pytorch/cpu/bezier_align.cpp +++ b/mmcv/ops/csrc/pytorch/cpu/bezier_align.cpp @@ -220,9 +220,9 @@ void BezierAlignForward(const int nthreads, const T *input, const T *rois, output[index] = output_val; } // for pw - } // for ph - } // for c - } // for n + } // for ph + } // for c + } // for n } template @@ -381,9 +381,9 @@ void BezierAlignBackward(const int nthreads, const T *grad_output, add(offset_grad_input + y_high * width + x_low, static_cast(g3)); add(offset_grad_input + y_high * width + x_high, static_cast(g4)); } // if - } // ix - } // iy - } // for + } // ix + } // iy + } // for } // BezierAlignBackward void BezierAlignForwardCPULauncher(Tensor input, Tensor rois, Tensor output, diff --git a/mmcv/ops/csrc/pytorch/cpu/roi_align.cpp b/mmcv/ops/csrc/pytorch/cpu/roi_align.cpp index d545390645..fc17a0b122 100644 --- a/mmcv/ops/csrc/pytorch/cpu/roi_align.cpp +++ b/mmcv/ops/csrc/pytorch/cpu/roi_align.cpp @@ -207,10 +207,10 @@ void ROIAlignForward(const int nthreads, const T* input, const T* rois, // We do average (integral) pooling inside a bin output[index] = output_val / count; } // if - } // for pw - } // for ph - } // for c - } // for n + } // for pw + } // for ph + } // for c + } // for n } template @@ -334,7 +334,7 @@ void ROIAlignBackward(const int nthreads, const T* grad_output, const T* rois, add(offset_grad_input + y_high * width + x_low, static_cast(g3)); add(offset_grad_input + y_high * width + x_high, static_cast(g4)); } // if - } // mode + } // mode } else if (pool_mode == 1) { // We do average (integral) pooling inside a bin // We use roi_bin_grid to sample the grid and mimic integral @@ -375,10 +375,10 @@ void ROIAlignBackward(const int nthreads, const T* grad_output, const T* rois, add(offset_grad_input + y_high * width + x_high, static_cast(g4)); } // if - } // ix - } // iy - } // mode - } // for + } // ix + } // iy + } // mode + } // for } // ROIAlignBackward void ROIAlignForwardCPULauncher(Tensor input, Tensor rois, Tensor output, diff --git a/mmcv/ops/csrc/pytorch/cpu/roi_align_rotated.cpp b/mmcv/ops/csrc/pytorch/cpu/roi_align_rotated.cpp index 8c849de0cb..6e09ed1517 100644 --- a/mmcv/ops/csrc/pytorch/cpu/roi_align_rotated.cpp +++ b/mmcv/ops/csrc/pytorch/cpu/roi_align_rotated.cpp @@ -206,9 +206,9 @@ void ROIAlignRotatedForward(const int nthreads, const T* input, output[index] = output_val; } // for pw - } // for ph - } // for c - } // for n + } // for ph + } // for c + } // for n } template @@ -366,9 +366,9 @@ void ROIAlignRotatedBackward( add(offset_grad_input + y_high * width + x_low, static_cast(g3)); add(offset_grad_input + y_high * width + x_high, static_cast(g4)); } // if - } // ix - } // iy - } // for + } // ix + } // iy + } // for } // ROIAlignRotatedBackward void ROIAlignRotatedForwardCPULauncher(Tensor input, Tensor rois, Tensor output, diff --git a/mmcv/ops/csrc/pytorch/cuda/filtered_lrelu.cu b/mmcv/ops/csrc/pytorch/cuda/filtered_lrelu.cu index cee9b26168..e1040fd681 100644 --- a/mmcv/ops/csrc/pytorch/cuda/filtered_lrelu.cu +++ b/mmcv/ops/csrc/pytorch/cuda/filtered_lrelu.cu @@ -100,8 +100,9 @@ void *choose_filtered_lrelu_act_kernel(void); //------------------------------------------------------------------------ // Helpers. -enum // Filter modes. -{ MODE_SUSD = 0, // Separable upsampling, separable downsampling. +enum // Filter modes. +{ + MODE_SUSD = 0, // Separable upsampling, separable downsampling. MODE_FUSD = 1, // Full upsampling, separable downsampling. MODE_SUFD = 2, // Separable upsampling, full downsampling. MODE_FUFD = 3, // Full upsampling, full downsampling. @@ -157,12 +158,11 @@ struct InternalType { #define MIN(A, B) ((A) < (B) ? (A) : (B)) #define MAX(A, B) ((A) > (B) ? (A) : (B)) -#define CEIL_DIV(A, B) \ - (((B) == 1) \ - ? (A) \ - : ((B) == 2) ? ((int)((A) + 1) >> 1) \ - : ((B) == 4) ? ((int)((A) + 3) >> 2) \ - : (((A) + ((A) > 0 ? (B)-1 : 0)) / (B))) +#define CEIL_DIV(A, B) \ + (((B) == 1) ? (A) \ + : ((B) == 2) ? ((int)((A) + 1) >> 1) \ + : ((B) == 4) ? ((int)((A) + 3) >> 2) \ + : (((A) + ((A) > 0 ? (B) - 1 : 0)) / (B))) // This works only up to blocks of size 256 x 256 and for all N that are powers // of two. @@ -333,22 +333,16 @@ static __global__ void filtered_lrelu_kernel(filtered_lrelu_kernel_params p) { const int szDownX = tileUpH * tileOutW; // Sizes for shared memory arrays. - const int s_buf0_size_base = - (filterMode == MODE_SUSD) - ? MAX(szIn, szUpXY) - : (filterMode == MODE_FUSD) - ? MAX(szIn, szDownX) - : (filterMode == MODE_SUFD) - ? MAX(szIn, szUpXY) - : (filterMode == MODE_FUFD) ? szIn : -1; - const int s_buf1_size_base = - (filterMode == MODE_SUSD) - ? MAX(szUpX, szDownX) - : (filterMode == MODE_FUSD) - ? szUpXY - : (filterMode == MODE_SUFD) - ? szUpX - : (filterMode == MODE_FUFD) ? szUpXY : -1; + const int s_buf0_size_base = (filterMode == MODE_SUSD) ? MAX(szIn, szUpXY) + : (filterMode == MODE_FUSD) ? MAX(szIn, szDownX) + : (filterMode == MODE_SUFD) ? MAX(szIn, szUpXY) + : (filterMode == MODE_FUFD) ? szIn + : -1; + const int s_buf1_size_base = (filterMode == MODE_SUSD) ? MAX(szUpX, szDownX) + : (filterMode == MODE_FUSD) ? szUpXY + : (filterMode == MODE_SUFD) ? szUpX + : (filterMode == MODE_FUFD) ? szUpXY + : -1; // Ensure U128 alignment. const int s_buf0_size = (s_buf0_size_base + 3) & ~3; @@ -980,17 +974,17 @@ static __global__ void filtered_lrelu_kernel(filtered_lrelu_kernel_params p) { #define X_LOOP(TAPY, PX) \ for (int sx = 0; sx < fuSize / up; sx++) { \ - v.x += a * (scalar_t)c_fu[(sx * up + (((PX)-0) & (up - 1))) + \ + v.x += a * (scalar_t)c_fu[(sx * up + (((PX) - 0) & (up - 1))) + \ (sy * up + (TAPY)) * MAX_FILTER_SIZE]; \ - v.z += b * (scalar_t)c_fu[(sx * up + (((PX)-0) & (up - 1))) + \ + v.z += b * (scalar_t)c_fu[(sx * up + (((PX) - 0) & (up - 1))) + \ (sy * up + (TAPY)) * MAX_FILTER_SIZE]; \ if ((PX) == 0) { \ a = b; \ b = s_tileIn[src0 + 2 + sx + sy * tileInW]; \ } \ - v.y += a * (scalar_t)c_fu[(sx * up + (((PX)-1) & (up - 1))) + \ + v.y += a * (scalar_t)c_fu[(sx * up + (((PX) - 1) & (up - 1))) + \ (sy * up + (TAPY)) * MAX_FILTER_SIZE]; \ - v.w += b * (scalar_t)c_fu[(sx * up + (((PX)-1) & (up - 1))) + \ + v.w += b * (scalar_t)c_fu[(sx * up + (((PX) - 1) & (up - 1))) + \ (sy * up + (TAPY)) * MAX_FILTER_SIZE]; \ if ((PX) == 1) { \ a = b; \ @@ -1447,7 +1441,7 @@ static __global__ void filtered_lrelu_act_kernel( s |= __shfl_xor(s, 4); s |= __shfl_xor(s, 8); #else - s |= __shfl_xor_sync(m, s, 1); // Distribute. + s |= __shfl_xor_sync(m, s, 1); // Distribute. s |= __shfl_xor_sync(m, s, 2); s |= __shfl_xor_sync(m, s, 4); s |= __shfl_xor_sync(m, s, 8); diff --git a/mmcv/ops/csrc/pytorch/cuda/upfirdn2d_kernel.cu b/mmcv/ops/csrc/pytorch/cuda/upfirdn2d_kernel.cu index 2f1ae0a683..63f6e11e94 100644 --- a/mmcv/ops/csrc/pytorch/cuda/upfirdn2d_kernel.cu +++ b/mmcv/ops/csrc/pytorch/cuda/upfirdn2d_kernel.cu @@ -225,9 +225,9 @@ static __global__ void upfirdn2d_kernel_small(upfirdn2d_kernel_params p) { scalar_t v = 0; if (inX >= 0 & inY >= 0 & inX < p.inSize.x & inY < p.inSize.y & c < p.inSize.z) - v = (scalar_t)( - (const T *)p.x)[inX * p.inStride.x + inY * p.inStride.y + - c * p.inStride.z + n * p.inStride.w]; + v = (scalar_t)((const T *) + p.x)[inX * p.inStride.x + inY * p.inStride.y + + c * p.inStride.z + n * p.inStride.w]; sx[relInY][relInX][relC] = v; } diff --git a/mmcv/ops/csrc/pytorch/npu/bbox_overlaps_npu.cpp b/mmcv/ops/csrc/pytorch/npu/bbox_overlaps_npu.cpp index 4a110ae07c..c93c0635cf 100644 --- a/mmcv/ops/csrc/pytorch/npu/bbox_overlaps_npu.cpp +++ b/mmcv/ops/csrc/pytorch/npu/bbox_overlaps_npu.cpp @@ -26,7 +26,7 @@ void bbox_overlaps_npu(const Tensor bboxes1, const Tensor bboxes2, Tensor ious, gtboxesFP32 = gtboxesFP32.to(at::kFloat); } c10::SmallVector iousSize = {gtboxesFP32.size(0), - bboxesFP32.size(0)}; + bboxesFP32.size(0)}; if (aligned) { iousSize = {gtboxesFP32.size(0), 1}; } diff --git a/mmcv/ops/csrc/pytorch/npu/box_iou_quadri_npu.cpp b/mmcv/ops/csrc/pytorch/npu/box_iou_quadri_npu.cpp index 6baf44f448..84e2c0a678 100644 --- a/mmcv/ops/csrc/pytorch/npu/box_iou_quadri_npu.cpp +++ b/mmcv/ops/csrc/pytorch/npu/box_iou_quadri_npu.cpp @@ -4,11 +4,10 @@ using namespace NPU_NAME_SPACE; using namespace std; void box_iou_quadri_impl(const Tensor boxes1, const Tensor boxes2, Tensor ious, - const int mode_flag, const bool aligned); + const int mode_flag, const bool aligned); void box_iou_quadri_npu(const Tensor boxes1, const Tensor boxes2, Tensor ious, - const int mode_flag, const bool aligned) { - + const int mode_flag, const bool aligned) { TORCH_CHECK(boxes1.size(1) == 8, "boxes1 must be 2D tensor (N, 8)"); TORCH_CHECK(boxes1.size(1) == 8, "boxes1 must be 2D tensor (N, 8)"); diff --git a/mmcv/ops/csrc/pytorch/npu/box_iou_rotated_npu.cpp b/mmcv/ops/csrc/pytorch/npu/box_iou_rotated_npu.cpp index d8b0bbaa67..5b229a3926 100644 --- a/mmcv/ops/csrc/pytorch/npu/box_iou_rotated_npu.cpp +++ b/mmcv/ops/csrc/pytorch/npu/box_iou_rotated_npu.cpp @@ -8,14 +8,13 @@ void box_iou_rotated_impl(const Tensor boxes1, const Tensor boxes2, Tensor ious, void box_iou_rotated_npu(const Tensor boxes1, const Tensor boxes2, Tensor ious, const int mode_flag, const bool aligned) { - TORCH_CHECK(boxes1.size(1) == 5, "boxes1 must be 2D tensor (N, 5)"); TORCH_CHECK(boxes1.size(1) == 5, "boxes1 must be 2D tensor (N, 5)"); auto trans = false; auto is_clockwise = false; EXEC_NPU_CMD(aclnnBoxesOverlapBev, boxes1, boxes2, trans, is_clockwise, - aligned, mode_flag, ious); + aligned, mode_flag, ious); return; } diff --git a/mmcv/ops/csrc/pytorch/npu/boxes_overlap_bev_npu.cpp b/mmcv/ops/csrc/pytorch/npu/boxes_overlap_bev_npu.cpp index 6bc6273083..0d97df6ad7 100644 --- a/mmcv/ops/csrc/pytorch/npu/boxes_overlap_bev_npu.cpp +++ b/mmcv/ops/csrc/pytorch/npu/boxes_overlap_bev_npu.cpp @@ -10,16 +10,17 @@ void iou3d_boxes_overlap_bev_forward_impl(const int num_a, const Tensor boxes_a, void iou3d_boxes_overlap_bev_forward_npu(const int num_a, const Tensor boxes_a, const int num_b, const Tensor boxes_b, Tensor ans_overlap) { + TORCH_CHECK(boxes_a.size(1) == 7, "boxes_a must be 2D tensor (N, 7)"); + TORCH_CHECK(boxes_b.size(1) == 7, "boxes_b must be 2D tensor (N, 7)"); - TORCH_CHECK(boxes_a.size(1) == 7, "boxes_a must be 2D tensor (N, 7)"); - TORCH_CHECK(boxes_b.size(1) == 7, "boxes_b must be 2D tensor (N, 7)"); - - auto trans = false; - auto is_clockwise = false; - auto aligned = false; - auto mode_flag = 2; - EXEC_NPU_CMD(aclnnBoxesOverlapBev, boxes_a, boxes_b, trans, is_clockwise, aligned, mode_flag, ans_overlap); - return; + auto trans = false; + auto is_clockwise = false; + auto aligned = false; + auto mode_flag = 2; + EXEC_NPU_CMD(aclnnBoxesOverlapBev, boxes_a, boxes_b, trans, is_clockwise, + aligned, mode_flag, ans_overlap); + return; } -REGISTER_NPU_IMPL(iou3d_boxes_overlap_bev_forward_impl, iou3d_boxes_overlap_bev_forward_npu); +REGISTER_NPU_IMPL(iou3d_boxes_overlap_bev_forward_impl, + iou3d_boxes_overlap_bev_forward_npu); diff --git a/mmcv/ops/csrc/pytorch/npu/points_in_box_npu_all.cpp b/mmcv/ops/csrc/pytorch/npu/points_in_box_npu_all.cpp index ae3befea63..e3cc9284cc 100644 --- a/mmcv/ops/csrc/pytorch/npu/points_in_box_npu_all.cpp +++ b/mmcv/ops/csrc/pytorch/npu/points_in_box_npu_all.cpp @@ -4,16 +4,17 @@ using namespace NPU_NAME_SPACE; using namespace std; void points_in_boxes_all_forward_impl_npu(int batch_size, int boxes_num, - int pts_num, const Tensor boxes, - const Tensor pts, - Tensor box_idx_of_points) { - c10::SmallVector output_size = {pts.size(0), pts.size(1), boxes.size(1)}; + int pts_num, const Tensor boxes, + const Tensor pts, + Tensor box_idx_of_points) { + c10::SmallVector output_size = {pts.size(0), pts.size(1), + boxes.size(1)}; auto boxes_trans = boxes.transpose(1, 2).contiguous(); EXEC_NPU_CMD(aclnnPointsInBoxAll, boxes_trans, pts, box_idx_of_points); } void points_in_boxes_all_forward_impl(int batch_size, int boxes_num, - int pts_num, const Tensor boxes, - const Tensor pts, - Tensor box_idx_of_points); + int pts_num, const Tensor boxes, + const Tensor pts, + Tensor box_idx_of_points); REGISTER_NPU_IMPL(points_in_boxes_all_forward_impl, points_in_boxes_all_forward_impl_npu); diff --git a/mmcv/ops/csrc/pytorch/npu/roi_align_rotated_v2_npu.cpp b/mmcv/ops/csrc/pytorch/npu/roi_align_rotated_v2_npu.cpp index b2ea93b261..f9fac97397 100644 --- a/mmcv/ops/csrc/pytorch/npu/roi_align_rotated_v2_npu.cpp +++ b/mmcv/ops/csrc/pytorch/npu/roi_align_rotated_v2_npu.cpp @@ -4,49 +4,41 @@ using namespace NPU_NAME_SPACE; using namespace std; void roi_align_rotated_v2_forward_npu(const Tensor input, Tensor rois_map, - Tensor output, - double spatial_scale, - int32_t sampling_ratio, - int32_t pooled_height, - int32_t pooled_width, - bool aligned, - bool clockwise) { + Tensor output, double spatial_scale, + int32_t sampling_ratio, + int32_t pooled_height, + int32_t pooled_width, bool aligned, + bool clockwise) { at::Tensor feature_map = input.permute({0, 2, 3, 1}).contiguous(); at::Tensor rois = rois_map.permute({1, 0}).contiguous(); - EXEC_NPU_CMD(aclnnRoiAlignRotatedV2, feature_map, rois, spatial_scale, sampling_ratio, pooled_height, pooled_width, aligned, clockwise, output); + EXEC_NPU_CMD(aclnnRoiAlignRotatedV2, feature_map, rois, spatial_scale, + sampling_ratio, pooled_height, pooled_width, aligned, clockwise, + output); } void roi_align_rotated_v2_forward_impl(const Tensor input, Tensor rois, - Tensor output, - double spatial_scale, - int32_t sampling_ratio, - int32_t pooled_height, - int32_t pooled_width, - bool aligned, - bool clockwise); + Tensor output, double spatial_scale, + int32_t sampling_ratio, + int32_t pooled_height, + int32_t pooled_width, bool aligned, + bool clockwise); -REGISTER_NPU_IMPL(roi_align_rotated_v2_forward_impl, roi_align_rotated_v2_forward_npu); +REGISTER_NPU_IMPL(roi_align_rotated_v2_forward_impl, + roi_align_rotated_v2_forward_npu); -void roi_align_rotated_v2_backward_npu(const Tensor input, Tensor rois, - Tensor grad_output, Tensor grad_input, - int32_t pooled_height, - int32_t pooled_width, - double spatial_scale, - int32_t sampling_ratio, - bool aligned, - bool clockwise) { +void roi_align_rotated_v2_backward_npu( + const Tensor input, Tensor rois, Tensor grad_output, Tensor grad_input, + int32_t pooled_height, int32_t pooled_width, double spatial_scale, + int32_t sampling_ratio, bool aligned, bool clockwise) { EXEC_NPU_CMD(aclnnRoiAlignRotatedGradV2, input, rois, grad_output, - pooled_height, pooled_width, spatial_scale, sampling_ratio, aligned, clockwise, - grad_input); + pooled_height, pooled_width, spatial_scale, sampling_ratio, + aligned, clockwise, grad_input); } -void roi_align_rotated_v2_backward_impl(const Tensor input, Tensor rois, - Tensor grad_output, Tensor grad_input, - int32_t pooled_height, - int32_t pooled_width, - double spatial_scale, - int32_t sampling_ratio, - bool aligned, - bool clockwise); +void roi_align_rotated_v2_backward_impl( + const Tensor input, Tensor rois, Tensor grad_output, Tensor grad_input, + int32_t pooled_height, int32_t pooled_width, double spatial_scale, + int32_t sampling_ratio, bool aligned, bool clockwise); -REGISTER_NPU_IMPL(roi_align_rotated_v2_backward_impl, roi_align_rotated_v2_backward_npu); +REGISTER_NPU_IMPL(roi_align_rotated_v2_backward_impl, + roi_align_rotated_v2_backward_npu); diff --git a/mmcv/ops/csrc/pytorch/npu/roipoint_pool3d_forward.cpp b/mmcv/ops/csrc/pytorch/npu/roipoint_pool3d_forward.cpp index 2fc645c62b..7d9e4b3c9d 100644 --- a/mmcv/ops/csrc/pytorch/npu/roipoint_pool3d_forward.cpp +++ b/mmcv/ops/csrc/pytorch/npu/roipoint_pool3d_forward.cpp @@ -19,7 +19,7 @@ void roipoint_pool3d_forward_impl_npu(int batch_size, int pts_num, at::Tensor pooled_features_trans = at::empty(features_trans_size, xyz.options()); c10::SmallVector empty_flag_size = {boxes3d.size(0), - boxes3d.size(1)}; + boxes3d.size(1)}; EXEC_NPU_CMD(aclnnRoipointPool3dForward, points_trans, point_features_trans, boxes3d, sampled_pts_num, pooled_features_trans, pooled_empty_flag); diff --git a/mmcv/ops/csrc/pytorch/pybind.cpp b/mmcv/ops/csrc/pytorch/pybind.cpp index c56c6437d1..864630f838 100644 --- a/mmcv/ops/csrc/pytorch/pybind.cpp +++ b/mmcv/ops/csrc/pytorch/pybind.cpp @@ -209,13 +209,15 @@ void roi_align_backward(Tensor grad_output, Tensor rois, Tensor argmax_y, int sampling_ratio, int pool_mode, bool aligned); void roi_align_rotated_v2_forward(Tensor input, Tensor rois, Tensor output, - double spatial_scale, int sampling_ratio, - int aligned_height, int aligned_width, - bool aligned, bool clockwise); + double spatial_scale, int sampling_ratio, + int aligned_height, int aligned_width, + bool aligned, bool clockwise); -void roi_align_rotated_v2_backward(Tensor input, Tensor rois, Tensor grad_output, Tensor grad_input, - int pooled_height, int pooled_width, double spatial_scale, - int sampling_ratio, bool aligned, bool clockwise); +void roi_align_rotated_v2_backward(Tensor input, Tensor rois, + Tensor grad_output, Tensor grad_input, + int pooled_height, int pooled_width, + double spatial_scale, int sampling_ratio, + bool aligned, bool clockwise); void roi_pool_forward(Tensor input, Tensor rois, Tensor output, Tensor argmax, int pooled_height, int pooled_width, float spatial_scale); @@ -804,13 +806,13 @@ PYBIND11_MODULE(TORCH_EXTENSION_NAME, m) { m.def("roi_align_rotated_v2_forward", &roi_align_rotated_v2_forward, "roi_align_rotated_v2_forward", py::arg("input"), py::arg("rois"), py::arg("output"), py::arg("spatial_scale"), py::arg("sampling_ratio"), - py::arg("pooled_height"), py::arg("pooled_width"), - py::arg("aligned"), py::arg("clockwise")); + py::arg("pooled_height"), py::arg("pooled_width"), py::arg("aligned"), + py::arg("clockwise")); m.def("roi_align_rotated_v2_backward", &roi_align_rotated_v2_backward, "roi_align_rotated_v2_backward", py::arg("input"), py::arg("rois"), py::arg("grad_output"), py::arg("grad_input"), py::arg("pooled_height"), - py::arg("pooled_width"), py::arg("spatial_scale"), py::arg("sampling_ratio"), - py::arg("aligned"), py::arg("clockwise")); + py::arg("pooled_width"), py::arg("spatial_scale"), + py::arg("sampling_ratio"), py::arg("aligned"), py::arg("clockwise")); m.def("dynamic_point_to_voxel_forward", &dynamic_point_to_voxel_forward, "dynamic_point_to_voxel_forward", py::arg("feats"), py::arg("coors"), py::arg("reduce_type")); diff --git a/mmcv/ops/csrc/pytorch/roi_align_rotated_v2.cpp b/mmcv/ops/csrc/pytorch/roi_align_rotated_v2.cpp index 7743775288..9b0a623530 100644 --- a/mmcv/ops/csrc/pytorch/roi_align_rotated_v2.cpp +++ b/mmcv/ops/csrc/pytorch/roi_align_rotated_v2.cpp @@ -3,35 +3,40 @@ #include "pytorch_device_registry.hpp" void roi_align_rotated_v2_forward_impl(Tensor input, Tensor rois, Tensor output, - double spatial_scale, int sampling_ratio, - int pooled_height, int pooled_width, - bool aligned, bool clockwise) { + double spatial_scale, int sampling_ratio, + int pooled_height, int pooled_width, + bool aligned, bool clockwise) { DISPATCH_DEVICE_IMPL(roi_align_rotated_v2_forward_impl, input, rois, output, - spatial_scale, sampling_ratio, pooled_height, pooled_width, - aligned, clockwise); + spatial_scale, sampling_ratio, pooled_height, + pooled_width, aligned, clockwise); } - void roi_align_rotated_v2_forward(Tensor input, Tensor rois, Tensor output, - double spatial_scale, int sampling_ratio, - int pooled_height, int pooled_width, - bool aligned, bool clockwise) { - roi_align_rotated_v2_forward_impl(input, rois, output, spatial_scale, sampling_ratio, - pooled_height, pooled_width, aligned, clockwise); + double spatial_scale, int sampling_ratio, + int pooled_height, int pooled_width, + bool aligned, bool clockwise) { + roi_align_rotated_v2_forward_impl(input, rois, output, spatial_scale, + sampling_ratio, pooled_height, pooled_width, + aligned, clockwise); } - -void roi_align_rotated_v2_backward_impl(Tensor input, Tensor rois, Tensor grad_output, Tensor grad_input, - int pooled_height, int pooled_width, double spatial_scale, - int sampling_ratio, bool aligned, bool clockwise) { - DISPATCH_DEVICE_IMPL(roi_align_rotated_v2_backward_impl, input, rois, grad_output, grad_input, - pooled_height, pooled_width, spatial_scale, sampling_ratio, aligned, clockwise); +void roi_align_rotated_v2_backward_impl(Tensor input, Tensor rois, + Tensor grad_output, Tensor grad_input, + int pooled_height, int pooled_width, + double spatial_scale, + int sampling_ratio, bool aligned, + bool clockwise) { + DISPATCH_DEVICE_IMPL(roi_align_rotated_v2_backward_impl, input, rois, + grad_output, grad_input, pooled_height, pooled_width, + spatial_scale, sampling_ratio, aligned, clockwise); } - -void roi_align_rotated_v2_backward(Tensor input, Tensor rois, Tensor grad_output, Tensor grad_input, - int pooled_height, int pooled_width, double spatial_scale, - int sampling_ratio, bool aligned, bool clockwise) { +void roi_align_rotated_v2_backward(Tensor input, Tensor rois, + Tensor grad_output, Tensor grad_input, + int pooled_height, int pooled_width, + double spatial_scale, int sampling_ratio, + bool aligned, bool clockwise) { roi_align_rotated_v2_backward_impl(input, rois, grad_output, grad_input, - pooled_height, pooled_width, spatial_scale, sampling_ratio, aligned, clockwise); + pooled_height, pooled_width, spatial_scale, + sampling_ratio, aligned, clockwise); } From 821a47ea4abf8a3e28d0fe44d03e624717eefa70 Mon Sep 17 00:00:00 2001 From: HAOCHENYE <21724054@zju.edu.cn> Date: Sat, 2 Nov 2024 03:31:30 +0800 Subject: [PATCH 3/4] use ACTIONS_ALLOW_USE_UNSECURE_NODE_VERSION --- .github/workflows/pr_stage_test.yml | 3 +++ 1 file changed, 3 insertions(+) diff --git a/.github/workflows/pr_stage_test.yml b/.github/workflows/pr_stage_test.yml index 8dfefeed7c..597049c81b 100644 --- a/.github/workflows/pr_stage_test.yml +++ b/.github/workflows/pr_stage_test.yml @@ -1,5 +1,8 @@ name: pr_stage_test +env: + ACTIONS_ALLOW_USE_UNSECURE_NODE_VERSION: true + on: pull_request: paths-ignore: From a3786374440a607054f7c65537a55cfe3be58299 Mon Sep 17 00:00:00 2001 From: HAOCHENYE <21724054@zju.edu.cn> Date: Sat, 2 Nov 2024 04:05:52 +0800 Subject: [PATCH 4/4] Add todo --- .pre-commit-config-zh-cn.yaml | 1 + 1 file changed, 1 insertion(+) diff --git a/.pre-commit-config-zh-cn.yaml b/.pre-commit-config-zh-cn.yaml index f705ec904a..1db6a40118 100644 --- a/.pre-commit-config-zh-cn.yaml +++ b/.pre-commit-config-zh-cn.yaml @@ -43,6 +43,7 @@ repos: - mdformat_frontmatter - linkify-it-py - repo: https://gitee.com/openmmlab/mirrors-docformatter + # TODO:https://github.com/PyCQA/docformatter/issues/289 rev: v1.3.1 hooks: - id: docformatter