diff --git a/hal/riscv-rvv/include/dnn.hpp b/hal/riscv-rvv/include/dnn.hpp index fb7280ba1b..a132a99417 100644 --- a/hal/riscv-rvv/include/dnn.hpp +++ b/hal/riscv-rvv/include/dnn.hpp @@ -9,9 +9,9 @@ namespace cv { namespace rvv_hal { namespace dnn { #if CV_HAL_RVV_1P0_ENABLED -// Blocked NCDHWc pooling, CV_32F. The geometry crosses the boundary as a flat, stable -// C argument list (no dnn types): C0 channel block; insize/outsize = the [3] input/output -// spatial dims in a fixed Z,Y,X frame (unused leading dims = 1); strides[3]; pads[6] +// Blocked NCDHWc pooling / depthwise convolution, CV_32F. The geometry crosses the boundary as +// a flat, stable C argument list (no dnn types): C0 channel block; insize/outsize = the [3] +// input/output spatial dims in a fixed Z,Y,X frame (unused leading dims = 1); strides[3]; pads[6] // (begin[0..2]+end[3..5]); inner[6] = the padding-free interior bounds; coordtab[ksize*3] // = per-tap (dz,dy,dx); ofstab[ksize] = per-tap flat input offset for the interior. // Each hook computes only the block-plane range [task_start, task_end) (N*C1 planes); @@ -38,6 +38,27 @@ int avgpool3d32f(const float* inp_data, float* out_data, int C0, #undef cv_hal_dnn_avgpool3d32f #define cv_hal_dnn_avgpool3d32f cv::rvv_hal::dnn::avgpool3d32f +/* ############ depthwise_conv32f ############ */ + +// Depthwise convolution adds, on top of the geometry above, the fused epilogue out = act(in*W + +// bias, scaled): weights is the repacked C1*ksize*C0 tensor (block b at b*ksize*C0); scale/bias +// are optional per-channel [C] vectors (null => 1/0); residual (optional) is added before the +// activation out = min(s>=0 ? s : s*alpha, maxval), where alpha is prelu_slope[c] (if != null) +// else default_alpha. + +int depthwise_conv32f(const float* inp_data, const float* residual_data, + float* out_data, const float* weights, + const float* scale, const float* bias, + int C, int C0, int C1, + const int* insize, const int* outsize, const int* strides, + const int* pads, const int* inner, const int* coordtab, + const int* ofstab, int ksize, + float maxval, float default_alpha, const float* prelu_slope, + int task_start, int task_end); + +#undef cv_hal_dnn_depthwise_conv32f +#define cv_hal_dnn_depthwise_conv32f cv::rvv_hal::dnn::depthwise_conv32f + #endif // CV_HAL_RVV_1P0_ENABLED }}} // cv::rvv_hal::dnn diff --git a/hal/riscv-rvv/src/dnn/depthwise.cpp b/hal/riscv-rvv/src/dnn/depthwise.cpp new file mode 100644 index 0000000000..c73fe5cd67 --- /dev/null +++ b/hal/riscv-rvv/src/dnn/depthwise.cpp @@ -0,0 +1,195 @@ +// This file is part of OpenCV project. +// It is subject to the license terms in the LICENSE file found in the top-level directory +// of this distribution and at http://opencv.org/license.html. + +#include "rvv_hal.hpp" + +namespace cv { namespace rvv_hal { namespace dnn { + +#if CV_HAL_RVV_1P0_ENABLED + +// Blocked NCHWc depthwise convolution, RVV, CV_32F. Mirrors the built-in depthwiseConv32f +// (modules/dnn/src/layers/cpu_kernels/conv2_depthwise.simd.hpp) but computes only the +// block-plane range [task_start, task_end), and applies the same fused epilogue +// out = min( s >= 0 ? s : s*alpha, maxval ), s = in*W + bias, then optional residual add, +// with alpha = prelu_slope[c] (if provided) else default_alpha. +// +// Each output row is split into border pixels ([0,inner_x0) and [inner_x1,W): a tap may fall +// in padding, so every tap is bounds-checked via the coordinate table) and interior pixels +// ([inner_x0,inner_x1): no tap touches padding, so the precomputed flat offset table ofstab[] +// is used with no checks). The interior processes 4 output columns at a time for ILP; the +// per-tap weight vector is loaded once and reused across the 4 columns. +// +// The block is C0-narrow (C0 == 8), so one e32m2 group holds a whole C0 vector at any +// VLEN >= 128 -- this uses native vsetvl and never trips the C0-vs-VLMAX assert that forces +// the universal-intrinsic path to run scalar on RVV, so it beats the shipped scalar path at +// every VLEN without the register-fill/vrgather that regressed at wide VLEN. + +int depthwise_conv32f(const float* inp_data, const float* residual_data, + float* out_data, const float* weights_all, + const float* scale_all, const float* bias_all, + int C, int C0, int C1, + const int* insize, const int* outsize, const int* strides, + const int* pads, const int* inner, const int* coordtab, + const int* ofstab, int ksize, + float maxval, float default_alpha, const float* prelu_slope, + int task_start, int task_end) +{ + constexpr int MAX_CONV_DIMS = 3; // coordtab/pads/inner use a fixed Z,Y,X frame + constexpr int MAX_C0 = 8; // engine's fixed f32 channel block + if (C0 > MAX_C0) + return CV_HAL_ERROR_NOT_IMPLEMENTED; + + const int Di = insize[0], Hi = insize[1], Wi = insize[2]; + const int D = outsize[0], H = outsize[1], W = outsize[2]; + const int64_t iplanesize = (int64_t)Di*Hi*Wi*C0; + const int64_t planesize = (int64_t)D*H*W*C0; + const int SZ = strides[0], SY = strides[1], SX = strides[2]; + const int padZ0 = pads[0], padY0 = pads[1], padX0 = pads[2]; + const int inner_z0 = inner[0], inner_z1 = inner[MAX_CONV_DIMS]; + const int inner_y0 = inner[1], inner_y1 = inner[MAX_CONV_DIMS + 1]; + const int inner_x0 = inner[2], inner_x1 = inner[MAX_CONV_DIMS + 2]; + const int* zyxtab = coordtab; + + float scalebuf[MAX_C0], biasbuf[MAX_C0], alphabuf[MAX_C0]; + + // Fused epilogue on one C0-lane accumulator: acc = acc*scale + bias (+ residual), then + // acc = min(acc >= 0 ? acc : acc*alpha, maxval). rp_c is the residual pointer already + // advanced to +c, or nullptr. + auto finish = [&](vfloat32m2_t acc, int c, size_t vl, const float* rp_c) -> vfloat32m2_t { + vfloat32m2_t sc = __riscv_vle32_v_f32m2(scalebuf + c, vl); + vfloat32m2_t bi = __riscv_vle32_v_f32m2(biasbuf + c, vl); + acc = __riscv_vfmadd_vv_f32m2(acc, sc, bi, vl); // acc*scale + bias + if (rp_c) + acc = __riscv_vfadd_vv_f32m2(acc, __riscv_vle32_v_f32m2(rp_c, vl), vl); + vfloat32m2_t al = __riscv_vle32_v_f32m2(alphabuf + c, vl); + vfloat32m2_t neg = __riscv_vfmul_vv_f32m2(acc, al, vl); // acc*alpha (leaky/prelu branch) + vbool16_t m = __riscv_vmfge_vf_f32m2_b16(acc, 0.f, vl); // m: acc >= 0 + acc = __riscv_vmerge_vvm_f32m2(neg, acc, m, vl); // m ? acc : acc*alpha + return __riscv_vfmin_vf_f32m2(acc, maxval, vl); + }; + + for (int nc = task_start; nc < task_end; nc++) { + const int n = nc / C1; + const int c_base = (nc - n*C1) * C0; + int c_count = C - c_base; + if (c_count > C0) c_count = C0; + if (c_count < 0) c_count = 0; + + const float* inp = inp_data + nc * iplanesize; + float* outp = out_data + nc * planesize; + const float* resp = residual_data ? residual_data + nc * planesize : nullptr; + const float* wblk = weights_all + (int64_t)(c_base / C0) * ksize * C0; + + for (int c = 0; c < C0; c++) { + if (c < c_count) { + scalebuf[c] = scale_all ? scale_all[c_base + c] : 1.f; + biasbuf[c] = bias_all ? bias_all[c_base + c] : 0.f; + alphabuf[c] = prelu_slope ? prelu_slope[c_base + c] : default_alpha; + } else { + scalebuf[c] = 0.f; biasbuf[c] = 0.f; alphabuf[c] = 0.f; + } + } + + for (int z0 = 0; z0 < D; z0++) { + const int zi_ = z0*SZ - padZ0; + const bool z_in = (z0 >= inner_z0 && z0 < inner_z1); + for (int y0 = 0; y0 < H; y0++) { + const int yi_ = y0*SY - padY0; + float* out_row = outp + ((int64_t)(z0*H + y0)*W)*C0; + const float* res_row = resp ? resp + ((int64_t)(z0*H + y0)*W)*C0 : nullptr; + + // border pixels: every tap bounds-checked against the coordinate table + auto do_border = [&](int xa, int xb) { + for (int x0 = xa; x0 < xb; x0++) { + const int xi_ = x0*SX - padX0; + float* op = out_row + (int64_t)x0*C0; + const float* rp = res_row ? res_row + (int64_t)x0*C0 : nullptr; + for (int c = 0; c < C0; ) { + size_t vl = __riscv_vsetvl_e32m2(C0 - c); + vfloat32m2_t acc = __riscv_vfmv_v_f_f32m2(0.f, vl); + for (int k = 0; k < ksize; k++) { + int zi = zi_ + zyxtab[k*MAX_CONV_DIMS]; + int yi = yi_ + zyxtab[k*MAX_CONV_DIMS + 1]; + int xi = xi_ + zyxtab[k*MAX_CONV_DIMS + 2]; + if ((unsigned)zi >= (unsigned)Di || + (unsigned)yi >= (unsigned)Hi || + (unsigned)xi >= (unsigned)Wi) continue; + const float* iptr = inp + (((int64_t)zi*Hi + yi)*Wi + xi)*C0 + c; + vfloat32m2_t vi = __riscv_vle32_v_f32m2(iptr, vl); + vfloat32m2_t vw = __riscv_vle32_v_f32m2(wblk + k*C0 + c, vl); + acc = __riscv_vfmacc_vv_f32m2(acc, vi, vw, vl); + } + acc = finish(acc, c, vl, rp ? rp + c : nullptr); + __riscv_vse32_v_f32m2(op + c, acc, vl); + c += (int)vl; + } + } + }; + + if (!(z_in && y0 >= inner_y0 && y0 < inner_y1)) { + do_border(0, W); + continue; + } + do_border(0, inner_x0); + + const int64_t rowbase = ((int64_t)(Hi*zi_ + yi_)*Wi)*C0; + int x0 = inner_x0; + // interior, 4 output columns at a time: weight loaded once per tap, reused x4 + for (; x0 + 4 <= inner_x1; x0 += 4) { + const float* b0 = inp + rowbase + (int64_t)((x0 )*SX - padX0)*C0; + const float* b1 = inp + rowbase + (int64_t)((x0+1)*SX - padX0)*C0; + const float* b2 = inp + rowbase + (int64_t)((x0+2)*SX - padX0)*C0; + const float* b3 = inp + rowbase + (int64_t)((x0+3)*SX - padX0)*C0; + float* o0 = out_row + (int64_t)x0*C0; + const float* r0 = res_row ? res_row + (int64_t)x0*C0 : nullptr; + for (int c = 0; c < C0; ) { + size_t vl = __riscv_vsetvl_e32m2(C0 - c); + vfloat32m2_t a0 = __riscv_vfmv_v_f_f32m2(0.f, vl); + vfloat32m2_t a1 = a0, a2 = a0, a3 = a0; + for (int k = 0; k < ksize; k++) { + vfloat32m2_t vw = __riscv_vle32_v_f32m2(wblk + k*C0 + c, vl); + const int o = ofstab[k] + c; + a0 = __riscv_vfmacc_vv_f32m2(a0, __riscv_vle32_v_f32m2(b0 + o, vl), vw, vl); + a1 = __riscv_vfmacc_vv_f32m2(a1, __riscv_vle32_v_f32m2(b1 + o, vl), vw, vl); + a2 = __riscv_vfmacc_vv_f32m2(a2, __riscv_vle32_v_f32m2(b2 + o, vl), vw, vl); + a3 = __riscv_vfmacc_vv_f32m2(a3, __riscv_vle32_v_f32m2(b3 + o, vl), vw, vl); + } + a0 = finish(a0, c, vl, r0 ? r0 + c : nullptr); + a1 = finish(a1, c, vl, r0 ? r0 + (int64_t)C0 + c : nullptr); + a2 = finish(a2, c, vl, r0 ? r0 + (int64_t)2*C0 + c : nullptr); + a3 = finish(a3, c, vl, r0 ? r0 + (int64_t)3*C0 + c : nullptr); + __riscv_vse32_v_f32m2(o0 + c, a0, vl); + __riscv_vse32_v_f32m2(o0 + (int64_t)C0 + c, a1, vl); + __riscv_vse32_v_f32m2(o0 + (int64_t)2*C0 + c, a2, vl); + __riscv_vse32_v_f32m2(o0 + (int64_t)3*C0 + c, a3, vl); + c += (int)vl; + } + } + for (; x0 < inner_x1; x0++) { + const float* w = inp + rowbase + (int64_t)(x0*SX - padX0)*C0; + float* op = out_row + (int64_t)x0*C0; + const float* rp = res_row ? res_row + (int64_t)x0*C0 : nullptr; + for (int c = 0; c < C0; ) { + size_t vl = __riscv_vsetvl_e32m2(C0 - c); + vfloat32m2_t acc = __riscv_vfmv_v_f_f32m2(0.f, vl); + for (int k = 0; k < ksize; k++) { + vfloat32m2_t vw = __riscv_vle32_v_f32m2(wblk + k*C0 + c, vl); + vfloat32m2_t vi = __riscv_vle32_v_f32m2(w + ofstab[k] + c, vl); + acc = __riscv_vfmacc_vv_f32m2(acc, vi, vw, vl); + } + acc = finish(acc, c, vl, rp ? rp + c : nullptr); + __riscv_vse32_v_f32m2(op + c, acc, vl); + c += (int)vl; + } + } + do_border(inner_x1, W); + } + } + } + return CV_HAL_ERROR_OK; +} + +#endif // CV_HAL_RVV_1P0_ENABLED + +}}} // cv::rvv_hal::dnn