Merge pull request #28585 from vpisarev:dnn_block_layout_v5

Block layout-based convolution in DNN #28585

merge together with https://github.com/opencv/opencv_extra/pull/1321

Some core parts of the new engine in DNN module have been revised substantially:

1. all tests seem to pass, except for `Test_Graph_Simplifier.ResizeSubgraph`, which has been disabled because it does not take the newly added `TransformLayoutLayer` into account. The test should be reworked perhaps.
1. convolution and related operations (maxpool/avgpool) now use so-called block layout (`DATA_LAYOUT_BLOCK`), where `NxCxHxW` tensors are represented  as `NxC1xHxWxC0`, where `C1=(C + C0-1)/C0` and `C0` is a power-of-two (usually 4, 8, 16 or 32).
1. graph is now pre-processed and `TransformLayoutLayer` is inserted to convert data from NCHW or NHWC layout to the block layout or vice versa. The transformations are done in a lazy way only when they are really needed. For example, in the whole Resnet only 2 transformations are performed.
1. transformer-based models and other models that do not use convolutions will run as usual, without going to block layout.
1. there is yet another graph preprocessing stage added that embeds constant weights/scale and bias into convolution and batch norm layers.
1. 'batchnorm', 'activation' and 'adding a residual' are now fused with convolution, just like in the old engine. That brings some noticeable acceleration.
1. optimized convolution kernels have been added.
     * depthwise convolution, as well as maxpool and avgpool support C0=4, 8, 16 etc. _as long as_  C0 is divisible by the number of fp32 lanes in a SIMD register of the target platform (e.g. on ARM with NEON there must be `C0 % 4 == 0`, on x64 with AVX2 `C0 % 8 == 0`).
     * non-depthwise convolution only supports C0=8 for now. C0=8 seems to be a sweetspot for ARM with NEON, x64 with AVX2 or RISC-V with RVV (with 128- or 256-bit registers). For some platforms with dedicated matrix accelerators C0=16 or even C0=32 might be more efficient, but we could add the respective kernels later.
     * only fp32 kernels have been added. fp16/bf16 kernels might be added a little later.

### Pull Request Readiness Checklist

See details at https://github.com/opencv/opencv/wiki/How_to_contribute#making-a-good-pull-request

- [x] I agree to contribute to the project under Apache 2 License.
- [x] To the best of my knowledge, the proposed patch is not based on a code under GPL or another license that is incompatible with OpenCV
- [x] The PR is proposed to the proper branch
- [ ] There is a reference to the original bug report and related work
- [x] There is accuracy test, performance test and test data in opencv_extra repository, if applicable
- [x] Patch to opencv_extra has the same branch name.
- [ ] The feature is well documented and sample code can be built with the project CMake

modules/core/include/opencv2/core/mat.hpp

@@ -152,14 +152,15 @@ struct CV_EXPORTS_W_SIMPLE MatShape
     int& operator [](size_t idx);
     Size operator()() const; // for compatibility with MatSize
 
+    CV_WRAP int channels() const; // returns the number of channels
+
     CV_WRAP bool hasSymbols() const; // negative elements in the shape may denote 'symbols' instead of actual values.
 
     // compute shape of the result with possible broadcasting
     CV_WRAP MatShape expand(const MatShape& another) const;
 
-    // convert shape to/from block layout
-    CV_WRAP MatShape toBlock(int C0) const;
-    CV_WRAP MatShape fromBlock(DataLayout newLayout) const;
+    // convert shape between layouts
+    CV_WRAP MatShape toLayout(DataLayout newLayout, int C0=0) const;
 
     size_t total() const; // returns the total number of elements in the tensor (including padding elements, i.e. the method ignores 'C' in the case of block layout). Returns 1 for scalar tensors. Returns 0 for empty shapes.
 

modules/core/src/matrix.cpp

@@ -13,7 +13,7 @@ std::string layoutToString(DataLayout layout)
         layout == DATA_LAYOUT_ND ? "ND" :
         layout == DATA_LAYOUT_NCHW ? "NCHW" :
         layout == DATA_LAYOUT_NHWC ? "NHWC" :
-        layout == DATA_LAYOUT_BLOCK ? "NC1HWC0" :
+        layout == DATA_LAYOUT_BLOCK ? "BLOCK" :
         layout == DATA_LAYOUT_NCDHW ? "NCDHW" :
         layout == DATA_LAYOUT_NDHWC ? "NDHWC" :
         layout == DATA_LAYOUT_PLANAR ? "PLANAR" :
@@ -335,39 +335,63 @@ bool MatShape::hasSymbols() const
     return false;
 }
 
-MatShape MatShape::toBlock(int C0) const
+int MatShape::channels() const
 {
-    CV_Assert(dims >= 3);
-    // C0 should be > 1 and be a power-of-2: 2, 4, 8, ...
-    CV_Assert(C0 > 1 && (C0 & (C0-1)) == 0);
-    CV_Assert(layout == DATA_LAYOUT_NCHW || layout == DATA_LAYOUT_NHWC);
-    int c_idx = layout == DATA_LAYOUT_NCHW ? 1 : dims-1;
-
-    MatShape newsize = *this;
-    newsize.layout = DATA_LAYOUT_BLOCK;
-    newsize.C = p[c_idx];
-    newsize.p[newsize.dims++] = C0;
-    newsize.p[c_idx] = (p[c_idx] + C0 - 1)/C0;
-
-    return newsize;
+    CV_Assert(layout == DATA_LAYOUT_BLOCK || layout == DATA_LAYOUT_NCHW || layout == DATA_LAYOUT_NHWC);
+    return layout == DATA_LAYOUT_BLOCK ? C : p[layout == DATA_LAYOUT_NCHW ? 1 : dims-1];
 }
 
-MatShape MatShape::fromBlock(DataLayout newLayout) const
+MatShape MatShape::toLayout(DataLayout newLayout, int C0) const
 {
-    CV_Assert(dims >= 4);
-    CV_Assert(layout == DATA_LAYOUT_BLOCK);
-    // C0 should be > 1 and be a power-of-2: 2, 4, 8, ...
-    int C0 = p[dims-1];
-    CV_Assert(C0 > 1 && (C0 & (C0-1)) == 0);
-    CV_Assert(p[1] == (C + C0-1)/C0);
-    CV_Assert(newLayout == DATA_LAYOUT_NCHW || newLayout == DATA_LAYOUT_NHWC);
-    int c_idx = newLayout == DATA_LAYOUT_NCHW ? 1 : dims-2;
+    CV_Assert(layout == DATA_LAYOUT_BLOCK || layout == DATA_LAYOUT_NCHW || layout == DATA_LAYOUT_NHWC);
+    CV_Assert(newLayout == DATA_LAYOUT_BLOCK || newLayout == DATA_LAYOUT_NCHW || newLayout == DATA_LAYOUT_NHWC);
 
     MatShape newsize = *this;
     newsize.layout = newLayout;
-    newsize.C = 0;
-    newsize.p[c_idx] = C;
-    newsize.dims--;
+
+    if (newLayout == DATA_LAYOUT_BLOCK) {
+        // any => BLOCK
+        CV_Assert_N(C0 > 1, (C0 & (C0-1)) == 0);
+        int Corig = channels();
+        newsize.C = Corig;
+
+        if (layout == DATA_LAYOUT_NHWC) {
+            for (int i = 2; i < dims; i++) {
+                newsize.p[i] = p[i-1];
+            }
+        }
+
+        newsize.dims += layout != DATA_LAYOUT_BLOCK;
+        newsize.p[1] = (Corig + C0 - 1)/C0;
+        newsize.p[newsize.dims-1] = C0;
+    } else if (layout == DATA_LAYOUT_BLOCK) {
+        // BLOCK => any (except for BLOCK, which is handled above)
+        newsize.C = 0;
+        if (newLayout == DATA_LAYOUT_NHWC) {
+            for (int i = 2; i < dims; i++) {
+                newsize.p[i-1] = p[i];
+            }
+        }
+        newsize.p[newLayout == DATA_LAYOUT_NCHW ? 1 : dims-2] = C;
+        newsize.dims--;
+    } else {
+        CV_Assert_N(C0 <= 1);
+
+        if (newLayout == layout)
+            return newsize;
+
+        // NHWC => NCHW
+        if (newLayout == DATA_LAYOUT_NCHW) {
+            for (int i = 2; i < dims; i++)
+                newsize.p[i] = p[i-1];
+            newsize.p[1] = p[dims-1];
+        } else {
+            // NCHW => NHWC
+            for (int i = 2; i < dims; i++)
+                newsize.p[i-1] = p[i];
+            newsize.p[dims-1] = p[1];
+        }
+    }
 
     return newsize;
 }
@@ -403,7 +427,7 @@ MatShape MatShape::expand(const MatShape& another) const
         int sz2 = i2 < 0 ? 1 : another.p[i2];
         CV_Assert(sz1 == sz2 || sz1 == 1 || sz2 == 1);
         // [TODO] handle symbolic shapes
-        result.p[i] = std::max(sz1, sz2);
+        result.p[i] = sz1 != 1 ? sz1 : sz2;
     }
     return result;
 }
@@ -943,6 +967,8 @@ void Mat::fit(const std::vector<int>& _shape, int _type)
 void Mat::fit(const MatShape& _shape, int _type)
 {
     fit(_shape.dims, _shape.p, _type);
+    size.layout = _shape.layout;
+    size.C = _shape.C;
 }
 
 void Mat::fit(std::initializer_list<int> _shape, int _type)
@@ -1129,6 +1155,8 @@ void Mat::create(const MatShape& _shape, int _type)
         return;
     }
     create(_shape.dims, _shape.p, _type);
+    size.layout = _shape.layout;
+    size.C = _shape.C;
 }
 
 void Mat::create(std::initializer_list<int> _shape, int _type)

modules/core/src/matrix_wrap.cpp

@@ -709,12 +709,52 @@ bool _InputArray::empty(int i) const
 
 MatShape _InputArray::shape(int i) const
 {
-    int sizes[CV_MAX_DIM];
-    int dims = sizend(sizes, i);
+    _InputArray::KindFlag k = kind();
+    MatShape shape;
 
-    if (dims == 0 && empty(i))
-        return MatShape();
-    return MatShape(dims, sizes);
+    if( k == NONE )
+        ;
+    else if( k == MAT )
+    {
+        CV_Assert( i < 0 );
+        const Mat& m = *(const Mat*)obj;
+        shape = m.size;
+    }
+    else if( k == UMAT )
+    {
+        CV_Assert( i < 0 );
+        const UMat& m = *(const UMat*)obj;
+        shape = m.size;
+    }
+    else if( k == STD_VECTOR_MAT && i >= 0 )
+    {
+        const std::vector<Mat>& vv = *(const std::vector<Mat>*)obj;
+        CV_Assert( (size_t)i < vv.size() );
+        const Mat& m = vv[i];
+        shape = m.size;
+    }
+    else if( k == STD_ARRAY_MAT && i >= 0 )
+    {
+        const Mat* vv = (const Mat*)obj;
+        CV_Assert( i < sz.height );
+        const Mat& m = vv[i];
+        shape = m.size;
+    }
+    else if( k == STD_VECTOR_UMAT && i >= 0 )
+    {
+        const std::vector<UMat>& vv = *(const std::vector<UMat>*)obj;
+        CV_Assert( (size_t)i < vv.size() );
+        const UMat& m = vv[i];
+        shape = m.size;
+    }
+    else
+    {
+        int sizes[CV_MAX_DIM];
+        int dims = sizend(sizes, i);
+        shape = MatShape(dims, sizes);
+    }
+
+    return shape;
 }
 
 bool _InputArray::sameSize(const _InputArray& arr) const

modules/core/src/umatrix.cpp

@@ -681,6 +681,8 @@ void UMat::create(const MatShape& _shape, int _type, UMatUsageFlags _usageFlags)
         release();
     } else {
         create(_shape.dims, _shape.p, _type, _usageFlags);
+        size.layout = _shape.layout;
+        size.C = _shape.C;
     }
 }
 
@@ -715,6 +717,8 @@ void UMat::fit(const std::vector<int>& _shape, int _type, UMatUsageFlags _usageF
 void UMat::fit(const MatShape& _shape, int _type, UMatUsageFlags _usageFlags)
 {
     fit(_shape.dims, _shape.p, _type, _usageFlags);
+    size.layout = _shape.layout;
+    size.C = _shape.C;
 }
 
 void UMat::fit(int _rows, int _cols, int _type, UMatUsageFlags _usageFlags)

modules/dnn/CMakeLists.txt

@@ -10,6 +10,8 @@ ocv_add_dispatched_file_force_all("layers/cpu_kernels/conv_block" AVX AVX2 NEON
 ocv_add_dispatched_file_force_all("layers/cpu_kernels/conv_depthwise" AVX AVX2 RVV LASX)
 ocv_add_dispatched_file("layers/cpu_kernels/conv_winograd_f63" AVX AVX2 NEON NEON_FP16)
 ocv_add_dispatched_file_force_all("layers/cpu_kernels/fast_gemm_kernels" AVX AVX2 NEON LASX)
+ocv_add_dispatched_file("layers/cpu_kernels/conv2_depthwise" AVX AVX2 NEON NEON_FP16)
+ocv_add_dispatched_file("layers/cpu_kernels/conv2_kernels" AVX AVX2 NEON NEON_FP16)
 
 ocv_add_module(dnn opencv_core opencv_imgproc WRAP python java objc js)
 

modules/dnn/include/opencv2/dnn/all_layers.hpp

@@ -357,6 +357,31 @@ CV__DNN_INLINE_NS_BEGIN
         static Ptr<BaseConvolutionLayer> create(const LayerParams& params);
     };
 
+    enum AutoPadding
+    {
+        AUTO_PAD_NONE = 0,
+        AUTO_PAD_SAME_UPPER = 1,
+        AUTO_PAD_SAME_LOWER = 2,
+        AUTO_PAD_VALID = 3
+    };
+
+    class CV_EXPORTS Conv2Layer : public Layer
+    {
+    public:
+        static Ptr<Conv2Layer> create(const LayerParams& params);
+        virtual void setWeights(InputArray weights, InputArray bias,
+                                int C0, int accuracy) = 0;
+        virtual bool fuseAddBias(InputArray bias) = 0;
+        virtual bool fuseBatchNorm(const Ptr<Layer>& bn) = 0;
+        virtual bool fuseActivation(const Ptr<Layer>& activ) = 0;
+        virtual bool fuseAddResidual(Arg residual) = 0;
+
+        std::vector<int> strides, dilations, pads;
+        int ngroups;
+        AutoPadding auto_pad;
+        bool ceil_mode;
+    };
+
     class CV_EXPORTS LRNLayer : public Layer
     {
     public:
@@ -444,6 +469,34 @@ CV__DNN_INLINE_NS_BEGIN
         static Ptr<PoolingLayerInt8> create(const LayerParams& params);
     };
 
+    class CV_EXPORTS AveragePoolLayer : public Layer
+    {
+    public:
+        std::vector<int> kernel_shape, strides, dilations, pads;
+        AutoPadding auto_pad;
+        bool ceil_mode;
+        bool count_include_pad;
+
+        static Ptr<AveragePoolLayer> create(const LayerParams& params);
+    };
+
+    class CV_EXPORTS MaxPoolLayer : public Layer
+    {
+    public:
+        std::vector<int> kernel_shape, strides, dilations, pads;
+        AutoPadding auto_pad;
+        bool ceil_mode;
+        int storage_order;
+
+        static Ptr<MaxPoolLayer> create(const LayerParams& params);
+    };
+
+    class CV_EXPORTS GlobalAveragePoolLayer : public Layer
+    {
+    public:
+        static Ptr<GlobalAveragePoolLayer> create(const LayerParams& params);
+    };
+
     class CV_EXPORTS ReduceLayer : public Layer
     {
     public:
@@ -831,12 +884,16 @@ CV__DNN_INLINE_NS_BEGIN
     class CV_EXPORTS ActivationLayer : public Layer
     {
     public:
-        virtual void forwardSlice(const float*, float*, int,
-                                  size_t, int, int) const {}
-        virtual void forwardSlice(const int*, const int*, int*, int,
-                                  size_t, int, int) const {}
-        virtual void forwardSlice(const int8_t*, const int8_t*, int8_t*, int,
-                                  size_t, int, int) const {}
+        virtual int getLayouts(const std::vector<DataLayout>& actualInputs,
+                                std::vector<DataLayout>& desiredInputs,
+                                const int requiredOutputs,
+                                std::vector<DataLayout>& outputs) const CV_OVERRIDE;
+        virtual void forwardSlice(const float* /*src*/, float* /*dst*/, int /*len*/,
+                                  size_t /*outPlaneSize*/, int /*cn0*/, int /*cn1*/) const {}
+        virtual void forwardSlice(const int* /*src*/, const int* /*lut*/, int* /*dst*/, int /*len*/,
+                                  size_t /*outPlaneSize*/, int /*cn0*/, int /*cn1*/) const {}
+        virtual void forwardSlice(const int8_t* /*src*/, const int8_t* /*lut*/, int8_t* /*dst*/, int /*len*/,
+                                  size_t /*outPlaneSize*/, int /*cn0*/, int /*cn1*/) const {}
     };
 
     class CV_EXPORTS ReLULayer : public ActivationLayer
@@ -1149,6 +1206,35 @@ CV__DNN_INLINE_NS_BEGIN
     class CV_EXPORTS NaryEltwiseLayer : public Layer
     {
     public:
+        enum class OPERATION
+        {
+            AND = 0,
+            EQUAL,
+            GREATER,
+            GREATER_EQUAL,
+            LESS,
+            LESS_EQUAL,
+            OR,
+            POW,
+            XOR,
+            BITSHIFT,
+            MAX,
+            MEAN,
+            MIN,
+            MOD,  // Integer Mod. Reminder's sign = Divisor's sign.
+            FMOD, // Floating-point Mod. Reminder's sign = Dividend's sign.
+            PROD,
+            SUB,
+            SUM,
+            ADD,
+            DIV,
+            WHERE,
+            BITWISE_AND,
+            BITWISE_OR,
+            BITWISE_XOR
+        };
+        OPERATION op;
+
         static Ptr<NaryEltwiseLayer> create(const LayerParams &params);
     };
 
@@ -1161,15 +1247,6 @@ CV__DNN_INLINE_NS_BEGIN
         static Ptr<BatchNormLayer> create(const LayerParams &params);
     };
 
-    class CV_EXPORTS BatchNorm2Layer : public Layer
-    {
-    public:
-        float epsilon;
-        bool useGlobalStats, hasWeights, hasBias;
-
-        static Ptr<BatchNorm2Layer> create(const LayerParams& params);
-    };
-
     class CV_EXPORTS BatchNormLayerInt8 : public BatchNormLayer
     {
     public:
@@ -1178,6 +1255,18 @@ CV__DNN_INLINE_NS_BEGIN
         static Ptr<BatchNormLayerInt8> create(const LayerParams &params);
     };
 
+    class CV_EXPORTS BatchNorm2Layer : public Layer
+    {
+    public:
+        float epsilon;
+        virtual bool freezeScaleBias() = 0;
+        virtual void getScaleBias(OutputArray scale, OutputArray bias) const = 0;
+        static void getScaleBias(InputArray scale, InputArray bias,
+                                 InputArray mean, InputArray variance, float eps,
+                                 OutputArray outscale, OutputArray outbias);
+        static Ptr<BatchNorm2Layer> create(const LayerParams &params);
+    };
+
     class CV_EXPORTS MaxUnpoolLayer : public Layer
     {
     public:
@@ -1468,6 +1557,14 @@ CV__DNN_INLINE_NS_BEGIN
         static Ptr<Tile2Layer> create(const LayerParams& params);
     };
 
+    class CV_EXPORTS TransformLayoutLayer : public Layer
+    {
+    public:
+        DataLayout layout;
+        int C0;
+        static Ptr<TransformLayoutLayer> create(const LayerParams& params);
+    };
+
     class CV_EXPORTS UniqueLayer : public Layer
     {
     public:

modules/dnn/include/opencv2/dnn/dnn.hpp

@@ -232,7 +232,7 @@ CV__DNN_INLINE_NS_BEGIN
         Arg();
         explicit Arg(int idx_);
         bool empty() const;
-        operator bool() const;
+        operator int() const;
         // idx > 0: the Arg is input or output argument of some operation inside inference graph
         // idx < 0: the Arg is input or output argument of a pattern
         // idx == 0: no/empty argument; used in operations where some of the inputs/outputs are optional.
@@ -460,6 +460,23 @@ CV__DNN_INLINE_NS_BEGIN
                               std::vector<MatType>&outputs,
                               std::vector<MatType>&internals) const;
 
+        // this is the method for Layer to express its attitude to the block layout
+        // or any other special form of layout. It takes
+        // layouts of the inputs and should return the desired layouts of
+        // inputs, as well as layouts of the outputs.
+        // By default, no mater what the actual inputs' layouts are,
+        // the desired inputs as well as outputs will get 'Unknown' layout values.
+        // It means that the layer can only handle non-block layout
+        // (depending on the model format, e.g. NCHW for ONNX or NHWC for TFLite)
+        // and will return tensors with non-block layout as well.
+        // Some layers could override this default behaviour:
+        // a) if they _can_ process block-layout data, like element-wise operations, or
+        // b) if they _need_ block-layout data, like convolution
+        virtual int getLayouts(const std::vector<DataLayout>& actualInputs,
+                                std::vector<DataLayout>& desiredInputs,
+                                const int requiredOutputs,
+                                std::vector<DataLayout>& outputs) const;
+
         virtual int64 getFLOPS(const std::vector<MatShape> &inputs,
                                const std::vector<MatShape> &outputs) const;
 

modules/dnn/include/opencv2/dnn/dnn.inl.hpp

@@ -436,7 +436,7 @@ inline Arg::Arg(int idx_) : idx(idx_) {}
 
 inline bool Arg::empty() const { return idx == 0; }
 
-inline Arg::operator bool() const { return idx != 0; }
+inline Arg::operator int() const { return idx; }
 
 inline bool operator == (const Arg& a, const Arg& b) { return a.idx == b.idx; }
 

modules/dnn/src/graph_block_layout.cpp

@@ -0,0 +1,231 @@
+// 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 "precomp.hpp"
+#include "net_impl.hpp"
+
+namespace cv { namespace dnn {
+CV__DNN_INLINE_NS_BEGIN
+
+using std::vector;
+using std::string;
+
+using PLayer = Ptr<Layer>;
+using PGraph = Ptr<Graph>;
+
+/* Inserts layout conversion operations (if needed) into the model graph and subgraphs.
+
+ Some of the operations
+ (let's call them 'operations of category B' or B-operations, 'B' stands for 'Block'),
+ most notably Convolution (including depthwise convolution), ConvTranspose, MaxPool and AveragePool,
+ can be computed more efficiently if data is represented in so-called block layout,
+ i.e. when 4D tensor NxCxHxW is represented in memory as 5D tensor NxC1xHxWxC0,
+ where all C channels of the original tensor are split into C1 groups of C0 channels each.
+ For each spatial location (y=y0, x=x0) each group of C0 channels is stored sequentially.
+ C1 is thus computed as following: C1 = (C + C0-1)/C0, where division is performed with truncation.
+
+ Some other operations (let's call them A-operations, 'A' stands for 'Any'),
+ most notably unary element-wise operations or special cases of binary
+ element-wise operations can be efficiently computed in any layout, including block layout.
+
+ Finally, all other operations (denoted as C-operations, 'C' for 'Casual' or 'Channels')
+ do not support block layout at all. The inputs should come in the original model format
+ (e.g. NCHW in the case of Onnx).
+
+ We want to transform the graph so that:
+ 1. B-operations always take the inputs in block layout. If not, the inputs must be converted to
+    block layout prior to B-operation. Note that only the first input should be converted
+    in the case of Convolution, convolution weights (if constant) are pre-processed separately.
+ 2. C-operations always take the inputs in non-block layout, e.g. NCHW. If some of the inputs are
+    stored in block layout, they must be converted from block layout prior to C-operation.
+ 3. the number of layout transformation operations is minimal,
+    i.e. we don't do transformations unless it's necessary.
+
+ Note that this graph transformation is applied after fusion, since inside a fused operation
+ (e.g. 'Convolution + Batch Norm + Activation + Adding Skip connection') we don't need to
+ transform layout. That is, we have to deal with less operations at this stage.
+*/
+
+struct BlockLayoutTransformer
+{
+    BlockLayoutTransformer(Net::Impl* netimpl_) : netimpl(netimpl_) {}
+
+    Net::Impl* netimpl;
+    vector<DataLayout> layouts; // layouts for each argument
+    vector<Arg> blockCache; // if an Arg needs to be converted to block layout and
+                            // if it's used by several operations,
+                            // then we reuse once transformed arg,
+                            // don't transform it several times
+    vector<Arg> nonblockCache; // another cache of non-block args
+    DataLayout defaultLayout;
+    int trlayoutIdx;
+
+    std::pair<Arg, PLayer> getProperArg(const Arg& arg, bool block, int defaultC0)
+    {
+        if (arg.empty() || (layouts[arg.idx] == DATA_LAYOUT_BLOCK) == block)
+            return {arg, PLayer()};
+        std::vector<Arg> *cache, *altCache;
+        if (block) {
+            cache = &blockCache;
+            altCache = &nonblockCache;
+        } else {
+            cache = &nonblockCache;
+            altCache = &blockCache;
+        }
+        Arg cached = cache->at(arg.idx);
+        if (!cached.empty())
+            return {cached, PLayer()};
+        size_t nargs = netimpl->args.size();
+        CV_Assert(layouts.size() == nargs &&
+                  cache->size() == nargs &&
+                  altCache->size() == nargs);
+
+        const ArgData& adata = netimpl->args.at(arg.idx);
+        const char* suffix = block ? "block" : "nonblock";
+
+        std::string newname = format("%s.%s", adata.name.c_str(), suffix);
+        int idx = 1;
+        for ( ; netimpl->haveArg(newname); idx++) {
+            newname = format("%s.%s%d", adata.name.c_str(), suffix, idx);
+        }
+
+        cached = netimpl->newArg(newname, DNN_ARG_TEMP);
+        CV_Assert(size_t(cached.idx) == nargs);
+
+        cache->at(arg.idx) = cached;
+        cache->push_back(cached);
+        altCache->push_back(arg);
+
+        LayerParams params;
+        params.name = format("trlayout.%d", trlayoutIdx++);
+        params.type = "TransformLayout";
+        params.set("layout", int(block ? DATA_LAYOUT_BLOCK : defaultLayout));
+        params.set("C0", int(defaultC0));
+        PLayer trlayer = TransformLayoutLayer::create(params);
+
+        trlayer->netimpl = netimpl;
+        trlayer->inputs = {arg};
+        trlayer->outputs = {cached};
+        layouts.push_back(block ? DATA_LAYOUT_BLOCK : defaultLayout);
+
+        return {cached, trlayer};
+    }
+
+    void transformGraph(PGraph& g)
+    {
+        const vector<PLayer>& currProg = g->prog();
+        int defaultC0 = netimpl->defaultC0;
+        vector<PLayer> newProg;
+        std::vector<Arg> newInputs;
+        std::vector<DataLayout> inputLayoutsOrig, inputLayoutsNew, outputLayouts;
+        size_t nchanges = 0;
+
+        for (const PLayer& layer: currProg) {
+            const vector<Arg>& inputs = layer->inputs;
+            const vector<Arg>& outputs = layer->outputs;
+            size_t ninputs = inputs.size(), noutputs = outputs.size();
+            std::string op_name = layer->type;
+            std::string name = layer->name;
+            vector<PGraph>* subgraphs = layer->subgraphs();
+            //std::cout << "name: " << name << ", op_name: " << op_name << ", inp0 layout: " << layoutToString(layouts[inputs[0].idx]) << "\n";
+
+            if (subgraphs) {
+                for (PGraph& subgraph: *subgraphs) {
+                    size_t nargsBefore = netimpl->args.size();
+                    CV_Assert_N(nargsBefore == blockCache.size(), nargsBefore == nonblockCache.size());
+                    transformGraph(subgraph);
+                    size_t nargsAfter = netimpl->args.size();
+                    CV_Assert_N(nargsAfter == blockCache.size(), nargsAfter == nonblockCache.size());
+                    // after transforming a subgraph we 'forget' all cached transformations,
+                    // because they need to be redone in the outer graph and/or other subgraphs.
+                    // however, we cannot remove the newly added arguments, because we need to properly
+                    // assign memory buffers for them etc.
+                    for (size_t i = nargsBefore; i < nargsAfter; i++) {
+                        Arg b = blockCache[i];
+                        Arg nb = nonblockCache[i];
+                        if (size_t(b.idx) < nargsBefore) {
+                            nonblockCache[b.idx] = Arg();
+                        }
+                        if (size_t(nb.idx) < nargsBefore) {
+                            blockCache[nb.idx] = Arg();
+                        }
+                    }
+                    nchanges += nargsAfter > nargsBefore;
+                }
+            }
+
+            inputLayoutsOrig.clear();
+            for (size_t i = 0; i < ninputs; i++) {
+                inputLayoutsOrig.push_back(layouts.at(inputs[i].idx));
+            }
+
+            layer->getLayouts(inputLayoutsOrig, inputLayoutsNew, int(noutputs), outputLayouts);
+            CV_Assert(inputLayoutsNew.size() == ninputs);
+            CV_Assert(outputLayouts.size() == noutputs);
+
+            newInputs.clear();
+            bool changedInputs = false;
+            for (size_t i = 0; i < ninputs; i++) {
+                Arg inp = inputs[i];
+                DataLayout prevLayout = inputLayoutsOrig[i];
+                DataLayout newLayout = inputLayoutsNew[i];
+                if (!inp.empty()) {
+                    if (prevLayout == DATA_LAYOUT_BLOCK) {
+                        blockCache.at(inp.idx) = inp;
+                    } else {
+                        nonblockCache.at(inp.idx) = inp;
+                    }
+                }
+                if (newLayout != prevLayout &&
+                    (newLayout == DATA_LAYOUT_BLOCK ||
+                     prevLayout == DATA_LAYOUT_BLOCK)) {
+                    auto p = getProperArg(inp, newLayout == DATA_LAYOUT_BLOCK, defaultC0);
+                    newInputs.push_back(p.first);
+                    if (p.second) {
+                        newProg.push_back(p.second);
+                    }
+                    changedInputs = true;
+                    nchanges++;
+                } else {
+                    newInputs.push_back(inp);
+                }
+            }
+
+            for (size_t i = 0; i < noutputs; i++) {
+                layouts.at(outputs[i].idx) = outputLayouts[i];
+            }
+
+            if (changedInputs) {
+                layer->inputs = newInputs;
+            }
+            newProg.push_back(layer);
+        }
+
+        if (nchanges > 0) {
+            g->setProg(newProg);
+        }
+    }
+
+    void transform()
+    {
+        size_t nargs = netimpl->args.size();
+        defaultLayout = netimpl->originalLayout;
+        trlayoutIdx = 0;
+
+        layouts.assign(nargs, DATA_LAYOUT_UNKNOWN);
+        blockCache.assign(nargs, Arg());
+        nonblockCache.assign(nargs, Arg());
+
+        transformGraph(netimpl->mainGraph);
+    }
+};
+
+void Net::Impl::useBlockLayout()
+{
+    BlockLayoutTransformer use_block_layout(this);
+    use_block_layout.transform();
+}
+
+CV__DNN_INLINE_NS_END
+}}

modules/dnn/src/graph_const_args.cpp

@@ -0,0 +1,107 @@
+// 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 "precomp.hpp"
+#include "net_impl.hpp"
+
+namespace cv { namespace dnn {
+CV__DNN_INLINE_NS_BEGIN
+
+using std::vector;
+using std::string;
+
+typedef std::pair<int, int> int_pair;
+typedef std::pair<int, Arg> int_arg_pair;
+
+struct ConstArgs
+{
+    ConstArgs(Net::Impl* netimpl_) : netimpl(netimpl_) {}
+
+    void process()
+    {
+        size_t nargs = netimpl->args.size();
+        netimpl->__tensors__.resize(nargs);
+        netimpl->useCounts(usecounts);
+        processGraph(netimpl->mainGraph);
+    }
+
+    void unuse(Arg inp)
+    {
+        CV_Assert(usecounts[inp.idx] > 0);
+        if (--usecounts[inp.idx] == 0 && netimpl->isConstArg(inp)) {
+            netimpl->__tensors__[inp.idx] = Mat(); // deallocate unused tensor
+        }
+    }
+
+    void processGraph(Ptr<Graph>& graph)
+    {
+        const std::vector<Ptr<Layer> >& prog = graph->prog();
+        size_t i, nops = prog.size();
+        std::vector<Arg> removed_args;
+        std::vector<Arg> saved_tail_inputs;
+
+        for (i = 0; i < nops; i++) {
+            const Ptr<Layer>& layer = prog[i];
+            Layer* layer_ptr = const_cast<Layer*>(layer.get());
+            std::vector<Ptr<Graph> >* subgraphs = layer->subgraphs();
+            if (subgraphs) {
+                for (Ptr<Graph>& g: *subgraphs) {
+                    processGraph(g);
+                }
+            }
+            const std::vector<Arg>& inputs = layer->inputs;
+            size_t j, ninputs = inputs.size();
+            if (ninputs == 1) {
+                continue;
+            }
+            bool tail_const = true, unuse_tail = false;
+            saved_tail_inputs.clear();
+            for (j = 1; j < ninputs; j++) {
+                Arg inp = inputs[j];
+                bool const_arg = netimpl->isConstArg(inp);
+                if (!const_arg)
+                    tail_const = false;
+                saved_tail_inputs.push_back(inp);
+            }
+
+            Conv2Layer* conv = dynamic_cast<Conv2Layer*>(layer_ptr);
+            BatchNorm2Layer* bn = dynamic_cast<BatchNorm2Layer*>(layer_ptr);
+            //ActivationLayer* activ = dynamic_cast<ActivationLayer*>(layer_ptr);
+
+            if (tail_const) {
+                if (conv) {
+                    // convolution with constant weights and bias
+                    conv->setWeights(netimpl->__tensors__[inputs[1]],
+                                     ninputs > 2 ? netimpl->__tensors__[inputs[2]] : Mat(),
+                                     netimpl->defaultC0, netimpl->accuracy);
+                    conv->inputs.resize(1);
+                    unuse_tail = true;
+                } else if (bn && bn->freezeScaleBias()) {
+                    // batch norm with constant parameters
+                    unuse_tail = true;
+                }/* else if (activ && dynamic_cast<ReLU6Layer>(activ)) {
+                    // [TODO] ...
+                    unuse_tail = true;
+                }*/
+            }
+
+            if (unuse_tail) {
+                for (Arg inp: saved_tail_inputs)
+                    unuse(inp);
+            }
+        }
+    }
+
+    Net::Impl* netimpl;
+    std::vector<int> usecounts;
+};
+
+void Net::Impl::constArgs()
+{
+    ConstArgs constargs(this);
+    constargs.process();
+}
+
+CV__DNN_INLINE_NS_END
+}}

modules/dnn/src/graph_fusion_basic.cpp

@@ -0,0 +1,182 @@
+// 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 "precomp.hpp"
+#include "net_impl.hpp"
+
+namespace cv { namespace dnn {
+CV__DNN_INLINE_NS_BEGIN
+
+using std::vector;
+using std::string;
+
+typedef std::pair<int, int> int_pair;
+typedef std::pair<int, Arg> int_arg_pair;
+
+struct ModelFusionBasic
+{
+    ModelFusionBasic(Net::Impl* netimpl_) : netimpl(netimpl_) {}
+
+    void fuse()
+    {
+        int i, niter = 10;
+        netimpl->useCounts(usecounts);
+        for (i = 0; i < niter; i++) {
+            bool fused_any = fuseGraph(netimpl->mainGraph);
+            if (!fused_any)
+                break;
+        }
+    }
+
+    template<typename _LayerType> _LayerType*
+    getLayer(std::vector<Ptr<Layer> >& newprog, int op_idx) const
+    {
+        return op_idx >= 0 ? dynamic_cast<_LayerType*>(newprog.at(op_idx).get()) : 0;
+    }
+
+    bool fuseGraph(Ptr<Graph>& graph)
+    {
+        vector<Arg> removed_args;
+        bool modified = false;
+        const std::vector<Ptr<Layer> >& prog = graph->prog();
+        size_t i, nargs = netimpl->args.size(), nops = prog.size();
+        std::vector<int> producer_of(nargs, -1);
+        std::vector<Ptr<Layer> > newprog;
+        std::vector<Arg> fused_inputs;
+
+        for (i = 0; i < nops; i++) {
+            const Ptr<Layer>& layer = prog[i];
+            Layer* layer_ptr = (Layer*)layer.get();
+            int fused_layer_idx = -1;
+            std::vector<Ptr<Graph> >* subgraphs = layer->subgraphs();
+            if (subgraphs) {
+                for (Ptr<Graph>& g: *subgraphs) {
+                    if (fuseGraph(g))
+                        modified = true;
+                }
+            }
+            const std::vector<Arg>& inputs = layer->inputs;
+            const std::vector<Arg>& outputs = layer->outputs;
+            size_t ninputs = inputs.size();
+            removed_args.clear();
+            fused_inputs.clear(); // leave it empty in the merge patterns below to re-use original fused node inputs as-is.
+
+            for(;;) {
+                BatchNorm2Layer* bn = dynamic_cast<BatchNorm2Layer*>(layer_ptr);
+                ActivationLayer* activ = dynamic_cast<ActivationLayer*>(layer_ptr);
+                NaryEltwiseLayer* elemwise = dynamic_cast<NaryEltwiseLayer*>(layer_ptr);
+
+                // merge convolution and batch norm
+                if (bn && ninputs == 1 &&
+                    usecounts.at(inputs[0].idx) == 1) {
+                    Arg bn_inp = inputs[0];
+                    int conv_layer_idx = producer_of.at(bn_inp.idx);
+                    Conv2Layer* conv = getLayer<Conv2Layer>(newprog, conv_layer_idx);
+                    if (conv) {
+                        bool ok = conv->fuseBatchNorm(layer);
+                        if (ok) {
+                            fused_layer_idx = conv_layer_idx;
+                            removed_args.push_back(bn_inp);
+                            break;
+                        }
+                    }
+                }
+
+                // merge residual 'add' into 'conv' node
+                if (elemwise && (elemwise->op == NaryEltwiseLayer::OPERATION::ADD ||
+                    elemwise->op == NaryEltwiseLayer::OPERATION::SUM) &&
+                    ninputs == 2) {
+
+                    int op0 = producer_of.at(inputs[0].idx);
+                    int op1 = producer_of.at(inputs[1].idx);
+
+                    if (op0 >= 0 && op1 >= 0) {
+                        int conv_layer_idx;
+                        Arg residual, conv_out;
+
+                        if (op0 > op1) { // choose the latter op to ensure that the other component is already computed
+                            conv_layer_idx = op0;
+                            conv_out = inputs[0];
+                            residual = inputs[1];
+                        } else {
+                            conv_layer_idx = op1;
+                            conv_out = inputs[1];
+                            residual = inputs[0];
+                        }
+
+                        Conv2Layer* conv = getLayer<Conv2Layer>(newprog, conv_layer_idx);
+                        if (conv && usecounts[conv_out.idx] == 1 &&
+                            conv->fuseAddResidual(residual)) {
+                            fused_layer_idx = conv_layer_idx;
+                            removed_args.push_back(conv_out);
+                            break;
+                        }
+                    }
+                }
+
+                // merge convolution and activation
+                if (activ && ninputs == 1 &&
+                    usecounts.at(inputs[0].idx) == 1) {
+                    Arg activ_inp = inputs[0];
+                    int conv_layer_idx = producer_of.at(activ_inp.idx);
+                    Conv2Layer* conv = getLayer<Conv2Layer>(newprog, conv_layer_idx);
+                    if (conv) {
+                        bool ok = conv->fuseActivation(layer);
+                        if (ok) {
+                            fused_layer_idx = conv_layer_idx;
+                            removed_args.push_back(activ_inp);
+                            break;
+                        }
+                    }
+                }
+                break;
+            }
+
+            if (fused_layer_idx >= 0) {
+                modified = true;
+                Layer* fused_layer = newprog[fused_layer_idx];
+                fused_layer->outputs = outputs;
+                for (Arg new_out: outputs)
+                    producer_of[new_out.idx] = fused_layer_idx;
+                for (Arg old_out: removed_args) {
+                    usecounts.at(old_out.idx) = 0;
+                    producer_of.at(old_out.idx) = -1;
+                }
+            } else {
+                for (auto out: outputs)
+                    producer_of[out.idx] = (int)newprog.size();
+                newprog.push_back(layer);
+            }
+        }
+
+        if (modified) {
+            size_t i, j = 0, newops = newprog.size();
+            for (i = 0; i < newops; i++) {
+                if (!newprog[i].empty()) {
+                    if (j < i)
+                        newprog[j] = newprog[i];
+                    j++;
+                }
+            }
+            newprog.resize(j);
+            //printf("fused some ops in graph %s. size before: %zu ops, size after: %zu ops\n",
+            //       graph->name().data(), nops, j);
+            graph->setProg(newprog);
+        }
+
+        return modified;
+    }
+
+    Net::Impl* netimpl;
+    vector<int> usecounts;
+};
+
+void Net::Impl::fuseBasic()
+{
+    ModelFusionBasic basicFusion(this);
+    basicFusion.fuse();
+}
+
+CV__DNN_INLINE_NS_END
+}}

modules/dnn/src/init.cpp

@@ -132,8 +132,11 @@ void initializeLayerFactory()
     CV_DNN_REGISTER_LAYER_CLASS(AffineGrid,     AffineGridLayer);
 
     CV_DNN_REGISTER_LAYER_CLASS(Convolution,    ConvolutionLayer);
+    CV_DNN_REGISTER_LAYER_CLASS(Conv2,          Conv2Layer);
     CV_DNN_REGISTER_LAYER_CLASS(Deconvolution,  DeconvolutionLayer);
     CV_DNN_REGISTER_LAYER_CLASS(Pooling,        PoolingLayer);
+    CV_DNN_REGISTER_LAYER_CLASS(MaxPool,        MaxPoolLayer);
+    CV_DNN_REGISTER_LAYER_CLASS(AveragePool,    AveragePoolLayer);
     CV_DNN_REGISTER_LAYER_CLASS(ROIPooling,     PoolingLayer);
     CV_DNN_REGISTER_LAYER_CLASS(PSROIPooling,   PoolingLayer);
     CV_DNN_REGISTER_LAYER_CLASS(Reduce,         ReduceLayer);

modules/dnn/src/layer.cpp

@@ -278,6 +278,16 @@ void Layer::getTypes(const std::vector<MatType>&inputs,
     internals.assign(requiredInternals, inputs[0]);
 }
 
+int Layer::getLayouts(const std::vector<DataLayout>& actualInputs,
+                       std::vector<DataLayout>& desiredInputs,
+                       const int requiredOutputs,
+                       std::vector<DataLayout>& outputs) const
+{
+    desiredInputs.assign(actualInputs.size(), DATA_LAYOUT_UNKNOWN);
+    outputs.assign(requiredOutputs, DATA_LAYOUT_UNKNOWN);
+    return 0;
+}
+
 int64 Layer::getFLOPS(const std::vector<MatShape>&,
                       const std::vector<MatShape>&) const
 {

modules/dnn/src/layers/avgpool_layer.cpp

@@ -0,0 +1,575 @@
+// 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 "../precomp.hpp"
+#include "layers_common.hpp"
+#include "../net_impl.hpp"
+#include "conv2_common.hpp"
+#include "opencv2/core/hal/intrin.hpp"
+
+namespace cv
+{
+namespace dnn
+{
+
+static void avgPool32f(const void* inp_, void* out_,
+                       const ConvState& cs, bool count_include_pad_)
+{
+    int NC1 = cs.inpshape[0]*cs.inpshape[1];
+
+    CV_Assert(cs.inpshape.layout == DATA_LAYOUT_BLOCK);
+    CV_Assert(cs.outshape.layout == DATA_LAYOUT_BLOCK);
+    CV_Assert(cs.inpshape.dims == cs.outshape.dims);
+
+    parallel_for_(Range(0, NC1), [&](const Range& r) {
+        constexpr int MAX_POOL_DIMS = ConvState::MAX_CONV_DIMS;
+
+        CV_Assert(cs.nspatialdims <= MAX_POOL_DIMS && MAX_POOL_DIMS == 3);
+
+        bool count_include_pad = count_include_pad_;
+        int sdims = cs.nspatialdims;
+        int nc0 = r.start, nc1 = r.end;
+        int C0 = cs.inpshape.back();
+        int Di = sdims > 2 ? cs.inpshape[sdims - 1] : 1;
+        int Hi = sdims > 1 ? cs.inpshape[sdims] : 1;
+        int Wi = cs.inpshape[sdims + 1];
+        int D = sdims > 2 ? cs.outshape[sdims - 1] : 1;
+        int H = sdims > 1 ? cs.outshape[sdims] : 1;
+        int W = cs.outshape[sdims + 1];
+        int iplanesize = Di*Hi*Wi*C0;
+        int planesize = D*H*W*C0;
+        int SZ = cs.strides[0], SY = cs.strides[1], SX = cs.strides[2];
+        int padZ0 = cs.pads[0], padY0 = cs.pads[1], padX0 = cs.pads[2];
+        int inner_z0 = cs.inner[0], inner_z1 = cs.inner[MAX_POOL_DIMS];
+        int inner_y0 = cs.inner[1], inner_y1 = cs.inner[MAX_POOL_DIMS + 1];
+        int inner_x0 = cs.inner[2], inner_x1 = cs.inner[MAX_POOL_DIMS + 2];
+        int ksize = (int)cs.ofstab.size();
+        const int* zyxtab = cs.coordtab.data();
+        const int* ofstab = cs.ofstab.data();
+
+        const float* inp = (const float*)inp_ + nc0*iplanesize;
+        float* out = (float*)out_ + nc0*planesize;
+        float iksize = 1.f/ksize;
+
+#if CV_SIMD || CV_SIMD_SCALABLE
+        int nlanes = VTraits<v_float32>::vlanes();
+        CV_Assert(C0 == nlanes || C0 == nlanes*2 || C0 % (nlanes*4) == 0);
+        v_float32 z = vx_setzero_f32();
+        v_float32 vscale0 = vx_setall_f32(iksize);
+#endif
+
+        for (int nc = nc0; nc < nc1; nc++, inp += iplanesize) {
+            for (int z0 = 0; z0 < D; z0++) {
+                int zi_ = z0*SZ - padZ0;
+                for (int y0 = 0; y0 < H; y0++, out += W*C0) {
+                    int x0 = 0;
+                    int x1 = z0 >= inner_z0 && z0 < inner_z1 &&
+                        y0 >= inner_y0 && y0 < inner_y1 ? inner_x0 : W;
+                    int yi_ = y0*SY - padY0;
+
+                #if !(CV_SIMD || CV_SIMD_SCALABLE)
+                    memset(out, 0, W*C0*sizeof(out[0]));
+                #endif
+
+                    for(;;) {
+                    #if CV_SIMD || CV_SIMD_SCALABLE
+                        if (nlanes == C0) {
+                            for (; x0 < x1; x0++) {
+                                int xi_ = x0*SX - padX0;
+                                v_float32 s0 = z;
+                                int nitems = 0;
+                                for (int k = 0; k < ksize; k++) {
+                                    int zi = zi_ + zyxtab[k*MAX_POOL_DIMS];
+                                    int yi = yi_ + zyxtab[k*MAX_POOL_DIMS+1];
+                                    int xi = xi_ + zyxtab[k*MAX_POOL_DIMS+2];
+                                    v_float32 v0;
+                                    if ((unsigned)zi >= (unsigned)Di ||
+                                        (unsigned)yi >= (unsigned)Hi ||
+                                        (unsigned)xi >= (unsigned)Wi)
+                                        continue;
+                                    v0 = vx_load(inp + ((zi*Hi + yi)*Wi + xi)*C0);
+                                    s0 = v_add(s0, v0);
+                                    nitems++;
+                                }
+                                s0 = v_mul(s0, count_include_pad ? vscale0 : vx_setall_f32(1.f/nitems));
+                                vx_store(out + x0*C0, s0);
+                            }
+                        } else {
+                            for (; x0 < x1; x0++) {
+                                int xi_ = x0*SX - padX0;
+                                for (int c = 0; c < C0; c += nlanes*2) {
+                                    v_float32 s0 = z, s1 = z;
+                                    int nitems = 0;
+                                    for (int k = 0; k < ksize; k++) {
+                                        int zi = zi_ + zyxtab[k*MAX_POOL_DIMS];
+                                        int yi = yi_ + zyxtab[k*MAX_POOL_DIMS+1];
+                                        int xi = xi_ + zyxtab[k*MAX_POOL_DIMS+2];
+                                        v_float32 v0, v1;
+                                        if ((unsigned)zi >= (unsigned)Di ||
+                                            (unsigned)yi >= (unsigned)Hi ||
+                                            (unsigned)xi >= (unsigned)Wi)
+                                            continue;
+                                        int ofs_k = ((zi*Hi + yi)*Wi + xi)*C0 + c;
+                                        v0 = vx_load(inp + ofs_k);
+                                        v1 = vx_load(inp + ofs_k + nlanes);
+                                        s0 = v_add(s0, v0);
+                                        s1 = v_add(s1, v1);
+                                        nitems++;
+                                    }
+                                    v_float32 vscale = count_include_pad ? vscale0 : vx_setall_f32(1.f/nitems);
+                                    s0 = v_mul(s0, vscale);
+                                    s1 = v_mul(s1, vscale);
+                                    vx_store(out + x0*C0 + c, s0);
+                                    vx_store(out + x0*C0 + c + nlanes, s1);
+                                }
+                            }
+                        }
+                    #else
+                        for (; x0 < x1; x0++) {
+                            int xi_ = x0*SX - padX0;
+                            int nitems = 0;
+                            for (int k = 0; k < ksize; k++) {
+                                int zi = zi_ + zyxtab[k*MAX_POOL_DIMS];
+                                int yi = yi_ + zyxtab[k*MAX_POOL_DIMS+1];
+                                int xi = xi_ + zyxtab[k*MAX_POOL_DIMS+2];
+                                if ((unsigned)zi >= (unsigned)Di ||
+                                    (unsigned)yi >= (unsigned)Hi ||
+                                    (unsigned)xi >= (unsigned)Wi)
+                                    continue;
+                                const float* inptr = inp + ((zi*Hi + yi)*Wi + xi)*C0;
+                                for (int c = 0; c < C0; c++)
+                                    out[x0*C0 + c] += inptr[c];
+                                nitems++;
+                            }
+                            float scale = count_include_pad ? iksize : 1.f/nitems;
+                            for (int c = 0; c < C0; c++)
+                                out[x0*C0 + c] *= scale;
+                        }
+                    #endif
+
+                        if (x0 == W)
+                            break;
+                        x1 = inner_x1;
+
+                    #if CV_SIMD || CV_SIMD_SCALABLE
+                        if (nlanes == C0) {
+                            for (; x0 < x1; x0++) {
+                                int xi_ = x0*SX - padX0;
+                                const float* inp_xi = inp + ((Hi*zi_ + yi_)*Wi + xi_)*C0;
+
+                                v_float32 s0 = vx_load(inp_xi + ofstab[0]);
+                                for (int k = 1; k < ksize; k++)
+                                    s0 = v_add(s0, vx_load(inp_xi + ofstab[k]));
+                                vx_store(out + x0*C0, v_mul(s0, vscale0));
+                            }
+                        } else if (nlanes*2 == C0) {
+                            for (; x0 < x1; x0++) {
+                                int xi_ = x0*SX - padX0;
+                                const float* inp_xi = inp + ((Hi*zi_ + yi_)*Wi + xi_)*C0;
+
+                                int ofs_k = ofstab[0];
+                                v_float32 s0 = vx_load(inp_xi + ofs_k);
+                                v_float32 s1 = vx_load(inp_xi + ofs_k + nlanes);
+                                for (int k = 1; k < ksize; k++) {
+                                    ofs_k = ofstab[k];
+                                    s0 = v_add(s0, vx_load(inp_xi + ofs_k));
+                                    s1 = v_add(s1, vx_load(inp_xi + ofs_k + nlanes));
+                                }
+                                s0 = v_mul(s0, vscale0);
+                                s1 = v_mul(s1, vscale0);
+                                vx_store(out + x0*C0, s0);
+                                vx_store(out + x0*C0 + nlanes, s1);
+                            }
+                        } else {
+                            for (; x0 < x1; x0++) {
+                                int xi_ = x0*SX - padX0;
+                                for (int c = 0; c < C0; c += nlanes*4) {
+                                    const float* inp_xi = inp + ((Hi*zi_ + yi_)*Wi + xi_)*C0 + c;
+
+                                    int ofs_k = ofstab[0];
+                                    v_float32 s0 = vx_load(inp_xi + ofs_k);
+                                    v_float32 s1 = vx_load(inp_xi + ofs_k + nlanes);
+                                    v_float32 s2 = vx_load(inp_xi + ofs_k + nlanes*2);
+                                    v_float32 s3 = vx_load(inp_xi + ofs_k + nlanes*3);
+                                    for (int k = 1; k < ksize; k++) {
+                                        ofs_k = ofstab[k];
+                                        s0 = v_add(s0, vx_load(inp_xi + ofs_k));
+                                        s1 = v_add(s1, vx_load(inp_xi + ofs_k + nlanes));
+                                        s2 = v_add(s2, vx_load(inp_xi + ofs_k + nlanes*2));
+                                        s3 = v_add(s3, vx_load(inp_xi + ofs_k + nlanes*3));
+                                    }
+                                    s0 = v_mul(s0, vscale0);
+                                    s1 = v_mul(s1, vscale0);
+                                    s2 = v_mul(s2, vscale0);
+                                    s3 = v_mul(s3, vscale0);
+                                    vx_store(out + x0*C0 + c, s0);
+                                    vx_store(out + x0*C0 + c + nlanes, s1);
+                                    vx_store(out + x0*C0 + c + nlanes*2, s2);
+                                    vx_store(out + x0*C0 + c + nlanes*3, s3);
+                                }
+                            }
+                        }
+                    #else
+                        for (; x0 < x1; x0++) {
+                            int xi_ = x0*SX - padX0;
+                            const float* inp_xi = inp + ((Hi*zi_ + yi_)*Wi + xi_)*C0;
+                            for (int k = 0; k < ksize; k++) {
+                                const float* inptr = inp_xi + ofstab[k];
+                                for (int c = 0; c < C0; c++)
+                                    out[x0*C0 + c] += inptr[c];
+                            }
+                            for (int c = 0; c < C0; c++)
+                                out[x0*C0 + c] *= iksize;
+                        }
+                    #endif
+                        x1 = W;
+                    }
+                }
+            }
+        }
+    });
+}
+
+// temporarily exclude fp16/bf16 versions,
+// since convolution and other layers don't support those types yet
+#if 0
+template<typename _Tp>
+static void avgPool16xf(const _Tp* inp_, _Tp* out_,
+                        const ConvState& cs, bool count_include_pad_)
+{
+    constexpr int MAX_POOL_DIMS = ConvState::MAX_CONV_DIMS;
+    int C0_ = cs.inpshape.back();
+    int NC = cs.inpshape[0]*cs.inpshape[1];
+    int nlanes_ = VTraits<v_float32>::vlanes();
+
+    CV_Assert(C0_ == nlanes_ || C0_ == nlanes_*2 || C0_ % (nlanes_*4) == 0);
+    CV_Assert(cs.nspatialdims <= MAX_POOL_DIMS && MAX_POOL_DIMS == 3);
+    CV_Assert(cs.inpshape.layout == DATA_LAYOUT_BLOCK);
+    CV_Assert(cs.outshape.layout == DATA_LAYOUT_BLOCK);
+    CV_Assert(cs.inpshape.dims == cs.outshape.dims);
+
+    parallel_for_(Range(0, NC), [&](const Range& r) {
+        bool count_include_pad = count_include_pad_;
+        int sdims = cs.nspatialdims;
+        int nc0 = r.start, nc1 = r.end;
+        int nlanes = nlanes_, C0 = cs.inpshape.back();
+        int Di = sdims > 2 ? cs.inpshape[sdims - 1] : 1;
+        int Hi = sdims > 1 ? cs.inpshape[sdims] : 1;
+        int Wi = cs.inpshape[sdims + 1];
+        int D = sdims > 2 ? cs.outshape[sdims - 1] : 1;
+        int H = sdims > 1 ? cs.outshape[sdims] : 1;
+        int W = cs.outshape[sdims + 1];
+        int iplanesize = Di*Hi*Wi*C0;
+        int planesize = D*H*W*C0;
+        int SZ = cs.strides[0], SY = cs.strides[1], SX = cs.strides[2];
+        int padZ0 = cs.pads[0], padY0 = cs.pads[1], padX0 = cs.pads[2];
+        int inner_z0 = cs.inner[0], inner_z1 = cs.inner[MAX_POOL_DIMS];
+        int inner_y0 = cs.inner[1], inner_y1 = cs.inner[MAX_POOL_DIMS + 1];
+        int inner_x0 = cs.inner[2], inner_x1 = cs.inner[MAX_POOL_DIMS + 2];
+        int ksize = (int)cs.ofstab.size();
+        const int* zyxtab = cs.coordtab.data();
+        const int* ofstab = cs.ofstab.data();
+
+        const _Tp* inp = (const _Tp*)inp_ + nc0*iplanesize;
+        _Tp* out = (_Tp*)out_ + nc0*planesize;
+        v_float32 z = vx_setzero_f32();
+        v_float32 vscale0 = vx_setall_f32(1.f/ksize);
+
+        for (int nc = nc0; nc < nc1; nc++, inp += iplanesize) {
+            for (int z0 = 0; z0 < D; z0++) {
+                int zi_ = z0*SZ - padZ0;
+                for (int y0 = 0; y0 < H; y0++, out += W*C0) {
+                    int x0 = 0;
+                    int x1 = z0 >= inner_z0 && z0 < inner_z1 &&
+                        y0 >= inner_y0 && y0 < inner_y1 ? inner_x0 : W;
+                    int yi_ = y0*SY - padY0;
+                    for(;;) {
+                        if (nlanes == C0) {
+                            for (; x0 < x1; x0++) {
+                                int xi_ = x0*SX - padX0;
+                                v_float32 s0 = z;
+                                int nitems = 0;
+                                for (int k = 0; k < ksize; k++) {
+                                    int zi = zi_ + zyxtab[k*MAX_POOL_DIMS];
+                                    int yi = yi_ + zyxtab[k*MAX_POOL_DIMS+1];
+                                    int xi = xi_ + zyxtab[k*MAX_POOL_DIMS+2];
+                                    v_float32 v0;
+                                    if ((unsigned)zi >= (unsigned)Di ||
+                                        (unsigned)yi >= (unsigned)Hi ||
+                                        (unsigned)xi >= (unsigned)Wi)
+                                        continue;
+                                    v0 = vx_load_expand(inp + ((zi*Hi + yi)*Wi + xi)*C0);
+                                    s0 = v_add(s0, v0);
+                                    nitems++;
+                                }
+                                s0 = v_mul(s0, count_include_pad ? vscale0 : vx_setall_f32(1.f/nitems));
+                                v_pack_store(out + x0*C0, s0);
+                            }
+                        } else {
+                            for (; x0 < x1; x0++) {
+                                int xi_ = x0*SX - padX0;
+                                for (int c = 0; c < C0; c += nlanes*2) {
+                                    v_float32 s0 = z, s1 = z;
+                                    int nitems = 0;
+                                    for (int k = 0; k < ksize; k++) {
+                                        int zi = zi_ + zyxtab[k*MAX_POOL_DIMS];
+                                        int yi = yi_ + zyxtab[k*MAX_POOL_DIMS+1];
+                                        int xi = xi_ + zyxtab[k*MAX_POOL_DIMS+2];
+                                        v_float32 v0, v1;
+                                        if ((unsigned)zi >= (unsigned)Di ||
+                                            (unsigned)yi >= (unsigned)Hi ||
+                                            (unsigned)xi >= (unsigned)Wi)
+                                            continue;
+                                        int ofs_k = ((zi*Hi + yi)*Wi + xi)*C0 + c;
+                                        v0 = vx_load_expand(inp + ofs_k);
+                                        v1 = vx_load_expand(inp + ofs_k + nlanes);
+                                        s0 = v_add(s0, v0);
+                                        s1 = v_add(s1, v1);
+                                    }
+                                    v_float32 vscale = count_include_pad ? vscale0 : vx_setall_f32(1.f/nitems);
+                                    s0 = v_mul(s0, vscale);
+                                    s1 = v_mul(s1, vscale);
+                                    v_pack_store(out + x0*C0 + c, s0);
+                                    v_pack_store(out + x0*C0 + c + nlanes, s1);
+                                }
+                            }
+                        }
+                        if (x0 == W)
+                            break;
+                        x1 = inner_x1;
+                        if (nlanes == C0) {
+                            for (; x0 < x1; x0++) {
+                                int xi_ = x0*SX - padX0;
+                                const _Tp* inp_xi = inp + ((Hi*zi_ + yi_)*Wi + xi_)*C0;
+
+                                v_float32 s0 = vx_load_expand(inp_xi + ofstab[0]);
+                                for (int k = 1; k < ksize; k++)
+                                    s0 = v_add(s0, vx_load_expand(inp_xi + ofstab[k]));
+                                v_pack_store(out + x0*C0, v_mul(s0, vscale0));
+                            }
+                        } else if (nlanes*2 == C0) {
+                            for (; x0 < x1; x0++) {
+                                int xi_ = x0*SX - padX0;
+                                const _Tp* inp_xi = inp + ((Hi*zi_ + yi_)*Wi + xi_)*C0;
+
+                                int ofs_k = ofstab[0];
+                                v_float32 s0 = vx_load_expand(inp_xi + ofs_k);
+                                v_float32 s1 = vx_load_expand(inp_xi + ofs_k + nlanes);
+                                for (int k = 1; k < ksize; k++) {
+                                    ofs_k = ofstab[k];
+                                    s0 = v_add(s0, vx_load_expand(inp_xi + ofs_k));
+                                    s1 = v_add(s1, vx_load_expand(inp_xi + ofs_k + nlanes));
+                                }
+                                s0 = v_mul(s0, vscale0);
+                                s1 = v_mul(s1, vscale0);
+                                v_pack_store(out + x0*C0, s0);
+                                v_pack_store(out + x0*C0 + nlanes, s1);
+                            }
+                        } else {
+                            for (; x0 < x1; x0++) {
+                                int xi_ = x0*SX - padX0;
+                                for (int c = 0; c < C0; c += nlanes*4) {
+                                    const _Tp* inp_xi = inp + ((Hi*zi_ + yi_)*Wi + xi_)*C0;
+
+                                    int ofs_k = ofstab[0];
+                                    v_float32 s0 = vx_load_expand(inp_xi + ofs_k);
+                                    v_float32 s1 = vx_load_expand(inp_xi + ofs_k + nlanes);
+                                    v_float32 s2 = vx_load_expand(inp_xi + ofs_k + nlanes*2);
+                                    v_float32 s3 = vx_load_expand(inp_xi + ofs_k + nlanes*3);
+                                    for (int k = 1; k < ksize; k++) {
+                                        ofs_k = ofstab[k];
+                                        s0 = v_add(s0, vx_load_expand(inp_xi + ofs_k));
+                                        s1 = v_add(s1, vx_load_expand(inp_xi + ofs_k + nlanes));
+                                        s2 = v_add(s2, vx_load_expand(inp_xi + ofs_k + nlanes*2));
+                                        s3 = v_add(s3, vx_load_expand(inp_xi + ofs_k + nlanes*3));
+                                    }
+                                    s0 = v_mul(s0, vscale0);
+                                    s1 = v_mul(s1, vscale0);
+                                    s2 = v_mul(s2, vscale0);
+                                    s3 = v_mul(s3, vscale0);
+                                    v_pack_store(out + x0*C0 + c, s0);
+                                    v_pack_store(out + x0*C0 + c + nlanes, s1);
+                                    v_pack_store(out + x0*C0 + c + nlanes*2, s2);
+                                    v_pack_store(out + x0*C0 + c + nlanes*3, s3);
+                                }
+                            }
+                        }
+                        x1 = W;
+                    }
+                }
+            }
+        }
+    });
+}
+
+static void avgPool16f(const void* inp_, void* out_,
+                       const ConvState& cs, bool countIncludePadding)
+{
+    avgPool16xf((const hfloat*)inp_, (hfloat*)out_, cs, countIncludePadding);
+}
+
+static void avgPool16bf(const void* inp_, void* out_,
+                        const ConvState& cs, bool countIncludePadding)
+{
+    avgPool16xf((const bfloat*)inp_, (bfloat*)out_, cs, countIncludePadding);
+}
+#endif
+
+typedef void (*AvgPoolFunc)(const void* inp, void* out,
+                            const ConvState& cs, bool countIncludePadding);
+
+class AveragePoolLayerImpl : public AveragePoolLayer
+{
+public:
+    AveragePoolLayerImpl(const LayerParams& params)
+    {
+        setParamsFrom(params);
+        auto_pad = getAutoPadding(params);
+        kernel_shape = params.getVector<int>("kernel_size");
+        strides = params.getVector<int>("stride");
+        dilations = params.getVector<int>("dilation");
+        pads = params.getVector<int>("pad");
+        ceil_mode = params.get<bool>("ceil_mode", false);
+        count_include_pad = params.get<bool>("count_include_pad", false);
+    }
+
+    virtual std::ostream& dumpAttrs(std::ostream& strm, int indent) const CV_OVERRIDE
+    {
+        prindent(strm, indent);
+        strm << "kernel_size: [";
+        for (size_t k = 0; k < kernel_shape.size(); k++)
+            strm << (k > 0 ? ", " : "") << kernel_shape[k];
+        strm << "],\n";
+
+        prindent(strm, indent);
+        strm << "dilation: [";
+        for (size_t k = 0; k < dilations.size(); k++)
+            strm << (k > 0 ? ", " : "") << dilations[k];
+        strm << "],\n";
+
+        prindent(strm, indent);
+        strm << "pad: [";
+        for (size_t k = 0; k < pads.size(); k++)
+            strm << (k > 0 ? ", " : "") << pads[k];
+        strm << "],\n";
+
+        prindent(strm, indent);
+        strm << "stride: [";
+        for (size_t k = 0; k < strides.size(); k++)
+            strm << (k > 0 ? ", " : "") << strides[k];
+        strm << "],\n";
+
+        return strm;
+    }
+
+    virtual int64_t getFLOPS(const std::vector<MatShape> &inputs,
+                             const std::vector<MatShape> &outputs) const CV_OVERRIDE
+    {
+        CV_Assert(inputs.size() == 1);
+        CV_Assert(outputs.size() == 1);
+        int ksize = 1;
+        for (auto sz: kernel_shape) ksize *= sz;
+        return (int64_t)(inputs[0].total()*ksize);
+    }
+
+    virtual void getTypes(const std::vector<MatType>& inptypes,
+                          const int, const int,
+                          std::vector<MatType>& outtypes,
+                          std::vector<MatType>& temptypes) const CV_OVERRIDE
+    {
+        int ninputs = (int)inptypes.size();
+        CV_Assert(ninputs == 1);
+
+        outtypes.assign(1, inptypes[0]);
+        temptypes.clear();
+    }
+
+    virtual bool getMemoryShapes(const std::vector<MatShape>& inpshapes,
+                                 const int,
+                                 std::vector<MatShape> &outshapes,
+                                 std::vector<MatShape> &tempshapes) const CV_OVERRIDE
+    {
+        size_t ninputs = inpshapes.size();
+        CV_Assert(ninputs == 1);
+
+        outshapes.assign(1, convInferShape(inpshapes[0], MatShape(),
+                                           kernel_shape, 0, strides, dilations,
+                                           pads, auto_pad, ceil_mode));
+        tempshapes.clear();
+        return true;
+    }
+
+    int getLayouts(const std::vector<DataLayout>& actualInputs,
+                    std::vector<DataLayout>& desiredInputs,
+                    const int requiredOutputs,
+                    std::vector<DataLayout>& outputs) const CV_OVERRIDE
+    {
+        CV_Assert(actualInputs.size() == 1u);
+        desiredInputs.assign(1, DATA_LAYOUT_BLOCK);
+        outputs.assign(requiredOutputs, DATA_LAYOUT_BLOCK);
+        return getNetImpl(this)->defaultC0;
+    }
+
+    void finalize(InputArrayOfArrays, OutputArrayOfArrays outputs_arr) CV_OVERRIDE
+    {
+    }
+
+    void forward(InputArrayOfArrays inputs_arr,
+                 OutputArrayOfArrays outputs_arr,
+                 OutputArrayOfArrays) CV_OVERRIDE
+    {
+        size_t ninputs = inputs_arr.total();
+        CV_Assert(ninputs == 1);
+
+        int inptype = inputs_arr.type(0);
+        MatShape inpshape = inputs_arr.shape(0);
+        MatShape outshape = convInferShape(inpshape, MatShape(),
+                                           kernel_shape, 0, strides, dilations,
+                                           pads, auto_pad, ceil_mode);
+        int outKind = outputs_arr.kind();
+        CV_Assert(outKind == _InputArray::STD_VECTOR_MAT ||
+                  outKind == _InputArray::STD_VECTOR_UMAT);
+
+        ConvState cs;
+        cs.initPooling(inpshape, outshape, kernel_shape, strides,
+                       dilations, pads, auto_pad, ceil_mode);
+
+        if (outKind == _InputArray::STD_VECTOR_MAT) {
+            Mat inp = inputs_arr.getMat(0);
+            std::vector<Mat>& outs = outputs_arr.getMatVecRef();
+            outs.resize(1);
+            outs[0].fit(outshape, inptype);
+            runOp(inp, outs[0], cs);
+        } else {
+            // [TODO] more efficient OpenCL implementation
+            Mat inp = inputs_arr.getMat(0);
+            std::vector<UMat>& outs = outputs_arr.getUMatVecRef();
+            outs.resize(1);
+            outs[0].fit(outshape, inptype);
+            Mat temp(outshape, inptype);
+            runOp(inp, temp, cs);
+            temp.copyTo(outs[0]);
+        }
+    }
+
+    void runOp(const Mat& inp, Mat& out, const ConvState& cs)
+    {
+        int inptype = inp.type();
+        AvgPoolFunc func =
+            inptype == CV_32F ? avgPool32f :
+            /*inptype == CV_16F ? avgPool16f :
+            inptype == CV_16BF ? avgPool16bf :*/
+            nullptr;
+
+        CV_Assert(func != nullptr && "AveragePool: unsupported data type");
+        func(inp.data, out.data, cs, count_include_pad);
+    }
+};
+
+Ptr<AveragePoolLayer> AveragePoolLayer::create(const LayerParams& params)
+{
+    return Ptr<AveragePoolLayer>(new AveragePoolLayerImpl(params));
+}
+
+}}

modules/dnn/src/layers/batch_norm2_layer.cpp

@@ -6,37 +6,351 @@
 
 #include "../precomp.hpp"
 #include "layers_common.hpp"
+#include "../net_impl.hpp"
 
 namespace cv {
 namespace dnn {
 
-class BatchNorm2LayerImpl CV_FINAL : public BatchNorm2Layer {
+/*
+    Implementation of BatchNormalization, as defined in ONNX specification:
+    https://onnx.ai/onnx/operators/onnx__BatchNormalization.html
+
+    Opset's 1 to 15 are covered.
+*/
+
+#undef CV_SIMD_ONLY
+#if CV_SIMD || CV_SIMD_SCALABLE
+#define CV_SIMD_ONLY(expr) expr
+#else
+#define CV_SIMD_ONLY(expr)
+#endif
+
+// out must be pre-allocated
+static void batchnorm(const Mat& inp, Mat& out, const Mat& scale,
+                      const Mat& bias, DataLayout defaultLayout)
+{
+    CV_Assert_N(inp.isContinuous(), out.isContinuous());
+    CV_Assert_N(scale.dims == 1, bias.dims == 1);
+    CV_Assert_N(scale.type() == CV_32F, bias.type() == CV_32F);
+
+    MatShape shape = inp.shape();
+    DataLayout layout = shape.layout;
+    if (layout == DATA_LAYOUT_UNKNOWN)
+        layout = defaultLayout;
+    CV_Assert(layout == DATA_LAYOUT_BLOCK ||
+              layout == DATA_LAYOUT_NCHW ||
+              layout == DATA_LAYOUT_NHWC);
+    int N = shape[0];
+    int C_ = layout == DATA_LAYOUT_BLOCK ? shape.C :
+             layout == DATA_LAYOUT_NCHW ? shape[1] : shape[shape.dims-1];
+    CV_Assert_N(scale.cols == C_, bias.cols == C_);
+    int C1_ = layout != DATA_LAYOUT_NHWC ? shape[1] : 1;
+    int C0_ = layout != DATA_LAYOUT_NCHW ? shape[shape.dims-1] : 1;
+    int type_ = inp.type();
+    CV_SIMD_ONLY(int vlanes_ = VTraits<v_float32>::vlanes());
+
+    size_t esz = inp.elemSize();
+
+    CV_Assert(type_ == CV_32F || type_ == CV_16F || type_ == CV_16BF);
+
+    CV_Assert(inp.shape() == out.shape());
+    CV_Assert(inp.isContinuous());
+    CV_Assert(out.isContinuous());
+
+    int planesize_ = 1;
+    for (int i = 1; i < shape.dims; i++) {
+        planesize_ *= shape[i];
+    }
+    planesize_ /= (C0_*C1_);
+
+    parallel_for_(Range(0, N*C1_), [&](const Range& r) {
+        int C0 = C0_, C1 = C1_, C = C_;
+        int planesize_C0 = planesize_*C0;
+        int type = type_;
+        CV_SIMD_ONLY(int vlanes = vlanes_;
+        constexpr int max_lanes = VTraits<v_float32>::max_nlanes;
+        constexpr int MAX_UNROLL = 4;
+        float scalebuf[max_lanes*MAX_UNROLL];
+        float biasbuf[max_lanes*MAX_UNROLL]);
+
+        for (int k = r.start; k < r.end; k++) {
+            int i = 0, n = k/C1, c1 = k % C1;
+            int c_start = c1*C0, c_end = std::min(C, c_start + C0), c_delta = c_end - c_start;
+            size_t ofs_k = (n*C1 + c1)*planesize_C0*esz;
+            const uchar* inptr_ = inp.data + ofs_k;
+            uchar* outptr_ = out.data + ofs_k;
+            const float* scaleptr = scale.ptr<float>() + c_start;
+            const float* biasptr = bias.ptr<float>() + c_start;
+
+            if (C0 == 1) {
+                // NCHW case
+                float scale_c = 0.f, bias_c = 0.f;
+                if (c_start < c_end) {
+                    scale_c = scaleptr[0];
+                    bias_c = biasptr[0];
+                }
+                CV_SIMD_ONLY(v_float32 vscale_c = vx_setall_f32(scale_c);
+                             v_float32 vbias_c = vx_setall_f32(bias_c));
+                if (type == CV_32F) {
+                    const float* inptr = (const float*)inptr_;
+                    float* outptr = (float*)outptr_;
+                    CV_SIMD_ONLY(for (; i <= planesize_C0 - vlanes*2; i += vlanes*2) {
+                        v_float32 x0 = vx_load(inptr + i);
+                        v_float32 x1 = vx_load(inptr + i + vlanes);
+                        x0 = v_fma(x0, vscale_c, vbias_c);
+                        x1 = v_fma(x1, vscale_c, vbias_c);
+                        v_store(outptr + i, x0);
+                        v_store(outptr + i + vlanes, x1);
+                    });
+                    for (; i < planesize_C0; i++) {
+                        outptr[i] = inptr[i]*scale_c + bias_c;
+                    }
+                } else if (type == CV_16F) {
+                    const hfloat* inptr = (const hfloat*)inptr_;
+                    hfloat* outptr = (hfloat*)outptr_;
+                    CV_SIMD_ONLY(for (; i <= planesize_C0 - vlanes*2; i += vlanes*2) {
+                        v_float32 x0 = vx_load_expand(inptr + i);
+                        v_float32 x1 = vx_load_expand(inptr + i + vlanes);
+                        x0 = v_fma(x0, vscale_c, vbias_c);
+                        x1 = v_fma(x1, vscale_c, vbias_c);
+                        v_pack_store(outptr + i, x0);
+                        v_pack_store(outptr + i + vlanes, x1);
+                    });
+                    for (; i < planesize_C0; i++) {
+                        outptr[i] = hfloat(float(inptr[i])*scale_c + bias_c);
+                    }
+                } else if (type == CV_16BF) {
+                    const bfloat* inptr = (const bfloat*)inptr_;
+                    bfloat* outptr = (bfloat*)outptr_;
+                    CV_SIMD_ONLY(for (; i <= planesize_C0 - vlanes*2; i += vlanes*2) {
+                        v_float32 x0 = vx_load_expand(inptr + i);
+                        v_float32 x1 = vx_load_expand(inptr + i + vlanes);
+                        x0 = v_fma(x0, vscale_c, vbias_c);
+                        x1 = v_fma(x1, vscale_c, vbias_c);
+                        v_pack_store(outptr + i, x0);
+                        v_pack_store(outptr + i + vlanes, x1);
+                    });
+                    for (; i < planesize_C0; i++) {
+                        outptr[i] = bfloat(float(inptr[i])*scale_c + bias_c);
+                    }
+                }
+            }
+        #if CV_SIMD || CV_SIMD_SCALABLE
+            /*
+                [TODO] support C0 == vlanes/2, maybe C0 == vlanes/4.
+                in this case, load everything into vsc0 and vb0, process
+                most part of the plane using vector code as if C0 == vlanes and
+                then process the tail using scalar code
+            */
+            else if (C0 == vlanes*4 || C0 == vlanes*2 || C0 == vlanes) {
+                // accelerated block layout case
+                int c = 0;
+                for (; c < c_delta; c++) {
+                    scalebuf[c] = scaleptr[c];
+                    biasbuf[c] = biasptr[c];
+                }
+                for (; c < MAX_UNROLL*vlanes; c++) {
+                    scalebuf[c] = biasbuf[c] = 0.f;
+                }
+                v_float32 vsc0, vsc1, vsc2, vsc3;
+                v_float32 vb0, vb1, vb2, vb3, vb4;
+                vsc0 = vx_load(scalebuf);
+                vsc1 = vx_load(scalebuf + vlanes);
+                vsc2 = vx_load(scalebuf + vlanes*2);
+                vsc3 = vx_load(scalebuf + vlanes*3);
+                vb0 = vx_load(biasbuf);
+                vb1 = vx_load(biasbuf + vlanes);
+                vb2 = vx_load(biasbuf + vlanes*2);
+                vb3 = vx_load(biasbuf + vlanes*3);
+                if (type == CV_32F) {
+                    const float* inptr = (const float*)inptr_;
+                    float* outptr = (float*)outptr_;
+                    if (C0 == vlanes*4) {
+                        for (; i < planesize_C0; i += C0) {
+                            v_float32 x0 = vx_load(inptr + i);
+                            v_float32 x1 = vx_load(inptr + i + vlanes);
+                            v_float32 x2 = vx_load(inptr + i + vlanes*2);
+                            v_float32 x3 = vx_load(inptr + i + vlanes*3);
+                            x0 = v_fma(x0, vsc0, vb0);
+                            x1 = v_fma(x1, vsc1, vb1);
+                            x2 = v_fma(x2, vsc2, vb2);
+                            x3 = v_fma(x3, vsc3, vb3);
+                            v_store(outptr + i, x0);
+                            v_store(outptr + i + vlanes, x1);
+                            v_store(outptr + i + vlanes*2, x2);
+                            v_store(outptr + i + vlanes*3, x3);
+                        }
+                    } else if (C0 == vlanes*2) {
+                        for (; i < planesize_C0; i += C0) {
+                            v_float32 x0 = vx_load(inptr + i);
+                            v_float32 x1 = vx_load(inptr + i + vlanes);
+                            x0 = v_fma(x0, vsc0, vb0);
+                            x1 = v_fma(x1, vsc1, vb1);
+                            v_store(outptr + i, x0);
+                            v_store(outptr + i + vlanes, x1);
+                        }
+                    } else {
+                        for (; i < planesize_C0; i += C0) {
+                            v_float32 x0 = vx_load(inptr + i);
+                            x0 = v_fma(x0, vsc0, vb0);
+                            v_store(outptr + i, x0);
+                        }
+                    }
+                } else if (type == CV_16F) {
+                    const hfloat* inptr = (const hfloat*)inptr_;
+                    hfloat* outptr = (hfloat*)outptr_;
+                    if (type == CV_32F) {
+                        if (C0 == vlanes*4) {
+                            for (; i < planesize_C0; i += C0) {
+                                v_float32 x0 = vx_load_expand(inptr + i);
+                                v_float32 x1 = vx_load_expand(inptr + i + vlanes);
+                                v_float32 x2 = vx_load_expand(inptr + i + vlanes*2);
+                                v_float32 x3 = vx_load_expand(inptr + i + vlanes*3);
+                                x0 = v_fma(x0, vsc0, vb0);
+                                x1 = v_fma(x1, vsc1, vb1);
+                                x2 = v_fma(x2, vsc2, vb2);
+                                x3 = v_fma(x3, vsc3, vb3);
+                                v_pack_store(outptr + i, x0);
+                                v_pack_store(outptr + i + vlanes, x1);
+                                v_pack_store(outptr + i + vlanes*2, x2);
+                                v_pack_store(outptr + i + vlanes*3, x3);
+                            }
+                        } else if (C0 == vlanes*2) {
+                            for (; i < planesize_C0; i += C0) {
+                                v_float32 x0 = vx_load_expand(inptr + i);
+                                v_float32 x1 = vx_load_expand(inptr + i + vlanes);
+                                x0 = v_fma(x0, vsc0, vb0);
+                                x1 = v_fma(x1, vsc1, vb1);
+                                v_pack_store(outptr + i, x0);
+                                v_pack_store(outptr + i + vlanes, x1);
+                            }
+                        } else {
+                            for (; i < planesize_C0; i += C0) {
+                                v_float32 x0 = vx_load_expand(inptr + i);
+                                x0 = v_fma(x0, vsc0, vb0);
+                                v_pack_store(outptr + i, x0);
+                            }
+                        }
+                    }
+                } else if (type == CV_16BF) {
+                    const bfloat* inptr = (const bfloat*)inptr_;
+                    bfloat* outptr = (bfloat*)outptr_;
+                    if (C0 == vlanes*4) {
+                        for (; i < planesize_C0; i += C0) {
+                            v_float32 x0 = vx_load_expand(inptr + i);
+                            v_float32 x1 = vx_load_expand(inptr + i + vlanes);
+                            v_float32 x2 = vx_load_expand(inptr + i + vlanes*2);
+                            v_float32 x3 = vx_load_expand(inptr + i + vlanes*3);
+                            x0 = v_fma(x0, vsc0, vb0);
+                            x1 = v_fma(x1, vsc1, vb1);
+                            x2 = v_fma(x2, vsc2, vb2);
+                            x3 = v_fma(x3, vsc3, vb3);
+                            v_pack_store(outptr + i, x0);
+                            v_pack_store(outptr + i + vlanes, x1);
+                            v_pack_store(outptr + i + vlanes*2, x2);
+                            v_pack_store(outptr + i + vlanes*3, x3);
+                        }
+                    } else if (C0 == vlanes*2) {
+                        for (; i < planesize_C0; i += C0) {
+                            v_float32 x0 = vx_load_expand(inptr + i);
+                            v_float32 x1 = vx_load_expand(inptr + i + vlanes);
+                            x0 = v_fma(x0, vsc0, vb0);
+                            x1 = v_fma(x1, vsc1, vb1);
+                            v_pack_store(outptr + i, x0);
+                            v_pack_store(outptr + i + vlanes, x1);
+                        }
+                    } else {
+                        for (; i < planesize_C0; i += C0) {
+                            v_float32 x0 = vx_load_expand(inptr + i);
+                            x0 = v_fma(x0, vsc0, vb0);
+                            v_pack_store(outptr + i, x0);
+                        }
+                    }
+                }
+            }
+        #endif
+            else {
+                // general block layout or NHWC case
+                if (type == CV_32F) {
+                    const float* inptr = (const float*)inptr_;
+                    float* outptr = (float*)outptr_;
+                    for (; i < planesize_C0; i += C0) {
+                        int c = 0;
+                        CV_SIMD_ONLY(for (; c <= c_delta - vlanes*2; c += vlanes*2) {
+                            v_float32 x0 = vx_load(inptr + i + c);
+                            v_float32 x1 = vx_load(inptr + i + c + vlanes);
+                            v_float32 sc0 = vx_load(scaleptr + c);
+                            v_float32 sc1 = vx_load(scaleptr + c + vlanes);
+                            v_float32 b0 = vx_load(biasptr + c);
+                            v_float32 b1 = vx_load(biasptr + c + vlanes);
+                            x0 = v_fma(x0, sc0, b0);
+                            x1 = v_fma(x1, sc1, b1);
+                            v_store(outptr + i + c, x0);
+                            v_store(outptr + i + c + vlanes, x1);
+                        });
+                        for (; c < c_delta; c++)
+                            outptr[i + c] = inptr[i + c]*scaleptr[c] + biasptr[c];
+                        for (; c < C0; c++)
+                            outptr[i + c] = 0.f;
+                    }
+                } else if (type == CV_16F) {
+                    const hfloat* inptr = (const hfloat*)inptr_;
+                    hfloat* outptr = (hfloat*)outptr_;
+                    hfloat z = hfloat(0.f);
+                    for (; i < planesize_C0; i += C0) {
+                        int c = 0;
+                        CV_SIMD_ONLY(for (; c <= c_delta - vlanes*2; c += vlanes*2) {
+                            v_float32 x0 = vx_load_expand(inptr + i + c);
+                            v_float32 x1 = vx_load_expand(inptr + i + c + vlanes);
+                            v_float32 sc0 = vx_load(scaleptr + c);
+                            v_float32 sc1 = vx_load(scaleptr + c + vlanes);
+                            v_float32 b0 = vx_load(biasptr + c);
+                            v_float32 b1 = vx_load(biasptr + c + vlanes);
+                            x0 = v_fma(x0, sc0, b0);
+                            x1 = v_fma(x1, sc1, b1);
+                            v_pack_store(outptr + i + c, x0);
+                            v_pack_store(outptr + i + c + vlanes, x1);
+                        });
+                        for (; c < c_delta; c++)
+                            outptr[i + c] = hfloat(float(inptr[i + c])*scaleptr[c] + biasptr[c]);
+                        for (; c < C0; c++)
+                            outptr[i + c] = z;
+                    }
+                } else if (type == CV_16BF) {
+                    const bfloat* inptr = (const bfloat*)inptr_;
+                    bfloat* outptr = (bfloat*)outptr_;
+                    bfloat z = bfloat(0.f);
+                    for (; i < planesize_C0; i += C0) {
+                        int c = 0;
+                        CV_SIMD_ONLY(for (; c <= c_delta - vlanes*2; c += vlanes*2) {
+                            v_float32 x0 = vx_load_expand(inptr + i + c);
+                            v_float32 x1 = vx_load_expand(inptr + i + c + vlanes);
+                            v_float32 sc0 = vx_load(scaleptr + c);
+                            v_float32 sc1 = vx_load(scaleptr + c + vlanes);
+                            v_float32 b0 = vx_load(biasptr + c);
+                            v_float32 b1 = vx_load(biasptr + c + vlanes);
+                            x0 = v_fma(x0, sc0, b0);
+                            x1 = v_fma(x1, sc1, b1);
+                            v_pack_store(outptr + i + c, x0);
+                            v_pack_store(outptr + i + c + vlanes, x1);
+                        });
+                        for (; c < c_delta; c++)
+                            outptr[i + c] = bfloat(float(inptr[i + c])*scaleptr[c] + biasptr[c]);
+                        for (; c < C0; c++)
+                            outptr[i + c] = z;
+                    }
+                }
+            }
+        }
+    }, (planesize_*C0_ > 1000000 ? N*C1_ : 1));
+}
+
+class BatchNorm2LayerImpl CV_FINAL : public BatchNorm2Layer
+{
 public:
     BatchNorm2LayerImpl(const LayerParams& params) {
         setParamsFrom(params);
-
-        epsilon = params.get<float>("epsilon", params.get<float>("eps", 1e-5f));
-        useGlobalStats = params.get<bool>("use_global_stats", true);
-        hasWeights = params.get<bool>("has_weight", false);
-        hasBias = params.get<bool>("has_bias", false);
-
-        if (blobs.size() >= 4) {
-            dynamicInputs = false;
-
-            const Mat& mean  = blobs[0];
-            const Mat& var   = blobs[1];
-            const Mat& scale = blobs[2];
-            const Mat& beta  = blobs[3];
-
-            weights_.create(scale.size(), CV_32F);
-            bias_.create(scale.size(), CV_32F);
-
-            cv::sqrt(var + epsilon, bias_);
-            cv::divide(scale, bias_, weights_);
-            bias_ = beta - mean.mul(weights_);
-        } else {
-            dynamicInputs = true;
-        }
+        epsilon = params.get<float>("epsilon", 1e-5);
     }
 
     bool supportBackend(int backendId) CV_OVERRIDE
@@ -44,121 +358,183 @@ public:
         return backendId == DNN_BACKEND_OPENCV;
     }
 
-    bool dynamicOutputShapes() const CV_OVERRIDE
+    MatShape getOutShape(const MatShape& inpShape) const
     {
-        return dynamicInputs;
+        return inpShape;
     }
 
-    bool getMemoryShapes(const std::vector<MatShape>& inputs,
-                                 const int requiredOutputs,
-                                 std::vector<MatShape>& outputs,
-                                 std::vector<MatShape>& internals) const CV_OVERRIDE
+    virtual bool getMemoryShapes(const std::vector<MatShape> &inputs,
+                         const int,
+                         std::vector<MatShape> &outputs,
+                         std::vector<MatShape> &internals) const CV_OVERRIDE
+    {
+        CV_Assert(!inputs.empty());
+        outputs.assign(1, getOutShape(inputs[0]));
+        internals.clear();
+        return true;
+    }
+
+    virtual void getTypes(const std::vector<MatType>& inputs,
+        const int requiredOutputs,
+        const int requiredInternals,
+        std::vector<MatType>& outputs,
+        std::vector<MatType>& internals) const CV_OVERRIDE
     {
         CV_Assert(!inputs.empty());
         outputs.assign(requiredOutputs, inputs[0]);
-        return false;
+        CV_Assert(requiredInternals == 0);
+        internals.clear();
     }
 
-    void getTypes(const std::vector<MatType>& inputs,
-                          const int requiredOutputs,
-                          const int requiredInternals,
-                          std::vector<MatType>& outputs,
-                          std::vector<MatType>& internals) const CV_OVERRIDE
+    int getLayouts(const std::vector<DataLayout>& actualInputs,
+                    std::vector<DataLayout>& desiredInputs,
+                    const int requiredOutputs,
+                    std::vector<DataLayout>& outputs) const CV_OVERRIDE
     {
-        CV_Assert(!inputs.empty());
-        outputs.assign(requiredOutputs, inputs[0]);
+        size_t ninputs = actualInputs.size();
+        CV_Assert(ninputs >= 1u);
+        desiredInputs.assign(ninputs, DATA_LAYOUT_UNKNOWN);
+        desiredInputs[0] = actualInputs[0];
+        outputs.assign(requiredOutputs, actualInputs[0]);
+        return 0;
     }
 
-    void forward(InputArrayOfArrays inputs_arr, OutputArrayOfArrays outputs_arr, OutputArrayOfArrays internals_arr) CV_OVERRIDE
+    virtual void finalize(InputArrayOfArrays, OutputArrayOfArrays outputs_arr) CV_OVERRIDE
     {
-        if (inputs_arr.depth() == CV_16F)
-        {
-            forward_fallback(inputs_arr, outputs_arr, internals_arr);
-            return;
+    }
+
+    virtual void forward(InputArrayOfArrays inputs_arr,
+                 OutputArrayOfArrays outputs_arr,
+                 OutputArrayOfArrays) CV_OVERRIDE
+    {
+        CV_TRACE_FUNCTION();
+        CV_TRACE_ARG_VALUE(name, "name", name.c_str());
+
+        size_t ninputs = inputs_arr.total(-1);
+        CV_Assert(ninputs > 0);
+
+        if (ninputs > 1) {
+            CV_Assert(ninputs == 5);
+            Mat scale_ = inputs_arr.getMat(1);
+            Mat bias_ = inputs_arr.getMat(2);
+            Mat mean_ = inputs_arr.getMat(3);
+            Mat var_ = inputs_arr.getMat(4);
+            BatchNorm2Layer::getScaleBias(scale_, bias_, mean_, var_, epsilon, scale, bias);
         }
 
-        std::vector<Mat> inputs;
-        inputs_arr.getMatVector(inputs);
+        MatShape inpShape = inputs_arr.shape(0);
+        int inpType = inputs_arr.type(0);
 
-        const Mat &X = inputs[0];
-        Mat Y;
-        Mat w, b;
+        MatShape outShape = getOutShape(inpShape);
+        int outKind = outputs_arr.kind();
 
-        if (dynamicInputs) {
-            CV_Assert(inputs.size() == 5);
+        CV_Assert(outKind == _InputArray::STD_VECTOR_MAT ||
+                  outKind == _InputArray::STD_VECTOR_UMAT);
 
-            const Mat& scale   = inputs[1];
-            const Mat& beta    = inputs[2];
-            const Mat& mean    = inputs[3];
-            const Mat& var     = inputs[4];
-
-            w.create(scale.size(), CV_32F);
-            b.create(scale.size(), CV_32F);
-
-            cv::sqrt(var + epsilon, b);
-            cv::divide(scale, b, w);
-            b = beta - mean.mul(w);
-        } else {
-            w = weights_;
-            b = bias_;
-        }
-
-        if (w.empty() || b.empty())
-             CV_Error(Error::StsBadArg, "BatchNorm2Layer: Weights not initialized");
-
-        MatShape outShape = shape(X);
-        auto kind = outputs_arr.kind();
-        if (kind == _InputArray::STD_VECTOR_MAT) {
+        if (outKind == _InputArray::STD_VECTOR_MAT) {
+            Mat inp = inputs_arr.getMat(0);
             std::vector<Mat>& outs = outputs_arr.getMatVecRef();
-            CV_Assert(outs.size() >= 1);
-            outs[0].fit(outShape, X.type());
-            Y = outs[0];
-        } else if (kind == _InputArray::STD_VECTOR_UMAT) {
-            std::vector<UMat>& uouts = outputs_arr.getUMatVecRef();
-            CV_Assert(uouts.size() >= 1);
-            uouts[0].fit(outShape, X.type());
-            Y = uouts[0].getMat(ACCESS_WRITE);
+            outs.resize(1);
+            outs[0].fit(outShape, inpType);
+            runOp(inp, outs[0]);
         } else {
-            CV_Error(Error::StsBadArg, "Unsupported output array kind");
+            // [TODO] more efficient OpenCL implementation
+            Mat inp = inputs_arr.getMat(0);
+            std::vector<UMat>& outs = outputs_arr.getUMatVecRef();
+            outs.resize(1);
+            outs[0].fit(outShape, inpType);
+            Mat temp(outShape, inpType);
+            runOp(inp, temp);
+            temp.copyTo(outs[0]);
         }
-
-        const int C = (X.dims >= 2) ? X.size[1] : 1;
-        const int N = X.size[0];
-        const size_t planeSize = X.total() / (N * C);
-
-        CV_Assert(w.total() == C);
-
-        parallel_for_(Range(0, N * C), [&](const Range& r) {
-            for (int i = r.start; i < r.end; ++i) {
-                int c = i % C;
-
-                float scale_val = w.ptr<float>()[c];
-                float shift_val = b.ptr<float>()[c];
-
-                const float* srcPtr = X.ptr<float>() + i * planeSize;
-                float* dstPtr = Y.ptr<float>() + i * planeSize;
-
-                int j = 0;
-#if CV_SIMD128
-                v_float32x4 v_scale = v_setall_f32(scale_val);
-                v_float32x4 v_shift = v_setall_f32(shift_val);
-                for (; j <= (int)planeSize - 4; j += 4) {
-                    v_float32x4 v_src = v_load(srcPtr + j);
-                    v_float32x4 v_dst = v_muladd(v_src, v_scale, v_shift);
-                    v_store(dstPtr + j, v_dst);
-                }
-#endif
-                for (; j < (int)planeSize; ++j) {
-                    dstPtr[j] = srcPtr[j] * scale_val + shift_val;
-                }
-            }
-        });
     }
-private:
-    bool dynamicInputs;
-    Mat weights_, bias_;
+
+    void runOp(const Mat& inp, Mat& out)
+    {
+        auto netimpl_ = getNetImpl(this);
+        batchnorm(inp, out, scale, bias, netimpl_->originalLayout);
+    }
+
+    virtual bool freezeScaleBias() CV_OVERRIDE
+    {
+        auto netimpl_ = getNetImpl(this);
+        size_t ninputs = inputs.size();
+        if (ninputs != 5)
+            return false;
+        if (!netimpl_->isConstArg(inputs[1]) ||
+            !netimpl_->isConstArg(inputs[2]) ||
+            !netimpl_->isConstArg(inputs[3]) ||
+            !netimpl_->isConstArg(inputs[4]))
+            return false;
+        Mat scale_ = netimpl_->argTensor(inputs[1]);
+        Mat bias_ = netimpl_->argTensor(inputs[2]);
+        Mat mean_ = netimpl_->argTensor(inputs[3]);
+        Mat var_ = netimpl_->argTensor(inputs[4]);
+        BatchNorm2Layer::getScaleBias(scale_, bias_, mean_, var_, epsilon, scale, bias);
+        inputs.resize(1);
+        return true;
+    }
+
+    virtual void getScaleBias(OutputArray scale_, OutputArray bias_) const CV_OVERRIDE
+    {
+        scale.copyTo(scale_);
+        bias.copyTo(bias_);
+    }
+
+    Mat scale, bias;
 };
 
+void BatchNorm2Layer::getScaleBias(InputArray scale_, InputArray bias_,
+                                   InputArray mean_, InputArray variance_, float eps,
+                                   OutputArray outscale_, OutputArray outbias_)
+{
+    Mat scale = scale_.getMat(), bias = bias_.getMat();
+    Mat mean = mean_.getMat(), var = variance_.getMat();
+
+    int sctype = scale.type(), btype = bias.type();
+    int mtype = mean.type(), vtype = var.type();
+
+    CV_Assert(sctype == CV_32F || sctype == CV_16F || sctype == CV_16BF);
+    CV_Assert(btype == CV_32F || btype == CV_16F || btype == CV_16BF);
+    CV_Assert(mtype == CV_32F || mtype == CV_16F || mtype == CV_16BF);
+    CV_Assert(vtype == CV_32F || vtype == CV_16F || vtype == CV_16BF);
+
+    CV_Assert_N(scale.dims == 1, bias.dims == 1, mean.dims == 1, var.dims == 1);
+    int C = scale.cols;
+    CV_Assert_N(bias.cols == C, mean.cols == C, var.cols == C);
+
+    Mat outscale(1, &C, CV_32F);
+    Mat outbias(1, &C, CV_32F);
+
+    const uchar* scdata = scale.data;
+    const uchar* bdata = bias.data;
+    const uchar* mdata = mean.data;
+    const uchar* vdata = var.data;
+    float* outsc = outscale.ptr<float>();
+    float* outb = outbias.ptr<float>();
+
+    #undef LOAD_AS_FLOAT
+    #define LOAD_AS_FLOAT(typ, ptr, i) \
+        (typ == CV_32F ? ((const float*)ptr)[i] : \
+         typ == CV_16F ? float(((const hfloat*)ptr)[i]) : \
+         float(((const bfloat*)ptr)[i]))
+
+    // ONNX documentation: Y = (X - input_mean) / sqrt(input_var + epsilon) * scale + B
+    for (int i = 0; i < C; i++) {
+        float sc = LOAD_AS_FLOAT(sctype, scdata, i);
+        float b = LOAD_AS_FLOAT(btype, bdata, i);
+        float m = LOAD_AS_FLOAT(mtype, mdata, i);
+        float v = LOAD_AS_FLOAT(vtype, vdata, i);
+
+        float outscval = sc/sqrtf(fabsf(v) + eps);
+        float outbval = b - m*outscval;
+        outsc[i] = outscval;
+        outb[i] = outbval;
+    }
+    outscale.copyTo(outscale_);
+    outbias.copyTo(outbias_);
+}
+
 Ptr<BatchNorm2Layer> BatchNorm2Layer::create(const LayerParams& params)
 {
     return makePtr<BatchNorm2LayerImpl>(params);

modules/dnn/src/layers/conv2_common.cpp

@@ -0,0 +1,391 @@
+// 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 "../precomp.hpp"
+#include "../net_impl.hpp"
+#include "layers_common.hpp"
+#include "conv2_common.hpp"
+#include <math.h>
+
+namespace cv { namespace dnn {
+CV__DNN_INLINE_NS_BEGIN
+
+std::string fastActivationToString(FastActivation fastActivation)
+{
+    return fastActivation == FAST_ACTIV_RELU ? "ReLU" :
+           fastActivation == FAST_ACTIV_LEAKY_RELU ? "LeakyReLU" :
+           fastActivation == FAST_ACTIV_PRELU ? "PReLU" :
+           fastActivation == FAST_ACTIV_CLIP ? "Clip" :
+           fastActivation == FAST_ACTIV_NONE ? "None" : format("unknown(%d)", int(fastActivation));
+}
+
+AutoPadding getAutoPadding(const LayerParams& params)
+{
+    std::string auto_pad = params.get<std::string>("auto_pad", "NOTSET");
+    std::string pad_mode = params.get<std::string>("pad_mode", "");
+    if (pad_mode == "SAME")
+        return AUTO_PAD_SAME_UPPER;
+    if (pad_mode == "VALID")
+        return AUTO_PAD_VALID;
+    if (auto_pad == "NOTSET")
+        return AUTO_PAD_NONE;
+    if (auto_pad == "SAME_UPPER")
+        return AUTO_PAD_SAME_UPPER;
+    if (auto_pad == "SAME_LOWER")
+        return AUTO_PAD_SAME_LOWER;
+    if (auto_pad != "VALID") {
+        CV_Error_(Error::StsBadArg, ("invalid auto_pad value '%s'", auto_pad.c_str()));
+    }
+    return AUTO_PAD_VALID;
+}
+
+// computes shape of the output tensor of convolution
+// (including depth-wise convolution), max pooling or average pooling operations
+// computes shape of the output tensor of convolution
+// (including depth-wise convolution), max pooling or average pooling operations
+MatShape convInferShape(const MatShape& inpShape, const MatShape& wshape,
+                        const std::vector<int>& kernelShape, int ngroups,
+                        const std::vector<int>& strides,
+                        const std::vector<int>& dilations,
+                        const std::vector<int>& pads,
+                        AutoPadding autoPad, bool ceilMode)
+{
+    bool blockLayout = true;
+    int ndims = inpShape.dims;
+    size_t nspatialdims = (size_t)(ndims - 2 - int(blockLayout));
+    MatShape outshape = inpShape;
+    int kshape[MatShape::MAX_DIMS];
+
+    if (!kernelShape.empty()) {
+        size_t kshape_size = kernelShape.size();
+        CV_Assert(kshape_size == nspatialdims || kshape_size == nspatialdims+2);
+        for (size_t i = 0; i < nspatialdims; i++)
+            kshape[i] = kernelShape[kshape_size - nspatialdims + i];
+    } else {
+        CV_Assert(!wshape.empty() && wshape.dims == nspatialdims + 2);
+        for (size_t i = 0; i < nspatialdims; i++)
+            kshape[i] = wshape[wshape.dims - nspatialdims + i];
+    }
+
+    if (ngroups == 0 || wshape.empty()) {
+        outshape[1] = inpShape[1];
+    } else if (blockLayout) {
+        int C0 = inpShape[ndims-1];
+        outshape[1] = (wshape[0] + C0 - 1)/C0;
+    } else {
+        outshape[1] = wshape[0];
+    }
+
+    CV_Assert(strides.empty() || strides.size() == nspatialdims);
+    CV_Assert(dilations.empty() || dilations.size() == nspatialdims);
+    CV_Assert(autoPad == AUTO_PAD_NONE || pads.empty());
+    CV_Assert(pads.empty() || pads.size() == nspatialdims*2);
+
+    for (size_t i = 0; i < nspatialdims; i++) {
+        int inpsz = inpShape[i+2], k_i = kshape[i];
+        int stride = strides.empty() ? 1 : strides[i];
+        int dilation = dilations.empty() ? 1 : dilations[i];
+        int outsz;
+        if (autoPad == AUTO_PAD_NONE || autoPad == AUTO_PAD_VALID) {
+            int pad = 0;
+            if (!pads.empty()) {
+                pad = pads[i] + pads[i + nspatialdims];
+            }
+            outsz = (inpsz + pad - 1 - dilation * (k_i - 1) + (ceilMode ? stride - 1 : 0)) / stride + 1;
+        } else {
+            if (ceilMode)
+                outsz = (inpsz + stride - 1)/stride;
+            else
+                outsz = (inpsz - 1)/stride + 1;
+        }
+        outshape[i + 2] = outsz;
+    }
+
+    if (blockLayout) {
+        outshape.C = ngroups == 0 || wshape.empty() ? inpShape.C : wshape[0];
+    } else {
+        outshape.C = 0;
+    }
+
+    return outshape;
+}
+
+static inline void getPadding(const std::vector<int>& pads,
+                              int dim, int nspatialdims, AutoPadding autoPad,
+                              int ksize, int& pad0, int& pad1)
+{
+    CV_Assert(0 <= dim && dim < nspatialdims);
+
+    if (autoPad == AUTO_PAD_NONE || autoPad == AUTO_PAD_VALID) {
+        if (!pads.empty()) {
+            pad0 = pads[dim];
+            pad1 = pads[dim + nspatialdims];
+        } else {
+            pad0 = pad1 = 0;
+        }
+    } else {
+        CV_Assert(autoPad == AUTO_PAD_SAME_LOWER || autoPad == AUTO_PAD_SAME_UPPER);
+        pad0 = pad1 = ksize/2;
+        if (pad0*2 == ksize) {
+            pad0 -= autoPad == AUTO_PAD_SAME_UPPER;
+            pad1 -= autoPad == AUTO_PAD_SAME_LOWER;
+        }
+    }
+}
+
+bool ConvState::sameShape(const ConvState& cs) const
+{
+    for (int i = 0; i < ConvState::MAX_CONV_DIMS; i++) {
+        if (kshape[i] != cs.kshape[i] ||
+            strides[i] != cs.strides[i] ||
+            dilations[i] != cs.dilations[i] ||
+            pads[i] != cs.pads[i] ||
+            pads[i + MAX_CONV_DIMS] != cs.pads[i + MAX_CONV_DIMS]) {
+            return false;
+        }
+    }
+    return inpshape == cs.inpshape && outshape == cs.outshape;
+}
+
+static MatShape getWpackShape(const MatShape& wshape, int ngroups, int C0)
+{
+    CV_Assert(wshape.dims >= 3);
+    int K = wshape[0], Cg = wshape[1];
+    int ksize = int(wshape.total())/(K*Cg);
+    CV_Assert_N(K % ngroups == 0);
+    int Kg = K / ngroups, K0 = C0;
+    int Kblk = (Kg + K0 - 1)/K0;
+    int C1Max = 0;
+    for (int g = 0; g < ngroups; ++g) {
+        int c_start = g * Cg;
+        int c00 = c_start & (C0 - 1);
+        int cblocks = (c00 + Cg + C0 - 1)/C0;
+        C1Max = std::max(C1Max, cblocks);
+    }
+    return MatShape({ngroups, Kblk, ksize, C1Max, C0*K0}, DATA_LAYOUT_UNKNOWN);
+}
+
+void ConvState::initConv(const MatShape& inpshape_,
+                         const MatShape& wshape_,
+                         const MatShape& outshape_,
+                         int ngroups_,
+                         const std::vector<int>& strides_,
+                         const std::vector<int>& dilations_,
+                         const std::vector<int>& pads_,
+                         AutoPadding autoPad, bool ceilMode,
+                         FastActivation fastActivation_,
+                         const std::vector<float>& activParams_)
+{
+    nspatialdims = wshape_.dims - 2;
+    CV_Assert(0 < nspatialdims && nspatialdims <= ConvState::MAX_CONV_DIMS);
+    CV_Assert(strides_.empty() || (strides_.size() == size_t(nspatialdims)));
+    CV_Assert(dilations_.empty() || (dilations_.size() == size_t(nspatialdims)));
+    CV_Assert(pads_.empty() || (pads_.size() == size_t(nspatialdims*2)));
+    CV_Assert(inpshape_.dims == outshape_.dims);
+    CV_Assert(inpshape_.dims == nspatialdims + 2 + int(inpshape_.layout == DATA_LAYOUT_BLOCK));
+    CV_Assert_N(inpshape_.layout == outshape_.layout,
+                inpshape_[0] == outshape_[0]);
+
+    inpshape = inpshape_;
+    outshape = outshape_;
+    ngroups = ngroups_;
+
+    int C = inpshape.channels();
+    int K = outshape.channels();
+
+    depthwise = ngroups == C && ngroups == K;
+
+    if (inpshape.layout == DATA_LAYOUT_BLOCK) {
+        CV_Assert(inpshape.back() == outshape.back());
+    }
+
+    fastActivation = fastActivation_;
+    activation = nullptr;
+    activParams = activParams_;
+
+    CV_Assert(wshape_[0] > 0 && wshape_[1] > 0);
+    for (int i = 0; i < MAX_CONV_DIMS; i++) {
+        kshape[i] = strides[i] = dilations[i] = 1;
+        pads[i] = pads[i + MAX_CONV_DIMS] = 0;
+        inner[i] = 0;
+        inner[i + MAX_CONV_DIMS] = 1;
+    }
+
+    for (int i = 0; i < nspatialdims; i++) {
+        int j = i + (MAX_CONV_DIMS - nspatialdims);
+        kshape[j] = wshape_[i+2];
+        CV_Assert(kshape[j] > 0);
+
+        strides[j] = strides_.empty() ? 1 : strides_[i];
+        dilations[j] = dilations_.empty() ? 1 : dilations_[i];
+
+        CV_Assert(strides[j] > 0);
+        CV_Assert(dilations[j] > 0);
+
+        int pad0, pad1;
+        getPadding(pads_, i, nspatialdims, autoPad, kshape[j], pad0, pad1);
+        CV_Assert_N(pad0 >= 0, pad1 >= 0);
+        pads[j] = pad0;
+        pads[j + MAX_CONV_DIMS] = pad1;
+
+        int inner0 = (pad0 + strides[j] - 1)/strides[j];
+        int inner1 = (inpshape[i+2] - (kshape[j] - 1)*dilations[j] + pad0)/strides[j];
+        inner1 += inner1*strides[j] - pad0 + (kshape[j] - 1)*dilations[j] < inpshape[i+2];
+        inner1 = std::min(inner1, outshape[i+2]);
+        if (inner0 >= inner1) {
+            inner0 = inner1 = outshape[i+2];
+        }
+        inner[j] = inner0;
+        inner[j + MAX_CONV_DIMS] = inner1;
+    }
+
+    initOfs();
+    if (!depthwise && inpshape.layout == DATA_LAYOUT_BLOCK) {
+        int C0 = inpshape.back();
+        wshape = getWpackShape(wshape_, ngroups, C0);
+
+        CV_Assert(wshape.dims == 5);
+    }
+}
+
+void repackConvWeights(const Mat& weights, Mat& Wpack, int outtype, int ngroups, int C0_)
+{
+    CV_Assert(weights.isContinuous());
+    CV_Assert_N(weights.type() == CV_32F, outtype == CV_32F);
+    CV_Assert(ngroups > 0);
+    CV_Assert((C0_ & (C0_ - 1)) == 0 && C0_ >= 4);
+
+    MatShape wshape = weights.shape();
+    CV_Assert(wshape.dims >= 3);
+
+    int K = wshape[0];
+    CV_Assert(K % ngroups == 0);
+
+    if (!Wpack.isContinuous()) {
+        Wpack.release();
+    }
+    MatShape wpackShape = getWpackShape(weights.shape(), ngroups, C0_);
+    Wpack.create(wpackShape, CV_32F);
+    Wpack.setZero();
+
+    parallel_for_(Range(0, K), [&](const Range& range) {
+        int Cg = wshape[1], Kg = K / ngroups;
+        int ksize = wpackShape[2], Kblk = wpackShape[1], C1Max = wpackShape[3];
+        int C0 = C0_, K0 = C0;
+        const float* wdata = weights.ptr<float>();
+        float* Wpackdata = Wpack.ptr<float>();
+
+        for (int k = range.start; k < range.end; ++k) {
+            int g = k / Kg;
+            int kin = k - g * Kg;
+            int kblk = kin / K0;
+            int k0   = kin & (K0 - 1);
+
+            int c_start = g * Cg;
+            int c00 = c_start & (C0 - 1);
+
+            for (int c = 0; c < Cg; ++c) {
+                int ch = c00 + c;
+                int c1  = ch / C0;
+                int c0  = ch & (C0 - 1);
+
+                const float* wptr = wdata + ((k * Cg + c) * ksize);
+                float* wpackptr = Wpackdata + (((g * Kblk + kblk) * ksize * C1Max + c1) * C0 + c0)*K0 + k0;
+                for (int i = 0; i < ksize; ++i) {
+                    wpackptr[i*(C1Max*C0*K0)] = wptr[i];
+                }
+            }
+        }
+    });
+}
+
+
+void ConvState::initPooling(const MatShape& inpshape_,
+                            const MatShape& outshape_,
+                            const std::vector<int>& kshape_,
+                            const std::vector<int>& strides_,
+                            const std::vector<int>& dilations_,
+                            const std::vector<int>& pads_,
+                            AutoPadding autoPad, bool ceilMode)
+{
+    nspatialdims = int(kshape_.size());
+    CV_Assert(0 < nspatialdims && nspatialdims <= ConvState::MAX_CONV_DIMS);
+    CV_Assert(strides_.empty() || (strides_.size() == size_t(nspatialdims)));
+    CV_Assert(dilations_.empty() || (dilations_.size() == size_t(nspatialdims)));
+    CV_Assert(pads_.empty() || (pads_.size() == size_t(nspatialdims*2)));
+    CV_Assert(inpshape_.layout == DATA_LAYOUT_BLOCK);
+    CV_Assert(inpshape_.dims == nspatialdims + 3);
+
+    int C = inpshape_.C;
+    inpshape = inpshape_;
+    outshape = outshape_;
+    ngroups = C;
+    depthwise = true;
+
+    for (int i = 0; i < MAX_CONV_DIMS; i++) {
+        kshape[i] = strides[i] = dilations[i] = 1;
+        pads[i] = pads[i + MAX_CONV_DIMS] = 0;
+        inner[i] = 0;
+        inner[i + MAX_CONV_DIMS] = 1;
+    }
+
+    for (int i = 0; i < nspatialdims; i++) {
+        int j = i + (MAX_CONV_DIMS - nspatialdims);
+        kshape[j] = kshape_[i];
+        CV_Assert(kshape[j] > 0);
+
+        strides[j] = strides_.empty() ? 1 : strides_[i];
+        dilations[j] = dilations_.empty() ? 1 : dilations_[i];
+
+        CV_Assert(strides[j] > 0);
+        CV_Assert(dilations[j] > 0);
+
+        int pad0, pad1;
+        getPadding(pads_, i, nspatialdims, autoPad, kshape[j], pad0, pad1);
+        CV_Assert_N(pad0 >= 0, pad1 >= 0);
+        pads[j] = pad0;
+        pads[j + MAX_CONV_DIMS] = pad1;
+
+        int inner0 = (pad0 + strides[j] - 1)/strides[j];
+        int inner1 = (inpshape[i+2] - (kshape[j] - 1)*dilations[j] + pad0)/strides[j];
+        inner1 += inner1*strides[j] - pad0 + (kshape[j] - 1)*dilations[j] < inpshape[i+2];
+        inner1 = std::min(inner1, outshape[i+2]);
+        if (inner0 >= inner1) {
+            inner0 = inner1 = outshape[i+2];
+        }
+        inner[j] = inner0;
+        inner[j + MAX_CONV_DIMS] = inner1;
+    }
+    initOfs();
+}
+
+void ConvState::initOfs()
+{
+    CV_Assert(MAX_CONV_DIMS == 3);
+    int sdims = nspatialdims;
+    int KD = kshape[0], KH = kshape[1], KW = kshape[2];
+    int DZ = dilations[0], DY = dilations[1], DX = dilations[2];
+    int Hi = sdims > 1 ? inpshape[sdims] : 1;
+    int Wi = inpshape[sdims + 1];
+    int ksize = KD*KH*KW;
+    int C0 = inpshape.back();
+    coordtab.resize(ksize*MAX_CONV_DIMS);
+    ofstab.resize(ksize);
+    for (int z = 0, k = 0; z < KD; z++) {
+        int dz = z*DZ;
+        for (int y = 0; y < KH; y++) {
+            int dy = y*DY;
+            for (int x = 0; x < KW; x++, k++) {
+                int dx = x*DX;
+                coordtab[k*MAX_CONV_DIMS] = dz;
+                coordtab[k*MAX_CONV_DIMS + 1] = dy;
+                coordtab[k*MAX_CONV_DIMS + 2] = dx;
+                ofstab[k] = ((dz*Hi + dy)*Wi + dx)*C0;
+            }
+        }
+    }
+}
+
+CV__DNN_INLINE_NS_END
+}}

modules/dnn/src/layers/conv2_common.hpp

@@ -0,0 +1,102 @@
+// 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.
+
+#ifndef __OPENCV_DNN_LAYERS_CONV2_COMMON_HPP__
+#define __OPENCV_DNN_LAYERS_CONV2_COMMON_HPP__
+
+#include <opencv2/dnn.hpp>
+#include <array>
+
+namespace cv
+{
+namespace dnn
+{
+CV__DNN_INLINE_NS_BEGIN
+
+// computes shape of the output tensor of convolution
+// (including depth-wise convolution), max pooling or average pooling operations
+MatShape convInferShape(const MatShape& inpshape, const MatShape& wshape,
+                        const std::vector<int>& kernel_shape, int ngroups,
+                        const std::vector<int>& strides,
+                        const std::vector<int>& dilations,
+                        const std::vector<int>& pads,
+                        AutoPadding auto_pad, bool ceil_mode);
+
+enum FastActivation {
+    FAST_ACTIV_NONE=0,
+    FAST_ACTIV_RELU,
+    FAST_ACTIV_LEAKY_RELU,
+    FAST_ACTIV_PRELU,
+    FAST_ACTIV_CLIP
+};
+
+std::string fastActivationToString(FastActivation fastActivation);
+
+typedef void (*ActivationFunc)(const void* input, void* output,
+                               size_t len, const float* params);
+
+struct ConvState
+{
+    enum { MAX_CONV_DIMS = 3 };
+    bool depthwise = true;
+    int ngroups, nspatialdims;
+    int kshape[MAX_CONV_DIMS];
+    int strides[MAX_CONV_DIMS];
+    int dilations[MAX_CONV_DIMS];
+    int pads[MAX_CONV_DIMS*2];
+    MatShape inpshape, outshape;
+    MatShape wshape; // (ngroups, Kblk, ksize, C1Max, C0*K0) in the case of non-depthwise convolution
+    int inner[MAX_CONV_DIMS*2];
+    std::vector<int> coordtab;
+    std::vector<int> ofstab;
+
+    FastActivation fastActivation = FAST_ACTIV_NONE;
+    ActivationFunc activation = nullptr;
+    std::vector<float> activParams;
+
+    std::ostream& dump(std::ostream& strm);
+    bool sameShape(const ConvState& cs) const;
+
+    void initConv(const MatShape& inpShape,
+                  const MatShape& wshape,
+                  const MatShape& outShape,
+                  int ngroups,
+                  const std::vector<int>& strides,
+                  const std::vector<int>& dilations,
+                  const std::vector<int>& pads,
+                  AutoPadding autoPad, bool ceilMode,
+                  FastActivation fastActivation,
+                  const std::vector<float>& activParams);
+
+    // initializes the structure of parameters for 1D/2D/3D
+    // depth-wise convolution, max pooling or average pooling
+    void initPooling(const MatShape& inpshape, const MatShape& outshape,
+                     const std::vector<int>& kernel_shape,
+                     const std::vector<int>& strides,
+                     const std::vector<int>& dilations,
+                     const std::vector<int>& pads,
+                     AutoPadding auto_pad, bool ceil_mode);
+
+    // internal-use method to initialize coordtab and ofstab.
+    // it's called from initConv and initPooling
+    void initOfs();
+};
+
+AutoPadding getAutoPadding(const LayerParams& params);
+
+typedef void (*ConvFunc)(const void* inp, const void* residual, void* out,
+                         const ConvState& cs, const void* weights,
+                         const float* scale, const float* bias);
+
+ConvFunc getConvFunc(int depth, int C0);
+ConvFunc getDepthwiseConvFunc(int depth);
+
+void repackDepthwiseConvWeights(const Mat& weights, Mat& Wpack, int outtype, int C0);
+void repackConvWeights(const Mat& weights, Mat& Wpack, int outtype, int ngroups, int C0);
+
+CV__DNN_INLINE_NS_END
+}
+}
+
+#endif

modules/dnn/src/layers/conv2_layer.cpp

@@ -0,0 +1,412 @@
+// 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 "../precomp.hpp"
+#include "../net_impl.hpp"
+#include "layers_common.hpp"
+#include "conv2_common.hpp"
+#include "opencv2/core/hal/intrin.hpp"
+
+namespace cv
+{
+namespace dnn
+{
+
+/*
+    Convolution layer, as defined in ONNX specification:
+    https://onnx.ai/onnx/operators/onnx__Conv.html
+
+    Opset's 1 to 22 are covered.
+*/
+
+class Conv2LayerImpl : public Conv2Layer
+{
+public:
+    Conv2LayerImpl(const LayerParams& params)
+    {
+        setParamsFrom(params);
+        auto_pad = getAutoPadding(params);
+        ceil_mode = params.get<bool>("ceil_mode", false);
+        strides = params.getVector<int>("stride");
+        dilations = params.getVector<int>("dilation");
+        pads = params.getVector<int>("pad");
+        ngroups = params.get<int>("group", 1);
+        fusedBatchNorm = false;
+        fastActivation = FAST_ACTIV_NONE;
+        addResidual = false;
+    }
+
+    virtual std::ostream& dumpAttrs(std::ostream& strm, int indent) const CV_OVERRIDE
+    {
+        prindent(strm, indent);
+        strm << "group: " << ngroups << ",\n";
+
+        if (!wshape0.empty()) {
+            prindent(strm, indent);
+            strm << "ksize: [";
+            for (int k = 0; k < wshape0.dims; k++)
+                strm << (k > 0 ? ", " : "") << wshape0[k];
+            strm << "],\n";
+        }
+
+        prindent(strm, indent);
+        strm << "stride: [";
+        for (size_t k = 0; k < strides.size(); k++)
+            strm << (k > 0 ? ", " : "") << strides[k];
+        strm << "],\n";
+
+        prindent(strm, indent);
+        strm << "dilation: [";
+        for (size_t k = 0; k < dilations.size(); k++)
+            strm << (k > 0 ? ", " : "") << dilations[k];
+        strm << "],\n";
+
+        prindent(strm, indent);
+        strm << "pad: [";
+        for (size_t k = 0; k < pads.size(); k++)
+            strm << (k > 0 ? ", " : "") << pads[k];
+        strm << "],\n";
+
+        if (fusedBatchNorm) {
+            prindent(strm, indent);
+            strm << "batch_norm: true,\n";
+        }
+
+        if (fastActivation != FAST_ACTIV_NONE || !activ.empty()) {
+            prindent(strm, indent);
+            strm << "fused_activation: " <<
+                (fastActivation != FAST_ACTIV_NONE ? fastActivationToString(fastActivation) :
+                 activ->type) << ",\n";
+        }
+
+        if (addResidual) {
+            prindent(strm, indent);
+            strm << "addResidual: true,\n";
+        }
+
+        if (activ) {
+            prindent(strm, indent);
+            strm << "activation: " << activ->name << ",\n";
+        }
+
+        return strm;
+    }
+
+    int inferType(int inptype0) const
+    {
+        return inptype0;
+    }
+
+    virtual void setWeights(InputArray weights_arr, InputArray bias_arr,
+                            int C0, int accuracy) CV_OVERRIDE
+    {
+        Mat weights_ = weights_arr.getMat();
+        Mat bias_ = bias_arr.getMat();
+        CV_Assert(!weights_.empty());
+        int wtype0 = weights_.type();
+        CV_Assert(wtype0 == CV_32F || wtype0 == CV_16F || wtype0 == CV_16BF);
+        CV_Assert(accuracy == -1 || accuracy == CV_32F);
+        int wtype = accuracy < 0 ? CV_32F : accuracy;
+
+        wshape0 = weights_.shape();
+        bool depthwise = ngroups == wshape0[0] && wshape0[1] == 1;
+
+        if (depthwise) {
+            repackDepthwiseConvWeights(weights_, weights, wtype, C0);
+        } else {
+            repackConvWeights(weights_, weights, wtype, ngroups, C0);
+        }
+
+        if (!bias_.empty()) {
+            CV_Assert(bias_.isContinuous() && bias_.total() == wshape0[0]);
+            bias_.convertTo(bias, CV_32F);
+        }
+    }
+
+    void fuseBatchNormWeights(const BatchNorm2Layer* bn)
+    {
+        Mat bn_scale, bn_bias;
+        bn->getScaleBias(bn_scale, bn_bias);
+
+        CV_Assert(bn_scale.isContinuous() && bn_bias.isContinuous());
+        CV_Assert(bn_scale.type() == CV_32F && bn_bias.type() == CV_32F);
+        CV_Assert(bn_scale.total() == bn_bias.total());
+        int K = (int)bn_scale.total();
+        CV_Assert(bias.empty() || (bias.type() == CV_32F && bias.total() == (size_t)K));
+        const float* bias_data = bias.data ? bias.ptr<float>() : nullptr;
+
+        fusedScale.fit(1, &K, CV_32F);
+        fusedBias.fit(1, &K, CV_32F);
+
+        const float* bn_scale_data = bn_scale.ptr<float>();
+        const float* bn_bias_data = bn_bias.ptr<float>();
+        float* fused_scale_data = fusedScale.ptr<float>();
+        float* fused_bias_data = fusedBias.ptr<float>();
+
+        // (sum(x*w) + bias)*bn_scale + bn_bias => sum(x*w)*fused_scale + fused_bias,
+        // where fused_scale = bn_scale and fused_bias = bias*bn_scale + bn_bias.
+        for (size_t i = 0; i < K; i++) {
+            fused_scale_data[i] = bn_scale_data[i];
+            fused_bias_data[i] = (bias_data ? bn_scale_data[i]*bias_data[i] : 0.f) + bn_bias_data[i];
+        }
+    }
+
+    virtual bool fuseBatchNorm(const Ptr<Layer>& bnlayer) override
+    {
+        BatchNorm2Layer* bn = dynamic_cast<BatchNorm2Layer*>(bnlayer.get());
+        if (fusedBatchNorm || !bn || bn->inputs.size() > 1)
+            return false;
+        fuseBatchNormWeights(bn);
+        fusedBatchNorm = true;
+        return true;
+    }
+
+    virtual bool fuseAddBias(InputArray arr) CV_OVERRIDE
+    {
+        if (inputs.size() > 1 || fusedBatchNorm || addResidual)
+            return false;
+        Mat new_bias = arr.getMat();
+        CV_Assert(new_bias.isContinuous() && new_bias.dims == 1);
+        if (new_bias.type() != CV_32F) {
+            Mat temp;
+            new_bias.convertTo(temp, CV_32F);
+            new_bias = temp;
+            CV_Assert(new_bias.type() == CV_32F);
+        }
+        if (!bias.empty()) {
+            CV_Assert(bias.shape() == new_bias.shape());
+            add(bias, new_bias, bias);
+        } else {
+            new_bias.copyTo(bias);
+        }
+        return true;
+    }
+
+    virtual bool fuseActivation(const Ptr<Layer>& activlayer) override
+    {
+        ActivationLayer* activ_ptr = dynamic_cast<ActivationLayer*>(activlayer.get());
+        if (!activ_ptr || fastActivation != FAST_ACTIV_NONE || !activ.empty())
+            return false;
+        ReLULayer* activRelu = dynamic_cast<ReLULayer*>(activ_ptr);
+        ReLU6Layer* activClip = dynamic_cast<ReLU6Layer*>(activ_ptr);
+        ChannelsPReLULayer* activPRelu = dynamic_cast<ChannelsPReLULayer*>(activ_ptr);
+        if (activRelu) {
+            float alpha = activRelu->negativeSlope;
+            if (alpha == 0.f) {
+                fastActivation = FAST_ACTIV_RELU;
+            } else {
+                fastActivation = FAST_ACTIV_LEAKY_RELU;
+                activParams = {alpha};
+            }
+        } else if (activClip && activClip->minValue == 0.f) {
+            fastActivation = FAST_ACTIV_CLIP;
+            activParams = {activClip->minValue, activClip->maxValue};
+        } else if (activPRelu && activPRelu->blobs.size() == 1) {
+            fastActivation = FAST_ACTIV_PRELU;
+            const Mat& slopes = activPRelu->blobs[0];
+            int slopesType = slopes.type();
+            CV_Assert_N((slopesType == CV_32F || slopesType == CV_16F || slopesType == CV_16BF),
+                        slopes.isContinuous());
+            int nslopes = int(slopes.total());
+            Mat(1, &nslopes, slopesType, (void*)slopes.data).convertTo(activParams, CV_32F);
+        } else {
+            //activ = activlayer;
+            return false;
+        }
+        return true;
+    }
+
+    virtual bool fuseAddResidual(Arg residual) CV_OVERRIDE
+    {
+        if (activ.empty() && fastActivation == FAST_ACTIV_NONE && !addResidual && residual.idx >= 0) {
+            addResidual = true;
+            inputs.push_back(residual);
+            return true;
+        }
+        return false;
+    }
+
+    virtual int64_t getFLOPS(const std::vector<MatShape>& inputs,
+                             const std::vector<MatShape>& outputs) const CV_OVERRIDE
+    {
+        CV_Assert(inputs.size() >= 1);
+        CV_Assert(outputs.size() == 1);
+        // probably, there should be a coefficient in the case of complex reduction functions
+        MatShape inpshape = inputs[0], wshape = inputs.size() > 1 ? inputs[1] : wshape0;
+        int C = inpshape[1]*inpshape.back();
+        size_t ksize = wshape.total();
+        return (int64_t)((inputs[0].total()/C)*ksize/ngroups);
+    }
+
+    virtual void getTypes(const std::vector<MatType>& inptypes,
+                          const int, const int,
+                          std::vector<MatType>& outtypes,
+                          std::vector<MatType>& temptypes) const CV_OVERRIDE
+    {
+        int ninputs = (int)inptypes.size();
+        CV_Assert(ninputs >= 1);
+
+        outtypes.assign(1, inferType(inptypes[0]));
+        temptypes.clear();
+    }
+
+    virtual bool getMemoryShapes(const std::vector<MatShape>& inpshapes,
+                                 const int,
+                                 std::vector<MatShape> &outshapes,
+                                 std::vector<MatShape> &tempshapes) const CV_OVERRIDE
+    {
+        size_t ninputs = inpshapes.size();
+        if (addResidual)
+            ninputs--;
+        CV_Assert(ninputs >= 1);
+
+        MatShape wshape = ninputs > 1 ? inpshapes[1] : wshape0;
+        outshapes.assign(1, convInferShape(inpshapes[0], wshape, emptyKernelShape,
+                                           ngroups, strides, dilations,
+                                           pads, auto_pad, ceil_mode));
+        tempshapes.clear();
+        return true;
+    }
+
+    int getLayouts(const std::vector<DataLayout>& actualInputs,
+                    std::vector<DataLayout>& desiredInputs,
+                    const int requiredOutputs,
+                    std::vector<DataLayout>& outputs) const CV_OVERRIDE
+    {
+        size_t ninputs = actualInputs.size();
+        CV_Assert(ninputs >= 1u && requiredOutputs == 1u);
+        desiredInputs = actualInputs;
+        desiredInputs[0] = DATA_LAYOUT_BLOCK;
+        for (size_t i = 1; i < ninputs; i++)
+            desiredInputs[i] = DATA_LAYOUT_UNKNOWN;
+        outputs.assign(requiredOutputs, DATA_LAYOUT_BLOCK);
+        return getNetImpl(this)->defaultC0;
+    }
+
+    void finalize(InputArrayOfArrays, OutputArrayOfArrays outputs_arr) CV_OVERRIDE
+    {
+    }
+
+    void forward(InputArrayOfArrays input_arrs,
+                 OutputArrayOfArrays output_arrs,
+                 OutputArrayOfArrays temp_arrs) CV_OVERRIDE
+    {
+        auto* netimpl_ = getNetImpl(this);
+        std::vector<Mat>* temp_mats = &temp_arrs.getMatVecRef();
+        temp_mats->resize(2);
+        int ninputs = (int)input_arrs.total();
+        CV_Assert(ninputs >= 1);
+        const Mat& inp = input_arrs.getMat(0);
+        Mat residual;
+        const void* resptr = nullptr;
+        int inptype = inp.type();
+        MatShape inpshape = inp.shape();
+        CV_Assert(inpshape.layout == DATA_LAYOUT_BLOCK);
+        CV_Assert(inp.isContinuous());
+
+        if (addResidual) {
+            residual = input_arrs.getMat(ninputs-1);
+            resptr = residual.data;
+            ninputs--;
+        }
+
+        bool dynamicWeights = false;
+        for (int i = 1; i < ninputs; i++) {
+            if (!netimpl_->isConstArg(inputs[i]))
+                dynamicWeights = true;
+        }
+        if (dynamicWeights || weights.empty()) {
+            setWeights(input_arrs.getMat(1), ninputs > 2 ? input_arrs.getMat(2) : Mat(),
+                       inpshape.back(), netimpl_->accuracy);
+        }
+
+        MatShape outshape = convInferShape(inpshape, wshape0, emptyKernelShape,
+                                           ngroups, strides, dilations,
+                                           pads, auto_pad, ceil_mode);
+        int outtype = inferType(inptype);
+        int C0 = inpshape.back();
+        int outkind = output_arrs.kind();
+        CV_Assert(outkind == _InputArray::STD_VECTOR_MAT ||
+                  outkind == _InputArray::STD_VECTOR_UMAT);
+
+        if (addResidual && (residual.size != outshape || residual.type() != outtype))
+        {
+            CV_Error(Error::StsBadArg,
+                    "residual added after convolution must have the same shape and the "
+                    "same type as the convolution output. If this error occurs, the only "
+                    "solution for now is to edit the model and add 'Expand' and/or 'Cast' "
+                    "operators to make the residual tensor match the convolution shape and type");
+        }
+
+        int nspatialdims = inpshape.dims - 3;
+        CV_Assert(wshape0.dims == nspatialdims+2);
+
+        if (inpshape != prevInpshape) {
+            cs.initConv(inpshape, wshape0, outshape, ngroups,
+                        strides, dilations, pads, auto_pad, ceil_mode,
+                        fastActivation, activParams);
+            prevInpshape = inpshape;
+        }
+
+        const float* scale_data = nullptr;
+        const float* bias_data = bias.ptr<float>();
+
+        if (fusedBatchNorm) {
+            scale_data = fusedScale.ptr<float>();
+            bias_data = fusedBias.ptr<float>();
+        }
+
+        std::vector<Mat>* outs = nullptr;
+        std::vector<UMat>* uouts = nullptr;
+        Mat out;
+
+        if (outkind == _InputArray::STD_VECTOR_MAT) {
+            outs = &output_arrs.getMatVecRef();
+            outs->resize(1);
+            outs->at(0).fit(outshape, outtype);
+            out = outs->at(0);
+        } else {
+            uouts = &output_arrs.getUMatVecRef();
+            uouts->resize(1);
+            uouts->at(0).fit(outshape, outtype);
+            out.fit(outshape, outtype);
+        }
+
+        const void* inptr = inp.data;
+        void* outptr = out.data;
+        const void* wptr = weights.data;
+
+        ConvFunc func = cs.depthwise ? getDepthwiseConvFunc(inptype) : getConvFunc(inptype, C0);
+        CV_Assert(func != nullptr);
+        func(inptr, resptr, outptr, cs, wptr, scale_data, bias_data);
+
+        if (uouts) {
+            out.copyTo(uouts->at(0));
+        }
+
+        if (dynamicWeights) {
+            // to keep memory footprint low in the case of
+            // very rare situation of dynamic convolution weights,
+            // we release temporarily allocated and reordered copy of the weights
+            weights.release();
+        }
+    }
+
+    std::vector<int> emptyKernelShape;
+    Ptr<Layer> activ, batchNorm;
+    Mat weights, bias, fusedScale, fusedBias;
+    MatShape wshape0, prevInpshape;
+    ConvState cs;
+    bool fusedBatchNorm;
+    FastActivation fastActivation;
+    std::vector<float> activParams;
+    bool addResidual;
+};
+
+Ptr<Conv2Layer> Conv2Layer::create(const LayerParams& params)
+{
+    return Ptr<Conv2Layer>(new Conv2LayerImpl(params));
+}
+
+}}

modules/dnn/src/layers/cpu_kernels/conv2_depthwise.cpp

@@ -0,0 +1,105 @@
+// 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 "../../precomp.hpp"
+#include "../../net_impl.hpp"
+#include "../conv2_common.hpp"
+#include "conv2_depthwise.simd.hpp"
+#include "layers/cpu_kernels/conv2_depthwise.simd_declarations.hpp"
+
+namespace cv { namespace dnn {
+CV__DNN_INLINE_NS_BEGIN
+
+ConvFunc getDepthwiseConvFunc(int depth)
+{
+    CV_CPU_DISPATCH(getDepthwiseConvFunc_, (depth), CV_CPU_DISPATCH_MODES_ALL);
+}
+
+template <typename InpT, typename OutT>
+static void repackDepthwiseWeightsBlock(const InpT* inpw, OutT* outw,
+                                        int ksize, int C0, int currC0)
+{
+    for (int xy = 0; xy < ksize; xy++, inpw++, outw += C0) {
+        for (int c0 = 0; c0 < currC0; c0++) {
+            outw[c0] = OutT(inpw[ksize*c0]);
+        }
+    }
+}
+
+// C x 1 x ksize ... => C1 x ksize x C0
+void repackDepthwiseConvWeights(const Mat& weights, Mat& wpack, int outtype_, int C0_)
+{
+    int inptype_ = weights.type();
+    MatShape wshape = weights.shape();
+    CV_Assert(wshape.dims >= 3);
+    CV_Assert(inptype_ == CV_32F || inptype_ == CV_16F || inptype_ == CV_16BF);
+    CV_Assert(outtype_ == CV_32F || outtype_ == CV_16F || outtype_ == CV_16BF);
+    int ksize_ = 1;
+    for (int i = 2; i < wshape.dims; i++)
+        ksize_ *= wshape[i];
+
+    int C1_ = (wshape[0] + C0_ - 1)/C0_;
+
+    if (!wpack.isContinuous())
+        wpack.release();
+
+    wpack.create(MatShape({C1_, ksize_, C0_}, DATA_LAYOUT_UNKNOWN), outtype_);
+
+    parallel_for_(Range(0, C1_), [&](const Range& r) {
+        int inptype = inptype_, outtype = outtype_;
+        size_t inpEsz = CV_ELEM_SIZE(inptype);
+        size_t outEsz = CV_ELEM_SIZE(outtype);
+        int C = wshape[0], C0 = C0_;
+        int ksize = ksize_;
+
+        for (int c1 = r.start; c1 < r.end; c1++) {
+            const uint8_t* inpw_ = (const uint8_t*)weights.data + c1*ksize*C0*inpEsz;
+            uint8_t* outw_ = (uint8_t*)wpack.data + c1*ksize*C0*outEsz;
+            int currC0 = std::min(C - c1*C0, C0);
+            if (currC0 < C0)
+                memset(outw_, 0, ksize*C0*outEsz);
+
+            if (inptype == CV_32F) {
+                const float* inpw = (const float*)inpw_;
+                if (outtype == CV_32F) {
+                    float* outw = (float*)outw_;
+                    repackDepthwiseWeightsBlock(inpw, outw, ksize, C0, currC0);
+                } else if (outtype == CV_16F) {
+                    hfloat* outw = (hfloat*)outw_;
+                    repackDepthwiseWeightsBlock(inpw, outw, ksize, C0, currC0);
+                } else if (outtype == CV_16BF) {
+                    bfloat* outw = (bfloat*)outw_;
+                    repackDepthwiseWeightsBlock(inpw, outw, ksize, C0, currC0);
+                }
+            } else if (inptype == CV_16F) {
+                const hfloat* inpw = (const hfloat*)inpw_;
+                if (outtype == CV_32F) {
+                    float* outw = (float*)outw_;
+                    repackDepthwiseWeightsBlock(inpw, outw, ksize, C0, currC0);
+                } else if (outtype == CV_16F) {
+                    hfloat* outw = (hfloat*)outw_;
+                    repackDepthwiseWeightsBlock(inpw, outw, ksize, C0, currC0);
+                } else if (outtype == CV_16BF) {
+                    bfloat* outw = (bfloat*)outw_;
+                    repackDepthwiseWeightsBlock(inpw, outw, ksize, C0, currC0);
+                }
+            } else if (inptype == CV_16BF) {
+                const bfloat* inpw = (const bfloat*)inpw_;
+                if (outtype == CV_32F) {
+                    float* outw = (float*)outw_;
+                    repackDepthwiseWeightsBlock(inpw, outw, ksize, C0, currC0);
+                } else if (outtype == CV_16F) {
+                    hfloat* outw = (hfloat*)outw_;
+                    repackDepthwiseWeightsBlock(inpw, outw, ksize, C0, currC0);
+                } else if (outtype == CV_16BF) {
+                    bfloat* outw = (bfloat*)outw_;
+                    repackDepthwiseWeightsBlock(inpw, outw, ksize, C0, currC0);
+                }
+            }
+        }
+    });
+}
+
+CV__DNN_INLINE_NS_END
+}}

modules/dnn/src/layers/cpu_kernels/conv2_depthwise.simd.hpp

@@ -0,0 +1,379 @@
+// 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 "../conv2_common.hpp"
+#include "opencv2/core/hal/intrin.hpp"
+
+// === dispatched calls (implemented here)
+
+namespace cv {
+namespace dnn {
+CV_CPU_OPTIMIZATION_NAMESPACE_BEGIN
+
+cv::dnn::ConvFunc getDepthwiseConvFunc_(int depth);
+
+CV_CPU_OPTIMIZATION_NAMESPACE_END
+}} // cv::dnn::
+
+// === implementation
+
+#ifndef CV_CPU_OPTIMIZATION_DECLARATIONS_ONLY
+
+namespace cv {
+namespace dnn {
+CV_CPU_OPTIMIZATION_NAMESPACE_BEGIN
+
+//
+// [TODO] add special branch for 3x3 depthwise convolution
+//
+static void depthwiseConv32f(const void* inp__, const void* residual__,
+                             void* out__, const ConvState& cs,
+                             const void* weights__, const float* scale__,
+                             const float* bias__)
+{
+    int NC1 = cs.inpshape[0]*cs.inpshape[1];
+
+    CV_Assert(cs.inpshape.layout == DATA_LAYOUT_BLOCK);
+    CV_Assert(cs.outshape.layout == DATA_LAYOUT_BLOCK);
+    CV_Assert(cs.inpshape.dims == cs.outshape.dims);
+
+    parallel_for_(Range(0, NC1), [&](const Range& range)
+    {
+        constexpr int MAX_CONV_DIMS = ConvState::MAX_CONV_DIMS;
+        constexpr int C0 = 8;
+
+        CV_Assert(cs.nspatialdims <= MAX_CONV_DIMS && MAX_CONV_DIMS == 3);
+        CV_Assert(C0 == cs.inpshape.back());
+
+        int sdims = cs.nspatialdims;
+        int C = cs.inpshape.C;
+        int C1 = cs.inpshape[1];
+        int Di = sdims > 2 ? cs.inpshape[sdims - 1] : 1;
+        int Hi = sdims > 1 ? cs.inpshape[sdims] : 1;
+        int Wi = cs.inpshape[sdims + 1];
+        int D = sdims > 2 ? cs.outshape[sdims - 1] : 1;
+        int H = sdims > 1 ? cs.outshape[sdims] : 1;
+        int W = cs.outshape[sdims + 1];
+        int iplanesize = Di*Hi*Wi*C0;
+        int planesize = D*H*W*C0;
+        int SZ = cs.strides[0], SY = cs.strides[1], SX = cs.strides[2];
+        int padZ0 = cs.pads[0], padY0 = cs.pads[1], padX0 = cs.pads[2];
+        int inner_z0 = cs.inner[0], inner_z1 = cs.inner[MAX_CONV_DIMS];
+        int inner_y0 = cs.inner[1], inner_y1 = cs.inner[MAX_CONV_DIMS + 1];
+        int inner_x0 = cs.inner[2], inner_x1 = cs.inner[MAX_CONV_DIMS + 2];
+        int ksize = (int)cs.ofstab.size();
+        const int* zyxtab = cs.coordtab.data();
+        const int* ofstab = cs.ofstab.data();
+        const float* scale_ = scale__;
+        const float* bias_ = bias__;
+
+        const float* inp = (const float*)inp__ + range.start*iplanesize;
+        float* out = (float*)out__ + range.start*planesize;
+        const float* residual = residual__ ? (const float*)residual__ + range.start*planesize : nullptr;
+
+        FastActivation fastActivation = cs.fastActivation;
+        const float* activParams = cs.activParams.data();
+        ActivationFunc activation = cs.activation;
+        float maxval = FLT_MAX, defaultAlpha = 0.f;
+        float scalebuf[C0], biasbuf[C0], alphabuf[C0];
+        if (fastActivation == FAST_ACTIV_CLIP) {
+            CV_Assert(cs.activParams.size() == 2u);
+            maxval = activParams[1];
+        } else if (fastActivation == FAST_ACTIV_RELU) {
+            CV_Assert(!activParams);
+        } else if (fastActivation == FAST_ACTIV_LEAKY_RELU) {
+            CV_Assert(cs.activParams.size() == 1u);
+            defaultAlpha = activParams[0];
+        } else if (fastActivation == FAST_ACTIV_PRELU) {
+            CV_Assert(cs.activParams.size() == size_t(C));
+        } else {
+            CV_Assert(fastActivation == FAST_ACTIV_NONE);
+            defaultAlpha = 1.f;
+        }
+
+    #if CV_SIMD || CV_SIMD_SCALABLE
+        v_float32 v_maxval = vx_setall_f32(maxval);
+        v_float32 z = vx_setzero_f32();
+        const int nlanes = VTraits<v_float32>::vlanes();
+        CV_Assert(C0 == nlanes || C0 == nlanes*2 || C0 % (nlanes*4) == 0);
+    #endif
+
+        for (int nc1 = range.start; nc1 < range.end; nc1++, inp += iplanesize) {
+            int n = nc1 / C1;
+            int c_base = (nc1 - n*C1)*C0;
+            int c_count = std::min(C0, C - c_base);
+            const float* weights = (const float*)weights__ + (c_base/C0)*ksize*C0;
+
+            {
+                int c = 0;
+                for (; c < c_count; c++) {
+                    scalebuf[c] = scale_ ? scale_[c_base + c] : 1.f;
+                    biasbuf[c] = bias_ ? bias_[c_base + c] : 0.f;
+                    alphabuf[c] = fastActivation == FAST_ACTIV_PRELU ? activParams[c_base + c] : defaultAlpha;
+                }
+                for (; c < C0; c++) {
+                    scalebuf[c] = 0.f;
+                    biasbuf[c] = 0.f;
+                    alphabuf[c] = 0.f;
+                }
+            }
+
+            for (int z0 = 0; z0 < D; z0++) {
+                int zi_ = z0*SZ - padZ0;
+                for (int y0 = 0; y0 < H; y0++, out += W*C0,
+                     residual += (residual ? W*C0 : 0)) {
+                    //int64_t x0 = 0, x1 = W;
+                    int x0 = 0;
+                    int x1 = z0 >= inner_z0 && z0 < inner_z1 &&
+                        y0 >= inner_y0 && y0 < inner_y1 ? inner_x0 : W;
+                    int yi_ = y0*SY - padY0;
+
+                #if !(CV_SIMD || CV_SIMD_SCALABLE)
+                    memset(out, 0, W*C0*sizeof(out[0]));
+                #endif
+
+                    for(;;) {
+                    #if CV_SIMD || CV_SIMD_SCALABLE
+                        if (nlanes == C0) {
+                            v_float32 sc0 = vx_load(scalebuf), b0 = vx_load(biasbuf);
+                            v_float32 alpha0 = vx_load(alphabuf);
+                            for (; x0 < x1; x0++) {
+                                int xi_ = x0*SX - padX0;
+                                v_float32 s0 = z;
+                                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];
+                                    v_float32 v0, w0;
+                                    if ((unsigned)zi >= (unsigned)Di ||
+                                        (unsigned)yi >= (unsigned)Hi ||
+                                        (unsigned)xi >= (unsigned)Wi)
+                                        continue;
+                                    v0 = vx_load(inp + ((zi*Hi + yi)*Wi + xi)*C0);
+                                    w0 = vx_load(weights + k*C0);
+                                    s0 = v_fma(v0, w0, s0);
+                                }
+                                s0 = v_fma(s0, sc0, b0);
+                                if (residual)
+                                    s0 = v_add(s0, vx_load(residual + x0*C0));
+                                s0 = v_min(v_select(v_ge(s0, z), s0, v_mul(s0, alpha0)), v_maxval);
+                                vx_store(out + x0*C0, s0);
+                            }
+                        } else {
+                            for (; x0 < x1; x0++) {
+                                int xi_ = x0*SX - padX0;
+                                for (int c = 0; c < C0; c += nlanes*2) {
+                                    v_float32 s0 = z, s1 = z;
+                                    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];
+                                        v_float32 v0, v1, w0, w1;
+                                        if ((unsigned)zi >= (unsigned)Di ||
+                                            (unsigned)yi >= (unsigned)Hi ||
+                                            (unsigned)xi >= (unsigned)Wi)
+                                            continue;
+                                        int ofs_k = ((zi*Hi + yi)*Wi + xi)*C0 + c;
+                                        int ofs_w = k*C0;
+                                        v0 = vx_load(inp + ofs_k);
+                                        v1 = vx_load(inp + ofs_k + nlanes);
+                                        w0 = vx_load(weights + ofs_w);
+                                        w1 = vx_load(weights + ofs_w + nlanes);
+                                        s0 = v_fma(v0, w0, s0);
+                                        s1 = v_fma(v1, w1, s1);
+                                    }
+                                    s0 = v_fma(s0, vx_load(scalebuf + c), vx_load(biasbuf + c));
+                                    s1 = v_fma(s1, vx_load(scalebuf + c + nlanes), vx_load(biasbuf + c + nlanes));
+                                    if (residual) {
+                                        s0 = v_add(s0, vx_load(residual + x0*C0 + c));
+                                        s1 = v_add(s1, vx_load(residual + x0*C0 + c + nlanes));
+                                    }
+                                    v_float32 alpha0 = vx_load(alphabuf + c);
+                                    v_float32 alpha1 = vx_load(alphabuf + c + nlanes);
+
+                                    s0 = v_min(v_select(v_ge(s0, z), s0, v_mul(s0, alpha0)), v_maxval);
+                                    s1 = v_min(v_select(v_ge(s1, z), s1, v_mul(s1, alpha1)), v_maxval);
+                                    vx_store(out + x0*C0 + c, s0);
+                                    vx_store(out + x0*C0 + c + nlanes, s1);
+                                }
+                            }
+                        }
+                    #else
+                        for (; x0 < x1; x0++) {
+                            int xi_ = x0*SX - padX0;
+                            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* inptr = inp + ((zi*Hi + yi)*Wi + xi)*C0;
+                                for (int c = 0; c < C0; c++)
+                                    out[x0*C0 + c] += inptr[c]*weights[k*C0 + c];
+                            }
+                        }
+                    #endif
+
+                        if (x0 == W)
+                            break;
+                        x1 = inner_x1;
+
+                    #if CV_SIMD || CV_SIMD_SCALABLE
+                        if (nlanes == C0) {
+                            v_float32 sc0 = vx_load(scalebuf), b0 = vx_load(biasbuf), alpha0 = vx_load(alphabuf);
+                            for (; x0 < x1; x0++) {
+                                int xi_ = x0*SX - padX0;
+                                const float* inp_xi = inp + ((Hi*zi_ + yi_)*Wi + xi_)*C0;
+
+                                v_float32 s0 = z;
+                                for (int k = 0; k < ksize; k++) {
+                                    v_float32 v0 = vx_load(inp_xi + ofstab[k]);
+                                    v_float32 w0 = vx_load(weights + k*C0);
+                                    s0 = v_fma(v0, w0, s0);
+                                }
+                                s0 = v_fma(s0, sc0, b0);
+                                if (residual)
+                                    s0 = v_add(s0, vx_load(residual + x0*C0));
+                                s0 = v_min(v_select(v_ge(s0, z), s0, v_mul(s0, alpha0)), v_maxval);
+                                vx_store(out + x0*C0, s0);
+                            }
+                        } else if (nlanes*2 == C0) {
+                            v_float32 sc0 = vx_load(scalebuf), sc1 = vx_load(scalebuf + nlanes);
+                            v_float32 b0 = vx_load(biasbuf), b1 = vx_load(biasbuf + nlanes);
+                            v_float32 alpha0 = vx_load(alphabuf), alpha1 = vx_load(alphabuf + nlanes);
+                            for (; x0 < x1; x0++) {
+                                int xi_ = x0*SX - padX0;
+                                const float* inp_xi = inp + ((Hi*zi_ + yi_)*Wi + xi_)*C0;
+
+                                v_float32 s0 = z, s1 = z;
+                                for (int k = 0; k < ksize; k++) {
+                                    int ofs_k = ofstab[k], ofs_w = k*C0;
+                                    v_float32 v0 = vx_load(inp_xi + ofs_k);
+                                    v_float32 v1 = vx_load(inp_xi + ofs_k + nlanes);
+                                    v_float32 w0 = vx_load(weights + ofs_w);
+                                    v_float32 w1 = vx_load(weights + ofs_w + nlanes);
+                                    s0 = v_fma(v0, w0, s0);
+                                    s1 = v_fma(v1, w1, s1);
+                                }
+                                s0 = v_fma(s0, sc0, b0);
+                                s1 = v_fma(s1, sc1, b1);
+                                if (residual) {
+                                    s0 = v_add(s0, vx_load(residual + x0*C0));
+                                    s1 = v_add(s1, vx_load(residual + x0*C0 + nlanes));
+                                }
+                                s0 = v_min(v_select(v_ge(s0, z), s0, v_mul(s0, alpha0)), v_maxval);
+                                s1 = v_min(v_select(v_ge(s1, z), s1, v_mul(s1, alpha1)), v_maxval);
+                                vx_store(out + x0*C0, s0);
+                                vx_store(out + x0*C0 + nlanes, s1);
+                            }
+                        } else {
+                            for (; x0 < x1; x0++) {
+                                int xi_ = x0*SX - padX0;
+                                for (int c = 0; c < C0; c += nlanes*4) {
+                                    const float* inp_xi = inp + ((Hi*zi_ + yi_)*Wi + xi_)*C0 + c;
+                                    v_float32 s0 = z, s1 = z, s2 = z, s3 = z;
+                                    for (int k = 0; k < ksize; k++) {
+                                        int ofs_k = ofstab[k], ofs_w = k*C0 + c;
+                                        v_float32 v0 = vx_load(inp_xi + ofs_k);
+                                        v_float32 v1 = vx_load(inp_xi + ofs_k + nlanes);
+                                        v_float32 v2 = vx_load(inp_xi + ofs_k + nlanes*2);
+                                        v_float32 v3 = vx_load(inp_xi + ofs_k + nlanes*3);
+                                        v_float32 w0 = vx_load(weights + ofs_w);
+                                        v_float32 w1 = vx_load(weights + ofs_w + nlanes);
+                                        v_float32 w2 = vx_load(weights + ofs_w + nlanes*2);
+                                        v_float32 w3 = vx_load(weights + ofs_w + nlanes*3);
+                                        s0 = v_fma(v0, w0, s0);
+                                        s1 = v_fma(v1, w1, s1);
+                                        s2 = v_fma(v2, w2, s2);
+                                        s3 = v_fma(v3, w3, s3);
+                                    }
+                                    s0 = v_fma(s0, vx_load(scalebuf + c),
+                                               vx_load(biasbuf + c));
+                                    s1 = v_fma(s1, vx_load(scalebuf + c + nlanes),
+                                               vx_load(biasbuf + c + nlanes));
+                                    s2 = v_fma(s2, vx_load(scalebuf + c + nlanes*2),
+                                               vx_load(biasbuf + c + nlanes*2));
+                                    s3 = v_fma(s3, vx_load(scalebuf + c + nlanes*3),
+                                               vx_load(biasbuf + c + nlanes*3));
+
+                                    if (residual) {
+                                        s0 = v_add(s0, vx_load(residual + x0*C0 + c));
+                                        s1 = v_add(s1, vx_load(residual + x0*C0 + c + nlanes));
+                                        s2 = v_add(s2, vx_load(residual + x0*C0 + c + nlanes*2));
+                                        s3 = v_add(s3, vx_load(residual + x0*C0 + c + nlanes*3));
+                                    }
+
+                                    v_float32 alpha0 = vx_load(alphabuf + c);
+                                    v_float32 alpha1 = vx_load(alphabuf + c + nlanes);
+                                    v_float32 alpha2 = vx_load(alphabuf + c + nlanes*2);
+                                    v_float32 alpha3 = vx_load(alphabuf + c + nlanes*3);
+
+                                    s0 = v_min(v_select(v_ge(s0, z), s0, v_mul(s0, alpha0)), v_maxval);
+                                    s1 = v_min(v_select(v_ge(s1, z), s1, v_mul(s1, alpha1)), v_maxval);
+                                    s2 = v_min(v_select(v_ge(s2, z), s2, v_mul(s2, alpha2)), v_maxval);
+                                    s3 = v_min(v_select(v_ge(s3, z), s3, v_mul(s3, alpha3)), v_maxval);
+                                    vx_store(out + x0*C0 + c, s0);
+                                    vx_store(out + x0*C0 + c + nlanes, s1);
+                                    vx_store(out + x0*C0 + c + nlanes*2, s2);
+                                    vx_store(out + x0*C0 + c + nlanes*3, s3);
+                                }
+                            }
+                        }
+                    #else
+                        for (; x0 < x1; x0++) {
+                            int xi_ = x0*SX - padX0;
+                            const float* inp_xi = inp + ((Hi*zi_ + yi_)*Wi + xi_)*C0;
+
+                            for (int k = 0; k < ksize; k++) {
+                                const float* inptr = inp_xi + ofstab[k];
+                                for (int c = 0; c < C0; c++)
+                                    out[x0*C0 + c] += inptr[c]*weights[k*C0 + c];
+                            }
+                        }
+                    #endif
+
+                        x1 = W;
+                    }
+
+                #if !(CV_SIMD || CV_SIMD_SCALABLE)
+                    if (residual) {
+                        for (int x = 0; x < W*C0; x += C0) {
+                            for (int c = 0; c < C0; c++) {
+                                float v = out[x + c]*scalebuf[c] + biasbuf[c] + residual[x + c];
+                                v = std::min(v*(v >= 0 ? 1.f : alphabuf[c]), maxval);
+                                out[x + c] = v;
+                            }
+                        }
+                    } else {
+                        for (int x = 0; x < W*C0; x += C0) {
+                            for (int c = 0; c < C0; c++) {
+                                float v = out[x + c]*scalebuf[c] + biasbuf[c];
+                                v = std::min(v*(v >= 0 ? 1.f : alphabuf[c]), maxval);
+                                out[x + c] = v;
+                            }
+                        }
+                    }
+                #endif
+                }
+
+                if (activation) {
+                    activation(out - planesize, out - planesize, planesize, activParams);
+                }
+            }
+        }
+    });
+}
+
+ConvFunc getDepthwiseConvFunc_(int depth)
+{
+    ConvFunc func = depth == CV_32F ? depthwiseConv32f : nullptr;
+    return func;
+}
+
+CV_CPU_OPTIMIZATION_NAMESPACE_END
+}}
+
+#endif

modules/dnn/src/layers/cpu_kernels/conv2_kernels.cpp

@@ -0,0 +1,20 @@
+// 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 "../../precomp.hpp"
+#include "../../net_impl.hpp"
+#include "../conv2_common.hpp"
+#include "conv2_kernels.simd.hpp"
+#include "layers/cpu_kernels/conv2_kernels.simd_declarations.hpp"
+
+namespace cv { namespace dnn {
+CV__DNN_INLINE_NS_BEGIN
+
+ConvFunc getConvFunc(int depth, int C0)
+{
+    CV_CPU_DISPATCH(getConvFunc_, (depth, C0), CV_CPU_DISPATCH_MODES_ALL);
+}
+
+CV__DNN_INLINE_NS_END
+}}

modules/dnn/src/layers/cpu_kernels/conv2_kernels.simd.hpp

@@ -0,0 +1,782 @@
+// 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 "../conv2_common.hpp"
+#include "opencv2/core/hal/intrin.hpp"
+
+// === dispatched calls (implemented here)
+
+namespace cv {
+namespace dnn {
+CV_CPU_OPTIMIZATION_NAMESPACE_BEGIN
+
+cv::dnn::ConvFunc getConvFunc_(int depth, int C0);
+
+CV_CPU_OPTIMIZATION_NAMESPACE_END
+}} // cv::dnn::
+
+// === implementation
+
+#ifndef CV_CPU_OPTIMIZATION_DECLARATIONS_ONLY
+
+namespace cv {
+namespace dnn {
+CV_CPU_OPTIMIZATION_NAMESPACE_BEGIN
+
+#define CONV_ENABLE_SIMD 1
+
+#undef CONV_ADD_NO_RESIDUAL2
+#define CONV_ADD_NO_RESIDUAL2(idx0, idx1) /* empty */
+
+#if (defined CV_NEON_AARCH64) && CV_NEON_AARCH64
+
+/////////////////////////// AARH64-optimized implementation /////////////////////////////
+
+#undef CONV_INIT_SUMS
+#define CONV_INIT_SUMS() \
+    float32x4_t zz = vdupq_n_f32(0.f); \
+    float32x4_t s0l = zz, s0h = zz, s1l = zz, s1h = zz; \
+    float32x4_t s2l = zz, s2h = zz, s3l = zz, s3h = zz; \
+    float32x4_t s4l = zz, s4h = zz, s5l = zz, s5h = zz; \
+    float32x4_t s6l = zz, s6h = zz, s7l = zz, s7h = zz; \
+    float32x4_t s8l = zz, s8h = zz, s9l = zz, s9h = zz
+
+#undef CONV_UPDATE_BLOCK
+#define CONV_UPDATE_BLOCK(w_ofs, lane) \
+    wl = vld1q_f32(wptr + w_ofs*K0 + 0); \
+    wh = vld1q_f32(wptr + w_ofs*K0 + 4); \
+    s0l = vfmaq_laneq_f32(s0l, wl, x0, lane); \
+    s0h = vfmaq_laneq_f32(s0h, wh, x0, lane); \
+    s1l = vfmaq_laneq_f32(s1l, wl, x1, lane); \
+    s1h = vfmaq_laneq_f32(s1h, wh, x1, lane); \
+    s2l = vfmaq_laneq_f32(s2l, wl, x2, lane); \
+    s2h = vfmaq_laneq_f32(s2h, wh, x2, lane); \
+    s3l = vfmaq_laneq_f32(s3l, wl, x3, lane); \
+    s3h = vfmaq_laneq_f32(s3h, wh, x3, lane); \
+    s4l = vfmaq_laneq_f32(s4l, wl, x4, lane); \
+    s4h = vfmaq_laneq_f32(s4h, wh, x4, lane); \
+    s5l = vfmaq_laneq_f32(s5l, wl, x5, lane); \
+    s5h = vfmaq_laneq_f32(s5h, wh, x5, lane); \
+    s6l = vfmaq_laneq_f32(s6l, wl, x6, lane); \
+    s6h = vfmaq_laneq_f32(s6h, wh, x6, lane); \
+    s7l = vfmaq_laneq_f32(s7l, wl, x7, lane); \
+    s7h = vfmaq_laneq_f32(s7h, wh, x7, lane); \
+    s8l = vfmaq_laneq_f32(s8l, wl, x8, lane); \
+    s8h = vfmaq_laneq_f32(s8h, wh, x8, lane); \
+    s9l = vfmaq_laneq_f32(s9l, wl, x9, lane); \
+    s9h = vfmaq_laneq_f32(s9h, wh, x9, lane)
+
+#undef CONV_UPDATE_LOOP_BODY
+#define CONV_UPDATE_LOOP_BODY() \
+    float32x4_t x0, x1, x2, x3, x4, x5, x6, x7, x8, x9; \
+    float32x4_t wl, wh; \
+    \
+    x0 = vld1q_f32(inptr[0]); \
+    x1 = vld1q_f32(inptr[1]); \
+    x2 = vld1q_f32(inptr[2]); \
+    x3 = vld1q_f32(inptr[3]); \
+    x4 = vld1q_f32(inptr[4]); \
+    x5 = vld1q_f32(inptr[5]); \
+    x6 = vld1q_f32(inptr[6]); \
+    x7 = vld1q_f32(inptr[7]); \
+    x8 = vld1q_f32(inptr[8]); \
+    x9 = vld1q_f32(inptr[9]); \
+    \
+    CONV_UPDATE_BLOCK(0, 0); \
+    CONV_UPDATE_BLOCK(1, 1); \
+    CONV_UPDATE_BLOCK(2, 2); \
+    CONV_UPDATE_BLOCK(3, 3); \
+    \
+    x0 = vld1q_f32(inptr[0] + 4); \
+    x1 = vld1q_f32(inptr[1] + 4); \
+    x2 = vld1q_f32(inptr[2] + 4); \
+    x3 = vld1q_f32(inptr[3] + 4); \
+    x4 = vld1q_f32(inptr[4] + 4); \
+    x5 = vld1q_f32(inptr[5] + 4); \
+    x6 = vld1q_f32(inptr[6] + 4); \
+    x7 = vld1q_f32(inptr[7] + 4); \
+    x8 = vld1q_f32(inptr[8] + 4); \
+    x9 = vld1q_f32(inptr[9] + 4); \
+    \
+    inptr[0] += inpstep[0]; inptr[1] += inpstep[1]; \
+    inptr[2] += inpstep[2]; inptr[3] += inpstep[3]; \
+    inptr[4] += inpstep[4]; inptr[5] += inpstep[5]; \
+    inptr[6] += inpstep[6]; inptr[7] += inpstep[7]; \
+    inptr[8] += inpstep[8]; inptr[9] += inpstep[9]; \
+    \
+    CONV_UPDATE_BLOCK(4, 0); \
+    CONV_UPDATE_BLOCK(5, 1); \
+    CONV_UPDATE_BLOCK(6, 2); \
+    CONV_UPDATE_BLOCK(7, 3)
+
+#undef CONV_START_FINALIZE_OUT
+#define CONV_START_FINALIZE_OUT() \
+    float32x4_t vscale_lo = vld1q_f32(scalebuf), vscale_hi = vld1q_f32(scalebuf + 4); \
+    float32x4_t vbias_lo = vld1q_f32(biasbuf), vbias_hi = vld1q_f32(biasbuf + 4); \
+    float32x4_t valpha_lo = vld1q_f32(alphabuf), valpha_hi = vld1q_f32(alphabuf + 4); \
+    float32x4_t vmaxval = vdupq_n_f32(maxval)
+
+#define CONV_ADD_RESIDUAL2(idx0, idx1) \
+    s##idx0##l = vaddq_f32(s##idx0##l, vld1q_f32(tmpbuf + idx0*K0)); \
+    s##idx0##h = vaddq_f32(s##idx0##h, vld1q_f32(tmpbuf + idx0*K0 + 4)); \
+    s##idx1##l = vaddq_f32(s##idx1##l, vld1q_f32(tmpbuf + idx1*K0)); \
+    s##idx1##h = vaddq_f32(s##idx1##h, vld1q_f32(tmpbuf + idx1*K0 + 4))
+
+#undef CONV_FINALIZE_OUT2
+#define CONV_FINALIZE_OUT2(idx0, idx1, add_residual2) \
+    s##idx0##l = vfmaq_f32(vbias_lo, s##idx0##l, vscale_lo); \
+    s##idx0##h = vfmaq_f32(vbias_hi, s##idx0##h, vscale_hi); \
+    s##idx1##l = vfmaq_f32(vbias_lo, s##idx1##l, vscale_lo); \
+    s##idx1##h = vfmaq_f32(vbias_hi, s##idx1##h, vscale_hi); \
+    add_residual2(idx0, idx1); \
+    s##idx0##l = vbslq_f32(vcgeq_f32(s##idx0##l, zz), s##idx0##l, vmulq_f32(s##idx0##l, valpha_lo)); \
+    s##idx0##h = vbslq_f32(vcgeq_f32(s##idx0##h, zz), s##idx0##h, vmulq_f32(s##idx0##h, valpha_hi)); \
+    s##idx1##l = vbslq_f32(vcgeq_f32(s##idx1##l, zz), s##idx1##l, vmulq_f32(s##idx1##l, valpha_lo)); \
+    s##idx1##h = vbslq_f32(vcgeq_f32(s##idx1##h, zz), s##idx1##h, vmulq_f32(s##idx1##h, valpha_hi)); \
+    s##idx0##l = vminq_f32(s##idx0##l, vmaxval); \
+    s##idx0##h = vminq_f32(s##idx0##h, vmaxval); \
+    s##idx1##l = vminq_f32(s##idx1##l, vmaxval); \
+    s##idx1##h = vminq_f32(s##idx1##h, vmaxval); \
+    vst1q_f32(outbuf + idx0*K0, s##idx0##l); \
+    vst1q_f32(outbuf + idx0*K0 + 4, s##idx0##h); \
+    vst1q_f32(outbuf + idx1*K0, s##idx1##l); \
+    vst1q_f32(outbuf + idx1*K0 + 4, s##idx1##h)
+
+#undef CONV_FINALIZE_OUT_ALL
+#define CONV_FINALIZE_OUT_ALL() \
+    CONV_START_FINALIZE_OUT(); \
+    if (resptr) { \
+        CONV_FINALIZE_OUT2(0, 1, CONV_ADD_RESIDUAL2); \
+        CONV_FINALIZE_OUT2(2, 3, CONV_ADD_RESIDUAL2); \
+        CONV_FINALIZE_OUT2(4, 5, CONV_ADD_RESIDUAL2); \
+        CONV_FINALIZE_OUT2(6, 7, CONV_ADD_RESIDUAL2); \
+        CONV_FINALIZE_OUT2(8, 9, CONV_ADD_RESIDUAL2); \
+    } else { \
+        CONV_FINALIZE_OUT2(0, 1, CONV_ADD_NO_RESIDUAL2); \
+        CONV_FINALIZE_OUT2(2, 3, CONV_ADD_NO_RESIDUAL2); \
+        CONV_FINALIZE_OUT2(4, 5, CONV_ADD_NO_RESIDUAL2); \
+        CONV_FINALIZE_OUT2(6, 7, CONV_ADD_NO_RESIDUAL2); \
+        CONV_FINALIZE_OUT2(8, 9, CONV_ADD_NO_RESIDUAL2); \
+    }
+
+#elif CV_SIMD256
+
+///////////// generic branch for arch's with 256-bit SIMD (however, it's tweaked for AVX2 with just 16 registers) ////////////////
+
+#undef CONV_INIT_SUMS
+#define CONV_INIT_SUMS() \
+    v_float32x8 zz = v256_setzero_f32(); \
+    v_float32x8 s0 = zz, s1 = zz, s2 = zz, s3 = zz, s4 = zz; \
+    v_float32x8 s5 = zz, s6 = zz, s7 = zz, s8 = zz, s9 = zz
+
+#undef CONV_UPDATE_BLOCK
+#define CONV_UPDATE_BLOCK2x8(ofs, idx0, idx1) \
+    x0 = v256_setall_f32(inptr[idx0][ofs]); \
+    x1 = v256_setall_f32(inptr[idx1][ofs]); \
+    s##idx0 = v_fma(x0, w0, s##idx0); \
+    s##idx1 = v_fma(x1, w0, s##idx1); \
+    x0 = v256_setall_f32(inptr[idx0][ofs+1]); \
+    x1 = v256_setall_f32(inptr[idx1][ofs+1]); \
+    s##idx0 = v_fma(x0, w1, s##idx0); \
+    s##idx1 = v_fma(x1, w1, s##idx1); \
+    x0 = v256_setall_f32(inptr[idx0][ofs+2]); \
+    x1 = v256_setall_f32(inptr[idx1][ofs+2]); \
+    s##idx0 = v_fma(x0, w2, s##idx0); \
+    s##idx1 = v_fma(x1, w2, s##idx1); \
+    x0 = v256_setall_f32(inptr[idx0][ofs+3]); \
+    x1 = v256_setall_f32(inptr[idx1][ofs+3]); \
+    s##idx0 = v_fma(x0, w3, s##idx0); \
+    s##idx1 = v_fma(x1, w3, s##idx1)
+
+#undef CONV_UPDATE_LOOP_BODY
+#define CONV_UPDATE_LOOP_BODY() \
+    v_float32x8 x0, x1; \
+    v_float32x8 w0, w1, w2, w3; \
+    \
+    w0 = v256_load(wptr + 0*K0); \
+    w1 = v256_load(wptr + 1*K0); \
+    w2 = v256_load(wptr + 2*K0); \
+    w3 = v256_load(wptr + 3*K0); \
+    \
+    CONV_UPDATE_BLOCK2x8(0, 0, 1); \
+    CONV_UPDATE_BLOCK2x8(0, 2, 3); \
+    CONV_UPDATE_BLOCK2x8(0, 4, 5); \
+    CONV_UPDATE_BLOCK2x8(0, 6, 7); \
+    CONV_UPDATE_BLOCK2x8(0, 8, 9); \
+    \
+    w0 = v256_load(wptr + 4*K0); \
+    w1 = v256_load(wptr + 5*K0); \
+    w2 = v256_load(wptr + 6*K0); \
+    w3 = v256_load(wptr + 7*K0); \
+    \
+    CONV_UPDATE_BLOCK2x8(4, 0, 1); \
+    CONV_UPDATE_BLOCK2x8(4, 2, 3); \
+    CONV_UPDATE_BLOCK2x8(4, 4, 5); \
+    CONV_UPDATE_BLOCK2x8(4, 6, 7); \
+    CONV_UPDATE_BLOCK2x8(4, 8, 9); \
+    \
+    inptr[0] += inpstep[0]; inptr[1] += inpstep[1]; \
+    inptr[2] += inpstep[2]; inptr[3] += inpstep[3]; \
+    inptr[4] += inpstep[4]; inptr[5] += inpstep[5]; \
+    inptr[6] += inpstep[6]; inptr[7] += inpstep[7]; \
+    inptr[8] += inpstep[8]; inptr[9] += inpstep[9]
+
+#undef CONV_START_FINALIZE_OUT
+#define CONV_START_FINALIZE_OUT() \
+    v_float32x8 vscale = v256_load(scalebuf); \
+    v_float32x8 vbias = v256_load(biasbuf); \
+    v_float32x8 valpha = v256_load(alphabuf); \
+    v_float32x8 vmaxval = v256_setall_f32(maxval)
+
+#define CONV_ADD_RESIDUAL2(idx0, idx1) \
+    s##idx0 = v_add(s##idx0, v256_load(tmpbuf + idx0*K0)); \
+    s##idx1 = v_add(s##idx1, v256_load(tmpbuf + idx1*K0))
+
+#undef CONV_FINALIZE_OUT2
+#define CONV_FINALIZE_OUT2(idx0, idx1, add_residual2) \
+    s##idx0 = v_fma(s##idx0, vscale, vbias); \
+    s##idx1 = v_fma(s##idx1, vscale, vbias); \
+    add_residual2(idx0, idx1); \
+    s##idx0 = v_select(v_ge(s##idx0, zz), s##idx0, v_mul(s##idx0, valpha)); \
+    s##idx1 = v_select(v_ge(s##idx1, zz), s##idx1, v_mul(s##idx1, valpha)); \
+    s##idx0 = v_min(s##idx0, vmaxval); \
+    s##idx1 = v_min(s##idx1, vmaxval); \
+    v_store(outbuf + idx0*K0, s##idx0); \
+    v_store(outbuf + idx1*K0, s##idx1)
+
+#undef CONV_FINALIZE_OUT_ALL
+#define CONV_FINALIZE_OUT_ALL() \
+    CONV_START_FINALIZE_OUT(); \
+    if (resptr) { \
+        CONV_FINALIZE_OUT2(0, 1, CONV_ADD_RESIDUAL2); \
+        CONV_FINALIZE_OUT2(2, 3, CONV_ADD_RESIDUAL2); \
+        CONV_FINALIZE_OUT2(4, 5, CONV_ADD_RESIDUAL2); \
+        CONV_FINALIZE_OUT2(6, 7, CONV_ADD_RESIDUAL2); \
+        CONV_FINALIZE_OUT2(8, 9, CONV_ADD_RESIDUAL2); \
+    } else { \
+        CONV_FINALIZE_OUT2(0, 1, CONV_ADD_NO_RESIDUAL2); \
+        CONV_FINALIZE_OUT2(2, 3, CONV_ADD_NO_RESIDUAL2); \
+        CONV_FINALIZE_OUT2(4, 5, CONV_ADD_NO_RESIDUAL2); \
+        CONV_FINALIZE_OUT2(6, 7, CONV_ADD_NO_RESIDUAL2); \
+        CONV_FINALIZE_OUT2(8, 9, CONV_ADD_NO_RESIDUAL2); \
+    }
+
+
+#elif CV_SIMD128
+
+/////////////////////////// generic branch for arch's with 128-bit SIMD /////////////////////////////
+
+#undef CONV_INIT_SUMS
+#define CONV_INIT_SUMS() \
+    v_float32x4 zz = v_setzero_f32(); \
+    v_float32x4 s0l = zz, s0h = zz, s1l = zz, s1h = zz; \
+    v_float32x4 s2l = zz, s2h = zz, s3l = zz, s3h = zz; \
+    v_float32x4 s4l = zz, s4h = zz, s5l = zz, s5h = zz; \
+    v_float32x4 s6l = zz, s6h = zz, s7l = zz, s7h = zz; \
+    v_float32x4 s8l = zz, s8h = zz, s9l = zz, s9h = zz
+
+#undef CONV_UPDATE_BLOCK
+#define CONV_UPDATE_BLOCK2x8(ofs, idx0, idx1) \
+    x0 = v_setall_f32(inptr[idx0][ofs]); \
+    x1 = v_setall_f32(inptr[idx1][ofs]); \
+    s##idx0##l = v_fma(x0, w0l, s##idx0##l); \
+    s##idx0##h = v_fma(x0, w0h, s##idx0##h); \
+    s##idx1##l = v_fma(x1, w0l, s##idx1##l); \
+    s##idx1##h = v_fma(x1, w0h, s##idx1##h); \
+    x0 = v_setall_f32(inptr[idx0][ofs+1]); \
+    x1 = v_setall_f32(inptr[idx1][ofs+1]); \
+    s##idx0##l = v_fma(x0, w1l, s##idx0##l); \
+    s##idx0##h = v_fma(x0, w1h, s##idx0##h); \
+    s##idx1##l = v_fma(x1, w1l, s##idx1##l); \
+    s##idx1##h = v_fma(x1, w1h, s##idx1##h); \
+    x0 = v_setall_f32(inptr[idx0][ofs+2]); \
+    x1 = v_setall_f32(inptr[idx1][ofs+2]); \
+    s##idx0##l = v_fma(x0, w2l, s##idx0##l); \
+    s##idx0##h = v_fma(x0, w2h, s##idx0##h); \
+    s##idx1##l = v_fma(x1, w2l, s##idx1##l); \
+    s##idx1##h = v_fma(x1, w2h, s##idx1##h); \
+    x0 = v_setall_f32(inptr[idx0][ofs+3]); \
+    x1 = v_setall_f32(inptr[idx1][ofs+3]); \
+    s##idx0##l = v_fma(x0, w3l, s##idx0##l); \
+    s##idx0##h = v_fma(x0, w3h, s##idx0##h); \
+    s##idx1##l = v_fma(x1, w3l, s##idx1##l); \
+    s##idx1##h = v_fma(x1, w3h, s##idx1##h)
+
+#undef CONV_UPDATE_LOOP_BODY
+#define CONV_UPDATE_LOOP_BODY() \
+    v_float32x4 x0, x1; \
+    v_float32x4 w0l, w0h, w1l, w1h, w2l, w2h, w3l, w3h; \
+    \
+    w0l = v_load(wptr + 0*K0); w0h = v_load(wptr + 0*K0 + 4); \
+    w1l = v_load(wptr + 1*K0); w1h = v_load(wptr + 1*K0 + 4); \
+    w2l = v_load(wptr + 2*K0); w2h = v_load(wptr + 2*K0 + 4); \
+    w3l = v_load(wptr + 3*K0); w3h = v_load(wptr + 3*K0 + 4); \
+    \
+    CONV_UPDATE_BLOCK2x8(0, 0, 1); \
+    CONV_UPDATE_BLOCK2x8(0, 2, 3); \
+    CONV_UPDATE_BLOCK2x8(0, 4, 5); \
+    CONV_UPDATE_BLOCK2x8(0, 6, 7); \
+    CONV_UPDATE_BLOCK2x8(0, 8, 9); \
+    \
+    w0l = v_load(wptr + 4*K0); w0h = v_load(wptr + 4*K0 + 4); \
+    w1l = v_load(wptr + 5*K0); w1h = v_load(wptr + 5*K0 + 4); \
+    w2l = v_load(wptr + 6*K0); w2h = v_load(wptr + 6*K0 + 4); \
+    w3l = v_load(wptr + 7*K0); w3h = v_load(wptr + 7*K0 + 4); \
+    \
+    CONV_UPDATE_BLOCK2x8(4, 0, 1); \
+    CONV_UPDATE_BLOCK2x8(4, 2, 3); \
+    CONV_UPDATE_BLOCK2x8(4, 4, 5); \
+    CONV_UPDATE_BLOCK2x8(4, 6, 7); \
+    CONV_UPDATE_BLOCK2x8(4, 8, 9); \
+    \
+    inptr[0] += inpstep[0]; inptr[1] += inpstep[1]; \
+    inptr[2] += inpstep[2]; inptr[3] += inpstep[3]; \
+    inptr[4] += inpstep[4]; inptr[5] += inpstep[5]; \
+    inptr[6] += inpstep[6]; inptr[7] += inpstep[7]; \
+    inptr[8] += inpstep[8]; inptr[9] += inpstep[9]
+
+#undef CONV_START_FINALIZE_OUT
+#define CONV_START_FINALIZE_OUT() \
+    v_float32x4 vscale_lo = v_load(scalebuf), vscale_hi = v_load(scalebuf + 4); \
+    v_float32x4 vbias_lo = v_load(biasbuf), vbias_hi = v_load(biasbuf + 4); \
+    v_float32x4 valpha_lo = v_load(alphabuf), valpha_hi = v_load(alphabuf + 4); \
+    v_float32x4 vmaxval = v_setall_f32(maxval)
+
+#define CONV_ADD_RESIDUAL2(idx0, idx1) \
+    s##idx0##l = v_add(s##idx0##l, v_load(tmpbuf + idx0*K0)); \
+    s##idx0##h = v_add(s##idx0##h, v_load(tmpbuf + idx0*K0 + 4)); \
+    s##idx1##l = v_add(s##idx1##l, v_load(tmpbuf + idx1*K0)); \
+    s##idx1##h = v_add(s##idx1##h, v_load(tmpbuf + idx1*K0 + 4))
+
+#undef CONV_FINALIZE_OUT2
+#define CONV_FINALIZE_OUT2(idx0, idx1, add_residual2) \
+    s##idx0##l = v_fma(s##idx0##l, vscale_lo, vbias_lo); \
+    s##idx0##h = v_fma(s##idx0##h, vscale_hi, vbias_hi); \
+    s##idx1##l = v_fma(s##idx1##l, vscale_lo, vbias_lo); \
+    s##idx1##h = v_fma(s##idx1##h, vscale_hi, vbias_hi); \
+    add_residual2(idx0, idx1); \
+    s##idx0##l = v_select(v_ge(s##idx0##l, zz), s##idx0##l, v_mul(s##idx0##l, valpha_lo)); \
+    s##idx0##h = v_select(v_ge(s##idx0##h, zz), s##idx0##h, v_mul(s##idx0##h, valpha_hi)); \
+    s##idx1##l = v_select(v_ge(s##idx1##l, zz), s##idx1##l, v_mul(s##idx1##l, valpha_lo)); \
+    s##idx1##h = v_select(v_ge(s##idx1##h, zz), s##idx1##h, v_mul(s##idx1##h, valpha_hi)); \
+    s##idx0##l = v_min(s##idx0##l, vmaxval); \
+    s##idx0##h = v_min(s##idx0##h, vmaxval); \
+    s##idx1##l = v_min(s##idx1##l, vmaxval); \
+    s##idx1##h = v_min(s##idx1##h, vmaxval); \
+    v_store(outbuf + idx0*K0, s##idx0##l); \
+    v_store(outbuf + idx0*K0 + 4, s##idx0##h); \
+    v_store(outbuf + idx1*K0, s##idx1##l); \
+    v_store(outbuf + idx1*K0 + 4, s##idx1##h)
+
+#undef CONV_FINALIZE_OUT_ALL
+#define CONV_FINALIZE_OUT_ALL() \
+    CONV_START_FINALIZE_OUT(); \
+    if (resptr) { \
+        CONV_FINALIZE_OUT2(0, 1, CONV_ADD_RESIDUAL2); \
+        CONV_FINALIZE_OUT2(2, 3, CONV_ADD_RESIDUAL2); \
+        CONV_FINALIZE_OUT2(4, 5, CONV_ADD_RESIDUAL2); \
+        CONV_FINALIZE_OUT2(6, 7, CONV_ADD_RESIDUAL2); \
+        CONV_FINALIZE_OUT2(8, 9, CONV_ADD_RESIDUAL2); \
+    } else { \
+        CONV_FINALIZE_OUT2(0, 1, CONV_ADD_NO_RESIDUAL2); \
+        CONV_FINALIZE_OUT2(2, 3, CONV_ADD_NO_RESIDUAL2); \
+        CONV_FINALIZE_OUT2(4, 5, CONV_ADD_NO_RESIDUAL2); \
+        CONV_FINALIZE_OUT2(6, 7, CONV_ADD_NO_RESIDUAL2); \
+        CONV_FINALIZE_OUT2(8, 9, CONV_ADD_NO_RESIDUAL2); \
+    }
+
+#else
+
+#undef CONV_ENABLE_SIMD
+
+#endif
+
+static void conv32fC8(const void* inp__, const void* residual__, void* out__,
+                      const ConvState& cs, const void* weights__,
+                      const float* scale__, const float* bias__)
+{
+    using FT = float;
+    const MatShape& inpshape = cs.inpshape;
+    const MatShape& outshape = cs.outshape;
+
+    CV_Assert_N(inpshape.layout == DATA_LAYOUT_BLOCK, outshape.layout == DATA_LAYOUT_BLOCK);
+
+    int K_ = outshape.channels();
+    int ndims_ = outshape.dims;
+    int N = outshape[0];
+    int D_ = ndims_ >= 6 ? outshape[ndims_ - 4] : 1;
+    int H_ = ndims_ >= 5 ? outshape[ndims_ - 3] : 1;
+    int W_ = outshape[ndims_-2];
+    int planeblocks_ = D_*H_*W_;
+    size_t outtotal = outshape.total();
+
+    int Kblk_ = cs.wshape[1];
+    int C1Max_ = cs.wshape[3];
+    int total_blocks = N * cs.ngroups * Kblk_;
+
+    if ((K_/cs.ngroups) % inpshape.back() != 0) {
+        // if there could be 'padding' channels in the output,
+        // clear the output before the parallel loop
+        // to make sure that all the padding channels are cleared.
+        memset(out__, 0, outtotal*sizeof(FT));
+    }
+
+    parallel_for_(Range(0, total_blocks), [&](const Range& range) {
+        constexpr int SPAT_BLOCK_SIZE = 10;
+        constexpr int C0shift = 3, K0shift = C0shift;
+        constexpr int C0 = 1 << C0shift, K0 = C0;
+
+        CV_Assert_N(inpshape.back() == C0, outshape.back() == K0);
+
+        const int C = inpshape.channels(), K = outshape.channels();
+        const int C1 = (C + C0 - 1)/C0, K1 = (K + K0 - 1)/K0;
+        const int ngroups = cs.ngroups, Kblk = Kblk_, C1Max = C1Max_;
+        const int Cg = C / ngroups;
+        const int Kg = K / ngroups;
+        int ksize = cs.wshape[2];
+        int ndims = ndims_;
+        int D = D_, H = H_, W = W_;
+        int Di = ndims >= 6 ? inpshape[ndims-4] : 1;
+        int Hi = ndims >= 5 ? inpshape[ndims-3] : 1;
+        int Wi = inpshape[ndims-2];
+        const int Sz = cs.strides[0], Sy = cs.strides[1], Sx = cs.strides[2];
+        const int padZ = cs.pads[0], padY = cs.pads[1], padX = cs.pads[2];
+        const float* scaleptr = (const float*)scale__;
+        const float* biasptr = (const float*)bias__;
+        const int* ofsZYX = cs.coordtab.data();
+        int planeblocks = planeblocks_;
+        int planesize = planeblocks*K0;
+        int iplanesize = Di*Hi*Wi*C0;
+
+    #ifdef CONV_ENABLE_SIMD
+        constexpr int MAX_CONV_DIMS = ConvState::MAX_CONV_DIMS;
+        int innerZ0 = cs.inner[0], innerZ1 = cs.inner[MAX_CONV_DIMS];
+        int innerY0 = cs.inner[1], innerY1 = cs.inner[MAX_CONV_DIMS+1];
+        int innerX0 = cs.inner[2], innerX1 = cs.inner[MAX_CONV_DIMS+2];
+        float zbuf[C0] = {};
+    #endif
+
+        FastActivation fastActivation = cs.fastActivation;
+        const float* activParams = cs.activParams.data();
+        ActivationFunc activation = cs.activation;
+        float maxval = FLT_MAX, defaultAlpha = 0.f;
+        float scalebuf[K0], biasbuf[K0], alphabuf[K0];
+        if (fastActivation == FAST_ACTIV_CLIP) {
+            CV_Assert(cs.activParams.size() == 2u);
+            maxval = activParams[1];
+        } else if (fastActivation == FAST_ACTIV_RELU) {
+            CV_Assert(!activParams);
+        } else if (fastActivation == FAST_ACTIV_LEAKY_RELU) {
+            CV_Assert(cs.activParams.size() == 1u);
+            defaultAlpha = activParams[0];
+        } else if (fastActivation == FAST_ACTIV_PRELU) {
+            CV_Assert(cs.activParams.size() == size_t(K));
+        } else {
+            CV_Assert(fastActivation == FAST_ACTIV_NONE);
+            defaultAlpha = 1.f;
+        }
+
+        // 1x1x1 convolution with (1,1,1) strides:
+        // flatten input/output tensors in this case to accelerate address computations
+        if (ksize == 1 && Sz*Sy*Sx == 1) {
+            W *= D*H;
+            Wi *= Di*Hi;
+            D = Di = H = Hi = 1;
+        #ifdef CONV_ENABLE_SIMD
+            innerZ1 = innerY1 = 1;
+            innerX1 = W;
+        #endif
+        }
+
+        for (int t = range.start; t < range.end; t++) {
+            const int p0 = 0, p1 = planeblocks;
+            const int n = t / (ngroups * Kblk);
+            const int rem = t - n * (ngroups * Kblk);
+            const int g = rem / Kblk;
+            const int kblk = rem - g * Kblk;
+
+            const int k_base = g * Kg + kblk * K0;
+            if (k_base >= K) continue;
+
+            const int k_count = std::min(std::min(K0, Kg - kblk*K0), K - k_base);
+            bool aligned_k = (k_base & (K0-1)) == 0 && k_count == K0;
+
+            const int c_start  = g * Cg;
+            const int c00      = c_start & (C0-1);
+            const int c1_start = c_start >> C0shift;
+            const int cblocks  = (c00 + Cg + C0 - 1) >> C0shift;
+            const float* inpbaseptr = (float*)inp__ + (n * C1 + c1_start) * iplanesize;
+            const float* wbaseptr = (float*)weights__ + (g*Kblk + kblk)*(ksize*C1Max*C0*K0);
+
+            {
+                int kk = 0;
+                for (; kk < k_count; kk++) {
+                    scalebuf[kk] = scaleptr ? scaleptr[k_base + kk] : 1.f;
+                    biasbuf[kk] = biasptr ? biasptr[k_base + kk] : 0.f;
+                    alphabuf[kk] = fastActivation == FAST_ACTIV_PRELU ? activParams[k_base + kk] : defaultAlpha;
+                }
+                for (; kk < K0; kk++) {
+                    scalebuf[kk] = 0.f;
+                    biasbuf[kk] = 0.f;
+                    alphabuf[kk] = 0.f;
+                }
+            }
+
+            float* outptr = (float*)out__ + n*(K1*planesize) + p0*K0;
+            const float* resptr = residual__ ? (float*)residual__ + n*(K1*planesize) + p0*K0 : nullptr;
+            float tmpbuf[SPAT_BLOCK_SIZE*K0] = {};
+            int p = p0;
+
+        #ifdef CONV_ENABLE_SIMD
+            for (; p < p1; p += SPAT_BLOCK_SIZE,
+                           outptr += SPAT_BLOCK_SIZE*K0)
+            {
+                Vec3i pt[SPAT_BLOCK_SIZE];
+                bool inner[SPAT_BLOCK_SIZE];
+
+                if (p + SPAT_BLOCK_SIZE > p1) {
+                    if (p == p0)
+                        break;
+                    int p_new = p1 - SPAT_BLOCK_SIZE;
+                    int dp = p_new - p;
+                    outptr += dp*K0;
+                    resptr += (resptr ? dp*K0 : 0);
+                    p = p_new;
+                }
+
+                if (resptr) {
+                    if (aligned_k) {
+                        memcpy(tmpbuf, resptr + k_base*planeblocks, SPAT_BLOCK_SIZE*K0*sizeof(FT));
+                    } else {
+                        for (int kk = 0; kk < k_count; ++kk) {
+                            const int k = k_base + kk;
+                            int kofs = (k >> K0shift) * planesize + (k & (K0-1));
+                            for (int j = 0; j < SPAT_BLOCK_SIZE; j++)
+                                tmpbuf[kk + j*K0] = resptr[kofs + j*K0];
+                        }
+                    }
+                    resptr += SPAT_BLOCK_SIZE*K0;
+                }
+
+                if ((p % W) + SPAT_BLOCK_SIZE <= W) {
+                    int zj = p / (H*W);
+                    int yxj = p - zj*(H*W);
+                    int yj = yxj / W;
+                    int x = yxj - yj*W;
+                    const bool zy_inner = (zj >= innerZ0 && zj < innerZ1) && (yj >= innerY0 && yj < innerY1);
+                    for (int j = 0; j < SPAT_BLOCK_SIZE; j++) {
+                        int xj = x + j;
+                        pt[j] = Vec3i(zj*Sz - padZ, yj*Sy - padY, xj*Sx - padX);
+                        inner[j] = zy_inner && (xj >= innerX0 && xj < innerX1);
+                    }
+                } else {
+                    for (int j = 0; j < SPAT_BLOCK_SIZE; j++) {
+                        int pj = p + j;
+                        int zj = pj / (H*W);
+                        int yxj = pj - zj*(H*W);
+                        int yj = yxj / W;
+                        int xj = yxj - yj*W;
+                        pt[j] = Vec3i(zj*Sz - padZ, yj*Sy - padY, xj*Sx - padX);
+                        inner[j] = (zj >= innerZ0 && zj < innerZ1) &&
+                                   (yj >= innerY0 && yj < innerY1) &&
+                                   (xj >= innerX0 && xj < innerX1);
+                    }
+                }
+
+                CONV_INIT_SUMS();
+
+                for (int i = 0; i < ksize; i++) {
+                    const float* inptr[SPAT_BLOCK_SIZE];
+                    int inpstep[SPAT_BLOCK_SIZE];
+                    for (int j = 0; j < SPAT_BLOCK_SIZE; j++) {
+                        Vec3i ptj = pt[j];
+                        int zij = ptj[0] + ofsZYX[i*3 + 0];
+                        int yij = ptj[1] + ofsZYX[i*3 + 1];
+                        int xij = ptj[2] + ofsZYX[i*3 + 2];
+                        if (inner[j] || ((((unsigned)zij < (unsigned)Di)&
+                                          ((unsigned)yij < (unsigned)Hi)&
+                                          ((unsigned)xij < (unsigned)Wi)) != 0)) {
+                            inptr[j] = inpbaseptr + (((zij * Hi) + yij) * Wi + xij) * C0;
+                            inpstep[j] = iplanesize;
+                        } else {
+                            inptr[j] = zbuf;
+                            inpstep[j] = 0;
+                        }
+                    }
+                    const float* wptr = wbaseptr + i*C1Max*K0*C0;
+
+                    for (int c1 = 0; c1 < cblocks; c1++, wptr += C0*K0) {
+                        CONV_UPDATE_LOOP_BODY();
+                    }
+                }
+
+                float* outbuf = aligned_k ? outptr + k_base*planeblocks : tmpbuf;
+                CONV_FINALIZE_OUT_ALL();
+
+                if (activation) {
+                    activation(outbuf, outbuf, SPAT_BLOCK_SIZE*K0, activParams);
+                }
+
+                if (!aligned_k) {
+                    for (int kk = 0; kk < k_count; ++kk) {
+                        const int k = k_base + kk;
+                        int kofs = (k >> K0shift) * planesize + (k & (K0-1));
+                        for (int j = 0; j < SPAT_BLOCK_SIZE; j++)
+                            outptr[kofs + j*K0] = tmpbuf[kk + j*K0];
+                    }
+                }
+            }
+        #endif
+
+            float resbuf[K0] = {};
+
+            for (; p < p1; p++, outptr += K0, resptr += (resptr ? K0 : 0))
+            {
+                int zj = p / (H*W);
+                int yxj = p - zj*(H*W);
+                int yj = yxj / W;
+                int xj = yxj - yj*W;
+                int zi_base = zj*Sz - padZ;
+                int yi_base = yj*Sy - padY;
+                int xi_base = xj*Sx - padX;
+
+            #if CV_SIMD256
+                v_float32x8 zz = v256_setzero_f32();
+                v_float32x8 s0 = zz;
+            #elif CV_SIMD128
+                v_float32x4 zz = v_setzero_f32();
+                v_float32x4 s0 = zz, s1 = zz;
+            #else
+                for (int kk = 0; kk < K0; kk++) {
+                    tmpbuf[kk] = 0.f;
+                }
+            #endif
+
+                if (resptr) {
+                    for (int kk = 0; kk < k_count; ++kk) {
+                        const int k = k_base + kk;
+                        int kofs = (k >> K0shift) * planesize + (k & (K0-1));
+                        resbuf[kk] = resptr[kofs];
+                    }
+                }
+
+                for (int i = 0; i < ksize; i++) {
+                    int zi = zi_base + ofsZYX[i*3 + 0];
+                    int yi = yi_base + ofsZYX[i*3 + 1];
+                    int xi = xi_base + ofsZYX[i*3 + 2];
+
+                    if ((((unsigned)zi >= (unsigned)Di) |
+                         ((unsigned)yi >= (unsigned)Hi) |
+                         ((unsigned)xi >= (unsigned)Wi)) != 0)
+                        continue;
+
+                    const float* inptr = inpbaseptr + (((zi * Hi) + yi) * Wi + xi) * C0;
+                    const float* wptr = wbaseptr + i*C1Max*K0*C0;
+
+                    for (int c1 = 0; c1 < cblocks; ++c1, inptr += iplanesize, wptr += K0*C0) {
+                    #if CV_SIMD256
+                        v_float32x8 w, x;
+                        #undef CONV_UPDATE_BLOCK1
+                        #define CONV_UPDATE_BLOCK1(ofs) \
+                            w = v256_load(wptr + ofs*K0); \
+                            x = v256_setall_f32(inptr[ofs]); \
+                            s0 = v_fma(x, w, s0)
+                        CONV_UPDATE_BLOCK1(0);
+                        CONV_UPDATE_BLOCK1(1);
+                        CONV_UPDATE_BLOCK1(2);
+                        CONV_UPDATE_BLOCK1(3);
+                        CONV_UPDATE_BLOCK1(4);
+                        CONV_UPDATE_BLOCK1(5);
+                        CONV_UPDATE_BLOCK1(6);
+                        CONV_UPDATE_BLOCK1(7);
+                    #elif CV_SIMD128
+                        v_float32x4 w0, w1, x;
+                        #undef CONV_UPDATE_BLOCK1
+                        #define CONV_UPDATE_BLOCK1(ofs) \
+                            w0 = v_load(wptr + ofs*K0); w1 = v_load(wptr + ofs*K0 + 4); \
+                            x = v_setall_f32(inptr[ofs]); \
+                            s0 = v_fma(x, w0, s0); s1 = v_fma(x, w1, s1)
+                        CONV_UPDATE_BLOCK1(0);
+                        CONV_UPDATE_BLOCK1(1);
+                        CONV_UPDATE_BLOCK1(2);
+                        CONV_UPDATE_BLOCK1(3);
+                        CONV_UPDATE_BLOCK1(4);
+                        CONV_UPDATE_BLOCK1(5);
+                        CONV_UPDATE_BLOCK1(6);
+                        CONV_UPDATE_BLOCK1(7);
+                    #else
+                        for (int c0 = 0; c0 < C0; ++c0) {
+                            const float xval = inptr[c0];
+                            for (int kk = 0; kk < K0; ++kk)
+                                tmpbuf[kk] += xval * wptr[c0*K0 + kk];
+                        }
+                    #endif
+                    }
+                }
+
+                float* outbuf = aligned_k ? outptr + k_base*planeblocks : tmpbuf;
+
+            #if CV_SIMD256
+                v_float32x8 vscale = v256_load(scalebuf);
+                v_float32x8 vbias = v256_load(biasbuf);
+                v_float32x8 valpha = v256_load(alphabuf);
+                v_float32x8 vmaxval = v256_setall_f32(maxval);
+
+                s0 = v_fma(s0, vscale, vbias);
+                s0 = v_add(s0, v256_load(resbuf));
+                s0 = v_select(v_ge(s0, zz), s0, v_mul(s0, valpha));
+                s0 = v_min(s0, vmaxval);
+                v_store(outbuf, s0);
+            #elif CV_SIMD128
+                v_float32x4 vscale_lo = v_load(scalebuf), vscale_hi = v_load(scalebuf + 4);
+                v_float32x4 vbias_lo = v_load(biasbuf), vbias_hi = v_load(biasbuf + 4);
+                v_float32x4 valpha_lo = v_load(alphabuf), valpha_hi = v_load(alphabuf + 4);
+                v_float32x4 vmaxval = v_setall_f32(maxval);
+
+                s0 = v_fma(s0, vscale_lo, vbias_lo);
+                s1 = v_fma(s1, vscale_hi, vbias_hi);
+                s0 = v_add(s0, v_load(resbuf));
+                s1 = v_add(s1, v_load(resbuf + 4));
+                s0 = v_select(v_ge(s0, zz), s0, v_mul(s0, valpha_lo));
+                s1 = v_select(v_ge(s1, zz), s1, v_mul(s1, valpha_hi));
+                s0 = v_min(s0, vmaxval);
+                s1 = v_min(s1, vmaxval);
+                v_store(outbuf, s0);
+                v_store(outbuf + 4, s1);
+            #else
+                for (int kk = 0; kk < K0; kk++) {
+                    float v = tmpbuf[kk]*scalebuf[kk] + biasbuf[kk] + resbuf[kk];
+                    v = std::min(v*(v >= 0 ? 1.f : alphabuf[kk]), maxval);
+                    outbuf[kk] = v;
+                }
+            #endif
+
+                if (activation) {
+                    activation(outbuf, outbuf, K0, activParams);
+                }
+
+                if (!aligned_k) {
+                    for (int kk = 0; kk < k_count; kk++) {
+                        const int k = k_base + kk;
+                        int kofs = (k >> K0shift) * planesize + (k & (K0-1));
+                        outptr[kofs] = tmpbuf[kk];
+                    }
+                }
+            }
+        }
+    });
+}
+
+cv::dnn::ConvFunc getConvFunc_(int depth, int C0)
+{
+    ConvFunc func = nullptr;
+    if (depth == CV_32F && C0 == 8) {
+        func = conv32fC8;
+    }
+    return func;
+}
+
+CV_CPU_OPTIMIZATION_NAMESPACE_END
+}}
+#endif // CV_CPU_OPTIMIZATION_DECLARATIONS_ONLY

modules/dnn/src/layers/elementwise_layers.cpp

@@ -93,6 +93,18 @@ using std::sin;
 using std::sinh;
 using std::tan;
 
+int ActivationLayer::getLayouts(const std::vector<DataLayout>& actualInputs,
+                                std::vector<DataLayout>& desiredInputs,
+                                const int requiredOutputs,
+                                std::vector<DataLayout>& outputs) const
+{
+    size_t ninputs = actualInputs.size();
+    CV_Assert(ninputs >= 1u);
+    desiredInputs = actualInputs;
+    outputs.assign(requiredOutputs, actualInputs[0]);
+    return 0;
+}
+
 struct PowerFunctor;
 
 template<typename Func>

modules/dnn/src/layers/maxpool_layer.cpp

@@ -0,0 +1,547 @@
+// 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 "../precomp.hpp"
+#include "layers_common.hpp"
+#include "../net_impl.hpp"
+#include "conv2_common.hpp"
+#include "opencv2/core/hal/intrin.hpp"
+
+namespace cv
+{
+namespace dnn
+{
+
+static void maxPool32f(const void* inp_, void* out_, const ConvState& cs)
+{
+    int NC1 = cs.inpshape[0]*cs.inpshape[1];
+
+    CV_Assert(cs.inpshape.layout == DATA_LAYOUT_BLOCK);
+    CV_Assert(cs.outshape.layout == DATA_LAYOUT_BLOCK);
+    CV_Assert(cs.inpshape.dims == cs.outshape.dims);
+
+    parallel_for_(Range(0, NC1), [&](const Range& r) {
+        constexpr int MAX_POOL_DIMS = ConvState::MAX_CONV_DIMS;
+
+        CV_Assert(cs.nspatialdims <= MAX_POOL_DIMS && MAX_POOL_DIMS == 3);
+
+        int sdims = cs.nspatialdims;
+        int nc0 = r.start, nc1 = r.end;
+        int C0 = cs.inpshape.back();
+        int Di = sdims > 2 ? cs.inpshape[sdims - 1] : 1;
+        int Hi = sdims > 1 ? cs.inpshape[sdims] : 1;
+        int Wi = cs.inpshape[sdims + 1];
+        int D = sdims > 2 ? cs.outshape[sdims - 1] : 1;
+        int H = sdims > 1 ? cs.outshape[sdims] : 1;
+        int W = cs.outshape[sdims + 1];
+        int iplanesize = Di*Hi*Wi*C0;
+        int planesize = D*H*W*C0;
+        int SZ = cs.strides[0], SY = cs.strides[1], SX = cs.strides[2];
+        int padZ0 = cs.pads[0], padY0 = cs.pads[1], padX0 = cs.pads[2];
+        int inner_z0 = cs.inner[0], inner_z1 = cs.inner[MAX_POOL_DIMS];
+        int inner_y0 = cs.inner[1], inner_y1 = cs.inner[MAX_POOL_DIMS + 1];
+        int inner_x0 = cs.inner[2], inner_x1 = cs.inner[MAX_POOL_DIMS + 2];
+        int ksize = (int)cs.ofstab.size();
+        const int* zyxtab = cs.coordtab.data();
+        const int* ofstab = cs.ofstab.data();
+
+        const float* inp = (const float*)inp_ + nc0*iplanesize;
+        float* out = (float*)out_ + nc0*planesize;
+        const float INITVAL = -FLT_MAX;
+
+    #if CV_SIMD || CV_SIMD_SCALABLE
+        int nlanes = VTraits<v_float32>::vlanes();
+        v_float32 s_min = vx_setall_f32(INITVAL);
+        CV_Assert(C0 == nlanes || C0 == nlanes*2 || C0 % (nlanes*4) == 0);
+    #endif
+
+        for (int nc = nc0; nc < nc1; nc++, inp += iplanesize) {
+            for (int z0 = 0; z0 < D; z0++) {
+                int zi_ = z0*SZ - padZ0;
+                for (int y0 = 0; y0 < H; y0++, out += W*C0) {
+                    int x0 = 0;
+                    int x1 = z0 >= inner_z0 && z0 < inner_z1 &&
+                        y0 >= inner_y0 && y0 < inner_y1 ? inner_x0 : W;
+                    int yi_ = y0*SY - padY0;
+
+                #if !(CV_SIMD || CV_SIMD_SCALABLE)
+                    for (int c = 0; c < C0*W; c++)
+                        out[c] = INITVAL;
+                #endif
+
+                    for(;;) {
+                    #if CV_SIMD || CV_SIMD_SCALABLE
+                        if (nlanes == C0) {
+                            for (; x0 < x1; x0++) {
+                                int xi_ = x0*SX - padX0;
+                                v_float32 s0 = s_min;
+                                for (int k = 0; k < ksize; k++) {
+                                    int zi = zi_ + zyxtab[k*MAX_POOL_DIMS];
+                                    int yi = yi_ + zyxtab[k*MAX_POOL_DIMS+1];
+                                    int xi = xi_ + zyxtab[k*MAX_POOL_DIMS+2];
+                                    v_float32 v0;
+                                    if ((unsigned)zi >= (unsigned)Di ||
+                                        (unsigned)yi >= (unsigned)Hi ||
+                                        (unsigned)xi >= (unsigned)Wi)
+                                        continue;
+                                    v0 = vx_load(inp + ((zi*Hi + yi)*Wi + xi)*C0);
+                                    s0 = v_max(s0, v0);
+                                }
+                                vx_store(out + x0*C0, s0);
+                            }
+                        } else {
+                            for (; x0 < x1; x0++) {
+                                int xi_ = x0*SX - padX0;
+                                for (int c = 0; c < C0; c += nlanes*2) {
+                                    v_float32 s0 = s_min, s1 = s_min;
+                                    for (int k = 0; k < ksize; k++) {
+                                        int zi = zi_ + zyxtab[k*MAX_POOL_DIMS];
+                                        int yi = yi_ + zyxtab[k*MAX_POOL_DIMS+1];
+                                        int xi = xi_ + zyxtab[k*MAX_POOL_DIMS+2];
+                                        v_float32 v0, v1;
+                                        if ((unsigned)zi >= (unsigned)Di ||
+                                            (unsigned)yi >= (unsigned)Hi ||
+                                            (unsigned)xi >= (unsigned)Wi)
+                                            continue;
+                                        int ofs_k = ((zi*Hi + yi)*Wi + xi)*C0 + c;
+                                        v0 = vx_load(inp + ofs_k);
+                                        v1 = vx_load(inp + ofs_k + nlanes);
+                                        s0 = v_max(s0, v0);
+                                        s1 = v_max(s1, v1);
+                                    }
+                                    vx_store(out + x0*C0 + c, s0);
+                                    vx_store(out + x0*C0 + c + nlanes, s1);
+                                }
+                            }
+                        }
+                    #else
+                        for (; x0 < x1; x0++) {
+                            int xi_ = x0*SX - padX0;
+                            for (int k = 0; k < ksize; k++) {
+                                int zi = zi_ + zyxtab[k*MAX_POOL_DIMS];
+                                int yi = yi_ + zyxtab[k*MAX_POOL_DIMS+1];
+                                int xi = xi_ + zyxtab[k*MAX_POOL_DIMS+2];
+                                if ((unsigned)zi >= (unsigned)Di ||
+                                    (unsigned)yi >= (unsigned)Hi ||
+                                    (unsigned)xi >= (unsigned)Wi)
+                                    continue;
+                                const float* inptr = inp + ((zi*Hi + yi)*Wi + xi)*C0;
+                                for (int c = 0; c < C0; c++)
+                                    out[x0*C0 + c] = std::max(out[x0*C0 + c], inptr[c]);
+                            }
+                        }
+                    #endif
+                        if (x0 == W)
+                            break;
+                        x1 = inner_x1;
+
+                    #if CV_SIMD || CV_SIMD_SCALABLE
+                        if (nlanes == C0) {
+                            for (; x0 < x1; x0++) {
+                                int xi_ = x0*SX - padX0;
+                                const float* inp_xi = inp + ((Hi*zi_ + yi_)*Wi + xi_)*C0;
+
+                                v_float32 s0 = vx_load(inp_xi + ofstab[0]);
+                                for (int k = 1; k < ksize; k++)
+                                    s0 = v_max(s0, vx_load(inp_xi + ofstab[k]));
+                                vx_store(out + x0*C0, s0);
+                            }
+                        } else if (nlanes*2 == C0) {
+                            for (; x0 < x1; x0++) {
+                                int xi_ = x0*SX - padX0;
+                                const float* inp_xi = inp + ((Hi*zi_ + yi_)*Wi + xi_)*C0;
+
+                                int ofs_k = ofstab[0];
+                                v_float32 s0 = vx_load(inp_xi + ofs_k);
+                                v_float32 s1 = vx_load(inp_xi + ofs_k + nlanes);
+                                for (int k = 1; k < ksize; k++) {
+                                    ofs_k = ofstab[k];
+                                    s0 = v_max(s0, vx_load(inp_xi + ofs_k));
+                                    s1 = v_max(s1, vx_load(inp_xi + ofs_k + nlanes));
+                                }
+                                vx_store(out + x0*C0, s0);
+                                vx_store(out + x0*C0 + nlanes, s1);
+                            }
+                        } else {
+                            for (; x0 < x1; x0++) {
+                                int xi_ = x0*SX - padX0;
+                                for (int c = 0; c < C0; c += nlanes*4) {
+                                    const float* inp_xi = inp + ((Hi*zi_ + yi_)*Wi + xi_)*C0 + c;
+
+                                    int ofs_k = ofstab[0];
+                                    v_float32 s0 = vx_load(inp_xi + ofs_k);
+                                    v_float32 s1 = vx_load(inp_xi + ofs_k + nlanes);
+                                    v_float32 s2 = vx_load(inp_xi + ofs_k + nlanes*2);
+                                    v_float32 s3 = vx_load(inp_xi + ofs_k + nlanes*3);
+                                    for (int k = 1; k < ksize; k++) {
+                                        ofs_k = ofstab[k];
+                                        s0 = v_max(s0, vx_load(inp_xi + ofs_k));
+                                        s1 = v_max(s1, vx_load(inp_xi + ofs_k + nlanes));
+                                        s2 = v_max(s2, vx_load(inp_xi + ofs_k + nlanes*2));
+                                        s3 = v_max(s3, vx_load(inp_xi + ofs_k + nlanes*3));
+                                    }
+                                    vx_store(out + x0*C0 + c, s0);
+                                    vx_store(out + x0*C0 + c + nlanes, s1);
+                                    vx_store(out + x0*C0 + c + nlanes*2, s2);
+                                    vx_store(out + x0*C0 + c + nlanes*3, s3);
+                                }
+                            }
+                        }
+                    #else
+                        for (; x0 < x1; x0++) {
+                            int xi_ = x0*SX - padX0;
+                            const float* inp_xi = inp + ((Hi*zi_ + yi_)*Wi + xi_)*C0;
+                            for (int k = 0; k < ksize; k++) {
+                                const float* inptr = inp_xi + ofstab[k];
+                                for (int c = 0; c < C0; c++)
+                                    out[x0*C0 + c] = std::max(out[x0*C0 + c], inptr[c]);
+                            }
+                        }
+                    #endif
+                        x1 = W;
+                    }
+                }
+            }
+        }
+    });
+}
+
+// temporarily exclude fp16/bf16 versions,
+// since convolution and other layers don't support those types yet
+#if 0
+template<typename _Tp>
+static void maxPool16xf(const _Tp* inp_, _Tp* out_, const ConvState& cs)
+{
+    constexpr int MAX_POOL_DIMS = ConvState::MAX_CONV_DIMS;
+    int C0_ = cs.inpshape.back();
+    int NC = cs.inpshape[0]*cs.inpshape[1];
+    int nlanes_ = VTraits<v_float32>::vlanes();
+
+    CV_Assert(C0_ == nlanes_ || C0_ == nlanes_*2 || C0_ % (nlanes_*4) == 0);
+    CV_Assert(cs.nspatialdims <= MAX_POOL_DIMS && MAX_POOL_DIMS == 3);
+    CV_Assert(cs.inpshape.layout == DATA_LAYOUT_BLOCK);
+    CV_Assert(cs.outshape.layout == DATA_LAYOUT_BLOCK);
+    CV_Assert(cs.inpshape.dims == cs.outshape.dims);
+
+    parallel_for_(Range(0, NC), [&](const Range& r) {
+        int sdims = cs.nspatialdims;
+        int nc0 = r.start, nc1 = r.end;
+        int nlanes = nlanes_, C0 = cs.inpshape.back();
+        int Di = sdims > 2 ? cs.inpshape[sdims - 1] : 1;
+        int Hi = sdims > 1 ? cs.inpshape[sdims] : 1;
+        int Wi = cs.inpshape[sdims + 1];
+        int D = sdims > 2 ? cs.outshape[sdims - 1] : 1;
+        int H = sdims > 1 ? cs.outshape[sdims] : 1;
+        int W = cs.outshape[sdims + 1];
+        int iplanesize = Di*Hi*Wi*C0;
+        int planesize = D*H*W*C0;
+        int SZ = cs.strides[0], SY = cs.strides[1], SX = cs.strides[2];
+        int padZ0 = cs.pads[0], padY0 = cs.pads[1], padX0 = cs.pads[2];
+        int inner_z0 = cs.inner[0], inner_z1 = cs.inner[MAX_POOL_DIMS];
+        int inner_y0 = cs.inner[1], inner_y1 = cs.inner[MAX_POOL_DIMS + 1];
+        int inner_x0 = cs.inner[2], inner_x1 = cs.inner[MAX_POOL_DIMS + 2];
+        int ksize = (int)cs.ofstab.size();
+        const int* zyxtab = cs.coordtab.data();
+        const int* ofstab = cs.ofstab.data();
+
+        const _Tp* inp = (const _Tp*)inp_ + nc0*iplanesize;
+        _Tp* out = (_Tp*)out_ + nc0*planesize;
+        v_float32 s_min = vx_setall_f32(-FLT_MAX);
+
+        for (int nc = nc0; nc < nc1; nc++, inp += iplanesize) {
+            for (int z0 = 0; z0 < D; z0++) {
+                int zi_ = z0*SZ - padZ0;
+                for (int y0 = 0; y0 < H; y0++, out += W*C0) {
+                    int x0 = 0;
+                    int x1 = z0 >= inner_z0 && z0 < inner_z1 &&
+                        y0 >= inner_y0 && y0 < inner_y1 ? inner_x0 : W;
+                    int yi_ = y0*SY - padY0;
+                    for(;;) {
+                        if (nlanes == C0) {
+                            for (; x0 < x1; x0++) {
+                                int xi_ = x0*SX - padX0;
+                                v_float32 s0 = s_min;
+                                for (int k = 0; k < ksize; k++) {
+                                    int zi = zi_ + zyxtab[k*MAX_POOL_DIMS];
+                                    int yi = yi_ + zyxtab[k*MAX_POOL_DIMS+1];
+                                    int xi = xi_ + zyxtab[k*MAX_POOL_DIMS+2];
+                                    v_float32 v0;
+                                    if ((unsigned)zi >= (unsigned)Di ||
+                                        (unsigned)yi >= (unsigned)Hi ||
+                                        (unsigned)xi >= (unsigned)Wi)
+                                        continue;
+                                    v0 = vx_load_expand(inp + ((zi*Hi + yi)*Wi + xi)*C0);
+                                    s0 = v_max(s0, v0);
+                                }
+                                v_pack_store(out + x0*C0, s0);
+                            }
+                        } else {
+                            for (; x0 < x1; x0++) {
+                                int xi_ = x0*SX - padX0;
+                                for (int c = 0; c < C0; c += nlanes*2) {
+                                    v_float32 s0 = s_min, s1 = s_min;
+                                    for (int k = 0; k < ksize; k++) {
+                                        int zi = zi_ + zyxtab[k*MAX_POOL_DIMS];
+                                        int yi = yi_ + zyxtab[k*MAX_POOL_DIMS+1];
+                                        int xi = xi_ + zyxtab[k*MAX_POOL_DIMS+2];
+                                        v_float32 v0, v1;
+                                        if ((unsigned)zi >= (unsigned)Di ||
+                                            (unsigned)yi >= (unsigned)Hi ||
+                                            (unsigned)xi >= (unsigned)Wi)
+                                            continue;
+                                        int ofs_k = ((zi*Hi + yi)*Wi + xi)*C0 + c;
+                                        v0 = vx_load_expand(inp + ofs_k);
+                                        v1 = vx_load_expand(inp + ofs_k + nlanes);
+                                        s0 = v_max(s0, v0);
+                                        s1 = v_max(s1, v1);
+                                    }
+                                    v_pack_store(out + x0*C0 + c, s0);
+                                    v_pack_store(out + x0*C0 + c + nlanes, s1);
+                                }
+                            }
+                        }
+                        if (x0 == W)
+                            break;
+                        x1 = inner_x1;
+                        if (nlanes == C0) {
+                            for (; x0 < x1; x0++) {
+                                int xi_ = x0*SX - padX0;
+                                const _Tp* inp_xi = inp + ((Hi*zi_ + yi_)*Wi + xi_)*C0;
+
+                                v_float32 s0 = vx_load_expand(inp_xi + ofstab[0]);
+                                for (int k = 1; k < ksize; k++)
+                                    s0 = v_max(s0, vx_load_expand(inp_xi + ofstab[k]));
+                                v_pack_store(out + x0*C0, s0);
+                            }
+                        } else if (nlanes*2 == C0) {
+                            for (; x0 < x1; x0++) {
+                                int xi_ = x0*SX - padX0;
+                                const _Tp* inp_xi = inp + ((Hi*zi_ + yi_)*Wi + xi_)*C0;
+
+                                int ofs_k = ofstab[0];
+                                v_float32 s0 = vx_load_expand(inp_xi + ofs_k);
+                                v_float32 s1 = vx_load_expand(inp_xi + ofs_k + nlanes);
+                                for (int k = 1; k < ksize; k++) {
+                                    ofs_k = ofstab[k];
+                                    s0 = v_max(s0, vx_load_expand(inp_xi + ofs_k));
+                                    s1 = v_max(s1, vx_load_expand(inp_xi + ofs_k + nlanes));
+                                }
+                                v_pack_store(out + x0*C0, s0);
+                                v_pack_store(out + x0*C0 + nlanes, s1);
+                            }
+                        } else {
+                            for (; x0 < x1; x0++) {
+                                int xi_ = x0*SX - padX0;
+                                for (int c = 0; c < C0; c += nlanes*4) {
+                                    const _Tp* inp_xi = inp + ((Hi*zi_ + yi_)*Wi + xi_)*C0;
+
+                                    int ofs_k = ofstab[0];
+                                    v_float32 s0 = vx_load_expand(inp_xi + ofs_k);
+                                    v_float32 s1 = vx_load_expand(inp_xi + ofs_k + nlanes);
+                                    v_float32 s2 = vx_load_expand(inp_xi + ofs_k + nlanes*2);
+                                    v_float32 s3 = vx_load_expand(inp_xi + ofs_k + nlanes*3);
+                                    for (int k = 1; k < ksize; k++) {
+                                        ofs_k = ofstab[k];
+                                        s0 = v_max(s0, vx_load_expand(inp_xi + ofs_k));
+                                        s1 = v_max(s1, vx_load_expand(inp_xi + ofs_k + nlanes));
+                                        s2 = v_max(s2, vx_load_expand(inp_xi + ofs_k + nlanes*2));
+                                        s3 = v_max(s3, vx_load_expand(inp_xi + ofs_k + nlanes*3));
+                                    }
+                                    v_pack_store(out + x0*C0 + c, s0);
+                                    v_pack_store(out + x0*C0 + c + nlanes, s1);
+                                    v_pack_store(out + x0*C0 + c + nlanes*2, s2);
+                                    v_pack_store(out + x0*C0 + c + nlanes*3, s3);
+                                }
+                            }
+                        }
+                        x1 = W;
+                    }
+                }
+            }
+        }
+    });
+}
+
+static void maxPool16f(const void* inp_, void* out_, const ConvState& cs)
+{
+    maxPool16xf((const hfloat*)inp_, (hfloat*)out_, cs);
+}
+
+static void maxPool16bf(const void* inp_, void* out_, const ConvState& cs)
+{
+    maxPool16xf((const bfloat*)inp_, (bfloat*)out_, cs);
+}
+#endif
+
+typedef void (*MaxPoolFunc)(const void* inp, void* out, const ConvState& cs);
+
+class MaxPoolLayerImpl : public MaxPoolLayer
+{
+public:
+    MaxPoolLayerImpl(const LayerParams& params)
+    {
+        setParamsFrom(params);
+        auto_pad = getAutoPadding(params);
+        kernel_shape = params.getVector<int>("kernel_size");
+        strides = params.getVector<int>("stride");
+        dilations = params.getVector<int>("dilation");
+        pads = params.getVector<int>("pad");
+        ceil_mode = params.get<bool>("ceil_mode", false);
+        storage_order = params.get<int>("storage_order", 0);
+    }
+
+    virtual std::ostream& dumpAttrs(std::ostream& strm, int indent) const CV_OVERRIDE
+    {
+        prindent(strm, indent);
+        strm << "kernel_size: [";
+        for (size_t k = 0; k < kernel_shape.size(); k++)
+            strm << (k > 0 ? ", " : "") << kernel_shape[k];
+        strm << "],\n";
+
+        prindent(strm, indent);
+        strm << "dilation: [";
+        for (size_t k = 0; k < dilations.size(); k++)
+            strm << (k > 0 ? ", " : "") << dilations[k];
+        strm << "],\n";
+
+        prindent(strm, indent);
+        strm << "pad: [";
+        for (size_t k = 0; k < pads.size(); k++)
+            strm << (k > 0 ? ", " : "") << pads[k];
+        strm << "],\n";
+
+        prindent(strm, indent);
+        strm << "stride: [";
+        for (size_t k = 0; k < strides.size(); k++)
+            strm << (k > 0 ? ", " : "") << strides[k];
+        strm << "],\n";
+
+        if (outputs.size() > 1u) {
+            prindent(strm, indent);
+            strm << "storage_order: " << storage_order << ",\n";
+        }
+        return strm;
+    }
+
+    virtual bool supportBackend(int backendId) CV_OVERRIDE
+    {
+        return backendId == DNN_BACKEND_OPENCV;
+    }
+
+    virtual int64_t getFLOPS(const std::vector<MatShape> &inputs,
+                           const std::vector<MatShape> &outputs) const CV_OVERRIDE
+    {
+        CV_Assert(inputs.size() == 1);
+        CV_Assert(outputs.size() == 1);
+        int ksize = 1;
+        for (auto sz: kernel_shape) ksize *= sz;
+        return (int64_t)(inputs[0].total()*ksize);
+    }
+
+    int inferType(int inptype0) const
+    {
+        return inptype0;
+    }
+
+    virtual void getTypes(const std::vector<MatType>& inptypes,
+                          const int, const int,
+                          std::vector<MatType>& outtypes,
+                          std::vector<MatType>& temptypes) const CV_OVERRIDE
+    {
+        int ninputs = (int)inptypes.size();
+        CV_Assert(ninputs == 1);
+
+        outtypes.assign(1, inferType(inptypes[0]));
+        temptypes.clear();
+    }
+
+    virtual bool getMemoryShapes(const std::vector<MatShape>& inpshapes,
+                                 const int,
+                                 std::vector<MatShape> &outshapes,
+                                 std::vector<MatShape> &tempshapes) const CV_OVERRIDE
+    {
+        CV_Assert(outputs.size() == 1u);
+        size_t ninputs = inpshapes.size();
+        CV_Assert(ninputs == 1);
+
+        outshapes.assign(1, convInferShape(inpshapes[0], MatShape(),
+                                           kernel_shape, 0, strides, dilations,
+                                           pads, auto_pad, ceil_mode));
+        tempshapes.clear();
+        return true;
+    }
+
+    int getLayouts(const std::vector<DataLayout>& actualInputs,
+                    std::vector<DataLayout>& desiredInputs,
+                    const int requiredOutputs,
+                    std::vector<DataLayout>& outputs) const CV_OVERRIDE
+    {
+        CV_Assert(actualInputs.size() == 1u);
+        desiredInputs.assign(1, DATA_LAYOUT_BLOCK);
+        outputs.assign(requiredOutputs, DATA_LAYOUT_BLOCK);
+        return getNetImpl(this)->defaultC0;
+    }
+
+    void finalize(InputArrayOfArrays, OutputArrayOfArrays outputs_arr) CV_OVERRIDE
+    {
+    }
+
+    void forward(InputArrayOfArrays inputs_arr,
+                 OutputArrayOfArrays outputs_arr,
+                 OutputArrayOfArrays) CV_OVERRIDE
+    {
+        size_t ninputs = inputs_arr.total();
+        CV_Assert(ninputs == 1);
+
+        int inptype = inputs_arr.type(0);
+        MatShape inpshape = inputs_arr.shape(0);
+        MatShape outshape = convInferShape(inpshape, MatShape(),
+                                           kernel_shape, 0, strides, dilations,
+                                           pads, auto_pad, ceil_mode);
+        int outKind = outputs_arr.kind();
+        CV_Assert(outKind == _InputArray::STD_VECTOR_MAT ||
+                  outKind == _InputArray::STD_VECTOR_UMAT);
+
+        ConvState cs;
+        cs.initPooling(inpshape, outshape, kernel_shape, strides,
+                       dilations, pads, auto_pad, ceil_mode);
+
+        if (outKind == _InputArray::STD_VECTOR_MAT) {
+            Mat inp = inputs_arr.getMat(0);
+            std::vector<Mat>& outs = outputs_arr.getMatVecRef();
+            outs.resize(1);
+            outs[0].fit(outshape, inptype);
+            runOp(inp, outs[0], cs);
+        } else {
+            // [TODO] more efficient OpenCL implementation
+            Mat inp = inputs_arr.getMat(0);
+            std::vector<UMat>& outs = outputs_arr.getUMatVecRef();
+            outs.resize(1);
+            outs[0].fit(outshape, inptype);
+            Mat temp(outshape, inptype);
+            runOp(inp, temp, cs);
+            temp.copyTo(outs[0]);
+        }
+    }
+
+    void runOp(const Mat& inp, Mat& out, const ConvState& cs)
+    {
+        int inptype = inp.type();
+        MaxPoolFunc func =
+            inptype == CV_32F ? maxPool32f :
+            /*inptype == CV_16F ? maxPool16f :
+            inptype == CV_16BF ? maxPool16bf :*/
+            nullptr;
+
+        CV_Assert(func != nullptr && "MaxPool: unsupported data type");
+        func(inp.data, out.data, cs);
+    }
+};
+
+Ptr<MaxPoolLayer> MaxPoolLayer::create(const LayerParams& params)
+{
+    return Ptr<MaxPoolLayer>(new MaxPoolLayerImpl(params));
+}
+
+}}

modules/dnn/src/layers/nary_eltwise_layers.cpp

@@ -3,6 +3,7 @@
 // of this distribution and at http://opencv.org/license.html.
 
 #include "../precomp.hpp"
+#include "../net_impl.hpp"
 #include "layers_common.hpp"
 #include "../op_cuda.hpp"
 #include "../op_cann.hpp"
@@ -170,39 +171,12 @@ class NaryEltwiseLayerImpl CV_FINAL : public NaryEltwiseLayer
 {
     NaryEltwiseHelper helper;
 public:
-    enum class OPERATION
-    {
-        AND = 0,
-        EQUAL,
-        GREATER,
-        GREATER_EQUAL,
-        LESS,
-        LESS_EQUAL,
-        OR,
-        POW,
-        XOR,
-        BITSHIFT,
-        MAX,
-        MEAN,
-        MIN,
-        MOD,  // Integer Mod. Reminder's sign = Divisor's sign.
-        FMOD, // Floating-point Mod. Reminder's sign = Dividend's sign.
-        PROD,
-        SUB,
-        SUM,
-        ADD,
-        DIV,
-        WHERE,
-        BITWISE_AND,
-        BITWISE_OR,
-        BITWISE_XOR,
-    } op;
+    std::string operation;
 
     NaryEltwiseLayerImpl(const LayerParams& params)
     {
         setParamsFrom(params);
-
-        String operation = toLowerCase(params.get<String>("operation", "sum"));
+        operation = toLowerCase(params.get<String>("operation", "sum"));
 
         if (operation == "equal")
             op = OPERATION::EQUAL;
@@ -256,6 +230,13 @@ public:
             CV_Error(cv::Error::StsBadArg, "Unknown operation type \"" + operation + "\"");
     }
 
+    virtual std::ostream& dumpAttrs(std::ostream& strm, int indent) const CV_OVERRIDE
+    {
+        prindent(strm, indent);
+        strm << "operation: \"" << operation << "\",\n";
+        return strm;
+    }
+
     virtual bool supportBackend(int backendId) CV_OVERRIDE
     {
 #ifdef HAVE_CANN
@@ -294,36 +275,46 @@ public:
         return backendId == DNN_BACKEND_OPENCV;
     }
 
-    static MatShape findCommonShape(std::vector<MatShape> shapes)
+    // [TODO] move it to MatShape
+    static MatShape findCommonShape(const std::vector<MatShape>& shapes)
     {
-        CV_Assert(!shapes.empty());
-        const size_t dim = std::max_element(shapes.begin(), shapes.end(),
-                                            [](const MatShape& a, const MatShape& b)
-                                            { return a.size() < b.size(); })->size();
+        size_t i, ninputs = shapes.size();
+        CV_Assert(ninputs > 0u);
 
-        for (auto& shape : shapes)
-        {
-            shape.insert(shape.begin(), dim - shape.size(), 1);
+        int C0 = shapes[0].C, dims0 = shapes[0].dims, maxdims = dims0;
+        bool constC = true;
+        bool allBlock = true;
+        bool constDims = true;
+        for (i = 0; i < ninputs; i++) {
+            const MatShape& inpShape = shapes[i];
+            int dims = inpShape.dims;
+            allBlock = allBlock && inpShape.layout == DATA_LAYOUT_BLOCK;
+            constC = constC && inpShape.C == C0;
+            constDims = constDims && dims == dims0;
+            maxdims = std::max(maxdims, dims);
         }
 
-        MatShape outShape(dim, 1);
-        for (size_t i = 0; i < dim; ++i)
+        MatShape outShape(maxdims, 1);
+        if (allBlock && constC && constDims) {
+            outShape.layout = DATA_LAYOUT_BLOCK;
+            outShape.C = C0;
+        }
+
+        for (i = 0; i < ninputs; i++)
         {
-            for (const auto& shape : shapes)
-            {
-                if (shape[i] != outShape[i])
-                {
-                    CV_Assert(shape[i] == 1 || outShape[i] == 1);
-                    if (outShape[i] == 1)
-                        outShape[i] = shape[i];
-                }
+            const MatShape& inpShape = shapes[i];
+            int dims = inpShape.dims, delta = maxdims - dims;
+            for (int j = 0; j < maxdims; j++) {
+                int inpsz = j < delta ? 1 : inpShape[j - delta];
+                int outsz = outShape[j];
+                CV_Assert(inpsz == outsz || inpsz == 1 || outsz == 1);
+                outShape[j] = inpsz != 1 ? inpsz : outsz;
             }
         }
 
         return outShape;
     }
 
-
     virtual void finalize(InputArrayOfArrays inputs_arr, OutputArrayOfArrays outputs_arr) CV_OVERRIDE {
         std::vector<Mat> inputs, outputs;
         inputs_arr.getMatVector(inputs);
@@ -428,6 +419,34 @@ public:
             outputs.assign(requiredOutputs, inputs[0]);
     }
 
+    int getLayouts(const std::vector<DataLayout>& actualInputs,
+                   std::vector<DataLayout>& desiredInputs,
+                   const int requiredOutputs,
+                   std::vector<DataLayout>& outputs) const CV_OVERRIDE
+    {
+        auto* netimpl_ = getNetImpl(this);
+        DataLayout defaultLayout = netimpl_->originalLayout;
+        size_t ninputs = actualInputs.size(), nblockInputs = 0;
+        CV_Assert(ninputs >= 1u);
+        for (size_t i = 0; i < ninputs; i++) {
+            DataLayout layout = actualInputs[i];
+            nblockInputs += layout == DATA_LAYOUT_BLOCK;
+        }
+
+        desiredInputs = actualInputs;
+        if (nblockInputs == ninputs) {
+            outputs.assign(requiredOutputs, DATA_LAYOUT_BLOCK);
+        } else {
+            if (nblockInputs < ninputs) {
+                for (size_t i = 0; i < ninputs; i++) {
+                    DataLayout layout = actualInputs[i];
+                    desiredInputs[i] = layout == DATA_LAYOUT_BLOCK ? defaultLayout : layout;
+                }
+            }
+            outputs.assign(requiredOutputs, DATA_LAYOUT_UNKNOWN);
+        }
+        return outputs[0] == DATA_LAYOUT_BLOCK ? netimpl_->defaultC0 : 0;
+    }
 
     template <typename T, typename RESULT_T, typename Functor>
     void binary_forward_impl(const Functor& op, int ndims, const std::vector<int>& shape,

modules/dnn/src/layers/transform_layout_layer.cpp

@@ -0,0 +1,411 @@
+// 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 "../precomp.hpp"
+#include "layers_common.hpp"
+#include "../net_impl.hpp"
+
+namespace cv
+{
+namespace dnn
+{
+
+static MatShape inferTransformLayoutShape(const MatShape& inpshape_,
+                                          DataLayout outlayout,
+                                          DataLayout defaultLayout,
+                                          int C0)
+{
+    MatShape inpshape = inpshape_;
+    if (inpshape.layout == DATA_LAYOUT_UNKNOWN) {
+        inpshape.layout = defaultLayout;
+    }
+
+    return inpshape.toLayout(outlayout, C0);
+}
+
+template<typename _Tp>
+static inline void transpose8x8(const _Tp* inp_, size_t istep,
+                                _Tp* out_, size_t ostep)
+{
+#if CV_SIMD128
+    if constexpr (sizeof(_Tp) == 4u) {
+        const uint32_t* inp = (const uint32_t*)inp_;
+        uint32_t* out = (uint32_t*)out_;
+        v_uint32x4 a0, a1, a2, a3, b0, b1, b2, b3;
+
+        a0 = v_load(inp + istep*0);
+        a1 = v_load(inp + istep*1);
+        a2 = v_load(inp + istep*2);
+        a3 = v_load(inp + istep*3);
+        v_transpose4x4(a0, a1, a2, a3, b0, b1, b2, b3);
+        v_store(out + ostep*0, b0);
+        v_store(out + ostep*1, b1);
+        v_store(out + ostep*2, b2);
+        v_store(out + ostep*3, b3);
+
+        a0 = v_load(inp + istep*0 + 4);
+        a1 = v_load(inp + istep*1 + 4);
+        a2 = v_load(inp + istep*2 + 4);
+        a3 = v_load(inp + istep*3 + 4);
+        v_transpose4x4(a0, a1, a2, a3, b0, b1, b2, b3);
+        v_store(out + ostep*4, b0);
+        v_store(out + ostep*5, b1);
+        v_store(out + ostep*6, b2);
+        v_store(out + ostep*7, b3);
+
+        a0 = v_load(inp + istep*4);
+        a1 = v_load(inp + istep*5);
+        a2 = v_load(inp + istep*6);
+        a3 = v_load(inp + istep*7);
+        v_transpose4x4(a0, a1, a2, a3, b0, b1, b2, b3);
+        v_store(out + ostep*0 + 4, b0);
+        v_store(out + ostep*1 + 4, b1);
+        v_store(out + ostep*2 + 4, b2);
+        v_store(out + ostep*3 + 4, b3);
+
+        a0 = v_load(inp + istep*4 + 4);
+        a1 = v_load(inp + istep*5 + 4);
+        a2 = v_load(inp + istep*6 + 4);
+        a3 = v_load(inp + istep*7 + 4);
+        v_transpose4x4(a0, a1, a2, a3, b0, b1, b2, b3);
+        v_store(out + ostep*4 + 4, b0);
+        v_store(out + ostep*5 + 4, b1);
+        v_store(out + ostep*6 + 4, b2);
+        v_store(out + ostep*7 + 4, b3);
+    } else
+#endif
+    {
+    for (int i = 0; i < 8; i++)
+        for (int j = 0; j < 8; j++)
+            out_[i * ostep + j] = inp_[j * istep + i];
+    }
+}
+
+template <typename _Tp>
+void transformLayoutInterleave_(const _Tp* inp_base, _Tp* out_base, int C, size_t len,
+                                int nc, int nzc, size_t dlen)
+{
+    size_t i = 0;
+    for (; i + 7u < dlen; i += 8u)
+    {
+        int c = 0;
+
+        for (; c + 7u < nzc; c += 8u) {
+            transpose8x8<_Tp>(inp_base + c * len + i, len, out_base + i * nc + c, nc);
+        }
+
+        for (; c < nzc; ++c) {
+            _Tp* outptr = out_base + i * nc + c;
+            const _Tp* inptr = inp_base + c * len + i;
+            outptr[0 * nc] = inptr[0];
+            outptr[1 * nc] = inptr[1];
+            outptr[2 * nc] = inptr[2];
+            outptr[3 * nc] = inptr[3];
+            outptr[4 * nc] = inptr[4];
+            outptr[5 * nc] = inptr[5];
+            outptr[6 * nc] = inptr[6];
+            outptr[7 * nc] = inptr[7];
+        }
+
+        for (; c < nc; ++c) {
+            _Tp* outptr = out_base + i * nc + c;
+            outptr[0 * nc] = (_Tp)0; outptr[1 * nc] = (_Tp)0; outptr[2 * nc] = (_Tp)0; outptr[3 * nc] = (_Tp)0;
+            outptr[4 * nc] = (_Tp)0; outptr[5 * nc] = (_Tp)0; outptr[6 * nc] = (_Tp)0; outptr[7 * nc] = (_Tp)0;
+        }
+    }
+    for (; i < dlen; ++i) {
+        _Tp* outptr = out_base + i * nc;
+        for (int c = 0; c < nc; ++c) {
+            outptr[c] = c < nzc ? inp_base[c*len + i] : (_Tp)0;
+        }
+    }
+}
+
+template <typename _Tp>
+void transformLayoutDeinterleave_(const _Tp* inp_base, _Tp* out_base, int C, size_t len,
+                                  int nc, int nzc, size_t dlen)
+{
+    size_t i = 0;
+    for (; i + 7u < dlen; i += 8u)
+    {
+        int c = 0;
+
+        for (; c + 7u < nzc; c += 8u)
+        {
+            transpose8x8<_Tp>(inp_base + i * nc + c, nc, out_base + c * len + i, len);
+        }
+
+        for (; c < nzc; ++c)
+        {
+            const _Tp* inptr = inp_base + i * nc + c;
+            _Tp* outptr = out_base + c * len + i;
+            outptr[0] = inptr[0 * nc];
+            outptr[1] = inptr[1 * nc];
+            outptr[2] = inptr[2 * nc];
+            outptr[3] = inptr[3 * nc];
+            outptr[4] = inptr[4 * nc];
+            outptr[5] = inptr[5 * nc];
+            outptr[6] = inptr[6 * nc];
+            outptr[7] = inptr[7 * nc];
+        }
+    }
+    for (; i < dlen; ++i)
+    {
+        const _Tp* inptr = inp_base + i * nc;
+        for (int c = 0; c < nzc; ++c) {
+            out_base[c*len + i] = inptr[c];
+        }
+    }
+}
+
+typedef void (*TransformLayoutFunc)(const void* inp, void* out, int C, size_t planesize,
+                                    int nc, int nzc, size_t dlen);
+
+#undef DECL_TRANSFORM_LAYOUT
+#define DECL_TRANSFORM_LAYOUT(suffix, _Tp) \
+static void transformLayoutInterleave_##suffix(const void* inp, void* out, int C, size_t planesize, \
+                                               int nc, int nzc, size_t dlen) \
+{ \
+    transformLayoutInterleave_((const _Tp*)inp, (_Tp*)out, C, planesize, nc, nzc, dlen); \
+} \
+static void transformLayoutDeinterleave_##suffix(const void* inp, void* out, int C, size_t planesize, \
+                                                 int nc, int nzc, size_t dlen) \
+{ \
+    transformLayoutDeinterleave_((const _Tp*)inp, (_Tp*)out, C, planesize, nc, nzc, dlen); \
+}
+
+DECL_TRANSFORM_LAYOUT(8u, uint8_t)
+DECL_TRANSFORM_LAYOUT(16u, uint16_t)
+DECL_TRANSFORM_LAYOUT(32u, uint32_t)
+DECL_TRANSFORM_LAYOUT(64u, uint64_t)
+
+TransformLayoutFunc getTransformLayoutFunc(DataLayout inplayout, DataLayout outlayout, size_t esz)
+{
+    if (inplayout == DATA_LAYOUT_NCHW &&
+        (outlayout == DATA_LAYOUT_BLOCK || outlayout == DATA_LAYOUT_NHWC)) {
+        return esz == 1u ? transformLayoutInterleave_8u :
+               esz == 2u ? transformLayoutInterleave_16u :
+               esz == 4u ? transformLayoutInterleave_32u :
+               esz == 8u ? transformLayoutInterleave_64u : nullptr;
+    }
+    if ((inplayout == DATA_LAYOUT_BLOCK || inplayout == DATA_LAYOUT_NHWC) &&
+        outlayout == DATA_LAYOUT_NCHW) {
+        return esz == 1u ? transformLayoutDeinterleave_8u :
+               esz == 2u ? transformLayoutDeinterleave_16u :
+               esz == 4u ? transformLayoutDeinterleave_32u :
+               esz == 8u ? transformLayoutDeinterleave_64u : nullptr;
+    }
+    return nullptr;
+}
+
+void transformLayout(const Mat& inp, Mat& out,
+                     DataLayout outlayout,
+                     DataLayout defaultLayout,
+                     int C0)
+{
+    CV_Assert(defaultLayout == DATA_LAYOUT_NCHW || defaultLayout == DATA_LAYOUT_NHWC);
+    CV_Assert(outlayout == DATA_LAYOUT_BLOCK || outlayout == DATA_LAYOUT_NCHW || outlayout == DATA_LAYOUT_NHWC);
+
+    MatShape inpshape = inp.size;
+    /*if (inpshape.layout == DATA_LAYOUT_NCHW &&
+        inpshape.dims == 4 && inpshape[1] == 272 && inpshape[2] == 14 && inpshape[3] == 14) {
+        putchar('.');
+    }*/
+
+    if (inpshape.layout == DATA_LAYOUT_UNKNOWN) {
+        inpshape.layout = defaultLayout;
+    }
+    DataLayout inplayout = inpshape.layout;
+    MatShape outshape = inferTransformLayoutShape(inpshape, outlayout, defaultLayout, C0);
+    out.fit(outshape, inp.type());
+
+    if (inp.empty())
+        return;
+
+    if (inplayout == outlayout) {
+        inp.copyTo(out);
+        return;
+    }
+
+    CV_Assert_N(inp.isContinuous(), out.isContinuous());
+
+    size_t esz = inp.elemSize();
+    TransformLayoutFunc kernel = getTransformLayoutFunc(inplayout, outlayout, esz);
+    CV_Assert(kernel != nullptr);
+
+    int N = inpshape[0];
+    int C = inpshape.channels();
+    C0 = inplayout == DATA_LAYOUT_BLOCK ? inpshape.back() : C0;
+    int C1 = (C + C0 - 1) / C0;
+
+    size_t planesize = 1;
+    int inp_sp0 = inplayout == DATA_LAYOUT_NHWC ? 1 : 2;
+    int inp_sp1 = inplayout == DATA_LAYOUT_NCHW ? inpshape.dims : inpshape.dims-1;
+    for (int i = inp_sp0; i < inp_sp1; i++) {
+        planesize *= (size_t)inpshape[i];
+    }
+
+    size_t total = N*C1*planesize*C0;
+    constexpr size_t min_elems_per_chunk = 1 << 17;
+    int nblocks = int((total + min_elems_per_chunk/2) / min_elems_per_chunk);
+    nblocks = std::clamp(nblocks, 1, 128);
+    nblocks = (nblocks + N*C1 - 1)/(N*C1);
+
+    parallel_for_(Range(0, N*C1*nblocks), [&](const Range& range)
+    {
+        int dchunk = 1u;
+        bool interleave = inplayout == DATA_LAYOUT_NCHW;
+        const uint8_t* inptr = (const uint8_t*)inp.data;
+        uint8_t* outptr = (uint8_t*)out.data;
+
+        for (int chunk = range.start; chunk < range.end; chunk += dchunk)
+        {
+            int n = chunk/(C1*nblocks);
+            int c1 = (chunk % (C1*nblocks))/nblocks;
+            int block = chunk % nblocks;
+            int nc = C0;
+            int nzc = std::min(nc, C - c1*C0);
+            dchunk = std::min(nblocks - block, range.end - chunk);
+            size_t block_start = block * planesize / nblocks;
+            size_t block_end = (block + dchunk) * planesize / nblocks;
+            size_t dlen = block_end - block_start;
+            size_t inpofs = ((n * C1 + c1) * planesize + block_start) * nc * esz;
+            size_t outofs = ((n * C + c1 * C0) * planesize + block_start) * esz;
+            if (interleave) {
+                std::swap(inpofs, outofs);
+            }
+
+            kernel(inptr + inpofs, outptr + outofs, C, planesize, nc, nzc, dlen);
+        }
+    });
+}
+
+class TransformLayoutLayerImpl : public TransformLayoutLayer
+{
+public:
+    TransformLayoutLayerImpl(const LayerParams& params)
+    {
+        setParamsFrom(params);
+        layout = (DataLayout)params.get<int>("layout");
+        C0 = params.get<int>("C0", 1);
+    }
+
+    virtual std::ostream& dumpAttrs(std::ostream& strm, int indent) const CV_OVERRIDE
+    {
+        prindent(strm, indent);
+        strm << "target_layout: \"" << layoutToString(layout) << "\",\n";
+
+        if (layout == DATA_LAYOUT_BLOCK) {
+            prindent(strm, indent);
+            strm << "C0: " << C0 << ",\n";
+        }
+        return strm;
+    }
+
+    virtual bool alwaysSupportInplace() const CV_OVERRIDE
+    {
+        return false;
+    }
+
+    virtual int64_t getFLOPS(const std::vector<MatShape> &inputs,
+                             const std::vector<MatShape> &outputs) const CV_OVERRIDE
+    {
+        CV_Assert(inputs.size() == 1);
+        CV_Assert(outputs.size() == 1);
+        // probably, there should be a coefficient in the case of complex reduction functions
+        return (int64_t)std::max(inputs[0].total(), outputs[0].total());
+    }
+
+    virtual void getTypes(const std::vector<MatType>& inptypes,
+                          const int, const int,
+                          std::vector<MatType>& outtypes,
+                          std::vector<MatType>& temptypes) const CV_OVERRIDE
+    {
+        int ninputs = (int)inptypes.size();
+        CV_Assert(ninputs == 1);
+
+        outtypes.assign(1, inptypes[0]);
+        temptypes.clear();
+    }
+
+    MatShape inferShape(const MatShape& inpshape_) const
+    {
+        return inferTransformLayoutShape(inpshape_, layout,
+                                         getNetImpl(this)->originalLayout, C0);
+    }
+
+    virtual bool getMemoryShapes(const std::vector<MatShape>& inpshapes,
+                                 const int,
+                                 std::vector<MatShape> &outshapes,
+                                 std::vector<MatShape> &tempshapes) const CV_OVERRIDE
+    {
+        size_t ninputs = inpshapes.size();
+        CV_Assert(ninputs == 1);
+
+        outshapes.assign(1, inferShape(inpshapes[0]));
+        tempshapes.clear();
+        return true;
+    }
+
+    void finalize(InputArrayOfArrays, OutputArrayOfArrays outputs_arr) CV_OVERRIDE
+    {
+    }
+
+    void forward(InputArrayOfArrays inputs_arr,
+                 OutputArrayOfArrays outputs_arr,
+                 OutputArrayOfArrays) CV_OVERRIDE
+    {
+        size_t ninputs = inputs_arr.total();
+        CV_Assert(ninputs == 1);
+
+        int inptype = inputs_arr.type(0);
+        MatShape inpshape = inputs_arr.shape(0);
+        MatShape outshape = inferShape(inpshape);
+        int outKind = outputs_arr.kind();
+        CV_Assert(outKind == _InputArray::STD_VECTOR_MAT ||
+                  outKind == _InputArray::STD_VECTOR_UMAT);
+
+        if (outKind == _InputArray::STD_VECTOR_MAT) {
+            Mat inp = inputs_arr.getMat(0);
+            std::vector<Mat>& outs = outputs_arr.getMatVecRef();
+            outs.resize(1);
+            outs[0].fit(outshape, inptype);
+            runOp(inp, outs[0]);
+        } else {
+            // [TODO] more efficient OpenCL implementation
+            Mat inp = inputs_arr.getMat(0);
+            std::vector<UMat>& outs = outputs_arr.getUMatVecRef();
+            outs.resize(1);
+            outs[0].fit(outshape, inptype);
+            Mat temp(outshape, inptype);
+            runOp(inp, temp);
+            temp.copyTo(outs[0]);
+        }
+    }
+
+    void runOp(const Mat& inp, Mat& out)
+    {
+        DataLayout origLayout = getNetImpl(this)->originalLayout;
+        transformLayout(inp, out, layout, origLayout, C0);
+#if 0
+        Mat temp;
+        transformLayout(out, temp, layout == DATA_LAYOUT_BLOCK ? origLayout : DATA_LAYOUT_BLOCK, origLayout, C0);
+        double err = norm(temp, inp, NORM_INF);
+        size_t i, N = inp.total();
+        const float* inpdata = inp.ptr<float>();
+        const float* tempdata = temp.ptr<float>();
+        for (i = 0; i < N; i++) {
+            CV_Assert_N(!cvIsNaN(inpdata[i]), !cvIsNaN(tempdata[i]));
+        }
+        CV_Assert(err == 0.);
+#endif
+    }
+};
+
+Ptr<TransformLayoutLayer> TransformLayoutLayer::create(const LayerParams& params)
+{
+    return Ptr<TransformLayoutLayer>(new TransformLayoutLayerImpl(params));
+}
+
+}}

modules/dnn/src/net_impl.cpp

@@ -67,6 +67,7 @@ Net::Impl::Impl()
     // onnx_opset = 0;
 
     accuracy = CV_32F;
+    defaultC0 = DEFAULT_C0;
     enableFP16 = haveFP16 = false;
     // FP16 is not ready yet in the new DNN engine
     // Ticket: https://github.com/opencv/opencv/issues/26196

modules/dnn/src/net_impl.hpp

@@ -103,6 +103,11 @@ struct Net::Impl : public detail::NetImplBase
     int globGraphIdx;
 
     int accuracy;
+    // if you change DEFAULT_C0/defaultC0, don't forget to update
+    // implementation of convolution, convTranspose,
+    // maxpool, avgpool, resize, pad ... where defaultC0 is accessed and used.
+    enum { DEFAULT_C0 = 8 };
+    int defaultC0;
     bool enableFP16, haveFP16;
     bool prepared; // need to rerun graph transformations/optimizations
     bool finalizeLayers; // need to initialize each layer
@@ -430,16 +435,24 @@ struct Net::Impl : public detail::NetImplBase
     std::ostream& dumpTypeShape(std::ostream& strm, int type, const MatShape& shape) const;
     std::ostream& dump(std::ostream& strm);
 
+    ///////////////// various graph transformations ///////////////////////
+
     // infers all types
     void inferTypes();
     // infers all shapes
     void inferShapes(bool symbolic);
     // sets certain buffer index for each intermediate argument (Arg)
     void assignBuffers();
-    //void useBlockLayout();
-    void fuse();
+    // fuse batch norm, add bias and activation to convolution
+    void fuseBasic();
+    // replace constant sub-expressions with their results
     void constFold();
+    // make some operations (activation, batch norm, convolution) unary if
+    // all their arguments except for the 1st one are constant.
     void constArgs();
+    // insert transformLayout operations where necessary;
+    // use block layout for convolution, pooling and some other operations where it matters
+    void useBlockLayout();
 
 };  // Net::Impl
 
@@ -456,5 +469,9 @@ Net readNetFromONNX2_ORT(const String& onnxFile);
 #endif
 
 CV__DNN_INLINE_NS_END
+
+void transformLayout(const Mat& inp, Mat& out,
+                     DataLayout outlayout, DataLayout defaultLayout, int C0);
+
 }}  // namespace cv::dnn
 #endif  // __OPENCV_DNN_SRC_NET_IMPL_HPP__

modules/dnn/src/net_impl2.cpp

@@ -325,7 +325,7 @@ public:
         prindent(strm, subindent);
         strm << "],\n";
         prindent(strm, subindent);
-        strm << "nodes: [\n";
+        strm << "layers: [\n";
         size_t nlayers = prog_.size();
         for (size_t i = 0; i < nlayers; i++) {
             prindent(strm, argindent);
@@ -498,12 +498,9 @@ void Net::Impl::prepareForInference()
 
     if (!prepared) {
         constFold();
-        //inferTypes();
-        //constArgs();
-        //inferShapes(true);
-        //fuse();
-        //useBlockLayout();
-        //inferShapes(true);
+        constArgs();
+        useBlockLayout();
+        fuseBasic();
         assignBuffers();
         totalLayers = updateGraphOfs(mainGraph, 0, true);
         prepared = true;
@@ -843,9 +840,13 @@ void Net::Impl::traceArg(std::ostream& strm_, const char* prefix, size_t i, Arg
     }
     strm_ << "\n  Layout: " << layoutToString(shape.layout) << "\n";
     if (dumpdata && !constArg) {
-        // [TODO] when we support block layout, block-layout tensor
-        // should be converted to the original layout before printing it
-        pprint(strm_, m, 0, PPRINT_CONTEXT, PPRINT_ALL_THRESHOLD, '[');
+        Mat temp;
+        if (m.size.layout == DATA_LAYOUT_BLOCK) {
+            transformLayout(m, temp, originalLayout, originalLayout, m.size.C);
+        } else {
+            temp = m;
+        }
+        pprint(strm_, temp, 0, PPRINT_CONTEXT, PPRINT_ALL_THRESHOLD, '[');
         strm_ << "\n";
     }
 }
@@ -1122,6 +1123,16 @@ void Net::Impl::forwardGraph(Ptr<Graph>& graph, InputArrayOfArrays inputs_,
             }
         }
 
+        size_t ntemps = tempMats.size();
+        scratchBufs.resize(std::max(ntemps, scratchBufs.size()));
+        for (size_t i = 0; i < ntemps; i++) {
+            size_t newtotal_i = tempMats[i].total()*tempMats[i].elemSize();
+            size_t total_i = scratchBufs[i].total()*scratchBufs[i].elemSize();
+            if (newtotal_i > total_i) {
+                scratchBufs[i] = tempMats[i];
+            }
+        }
+
         timestamp = getTickCount() - timestamp;
         layersTimings[opidx + graph_ofs + 1] += timestamp;
 
@@ -1141,8 +1152,12 @@ void Net::Impl::forwardGraph(Ptr<Graph>& graph, InputArrayOfArrays inputs_,
         Arg out = gr_outputs[i];
         const Mat& outm = argTensor(out);
         if (isMainGraph) {
-            outputsVec[i].fit(outm.shape(), outm.type());
-            outm.copyTo(outputsVec[i]);
+            if (outm.size.layout == DATA_LAYOUT_BLOCK) {
+                transformLayout(outm, outputsVec[i], originalLayout, originalLayout, outm.size.C);
+            } else {
+                outputsVec[i].fit(outm.shape(), outm.type());
+                outm.copyTo(outputsVec[i]);
+            }
         } else {
             outputsVec[i] = outm;
         }

modules/dnn/src/onnx/onnx_importer2.cpp

@@ -997,19 +997,6 @@ void ONNXImporter2::parseArgMinMax(LayerParams& layerParams, const opencv_onnx::
     addLayer(layerParams, node_proto);
 }
 
-static void setCeilMode(LayerParams& layerParams)
-{
-    // auto_pad attribute is deprecated and uses ceil
-    if (layerParams.has("pad_mode"))
-    {
-        layerParams.set("ceil_mode", true);
-    }
-    else if (!layerParams.has("ceil_mode"))
-    {
-        layerParams.set("ceil_mode", false);
-    }
-}
-
 void ONNXImporter2::parseMaxUnpool(LayerParams& layerParams, const opencv_onnx::NodeProto& node_proto)
 {
     layerParams.type = "MaxUnpool";
@@ -1047,19 +1034,17 @@ void ONNXImporter2::parseMaxUnpool(LayerParams& layerParams, const opencv_onnx::
 
 void ONNXImporter2::parseMaxPool(LayerParams& layerParams, const opencv_onnx::NodeProto& node_proto)
 {
-    int depth = layerParams.get<int>("depth", CV_32F);
-    layerParams.type = (depth == CV_8S) ? "PoolingInt8" : "Pooling";
-    layerParams.set("pool", "MAX");
-    setCeilMode(layerParams);
+    CV_CheckEQ(node_outputs.size(), 1u, "the new engine does not support MaxPool with 2 outputs yet");
+    layerParams.type = "MaxPool";
     addLayer(layerParams, node_proto);
 }
 
 void ONNXImporter2::parseAveragePool(LayerParams& layerParams, const opencv_onnx::NodeProto& node_proto)
 {
-    layerParams.type = "Pooling";
-    layerParams.set("pool", "AVE");
-    setCeilMode(layerParams);
-    layerParams.set("ave_pool_padded_area", framework_name == "pytorch");
+    layerParams.type = "AveragePool";
+    if (!layerParams.has("count_include_pad")) {
+        layerParams.set("count_include_pad", framework_name == "pytorch");
+    }
     addLayer(layerParams, node_proto);
 }
 
@@ -1341,31 +1326,7 @@ void ONNXImporter2::parseInstanceNormalization(LayerParams& layerParams, const o
 
 void ONNXImporter2::parseBatchNormalization(LayerParams& layerParams, const opencv_onnx::NodeProto& node_proto)
 {
-    if (node_proto.input_size() != 5)
-        CV_Error(Error::StsNotImplemented, "Expected input, scale, bias, mean and var");
-
     layerParams.type = "BatchNorm2";
-
-    float eps = 1e-5f;
-    for (int i = 0; i < node_proto.attribute_size(); i++) {
-        const opencv_onnx::AttributeProto& attr = node_proto.attribute(i);
-        if (attr.name() == "epsilon") eps = attr.f();
-    }
-    layerParams.set("epsilon", eps);
-
-    bool isStatic = net.isConstArg(node_inputs[1]) && // scale
-                    net.isConstArg(node_inputs[2]) && // bias
-                    net.isConstArg(node_inputs[3]) && // mean
-                    net.isConstArg(node_inputs[4]);   // var
-
-    if (isStatic) {
-        layerParams.blobs.resize(4);
-        layerParams.blobs[0] = net.argTensor(node_inputs[3]); // mean
-        layerParams.blobs[1] = net.argTensor(node_inputs[4]); // var
-        layerParams.blobs[2] = net.argTensor(node_inputs[1]); // scale
-        layerParams.blobs[3] = net.argTensor(node_inputs[2]); // bias
-    }
-
     addLayer(layerParams, node_proto);
 }
 
@@ -1408,18 +1369,8 @@ void ONNXImporter2::parseConv(LayerParams& layerParams, const opencv_onnx::NodeP
 {
     int n_inputs = node_proto.input_size();
     CV_Assert(2 <= n_inputs && n_inputs <= 3);
-    layerParams.type = "Convolution";
-
-    if (net.isConstArg(node_inputs[1]) && (n_inputs == 2 || net.isConstArg(node_inputs[2]))) {
-        Mat weights = net.argTensor(node_inputs[1]);
-        layerParams.blobs.push_back(weights);
-        if (n_inputs > 2) {
-            Mat bias = net.argTensor(node_inputs[2]);
-            layerParams.blobs.push_back(bias);
-        }
-        n_inputs = 1;
-    }
-    addLayer(layerParams, node_proto, n_inputs);
+    layerParams.type = "Conv2";
+    addLayer(layerParams, node_proto);
 }
 
 void ONNXImporter2::parseConvTranspose(LayerParams& layerParams, const opencv_onnx::NodeProto& node_proto)
@@ -2843,6 +2794,7 @@ Net readNetFromONNX2(const String& onnxFile)
     Net net = importer.parseFile(onnxFile.c_str());
     if (net.getMainGraph()) {
         net.getImpl()->modelFileName = onnxFile;
+        //net.setTracingMode(DNN_TRACE_ALL);
     }
     return net;
 }

modules/dnn/test/test_backends.cpp

@@ -481,7 +481,7 @@ TEST_P(DNNTestNetwork, FastNeuralStyle_eccv16)
     Mat img = imread(findDataFile("dnn/googlenet_1.png"));
     Mat inp = blobFromImage(img, 1.0, Size(224, 224), Scalar(0.0, 0.0, 0.0), true, false);
     // Output image has values in range [0.0, 255.0].
-    float l1 = 5e-4, lInf = 1e-2;
+    float l1 = 6e-4, lInf = 1e-2;
     if (target == DNN_TARGET_MYRIAD)
     {
         l1 = 0.4;

modules/dnn/test/test_graph_simplifier.cpp

@@ -57,7 +57,7 @@ TEST_F(Test_Graph_Simplifier, LayerNormNoFusionSubGraph) {
     test("layer_norm_no_fusion", std::vector<std::string>{"NaryEltwise", "Reduce", "Sqrt"});
 }
 
-TEST_F(Test_Graph_Simplifier, ResizeSubgraph) {
+TEST_F(Test_Graph_Simplifier, DISABLED_ResizeSubgraph) {
     /* Test for 6 subgraphs:
         - GatherCastSubgraph
         - MulCastSubgraph

modules/dnn/test/test_model.cpp

@@ -838,4 +838,88 @@ TEST_P(Test_Model, TextDetectionByEAST)
 
 INSTANTIATE_TEST_CASE_P(/**/, Test_Model, dnnBackendsAndTargets());
 
+static void topK(const Mat& probs, std::vector<std::pair<int, float> >& result, int K)
+{
+    CV_Assert(probs.type() == CV_32F);
+    CV_Assert(probs.dims == 2 && probs.rows == 1);
+    int N = int(probs.total());
+    K = std::min(K, N);
+    std::vector<std::pair<float, int> > pairs(N);
+    for (int i = 0; i < N; i++) {
+        pairs[i] = {-probs.at<float>(i), i};
+    }
+    std::partial_sort(pairs.begin(), pairs.begin() + K, pairs.end());
+    result.resize(K);
+    for (int i = 0; i < K; i++) {
+        result[i] = {pairs[i].second, -pairs[i].first};
+    }
+}
+
+typedef testing::TestWithParam<Target> Reproducibility_ResNet50_ONNX;
+TEST_P(Reproducibility_ResNet50_ONNX, Accuracy)
+{
+    Target targetId = GetParam();
+    applyTestTag(targetId == DNN_TARGET_CPU ? CV_TEST_TAG_MEMORY_512MB : CV_TEST_TAG_MEMORY_1GB);
+    ASSERT_TRUE(ocl::useOpenCL() || targetId == DNN_TARGET_CPU || targetId == DNN_TARGET_CPU_FP16);
+
+    std::string modelname = _tf("onnx/models/resnet50v1.onnx", false);
+    Net net = readNetFromONNX(modelname);
+
+    net.setPreferableBackend(DNN_BACKEND_OPENCV);
+    net.setPreferableTarget(targetId);
+
+    if (targetId == DNN_TARGET_CPU_FP16)
+        net.enableWinograd(false);
+
+    //net.dumpToStream(std::cout);
+    //net.setTracingMode(DNN_TRACE_ALL);
+
+    std::string imgname = _tf("sqcat.png");
+    Mat image = imread(imgname);
+    Mat input = blobFromImage(image, 0.017, Size(224,224),
+                              Scalar(103.939, 116.779, 123.68),
+                              false, true, CV_32F);
+    ASSERT_TRUE(!input.empty());
+
+    Mat out;
+    double min_t = 0;
+    const int niters =
+#ifdef _DEBUG
+        1;
+#else
+        30;
+#endif
+
+    for (int i = 0; i < niters; i++) {
+        double t = (double)getTickCount();
+        net.setInput(input);
+        out = net.forward();
+        t = (double)getTickCount() - t;
+        min_t = i == 0 ? t : std::min(min_t, t);
+    }
+    printf("run time = %.2fms\n", min_t*1000./getTickFrequency());
+
+    std::vector<std::pair<int, float> > ref = {{285, 10.13}, {287, 9.68}, {283, 8.83}, {278, 8.56}, {279, 8.34}};
+    std::vector<std::pair<int, float> > res;
+    const int K = 5;
+
+    topK(out, res, K);
+    const float eps = 0.15f;
+
+    ASSERT_EQ(int(res.size()), K);
+
+    std::vector<int> reflabels(K), reslabels(K);
+    for (int i = 0; i < K; i++) {
+        reflabels[i] = ref[i].first;
+        reslabels[i] = res[i].first;
+    }
+    ASSERT_EQ(reflabels, reslabels);
+
+    for (int i = 0; i < K; i++) {
+        EXPECT_NEAR(ref[i].second, res[i].second, eps);
+    }
+}
+INSTANTIATE_TEST_CASE_P(/**/, Reproducibility_ResNet50_ONNX,
+                        testing::ValuesIn(getAvailableTargets(DNN_BACKEND_OPENCV)));
+
 }} // namespace

modules/dnn/test/test_onnx_conformance.cpp

@@ -1986,11 +1986,12 @@ TEST_P(Test_ONNX_conformance, Layer_Test)
     std::vector<Mat> ref_outputs;
 
     std::string prefix = cv::format("dnn/onnx/conformance/node/%s", test_case.name);
+    std::string model_path;
 
     Net net;
     try
     {
-        std::string model_path = findDataFile(prefix + "/model.onnx");
+        model_path = findDataFile(prefix + "/model.onnx");
 
         std::string test_data_dir = cv::utils::fs::join(cv::utils::fs::getParent(model_path), "test_data_set_0");
         std::vector<cv::String> inputFiles, outputFiles;
@@ -2097,6 +2098,7 @@ TEST_P(Test_ONNX_conformance, Layer_Test)
     std::vector<Mat> outputs;
     try
     {
+        //net.setTracingMode(DNN_TRACE_ALL);
         net.forward(outputs, layerNames);
     }
     catch (...)
@@ -2118,6 +2120,14 @@ TEST_P(Test_ONNX_conformance, Layer_Test)
         {
             if (ref_outputs.size() == 1)
             {
+                /*std::cout << "\n-------------------------------------------\nreference tensor:\n";
+                pprint(std::cout, ref_outputs[0], 0, 3, 100, '[');
+                std::cout << "\n";
+                std::cout << "\n-------------------------------------------\nabsdiff:\n";
+                Mat temp;
+                absdiff(ref_outputs[0], outputs[0], temp);
+                pprint(std::cout, temp, 0, 3, 100, '[');
+                std::cout << "\n";*/
                 // probably we found random unconnected layers.
                 normAssert(ref_outputs[0], outputs[0], "", default_l1, default_lInf);
             }

modules/objdetect/test/test_face.cpp

@@ -190,6 +190,8 @@ TEST(Objdetect_face_recognition, regression)
         Mat faces;
         faceDetector->detect(image, faces);
 
+        ASSERT_EQ(faces.rows, 1);
+
         Mat aligned_face;
         faceRecognizer->alignCrop(image, faces.row(0), aligned_face);