From 09ed18d2802d892755ff168d9b6677c462ae4ec2 Mon Sep 17 00:00:00 2001 From: MengqingCao Date: Wed, 21 Feb 2024 10:33:09 +0800 Subject: [PATCH] Remove redundant code 1. remove threshold_opencv_kernel 2. typo ASCENDC_KERNELS_H 3. add ALIGN_UP macro --- .../cannops/ascendc_kernels/threshold_opencv_kernel.cpp | 8 -------- modules/cannops/include/opencv2/ascendc_kernels.hpp | 2 +- modules/cannops/src/element_operations.cpp | 2 +- modules/cannops/src/precomp.hpp | 1 + 4 files changed, 3 insertions(+), 10 deletions(-) diff --git a/modules/cannops/ascendc_kernels/threshold_opencv_kernel.cpp b/modules/cannops/ascendc_kernels/threshold_opencv_kernel.cpp index ffab30ebd54..7fa1867c8b1 100644 --- a/modules/cannops/ascendc_kernels/threshold_opencv_kernel.cpp +++ b/modules/cannops/ascendc_kernels/threshold_opencv_kernel.cpp @@ -377,11 +377,3 @@ extern "C" __global__ __aicore__ void threshold_opencv(GM_ADDR tilingGM, GM_ADDR // Clear tiling GM cache manually. (cce compiler bug) dcci(tilingGM, 1); } - -#ifndef __CCE_KT_TEST__ -void threshold_opencv_kernel(uint32_t blockDim, void* l2ctrl, void* stream, uint8_t* tiling, - uint8_t* x, uint8_t* y) -{ - threshold_opencv<<>>(tiling, x, y); -} -#endif diff --git a/modules/cannops/include/opencv2/ascendc_kernels.hpp b/modules/cannops/include/opencv2/ascendc_kernels.hpp index 714b6460fd7..b030920a62f 100644 --- a/modules/cannops/include/opencv2/ascendc_kernels.hpp +++ b/modules/cannops/include/opencv2/ascendc_kernels.hpp @@ -1,5 +1,5 @@ #ifndef ASCENDC_KERNELS_H -#define KERNEL_TILINASCENDC_KERNELS_HG_H +#define ASCENDC_KERNELS_H #include "../../ascendc_kernels/kernel_tiling_types.h" #include "aclrtlaunch_threshold_opencv.h" diff --git a/modules/cannops/src/element_operations.cpp b/modules/cannops/src/element_operations.cpp index 48d9edb596b..cacf6e6cff1 100644 --- a/modules/cannops/src/element_operations.cpp +++ b/modules/cannops/src/element_operations.cpp @@ -444,7 +444,7 @@ double threshold(const AscendMat& src, AscendMat& dst, double thresh, double max tiling.thresh = thresh; // AscendMat memory will be align to 32B, it's safe to set totalLengh a little bigger. size_t totalBytes = src.rows * src.cols * src.channels(); - tiling.totalLength = ((totalBytes + 32) & ~31); + tiling.totalLength = ALIGN_UP(totalBytes, 32); tiling.threshType = type; tiling.dtype = src.depth(); diff --git a/modules/cannops/src/precomp.hpp b/modules/cannops/src/precomp.hpp index 53ed398fde3..8aadaf4d8de 100644 --- a/modules/cannops/src/precomp.hpp +++ b/modules/cannops/src/precomp.hpp @@ -11,5 +11,6 @@ #include "opencv2/cann_interface.hpp" #include "opencv2/cann_private.hpp" #include "opencv2/ascendc_kernels.hpp" +#define ALIGN_UP(num, align) (((num) + (align) - 1) & ~((align) - 1)) #endif /* __OPENCV_PRECOMP_H__ */