1
0
mirror of https://github.com/opencv/opencv.git synced 2026-07-29 23:33:05 +04:00

Merge pull request #26271 from fengyuentau:imgproc/warpperspective_opt

imgproc: add optimized warpPerspective linear kernels for inputs of type CV_8U/16U/32F+C1/C3/C4

Merge with https://github.com/opencv/opencv_extra/pull/1214

## Performance

### Apple Mac Mini (M2, 16GB memory)

```
Geometric mean (ms)

                                      Name of Test                                        base  patch   patch   
                                                                                                          vs    
                                                                                                         base   
                                                                                                      (x-factor)
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 8UC1)      0.397 0.119    3.34   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 16UC1)     0.412 0.155    2.65   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 32FC1)     0.382 0.134    2.85   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 8UC3)      0.562 0.201    2.80   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 16UC3)     0.580 0.279    2.07   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 32FC3)     0.477 0.269    1.78   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 8UC4)      0.536 0.221    2.43   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 16UC4)     0.662 0.328    2.02   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 32FC4)     0.511 0.325    1.57   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 8UC1)     0.433 0.171    2.54   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 16UC1)    0.452 0.204    2.21   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 32FC1)    0.410 0.180    2.27   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 8UC3)     0.624 0.243    2.57   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 16UC3)    0.636 0.331    1.92   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 32FC3)    0.511 0.315    1.62   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 8UC4)     0.604 0.281    2.15   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 16UC4)    0.712 0.393    1.81   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 32FC4)    0.552 0.368    1.50   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 8UC1)     0.768 0.214    3.58   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 16UC1)    0.743 0.260    2.86   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 32FC1)    0.722 0.235    3.07   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 8UC3)     1.091 0.333    3.28   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 16UC3)    1.035 0.453    2.29   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 32FC3)    0.955 0.442    2.16   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 8UC4)     1.097 0.364    3.01   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 16UC4)    1.141 0.547    2.09   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 32FC4)    1.015 0.591    1.72   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 8UC1)    1.012 1.006    1.01   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 16UC1)   0.996 1.060    0.94   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 32FC1)   0.930 0.993    0.94   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 8UC3)    1.560 1.260    1.24   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 16UC3)   1.374 1.410    0.97   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 32FC3)   1.212 1.292    0.94   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 8UC4)    1.702 1.354    1.26   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 16UC4)   1.554 1.522    1.02   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 32FC4)   1.342 1.435    0.94   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 8UC1)    1.561 0.364    4.29   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 16UC1)   1.444 0.406    3.56   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 32FC1)   1.423 0.394    3.61   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 8UC3)    2.177 0.533    4.08   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 16UC3)   2.006 0.689    2.91   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 32FC3)   1.907 0.688    2.77   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 8UC4)    2.213 0.581    3.81   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 16UC4)   2.238 0.810    2.76   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 32FC4)   2.072 1.055    1.96   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 8UC1)   2.201 2.908    0.76   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 16UC1)  2.108 2.951    0.71   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 32FC1)  1.997 2.840    0.70   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 8UC3)   3.444 3.293    1.05   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 16UC3)  2.889 3.417    0.85   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 32FC3)  2.671 3.354    0.80   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 8UC4)   3.765 3.767    1.00   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 16UC4)  3.247 3.962    0.82   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 32FC4)  2.993 3.669    0.82   
```

### Desktop (i7-12700K, 64GB memory)

```
Geometric mean (ms)

                                      Name of Test                                        base  patch   patch   
                                                                                                          vs    
                                                                                                         base   
                                                                                                      (x-factor)
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 8UC1)      0.274 0.076    3.62   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 16UC1)     0.299 0.058    5.14   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 32FC1)     0.299 0.043    6.90   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 8UC3)      0.330 0.115    2.87   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 16UC3)     0.480 0.109    4.39   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 32FC3)     0.608 0.180    3.37   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 8UC4)      0.317 0.143    2.21   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 16UC4)     0.704 0.139    5.07   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 32FC4)     0.508 0.141    3.60   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 8UC1)     0.293 0.064    4.57   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 16UC1)    0.311 0.061    5.07   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 32FC1)    0.299 0.057    5.29   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 8UC3)     0.373 0.135    2.75   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 16UC3)    0.501 0.129    3.87   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 32FC3)    0.403 0.123    3.26   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 8UC4)     0.372 0.163    2.28   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 16UC4)    0.582 0.161    3.63   
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 32FC4)    0.459 0.152    3.03   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 8UC1)     0.558 0.099    5.63   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 16UC1)    0.607 0.098    6.20   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 32FC1)    0.599 0.090    6.65   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 8UC3)     0.636 0.198    3.22   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 16UC3)    0.806 0.187    4.31   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 32FC3)    1.329 0.227    5.85   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 8UC4)     0.643 0.238    2.70   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 16UC4)    1.401 0.233    6.02   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 32FC4)    1.083 0.229    4.72   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 8UC1)    0.682 0.358    1.91   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 16UC1)   0.695 0.350    1.99   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 32FC1)   0.666 0.334    2.00   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 8UC3)    0.895 0.502    1.78   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 16UC3)   1.035 0.492    2.11   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 32FC3)   0.924 0.466    1.98   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 8UC4)    0.969 0.551    1.76   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 16UC4)   1.201 0.550    2.18   
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 32FC4)   0.948 0.544    1.74   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 8UC1)    1.018 0.174    5.84   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 16UC1)   0.973 0.173    5.63   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 32FC1)   1.002 0.164    6.13   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 8UC3)    1.100 0.297    3.71   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 16UC3)   1.197 0.278    4.30   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 32FC3)   3.108 0.296   10.49   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 8UC4)    1.086 0.340    3.20   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 16UC4)   2.987 0.336    8.88   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 32FC4)   2.249 0.835    2.69   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 8UC1)   1.330 1.007    1.32   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 16UC1)  1.352 0.974    1.39   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 32FC1)  1.241 0.933    1.33   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 8UC3)   1.896 1.287    1.47   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 16UC3)  2.071 1.288    1.61   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 32FC3)  1.870 1.211    1.54   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 8UC4)   2.059 1.362    1.51   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 16UC4)  2.366 1.395    1.70   
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 32FC4)  1.859 1.416    1.31   
```

### Khadas VIM3 (A311D, 4xA73+2xA53, no fp16 vector intrinsics support, 4GB memory)

```
Geometric mean (ms)

                                      Name of Test                                         base  patch    patch
                                                                                                            vs
                                                                                                           base
                                                                                                        (x-factor)
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 8UC1)      2.543  0.702     3.62
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 16UC1)     3.175  0.727     4.37
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 32FC1)     2.877  0.777     3.70
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 8UC3)      4.059  1.192     3.41
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 16UC3)     4.689  1.642     2.86
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 32FC3)     4.071  2.064     1.97
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 8UC4)      3.897  1.501     2.60
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 16UC4)     5.485  2.106     2.60
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_CONSTANT, 32FC4)     4.611  2.938     1.57
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 8UC1)     2.717  0.912     2.98
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 16UC1)    3.426  0.885     3.87
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 32FC1)    3.009  0.979     3.07
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 8UC3)     4.409  1.488     2.96
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 16UC3)    5.236  1.901     2.75
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 32FC3)    4.445  2.232     1.99
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 8UC4)     4.400  1.816     2.42
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 16UC4)    6.211  2.390     2.60
WarpPerspective::TestWarpPerspective::(640x480, INTER_LINEAR, BORDER_REPLICATE, 32FC4)    4.779  3.154     1.52
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 8UC1)     5.487  1.599     3.43
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 16UC1)    6.589  1.652     3.99
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 32FC1)    4.916  1.779     2.76
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 8UC3)     7.676  2.465     3.11
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 16UC3)    8.783  3.020     2.91
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 32FC3)    8.468  4.314     1.96
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 8UC4)     7.670  2.944     2.60
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 16UC4)    9.364  3.871     2.42
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_CONSTANT, 32FC4)    9.297  5.770     1.61
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 8UC1)    6.809  5.359     1.27
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 16UC1)   9.010  4.745     1.90
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 32FC1)   8.501  4.712     1.80
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 8UC3)    10.652 7.345     1.45
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 16UC3)   12.319 7.647     1.61
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 32FC3)   10.466 7.849     1.33
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 8UC4)    11.659 8.226     1.42
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 16UC4)   13.157 8.825     1.49
WarpPerspective::TestWarpPerspective::(1280x720, INTER_LINEAR, BORDER_REPLICATE, 32FC4)   11.557 9.869     1.17
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 8UC1)    14.773 3.081     4.79
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 16UC1)   14.971 3.135     4.78
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 32FC1)   14.795 3.321     4.45
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 8UC3)    20.823 4.319     4.82
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 16UC3)   22.128 5.029     4.40
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 32FC3)   22.583 8.036     2.81
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 8UC4)    20.141 5.018     4.01
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 16UC4)   23.505 6.132     3.83
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_CONSTANT, 32FC4)   20.226 10.050    2.01
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 8UC1)   18.904 15.189    1.24
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 16UC1)  22.749 12.979    1.75
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 32FC1)  19.685 12.981    1.52
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 8UC3)   29.636 19.974    1.48
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 16UC3)  36.266 19.563    1.85
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 32FC3)  30.124 19.434    1.55
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 8UC4)   34.290 21.998    1.56
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 16UC4)  41.765 21.705    1.92
WarpPerspective::TestWarpPerspective::(1920x1080, INTER_LINEAR, BORDER_REPLICATE, 32FC4)  27.767 22.838    1.22
```

### 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
- [x] 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
      Patch to opencv_extra has the same branch name.
- [x] The feature is well documented and sample code can be built with the project CMake
This commit is contained in:
Yuantao Feng
2024-10-15 16:13:41 +08:00
committed by GitHub
parent 3c627b0a97
commit 12738deaef
7 changed files with 1935 additions and 777 deletions
+3 -1
View File
@@ -2508,6 +2508,7 @@ optional flag #WARP_INVERSE_MAP, that sets M as the inverse transformation (
\f$\texttt{dst}\rightarrow\texttt{src}\f$ ).
@param borderMode pixel extrapolation method (#BORDER_CONSTANT or #BORDER_REPLICATE).
@param borderValue value used in case of a constant border; by default, it equals 0.
@param hint Implementation modfication flags. See #AlgorithmHint
@sa warpAffine, resize, remap, getRectSubPix, perspectiveTransform
*/
@@ -2515,7 +2516,8 @@ CV_EXPORTS_W void warpPerspective( InputArray src, OutputArray dst,
InputArray M, Size dsize,
int flags = INTER_LINEAR,
int borderMode = BORDER_CONSTANT,
const Scalar& borderValue = Scalar());
const Scalar& borderValue = Scalar(),
AlgorithmHint hint = cv::ALGO_HINT_DEFAULT);
/** @brief Applies a generic geometrical transformation to an image.
+23 -8
View File
@@ -13,7 +13,7 @@ CV_ENUM(InterTypeExtended, INTER_NEAREST, INTER_LINEAR, WARP_RELATIVE_MAP)
CV_ENUM(RemapMode, HALF_SIZE, UPSIDE_DOWN, REFLECTION_X, REFLECTION_BOTH)
typedef TestBaseWithParam< tuple<Size, InterType, BorderMode, MatType> > TestWarpAffine;
typedef TestBaseWithParam< tuple<Size, InterType, BorderMode, int> > TestWarpPerspective;
typedef TestBaseWithParam< tuple<Size, InterType, BorderMode, MatType> > TestWarpPerspective;
typedef TestBaseWithParam< tuple<Size, InterType, BorderMode, MatType> > TestWarpPerspectiveNear_t;
typedef TestBaseWithParam< tuple<MatType, Size, InterTypeExtended, BorderMode, RemapMode> > TestRemap;
@@ -67,22 +67,37 @@ PERF_TEST_P( TestWarpPerspective, WarpPerspective,
Values( szVGA, sz720p, sz1080p ),
InterType::all(),
BorderMode::all(),
Values(1, 3, 4)
Values(CV_8UC3, CV_16UC3, CV_32FC3, CV_8UC1, CV_16UC1, CV_32FC1, CV_8UC4, CV_16UC4, CV_32FC4)
)
)
{
Size sz, szSrc(512, 512);
int borderMode, interType, channels;
int type, borderMode, interType;
sz = get<0>(GetParam());
interType = get<1>(GetParam());
borderMode = get<2>(GetParam());
channels = get<3>(GetParam());
type = get<3>(GetParam());
Scalar borderColor = Scalar::all(150);
Mat src(szSrc, CV_8UC(channels)), dst(sz, CV_8UC(channels));
cvtest::fillGradient<uint8_t>(src);
if(borderMode == BORDER_CONSTANT) cvtest::smoothBorder<uint8_t>(src, borderColor, 1);
Mat src(szSrc, type), dst(sz, type);
switch (src.depth()) {
case CV_8U: {
cvtest::fillGradient<uint8_t>(src);
if(borderMode == BORDER_CONSTANT) cvtest::smoothBorder<uint8_t>(src, borderColor, 1);
break;
}
case CV_16U: {
cvtest::fillGradient<uint16_t>(src);
if(borderMode == BORDER_CONSTANT) cvtest::smoothBorder<uint16_t>(src, borderColor, 1);
break;
}
case CV_32F: {
cvtest::fillGradient<float>(src);
if(borderMode == BORDER_CONSTANT) cvtest::smoothBorder<float>(src, borderColor, 1);
break;
}
}
Mat rotMat = getRotationMatrix2D(Point2f(src.cols/2.f, src.rows/2.f), 30., 2.2);
Mat warpMat(3, 3, CV_64FC1);
for(int r=0; r<2; r++)
+59 -7
View File
@@ -2278,6 +2278,7 @@ static bool ocl_warpTransform_cols4(InputArray _src, OutputArray _dst, InputArra
if ( !dev.isIntel() || !(type == CV_8UC1) ||
!(dtype == CV_8UC1) || !(_dst.cols() % 4 == 0) ||
(op_type == OCL_OP_PERSPECTIVE && interpolation == INTER_LINEAR && (cn == 1 || cn == 3 || cn == 4)) ||
!(borderType == cv::BORDER_CONSTANT &&
(interpolation == cv::INTER_NEAREST || interpolation == cv::INTER_LINEAR || interpolation == cv::INTER_CUBIC)))
return false;
@@ -2370,7 +2371,9 @@ static bool ocl_warpTransform(InputArray _src, OutputArray _dst, InputArray _M0,
const char * const kernelName = op_type == OCL_OP_AFFINE ? "warpAffine" : "warpPerspective";
int scalarcn = cn == 3 ? 4 : cn;
bool is32f = !dev.isAMD() && (interpolation == INTER_CUBIC || interpolation == INTER_LINEAR) && op_type == OCL_OP_AFFINE;
bool is32f = op_type == OCL_OP_AFFINE ?
/* Affine*/ !dev.isAMD() && (interpolation == INTER_CUBIC || interpolation == INTER_LINEAR) :
/* Perspective*/ interpolation == INTER_LINEAR;
int wdepth = interpolation == INTER_NEAREST ? depth : std::max(is32f ? CV_32F : CV_32S, depth);
int sctype = CV_MAKETYPE(wdepth, scalarcn);
@@ -2385,8 +2388,7 @@ static bool ocl_warpTransform(InputArray _src, OutputArray _dst, InputArray _M0,
ocl::typeToStr(CV_MAT_DEPTH(type)),
ocl::typeToStr(sctype), cn, rowsPerWI);
}
else
{
else {
char cvt[2][50];
opts = format("-D INTER_%s -D T=%s -D T1=%s -D ST=%s -D WT=%s -D SRC_DEPTH=%d"
" -D CONVERT_TO_WT=%s -D CONVERT_TO_T=%s%s -D CT=%s -D CN=%d -D ROWS_PER_WI=%d",
@@ -3256,12 +3258,58 @@ private:
namespace hal {
void warpPerspective(int src_type,
static void warpPerspective(int src_type,
const uchar * src_data, size_t src_step, int src_width, int src_height,
uchar * dst_data, size_t dst_step, int dst_width, int dst_height,
const double M[9], int interpolation, int borderType, const double borderValue[4])
const double M[9], int interpolation, int borderType, const double borderValue[4], AlgorithmHint hint)
{
CALL_HAL(warpPerspective, cv_hal_warpPerspective, src_type, src_data, src_step, src_width, src_height, dst_data, dst_step, dst_width, dst_height, M, interpolation, borderType, borderValue);
if (interpolation == INTER_LINEAR) {
switch (src_type) {
case CV_8UC1: {
if (hint == cv::ALGO_HINT_APPROX) {
CV_CPU_DISPATCH(warpPerspectiveLinearApproxInvoker_8UC1, (src_data, src_step, src_height, src_width, dst_data, dst_step, dst_height, dst_width, M, borderType, borderValue), CV_CPU_DISPATCH_MODES_ALL);
} else {
CV_CPU_DISPATCH(warpPerspectiveLinearInvoker_8UC1, (src_data, src_step, src_height, src_width, dst_data, dst_step, dst_height, dst_width, M, borderType, borderValue), CV_CPU_DISPATCH_MODES_ALL);
}
}
case CV_8UC3: {
if (hint == cv::ALGO_HINT_APPROX) {
CV_CPU_DISPATCH(warpPerspectiveLinearApproxInvoker_8UC3, (src_data, src_step, src_height, src_width, dst_data, dst_step, dst_height, dst_width, M, borderType, borderValue), CV_CPU_DISPATCH_MODES_ALL);
} else {
CV_CPU_DISPATCH(warpPerspectiveLinearInvoker_8UC3, (src_data, src_step, src_height, src_width, dst_data, dst_step, dst_height, dst_width, M, borderType, borderValue), CV_CPU_DISPATCH_MODES_ALL);
}
}
case CV_8UC4: {
if (hint == cv::ALGO_HINT_APPROX) {
CV_CPU_DISPATCH(warpPerspectiveLinearApproxInvoker_8UC4, (src_data, src_step, src_height, src_width, dst_data, dst_step, dst_height, dst_width, M, borderType, borderValue), CV_CPU_DISPATCH_MODES_ALL);
} else {
CV_CPU_DISPATCH(warpPerspectiveLinearInvoker_8UC4, (src_data, src_step, src_height, src_width, dst_data, dst_step, dst_height, dst_width, M, borderType, borderValue), CV_CPU_DISPATCH_MODES_ALL);
}
}
case CV_16UC1: {
CV_CPU_DISPATCH(warpPerspectiveLinearInvoker_16UC1, ((const uint16_t*)src_data, src_step, src_height, src_width, (uint16_t*)dst_data, dst_step, dst_height, dst_width, M, borderType, borderValue), CV_CPU_DISPATCH_MODES_ALL);
}
case CV_16UC3: {
CV_CPU_DISPATCH(warpPerspectiveLinearInvoker_16UC3, ((const uint16_t*)src_data, src_step, src_height, src_width, (uint16_t*)dst_data, dst_step, dst_height, dst_width, M, borderType, borderValue), CV_CPU_DISPATCH_MODES_ALL);
}
case CV_16UC4: {
CV_CPU_DISPATCH(warpPerspectiveLinearInvoker_16UC4, ((const uint16_t*)src_data, src_step, src_height, src_width, (uint16_t*)dst_data, dst_step, dst_height, dst_width, M, borderType, borderValue), CV_CPU_DISPATCH_MODES_ALL);
}
case CV_32FC1: {
CV_CPU_DISPATCH(warpPerspectiveLinearInvoker_32FC1, ((const float*)src_data, src_step, src_height, src_width, (float*)dst_data, dst_step, dst_height, dst_width, M, borderType, borderValue), CV_CPU_DISPATCH_MODES_ALL);
}
case CV_32FC3: {
CV_CPU_DISPATCH(warpPerspectiveLinearInvoker_32FC3, ((const float*)src_data, src_step, src_height, src_width, (float*)dst_data, dst_step, dst_height, dst_width, M, borderType, borderValue), CV_CPU_DISPATCH_MODES_ALL);
}
case CV_32FC4: {
CV_CPU_DISPATCH(warpPerspectiveLinearInvoker_32FC4, ((const float*)src_data, src_step, src_height, src_width, (float*)dst_data, dst_step, dst_height, dst_width, M, borderType, borderValue), CV_CPU_DISPATCH_MODES_ALL);
}
// no default
}
}
Mat src(Size(src_width, src_height), src_type, const_cast<uchar*>(src_data), src_step);
Mat dst(Size(dst_width, dst_height), src_type, dst_data, dst_step);
@@ -3342,10 +3390,14 @@ void warpPerspectiveBlockline(const double *M, short* xy, short* alpha, double X
} // cv::
void cv::warpPerspective( InputArray _src, OutputArray _dst, InputArray _M0,
Size dsize, int flags, int borderType, const Scalar& borderValue )
Size dsize, int flags, int borderType, const Scalar& borderValue,
AlgorithmHint hint )
{
CV_INSTRUMENT_REGION();
if (hint == cv::ALGO_HINT_DEFAULT)
hint = cv::getDefaultAlgorithmHint();
CV_Assert( _src.total() > 0 );
CV_OCL_RUN(_src.dims() <= 2 && _dst.isUMat() &&
@@ -3436,7 +3488,7 @@ void cv::warpPerspective( InputArray _src, OutputArray _dst, InputArray _M0,
invert(matM, matM);
hal::warpPerspective(src.type(), src.data, src.step, src.cols, src.rows, dst.data, dst.step, dst.cols, dst.rows,
matM.ptr<double>(), interpolation, borderType, borderValue.val);
matM.ptr<double>(), interpolation, borderType, borderValue.val, hint);
}
+11 -25
View File
@@ -122,16 +122,14 @@ __kernel void warpPerspective(__global const uchar * srcptr, int src_step, int s
if (dx < dst_cols && dy < dst_rows)
{
CT X0 = M[0] * dx + M[1] * dy + M[2];
CT Y0 = M[3] * dx + M[4] * dy + M[5];
CT W = M[6] * dx + M[7] * dy + M[8];
W = W != 0.0f ? INTER_TAB_SIZE / W : 0.0f;
int X = rint(X0 * W), Y = rint(Y0 * W);
float W = M[6] * dx + M[7] * dy + M[8];
float X0 = (M[0] * dx + M[1] * dy + M[2]) / W;
float Y0 = (M[3] * dx + M[4] * dy + M[5]) / W;
short sx = convert_short_sat(X >> INTER_BITS);
short sy = convert_short_sat(Y >> INTER_BITS);
short ay = (short)(Y & (INTER_TAB_SIZE - 1));
short ax = (short)(X & (INTER_TAB_SIZE - 1));
int sx = convert_short_rtn(X0);
int sy = convert_short_rtn(Y0);
float ay = Y0 - (CT)sy;
float ax = X0 - (CT)sx;
WT v0 = (sx >= 0 && sx < src_cols && sy >= 0 && sy < src_rows) ?
CONVERT_TO_WT(loadpix(srcptr + mad24(sy, src_step, src_offset + sx * pixsize))) : scalar;
@@ -142,24 +140,12 @@ __kernel void warpPerspective(__global const uchar * srcptr, int src_step, int s
WT v3 = (sx+1 >= 0 && sx+1 < src_cols && sy+1 >= 0 && sy+1 < src_rows) ?
CONVERT_TO_WT(loadpix(srcptr + mad24(sy+1, src_step, src_offset + (sx+1) * pixsize))) : scalar;
float taby = 1.f/INTER_TAB_SIZE*ay;
float tabx = 1.f/INTER_TAB_SIZE*ax;
int dst_index = mad24(dy, dst_step, dst_offset + dx * pixsize);
#if SRC_DEPTH <= 4
int itab0 = convert_short_sat_rte( (1.0f-taby)*(1.0f-tabx) * INTER_REMAP_COEF_SCALE );
int itab1 = convert_short_sat_rte( (1.0f-taby)*tabx * INTER_REMAP_COEF_SCALE );
int itab2 = convert_short_sat_rte( taby*(1.0f-tabx) * INTER_REMAP_COEF_SCALE );
int itab3 = convert_short_sat_rte( taby*tabx * INTER_REMAP_COEF_SCALE );
WT val = v0 * itab0 + v1 * itab1 + v2 * itab2 + v3 * itab3;
storepix(CONVERT_TO_T((val + (1 << (INTER_REMAP_COEF_BITS-1))) >> INTER_REMAP_COEF_BITS), dstptr + dst_index);
#else
float tabx2 = 1.0f - tabx, taby2 = 1.0f - taby;
WT val = v0 * tabx2 * taby2 + v1 * tabx * taby2 + v2 * tabx2 * taby + v3 * tabx * taby;
storepix(CONVERT_TO_T(val), dstptr + dst_index);
#endif
v0 = fma(v1 - v0, ax, v0);
v2 = fma(v3 - v2, ax, v2);
v0 = fma(v2 - v0, ay, v0);
storepix(CONVERT_TO_T(v0), dstptr + dst_index);
}
}
+199 -29
View File
@@ -57,6 +57,59 @@
#define CV_WARP_LINEAR_VECTOR_SHUFFLE_ALLWITHIN(CN, DEPTH) \
CV_WARP_LINEAR_VECTOR_SHUFFLE_ALLWITHIN_##DEPTH(CN)
// Shuffle (ARM NEON)
#define CV_WARP_LINEAR_VECTOR_SHUFFLE_ALLWITHIN_NEON_U8_LOAD() \
uint8x8x4_t t00 = { \
vld1_u8(src + addr[0]), \
vld1_u8(src + addr[1]), \
vld1_u8(src + addr[2]), \
vld1_u8(src + addr[3]) \
}; \
uint8x8x4_t t01 = { \
vld1_u8(src + addr[4]), \
vld1_u8(src + addr[5]), \
vld1_u8(src + addr[6]), \
vld1_u8(src + addr[7]) \
}; \
uint8x8x4_t t10 = { \
vld1_u8(src + addr[0] + srcstep), \
vld1_u8(src + addr[1] + srcstep), \
vld1_u8(src + addr[2] + srcstep), \
vld1_u8(src + addr[3] + srcstep) \
}; \
uint8x8x4_t t11 = { \
vld1_u8(src + addr[4] + srcstep), \
vld1_u8(src + addr[5] + srcstep), \
vld1_u8(src + addr[6] + srcstep), \
vld1_u8(src + addr[7] + srcstep) \
}; \
uint32x2_t p00_, p01_, p10_, p11_;
#define CV_WARP_LINEAR_VECTOR_SHUFFLE_ALLWITHIN_NEON_U8_TRN(coords, cn) \
p00_ = vreinterpret_u32_u8(vtbl4_u8(t00, coords)); \
p01_ = vreinterpret_u32_u8(vtbl4_u8(t01, coords)); \
p10_ = vreinterpret_u32_u8(vtbl4_u8(t10, coords)); \
p11_ = vreinterpret_u32_u8(vtbl4_u8(t11, coords)); \
p00##cn = vreinterpret_u8_u32(vtrn1_u32(p00_, p01_)); \
p01##cn = vreinterpret_u8_u32(vtrn2_u32(p00_, p01_)); \
p10##cn = vreinterpret_u8_u32(vtrn1_u32(p10_, p11_)); \
p11##cn = vreinterpret_u8_u32(vtrn2_u32(p10_, p11_));
#define CV_WARP_LINEAR_VECTOR_SHUFFLE_ALLWITHIN_NEON_U8_C1() \
CV_WARP_LINEAR_VECTOR_SHUFFLE_ALLWITHIN_NEON_U8_LOAD() \
CV_WARP_LINEAR_VECTOR_SHUFFLE_ALLWITHIN_NEON_U8_TRN(grays, g)
#define CV_WARP_LINEAR_VECTOR_SHUFFLE_ALLWITHIN_NEON_U8_C3() \
CV_WARP_LINEAR_VECTOR_SHUFFLE_ALLWITHIN_NEON_U8_LOAD() \
CV_WARP_LINEAR_VECTOR_SHUFFLE_ALLWITHIN_NEON_U8_TRN(reds, r) \
CV_WARP_LINEAR_VECTOR_SHUFFLE_ALLWITHIN_NEON_U8_TRN(greens, g) \
CV_WARP_LINEAR_VECTOR_SHUFFLE_ALLWITHIN_NEON_U8_TRN(blues, b)
#define CV_WARP_LINEAR_VECTOR_SHUFFLE_ALLWITHIN_NEON_U8_C4() \
CV_WARP_LINEAR_VECTOR_SHUFFLE_ALLWITHIN_NEON_U8_LOAD() \
CV_WARP_LINEAR_VECTOR_SHUFFLE_ALLWITHIN_NEON_U8_TRN(reds, r) \
CV_WARP_LINEAR_VECTOR_SHUFFLE_ALLWITHIN_NEON_U8_TRN(greens, g) \
CV_WARP_LINEAR_VECTOR_SHUFFLE_ALLWITHIN_NEON_U8_TRN(blues, b) \
CV_WARP_LINEAR_VECTOR_SHUFFLE_ALLWITHIN_NEON_U8_TRN(alphas, a)
#define CV_WARP_LINEAR_VECTOR_SHUFFLE_ALLWITHIN_NEON_U8(CN) \
CV_WARP_LINEAR_VECTOR_SHUFFLE_ALLWITHIN_NEON_U8_##CN()
// Shuffle (not all pixels within image)
#define CV_WARP_LINEAR_VECTOR_SHUFFLE_STORE_CONSTANT_BORDER_8UC1() \
@@ -190,6 +243,25 @@
CV_WARP_LINEAR_VECTOR_FETCH_PIXEL_##CN(1, 1, uf*3); \
}
// Shuffle (not all pixels within image) (ARM NEON)
#define CV_WARP_LINEAR_VECTOR_SHUFFLE_NOTALLWITHIN_NEON_U8_LOAD(cn, offset)\
p00##cn = vld1_u8(pixbuf + offset); \
p01##cn = vld1_u8(pixbuf + offset + 8); \
p10##cn = vld1_u8(pixbuf + offset + 16); \
p11##cn = vld1_u8(pixbuf + offset + 24);
#define CV_WARP_LINEAR_VECTOR_SHUFFLE_NOTALLWITHIN_NEON_U8_C1() \
CV_WARP_LINEAR_VECTOR_SHUFFLE_NOTALLWITHIN_NEON_U8_LOAD(g, 0)
#define CV_WARP_LINEAR_VECTOR_SHUFFLE_NOTALLWITHIN_NEON_U8_C3() \
CV_WARP_LINEAR_VECTOR_SHUFFLE_NOTALLWITHIN_NEON_U8_LOAD(r, 0) \
CV_WARP_LINEAR_VECTOR_SHUFFLE_NOTALLWITHIN_NEON_U8_LOAD(g, 32) \
CV_WARP_LINEAR_VECTOR_SHUFFLE_NOTALLWITHIN_NEON_U8_LOAD(b, 64)
#define CV_WARP_LINEAR_VECTOR_SHUFFLE_NOTALLWITHIN_NEON_U8_C4() \
CV_WARP_LINEAR_VECTOR_SHUFFLE_NOTALLWITHIN_NEON_U8_LOAD(r, 0) \
CV_WARP_LINEAR_VECTOR_SHUFFLE_NOTALLWITHIN_NEON_U8_LOAD(g, 32) \
CV_WARP_LINEAR_VECTOR_SHUFFLE_NOTALLWITHIN_NEON_U8_LOAD(b, 64) \
CV_WARP_LINEAR_VECTOR_SHUFFLE_NOTALLWITHIN_NEON_U8_LOAD(a, 96)
#define CV_WARP_LINEAR_VECTOR_SHUFFLE_NOTALLWITHIN_NEON_U8(CN) \
CV_WARP_LINEAR_VECTOR_SHUFFLE_NOTALLWITHIN_NEON_U8_##CN()
// Load pixels for linear interpolation (uint8_t -> int16_t)
#define CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U8S16(cn, i) \
@@ -211,6 +283,26 @@
#define CV_WARP_LINEAR_VECTOR_INTER_LOAD_U8S16(CN) \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_U8S16_##CN();
// Load pixels for linear interpolation (uint8_t -> int16_t) (ARM NEON)
#define CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U8S16_NEON(cn) \
v_int16 f00##cn = v_reinterpret_as_s16(v_uint16(vmovl_u8(p00##cn))), \
f01##cn = v_reinterpret_as_s16(v_uint16(vmovl_u8(p01##cn))), \
f10##cn = v_reinterpret_as_s16(v_uint16(vmovl_u8(p10##cn))), \
f11##cn = v_reinterpret_as_s16(v_uint16(vmovl_u8(p11##cn)));
#define CV_WARP_LINEAR_VECTOR_INTER_LOAD_U8S16_NEON_C1() \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U8S16_NEON(g)
#define CV_WARP_LINEAR_VECTOR_INTER_LOAD_U8S16_NEON_C3() \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U8S16_NEON(r) \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U8S16_NEON(g) \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U8S16_NEON(b)
#define CV_WARP_LINEAR_VECTOR_INTER_LOAD_U8S16_NEON_C4() \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U8S16_NEON(r) \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U8S16_NEON(g) \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U8S16_NEON(b) \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U8S16_NEON(a)
#define CV_WARP_LINEAR_VECTOR_INTER_LOAD_U8S16_NEON(CN) \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_U8S16_NEON_##CN();
// Load pixels for linear interpolation (uint16_t -> uint16_t)
#define CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U16(cn, i) \
v_uint16 f00##cn = vx_load(pixbuf + uf * i), \
@@ -232,44 +324,44 @@
CV_WARP_LINEAR_VECTOR_INTER_LOAD_U16_##CN();
// Load pixels for linear interpolation (int16_t -> float)
#define CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_S16F32(cn) \
#define CV_WARP_LINEAR_VECTOR_INTER_CONVERT_CN_S16F32(cn) \
v_float32 f00##cn##l = v_cvt_f32(v_expand_low(f00##cn)), f00##cn##h = v_cvt_f32(v_expand_high(f00##cn)), \
f01##cn##l = v_cvt_f32(v_expand_low(f01##cn)), f01##cn##h = v_cvt_f32(v_expand_high(f01##cn)), \
f10##cn##l = v_cvt_f32(v_expand_low(f10##cn)), f10##cn##h = v_cvt_f32(v_expand_high(f10##cn)), \
f11##cn##l = v_cvt_f32(v_expand_low(f11##cn)), f11##cn##h = v_cvt_f32(v_expand_high(f11##cn));
#define CV_WARP_LINEAR_VECTOR_INTER_LOAD_S16F32_C1() \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_S16F32(g)
#define CV_WARP_LINEAR_VECTOR_INTER_LOAD_S16F32_C3() \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_S16F32(r) \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_S16F32(g) \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_S16F32(b)
#define CV_WARP_LINEAR_VECTOR_INTER_LOAD_S16F32_C4() \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_S16F32(r) \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_S16F32(g) \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_S16F32(b) \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_S16F32(a)
#define CV_WARP_LINEAR_VECTOR_INTER_LOAD_S16F32(CN) \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_S16F32_##CN()
#define CV_WARP_LINEAR_VECTOR_INTER_CONVERT_S16F32_C1() \
CV_WARP_LINEAR_VECTOR_INTER_CONVERT_CN_S16F32(g)
#define CV_WARP_LINEAR_VECTOR_INTER_CONVERT_S16F32_C3() \
CV_WARP_LINEAR_VECTOR_INTER_CONVERT_CN_S16F32(r) \
CV_WARP_LINEAR_VECTOR_INTER_CONVERT_CN_S16F32(g) \
CV_WARP_LINEAR_VECTOR_INTER_CONVERT_CN_S16F32(b)
#define CV_WARP_LINEAR_VECTOR_INTER_CONVERT_S16F32_C4() \
CV_WARP_LINEAR_VECTOR_INTER_CONVERT_CN_S16F32(r) \
CV_WARP_LINEAR_VECTOR_INTER_CONVERT_CN_S16F32(g) \
CV_WARP_LINEAR_VECTOR_INTER_CONVERT_CN_S16F32(b) \
CV_WARP_LINEAR_VECTOR_INTER_CONVERT_CN_S16F32(a)
#define CV_WARP_LINEAR_VECTOR_INTER_CONVERT_S16F32(CN) \
CV_WARP_LINEAR_VECTOR_INTER_CONVERT_S16F32_##CN()
// Load pixels for linear interpolation (uint16_t -> float)
#define CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U16F32(cn) \
#define CV_WARP_LINEAR_VECTOR_INTER_CONVERT_CN_U16F32(cn) \
v_float32 f00##cn##l = v_cvt_f32(v_reinterpret_as_s32(v_expand_low(f00##cn))), f00##cn##h = v_cvt_f32(v_reinterpret_as_s32(v_expand_high(f00##cn))), \
f01##cn##l = v_cvt_f32(v_reinterpret_as_s32(v_expand_low(f01##cn))), f01##cn##h = v_cvt_f32(v_reinterpret_as_s32(v_expand_high(f01##cn))), \
f10##cn##l = v_cvt_f32(v_reinterpret_as_s32(v_expand_low(f10##cn))), f10##cn##h = v_cvt_f32(v_reinterpret_as_s32(v_expand_high(f10##cn))), \
f11##cn##l = v_cvt_f32(v_reinterpret_as_s32(v_expand_low(f11##cn))), f11##cn##h = v_cvt_f32(v_reinterpret_as_s32(v_expand_high(f11##cn)));
#define CV_WARP_LINEAR_VECTOR_INTER_LOAD_U16F32_C1() \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U16F32(g)
#define CV_WARP_LINEAR_VECTOR_INTER_LOAD_U16F32_C3() \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U16F32(r) \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U16F32(g) \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U16F32(b)
#define CV_WARP_LINEAR_VECTOR_INTER_LOAD_U16F32_C4() \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U16F32(r) \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U16F32(g) \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U16F32(b) \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U16F32(a)
#define CV_WARP_LINEAR_VECTOR_INTER_LOAD_U16F32(CN) \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_U16F32_##CN()
#define CV_WARP_LINEAR_VECTOR_INTER_CONVERT_U16F32_C1() \
CV_WARP_LINEAR_VECTOR_INTER_CONVERT_CN_U16F32(g)
#define CV_WARP_LINEAR_VECTOR_INTER_CONVERT_U16F32_C3() \
CV_WARP_LINEAR_VECTOR_INTER_CONVERT_CN_U16F32(r) \
CV_WARP_LINEAR_VECTOR_INTER_CONVERT_CN_U16F32(g) \
CV_WARP_LINEAR_VECTOR_INTER_CONVERT_CN_U16F32(b)
#define CV_WARP_LINEAR_VECTOR_INTER_CONVERT_U16F32_C4() \
CV_WARP_LINEAR_VECTOR_INTER_CONVERT_CN_U16F32(r) \
CV_WARP_LINEAR_VECTOR_INTER_CONVERT_CN_U16F32(g) \
CV_WARP_LINEAR_VECTOR_INTER_CONVERT_CN_U16F32(b) \
CV_WARP_LINEAR_VECTOR_INTER_CONVERT_CN_U16F32(a)
#define CV_WARP_LINEAR_VECTOR_INTER_CONVERT_U16F32(CN) \
CV_WARP_LINEAR_VECTOR_INTER_CONVERT_U16F32_##CN()
// Load pixels for linear interpolation (float -> float)
#define CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_F32(cn, i) \
@@ -291,8 +383,27 @@
#define CV_WARP_LINEAR_VECTOR_INTER_LOAD_F32(CN) \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_F32_##CN()
// Load pixels for linear interpolation (uint8_t -> float16)
#define CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U8F16(cn) \
v_float16 f00##cn = v_float16(vcvtq_f16_u16(vmovl_u8(p00##cn))), \
f01##cn = v_float16(vcvtq_f16_u16(vmovl_u8(p01##cn))), \
f10##cn = v_float16(vcvtq_f16_u16(vmovl_u8(p10##cn))), \
f11##cn = v_float16(vcvtq_f16_u16(vmovl_u8(p11##cn)));
#define CV_WARP_LINEAR_VECTOR_INTER_LOAD_U8F16_C1() \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U8F16(g)
#define CV_WARP_LINEAR_VECTOR_INTER_LOAD_U8F16_C3() \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U8F16(r) \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U8F16(g) \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U8F16(b)
#define CV_WARP_LINEAR_VECTOR_INTER_LOAD_U8F16_C4() \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U8F16(r) \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U8F16(g) \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U8F16(b) \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_CN_U8F16(a)
#define CV_WARP_LINEAR_VECTOR_INTER_LOAD_U8F16(CN) \
CV_WARP_LINEAR_VECTOR_INTER_LOAD_U8F16_##CN()
// Linear interpolation calculation
// Linear interpolation calculation (F32)
#define CV_WARP_LINEAR_VECTOR_INTER_CALC_ALPHA_F32(cn) \
f00##cn##l = v_fma(alphal, v_sub(f01##cn##l, f00##cn##l), f00##cn##l); f00##cn##h = v_fma(alphah, v_sub(f01##cn##h, f00##cn##h), f00##cn##h); \
f10##cn##l = v_fma(alphal, v_sub(f11##cn##l, f10##cn##l), f10##cn##l); f10##cn##h = v_fma(alphah, v_sub(f11##cn##h, f10##cn##h), f10##cn##h);
@@ -328,6 +439,42 @@
CV_WARP_LINEAR_VECTOR_INTER_CALC_ALPHA_F32_##CN() \
CV_WARP_LINEAR_VECTOR_INTER_CALC_BETA_F32_##CN()
// Linear interpolation calculation (F16)
#define CV_WARP_LINEAR_VECTOR_INTER_CALC_ALPHA_F16(cn) \
f00##cn = v_fma(alpha, v_sub(f01##cn, f00##cn), f00##cn); \
f10##cn = v_fma(alpha, v_sub(f11##cn, f10##cn), f10##cn);
#define CV_WARP_LINEAR_VECTOR_INTER_CALC_ALPHA_F16_C1() \
CV_WARP_LINEAR_VECTOR_INTER_CALC_ALPHA_F16(g)
#define CV_WARP_LINEAR_VECTOR_INTER_CALC_ALPHA_F16_C3() \
CV_WARP_LINEAR_VECTOR_INTER_CALC_ALPHA_F16(r) \
CV_WARP_LINEAR_VECTOR_INTER_CALC_ALPHA_F16(g) \
CV_WARP_LINEAR_VECTOR_INTER_CALC_ALPHA_F16(b)
#define CV_WARP_LINEAR_VECTOR_INTER_CALC_ALPHA_F16_C4() \
CV_WARP_LINEAR_VECTOR_INTER_CALC_ALPHA_F16(r) \
CV_WARP_LINEAR_VECTOR_INTER_CALC_ALPHA_F16(g) \
CV_WARP_LINEAR_VECTOR_INTER_CALC_ALPHA_F16(b) \
CV_WARP_LINEAR_VECTOR_INTER_CALC_ALPHA_F16(a)
#define CV_WARP_LINEAR_VECTOR_INTER_CALC_BETA_F16(cn) \
f00##cn = v_fma(beta, v_sub(f10##cn, f00##cn), f00##cn);
#define CV_WARP_LINEAR_VECTOR_INTER_CALC_BETA_F16_C1() \
CV_WARP_LINEAR_VECTOR_INTER_CALC_BETA_F16(g)
#define CV_WARP_LINEAR_VECTOR_INTER_CALC_BETA_F16_C3() \
CV_WARP_LINEAR_VECTOR_INTER_CALC_BETA_F16(r) \
CV_WARP_LINEAR_VECTOR_INTER_CALC_BETA_F16(g) \
CV_WARP_LINEAR_VECTOR_INTER_CALC_BETA_F16(b)
#define CV_WARP_LINEAR_VECTOR_INTER_CALC_BETA_F16_C4() \
CV_WARP_LINEAR_VECTOR_INTER_CALC_BETA_F16(r) \
CV_WARP_LINEAR_VECTOR_INTER_CALC_BETA_F16(g) \
CV_WARP_LINEAR_VECTOR_INTER_CALC_BETA_F16(b) \
CV_WARP_LINEAR_VECTOR_INTER_CALC_BETA_F16(a)
#define CV_WARP_LINEAR_VECTOR_INTER_CALC_F16(CN) \
v_float16 alpha = v_cvt_f16(src_x0, src_x1), \
beta = v_cvt_f16(src_y0, src_y1); \
CV_WARP_LINEAR_VECTOR_INTER_CALC_ALPHA_F16_##CN() \
CV_WARP_LINEAR_VECTOR_INTER_CALC_BETA_F16_##CN()
// Store
#define CV_WARP_LINEAR_VECTOR_INTER_STORE_F32U8_C1() \
@@ -385,3 +532,26 @@
v_store_interleave(dstptr + x*4 + vlanes_32*4, f00rh, f00gh, f00bh, f00ah);
#define CV_WARP_LINEAR_VECTOR_INTER_STORE_F32F32(CN) \
CV_WARP_LINEAR_VECTOR_INTER_STORE_F32F32_##CN()
#define CV_WARP_LINEAR_VECTOR_INTER_STORE_F16U8_C1() \
uint8x8_t result = { \
vqmovun_s16(vcvtnq_s16_f16(f00g.val)), \
}; \
vst1_u8(dstptr + x, result);
#define CV_WARP_LINEAR_VECTOR_INTER_STORE_F16U8_C3() \
uint8x8x3_t result = { \
vqmovun_s16(vcvtnq_s16_f16(f00r.val)), \
vqmovun_s16(vcvtnq_s16_f16(f00g.val)), \
vqmovun_s16(vcvtnq_s16_f16(f00b.val)), \
}; \
vst3_u8(dstptr + x*3, result);
#define CV_WARP_LINEAR_VECTOR_INTER_STORE_F16U8_C4() \
uint8x8x4_t result = { \
vqmovun_s16(vcvtnq_s16_f16(f00r.val)), \
vqmovun_s16(vcvtnq_s16_f16(f00g.val)), \
vqmovun_s16(vcvtnq_s16_f16(f00b.val)), \
vqmovun_s16(vcvtnq_s16_f16(f00a.val)), \
}; \
vst4_u8(dstptr + x*4, result);
#define CV_WARP_LINEAR_VECTOR_INTER_STORE_F16U8(CN) \
CV_WARP_LINEAR_VECTOR_INTER_STORE_F16U8_##CN()
File diff suppressed because it is too large Load Diff
+68 -5
View File
@@ -1042,16 +1042,16 @@ protected:
virtual void run_func();
virtual void run_reference_func();
template<int channels, typename T>
void newLinear(int x, float sx, float sy, const T *srcptr_, T *dstptr, int srccols, int srcrows, size_t srcstep,
const T *bval, int borderType_x, int borderType_y);
Mat M;
private:
void warpAffine(const Mat&, Mat&);
template<typename T>
void newWarpAffine(const Mat&, Mat&, const Mat&);
template<int channels, typename T>
void newLinear(int x, float sx, float sy, const T *srcptr_, T *dstptr, int srccols, int srcrows, size_t srcstep,
const T *bval, int borderType_x, int borderType_y);
};
CV_WarpAffine_Test::CV_WarpAffine_Test() :
@@ -1336,6 +1336,9 @@ protected:
private:
void warpPerspective(const Mat&, Mat&);
template<typename T>
void newWarpPerspective(const Mat&, Mat&, const Mat&);
};
CV_WarpPerspective_Test::CV_WarpPerspective_Test() :
@@ -1369,7 +1372,8 @@ void CV_WarpPerspective_Test::generate_test_data()
void CV_WarpPerspective_Test::run_func()
{
cv::warpPerspective(src, dst, M, dst.size(), interpolation, borderType, borderValue);
cv::warpPerspective(src, dst, M, dst.size(), interpolation, borderType, borderValue, cv::ALGO_HINT_APPROX);
// cv::warpPerspective(src, dst, M, dst.size(), interpolation, borderType, borderValue);
}
float CV_WarpPerspective_Test::get_success_error_level(int _interpolation, int _depth) const
@@ -1384,6 +1388,54 @@ void CV_WarpPerspective_Test::run_reference_func()
tmp.convertTo(reference_dst, reference_dst.depth());
}
template<typename T>
void CV_WarpPerspective_Test::newWarpPerspective(const Mat &_src, Mat &_dst, const Mat &tM)
{
int num_channels = _dst.channels();
CV_CheckTrue(num_channels == 1 || num_channels == 3 || num_channels == 4, "");
auto *srcptr_ = _src.ptr<const T>();
auto *dstptr_ = _dst.ptr<T>();
size_t srcstep = _src.step/sizeof(T), dststep = _dst.step/sizeof(T);
int srccols = _src.cols, srcrows = _src.rows;
int dstcols = _dst.cols, dstrows = _dst.rows;
Mat tmp;
tM.convertTo(tmp, CV_32F);
auto *_M = tmp.ptr<const float>();
T bval[] = {
saturate_cast<T>(borderValue[0]),
saturate_cast<T>(borderValue[1]),
saturate_cast<T>(borderValue[2]),
saturate_cast<T>(borderValue[3]),
};
int borderType_x = borderType != BORDER_CONSTANT &&
borderType != BORDER_TRANSPARENT &&
srccols <= 1 ? BORDER_REPLICATE : borderType;
int borderType_y = borderType != BORDER_CONSTANT &&
borderType != BORDER_TRANSPARENT &&
srcrows <= 1 ? BORDER_REPLICATE : borderType;
for (int y = 0; y < dstrows; y++) {
T* dstptr = dstptr_ + y*dststep;
for (int x = 0; x < dstcols; x++) {
float w = x*_M[6] + y*_M[7] + _M[8];
float sx = (x*_M[0] + y*_M[1] + _M[2]) / w;
float sy = (x*_M[3] + y*_M[4] + _M[5]) / w;
if (num_channels == 3) {
newLinear<3>(x, sx, sy, srcptr_, dstptr, srccols, srcrows, srcstep, bval, borderType_x, borderType_y);
} else if (num_channels == 4) {
newLinear<4>(x, sx, sy, srcptr_, dstptr, srccols, srcrows, srcstep, bval, borderType_x, borderType_y);
} else {
newLinear<1>(x, sx, sy, srcptr_, dstptr, srccols, srcrows, srcstep, bval, borderType_x, borderType_y);
}
}
}
}
void CV_WarpPerspective_Test::warpPerspective(const Mat& _src, Mat& _dst)
{
Size ssize = _src.size(), dsize = _dst.size();
@@ -1410,6 +1462,17 @@ void CV_WarpPerspective_Test::warpPerspective(const Mat& _src, Mat& _dst)
if (inter == INTER_AREA)
inter = INTER_LINEAR;
if (inter == INTER_LINEAR) {
int dst_depth = _dst.depth(), dst_channels = _dst.channels();
if (dst_depth == CV_8U && (dst_channels == 1 || dst_channels == 3 || dst_channels == 4)) {
return newWarpPerspective<uint8_t>(_src, _dst, M);
} else if (dst_depth == CV_16U && (dst_channels == 1 || dst_channels == 3 || dst_channels == 4)) {
return newWarpPerspective<uint16_t>(_src, _dst, M);
} else if (dst_depth == CV_32F && (dst_channels == 1 || dst_channels == 3 || dst_channels == 4)) {
return newWarpPerspective<float>(_src, _dst, M);
}
}
mapx.create(dsize, CV_16SC2);
if (inter != INTER_NEAREST)
mapy.create(dsize, CV_16SC1);