Compute Library
 21.02
NERemapKernel.cpp
Go to the documentation of this file.
1 /*
2  * Copyright (c) 2017-2020 Arm Limited.
3  *
4  * SPDX-License-Identifier: MIT
5  *
6  * Permission is hereby granted, free of charge, to any person obtaining a copy
7  * of this software and associated documentation files (the "Software"), to
8  * deal in the Software without restriction, including without limitation the
9  * rights to use, copy, modify, merge, publish, distribute, sublicense, and/or
10  * sell copies of the Software, and to permit persons to whom the Software is
11  * furnished to do so, subject to the following conditions:
12  *
13  * The above copyright notice and this permission notice shall be included in all
14  * copies or substantial portions of the Software.
15  *
16  * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
17  * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
18  * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
19  * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
20  * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
21  * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
22  * SOFTWARE.
23  */
25 
26 #include "arm_compute/core/Error.h"
36 
37 #include <arm_neon.h>
38 #include <cstddef>
39 #include <cstdint>
40 
41 using namespace arm_compute;
42 
43 namespace arm_compute
44 {
45 class Coordinates;
46 } // namespace arm_compute
47 
48 namespace
49 {
50 inline int32x4_t offset_nearest_interpolation(const float *mapx_ptr, const float *mapy_ptr, const float32x4_t &width, const float32x4_t &height, const int32x4_t &stride)
51 {
52  const float32x4_t lowerxy = vdupq_n_f32(-1.f);
53 
54  float32x4_t x = vld1q_f32(mapx_ptr);
55  float32x4_t y = vld1q_f32(mapy_ptr);
56 
57  // Clamp x coordinates
58  x = vmaxq_f32(lowerxy, vminq_f32(x, width));
59  y = vmaxq_f32(lowerxy, vminq_f32(y, height));
60 
61  const int32x4_t x_s32 = vcvtq_s32_f32(x);
62  const int32x4_t y_s32 = vcvtq_s32_f32(y);
63 
64  return vmlaq_s32(x_s32, y_s32, stride);
65 }
66 
67 } // namespace
68 
70  : _func(nullptr), _input(nullptr), _output(nullptr), _map_x(nullptr), _map_y(nullptr)
71 {
72 }
73 
75 {
76  return BorderSize(1);
77 }
78 
79 void NERemapKernel::configure(const ITensor *input, const ITensor *map_x, const ITensor *map_y, ITensor *output, InterpolationPolicy policy)
80 {
85 
86  _input = input;
87  _output = output;
88  _map_x = map_x;
89  _map_y = map_y;
90 
91  switch(policy)
92  {
94  {
95  _func = &NERemapKernel::remap_nearest;
96  break;
97  }
99  {
101  break;
102  }
103  default:
104  ARM_COMPUTE_ERROR("Unsupported interpolation mode");
105  break;
106  }
107 
108  constexpr unsigned int num_elems_processed_per_iteration = 16;
109 
110  // Configure kernel window
111  Window win = calculate_max_window(*output->info(), Steps(num_elems_processed_per_iteration));
112 
113  const int total_right = ceil_to_multiple(input->info()->dimension(0), num_elems_processed_per_iteration);
114  const int access_right = total_right + (((total_right - input->info()->dimension(0)) == 0) ? border_size().right : 0);
115 
116  AccessWindowStatic input_access(input->info(), -border_size().left, -border_size().top, access_right, input->info()->dimension(1) + border_size().bottom);
117 
121 
122  update_window_and_padding(win, input_access, mapx_access, mapy_access, output_access);
123 
124  output_access.set_valid_region(win, ValidRegion(Coordinates(), output->info()->tensor_shape()));
125 
126  INEKernel::configure(win);
127 }
128 
129 void NERemapKernel::remap_nearest(const Window &window)
130 {
131  // Don't increment in X and Y direction for the input tensor
132  // A pointer to the start of this plane is needed as base for the precomputed offsets
133  Window win_in(window);
134  win_in.set(Window::DimX, Window::Dimension(0, 0, 0));
135  win_in.set(Window::DimY, Window::Dimension(0, 0, 0));
136 
137  Iterator in(_input, win_in);
138  Iterator out(_output, window);
139  Iterator mapx(_map_x, window);
140  Iterator mapy(_map_y, window);
141 
142  const float32x4_t width = vdupq_n_f32(static_cast<float>(_input->info()->dimension(0)));
143  const float32x4_t height = vdupq_n_f32(static_cast<float>(_input->info()->dimension(1)));
144  const int32x4_t in_stride = vdupq_n_s32(static_cast<int32_t>(_input->info()->strides_in_bytes()[1]));
145 
146  execute_window_loop(window, [&](const Coordinates &)
147  {
148  const auto mapx_ptr = reinterpret_cast<const float *>(mapx.ptr());
149  const auto mapy_ptr = reinterpret_cast<const float *>(mapy.ptr());
150  const uint8_t *in_ptr = in.ptr();
151 
152  const int32x4_t offset0 = offset_nearest_interpolation(mapx_ptr + 0, mapy_ptr + 0, width, height, in_stride);
153  const int32x4_t offset1 = offset_nearest_interpolation(mapx_ptr + 4, mapy_ptr + 4, width, height, in_stride);
154  const int32x4_t offset2 = offset_nearest_interpolation(mapx_ptr + 8, mapy_ptr + 8, width, height, in_stride);
155  const int32x4_t offset3 = offset_nearest_interpolation(mapx_ptr + 12, mapy_ptr + 12, width, height, in_stride);
156 
157  uint8x16_t tmp = vdupq_n_u8(0);
158  tmp = vsetq_lane_u8(in_ptr[vgetq_lane_s32(offset0, 0)], tmp, 0);
159  tmp = vsetq_lane_u8(in_ptr[vgetq_lane_s32(offset0, 1)], tmp, 1);
160  tmp = vsetq_lane_u8(in_ptr[vgetq_lane_s32(offset0, 2)], tmp, 2);
161  tmp = vsetq_lane_u8(in_ptr[vgetq_lane_s32(offset0, 3)], tmp, 3);
162  tmp = vsetq_lane_u8(in_ptr[vgetq_lane_s32(offset1, 0)], tmp, 4);
163  tmp = vsetq_lane_u8(in_ptr[vgetq_lane_s32(offset1, 1)], tmp, 5);
164  tmp = vsetq_lane_u8(in_ptr[vgetq_lane_s32(offset1, 2)], tmp, 6);
165  tmp = vsetq_lane_u8(in_ptr[vgetq_lane_s32(offset1, 3)], tmp, 7);
166  tmp = vsetq_lane_u8(in_ptr[vgetq_lane_s32(offset2, 0)], tmp, 8);
167  tmp = vsetq_lane_u8(in_ptr[vgetq_lane_s32(offset2, 1)], tmp, 9);
168  tmp = vsetq_lane_u8(in_ptr[vgetq_lane_s32(offset2, 2)], tmp, 10);
169  tmp = vsetq_lane_u8(in_ptr[vgetq_lane_s32(offset2, 3)], tmp, 11);
170  tmp = vsetq_lane_u8(in_ptr[vgetq_lane_s32(offset3, 0)], tmp, 12);
171  tmp = vsetq_lane_u8(in_ptr[vgetq_lane_s32(offset3, 1)], tmp, 13);
172  tmp = vsetq_lane_u8(in_ptr[vgetq_lane_s32(offset3, 2)], tmp, 14);
173  tmp = vsetq_lane_u8(in_ptr[vgetq_lane_s32(offset3, 3)], tmp, 15);
174  vst1q_u8(out.ptr(), tmp);
175  },
176  in, out, mapx, mapy);
177 }
178 
179 void NERemapKernel::remap_bilinear(const Window &window)
180 {
181  using namespace scale_helpers;
182 
183  // Don't increment in X and Y direction for the input tensor
184  // A pointer to the start of this plane is needed as base for the precomputed offsets
185  Window win_in(window);
186  win_in.set(Window::DimX, Window::Dimension(0, 0, 0));
187  win_in.set(Window::DimY, Window::Dimension(0, 0, 0));
188 
189  Iterator in(_input, win_in);
190  Iterator out(_output, window);
191  Iterator mapx(_map_x, window);
192  Iterator mapy(_map_y, window);
193 
194  const size_t width = _input->info()->dimension(0);
195  const size_t height = _input->info()->dimension(1);
196  const size_t in_stride = _input->info()->strides_in_bytes()[1];
197 
198  execute_window_loop(window, [&](const Coordinates &)
199  {
200  const auto mapx_ptr = reinterpret_cast<float *>(mapx.ptr());
201  const auto mapy_ptr = reinterpret_cast<float *>(mapy.ptr());
202  const uint8_t *in_ptr = in.ptr();
203 
204  uint8x8_t tmp0 = vdup_n_u8(0);
205  tmp0 = vset_lane_u8(pixel_bilinear_c1_clamp(in_ptr, in_stride, width, height, mapx_ptr[0], mapy_ptr[0]), tmp0, 0);
206  tmp0 = vset_lane_u8(pixel_bilinear_c1_clamp(in_ptr, in_stride, width, height, mapx_ptr[1], mapy_ptr[1]), tmp0, 1);
207  tmp0 = vset_lane_u8(pixel_bilinear_c1_clamp(in_ptr, in_stride, width, height, mapx_ptr[2], mapy_ptr[2]), tmp0, 2);
208  tmp0 = vset_lane_u8(pixel_bilinear_c1_clamp(in_ptr, in_stride, width, height, mapx_ptr[3], mapy_ptr[3]), tmp0, 3);
209  tmp0 = vset_lane_u8(pixel_bilinear_c1_clamp(in_ptr, in_stride, width, height, mapx_ptr[4], mapy_ptr[4]), tmp0, 4);
210  tmp0 = vset_lane_u8(pixel_bilinear_c1_clamp(in_ptr, in_stride, width, height, mapx_ptr[5], mapy_ptr[5]), tmp0, 5);
211  tmp0 = vset_lane_u8(pixel_bilinear_c1_clamp(in_ptr, in_stride, width, height, mapx_ptr[6], mapy_ptr[6]), tmp0, 6);
212  tmp0 = vset_lane_u8(pixel_bilinear_c1_clamp(in_ptr, in_stride, width, height, mapx_ptr[7], mapy_ptr[7]), tmp0, 7);
213 
214  uint8x8_t tmp1 = vdup_n_u8(0);
215  tmp1 = vset_lane_u8(pixel_bilinear_c1_clamp(in_ptr, in_stride, width, height, mapx_ptr[8], mapy_ptr[8]), tmp1, 0);
216  tmp1 = vset_lane_u8(pixel_bilinear_c1_clamp(in_ptr, in_stride, width, height, mapx_ptr[9], mapy_ptr[9]), tmp1, 1);
217  tmp1 = vset_lane_u8(pixel_bilinear_c1_clamp(in_ptr, in_stride, width, height, mapx_ptr[10], mapy_ptr[10]), tmp1, 2);
218  tmp1 = vset_lane_u8(pixel_bilinear_c1_clamp(in_ptr, in_stride, width, height, mapx_ptr[11], mapy_ptr[11]), tmp1, 3);
219  tmp1 = vset_lane_u8(pixel_bilinear_c1_clamp(in_ptr, in_stride, width, height, mapx_ptr[12], mapy_ptr[12]), tmp1, 4);
220  tmp1 = vset_lane_u8(pixel_bilinear_c1_clamp(in_ptr, in_stride, width, height, mapx_ptr[13], mapy_ptr[13]), tmp1, 5);
221  tmp1 = vset_lane_u8(pixel_bilinear_c1_clamp(in_ptr, in_stride, width, height, mapx_ptr[14], mapy_ptr[14]), tmp1, 6);
222  tmp1 = vset_lane_u8(pixel_bilinear_c1_clamp(in_ptr, in_stride, width, height, mapx_ptr[15], mapy_ptr[15]), tmp1, 7);
223 
224  vst1q_u8(out.ptr(), vcombine_u8(tmp0, tmp1));
225  },
226  in, out, mapx, mapy);
227 }
228 
229 void NERemapKernel::run(const Window &window, const ThreadInfo &info)
230 {
231  ARM_COMPUTE_UNUSED(info);
234  ARM_COMPUTE_ERROR_ON(_func == nullptr);
235 
236  (this->*_func)(window);
237 }
unsigned int top
top of the border
Definition: Types.h:375
Window calculate_max_window(const ValidRegion &valid_region, const Steps &steps, bool skip_border, BorderSize border_size)
InterpolationPolicy
Interpolation method.
Definition: Types.h:392
const Window & window() const
The maximum window the kernel can be executed on.
Definition: IKernel.cpp:28
NERemapKernel()
Default constructor.
virtual size_t dimension(size_t index) const =0
Return the size of the requested dimension.
Container for 2D border size.
Definition: Types.h:273
#define ARM_COMPUTE_ERROR(msg)
Print the given message then throw an std::runtime_error.
Definition: Error.h:352
1 channel, 1 U8 per channel
1 channel, 1 F32 per channel
Output values are defined by bilinear interpolation between the pixels.
#define ARM_COMPUTE_ERROR_ON(cond)
If the condition is true then an error message is printed and an exception thrown.
Definition: Error.h:466
Describe one of the image&#39;s dimensions with a start, end and step.
Definition: Window.h:77
unsigned int bottom
bottom of the border
Definition: Types.h:377
Output values are defined to match the source pixel whose center is nearest to the sample position...
Interface for Neon tensor.
Definition: ITensor.h:36
Copyright (c) 2017-2021 Arm Limited.
Implementation of a static rectangular access pattern.
void configure(const ITensor *input, const ITensor *map_x, const ITensor *map_y, ITensor *output, InterpolationPolicy policy)
Initialize the kernel&#39;s input, output and border mode.
static constexpr size_t DimX
Alias for dimension 0 also known as X dimension.
Definition: Window.h:43
bool update_window_and_padding(Window &win, Ts &&... patterns)
Update window and padding size for each of the access patterns.
Definition: WindowHelpers.h:46
#define ARM_COMPUTE_UNUSED(...)
To avoid unused variables warnings.
Definition: Error.h:152
virtual const TensorShape & tensor_shape() const =0
Size for each dimension of the tensor.
auto ceil_to_multiple(S value, T divisor) -> decltype(((value+divisor - 1)/divisor) *divisor)
Computes the smallest number larger or equal to value that is a multiple of divisor.
Definition: Utils.h:71
void run(const Window &window, const ThreadInfo &info) override
Execute the kernel on the passed window.
Class to describe a number of elements in each dimension.
Definition: Steps.h:40
Coordinates of an item.
Definition: Coordinates.h:37
Implementation of a row access pattern.
virtual ITensorInfo * info() const =0
Interface to be implemented by the child class to return the tensor&#39;s metadata.
constexpr uint8_t * ptr() const
Return a pointer to the current pixel.
Definition: Helpers.inl:139
void set(size_t dimension, const Dimension &dim)
Set the values of a given dimension.
Definition: Window.inl:49
BorderSize border_size() const override
The size of the border for that kernel.
unsigned int left
left of the border
Definition: Types.h:378
unsigned int right
right of the border
Definition: Types.h:376
#define ARM_COMPUTE_ERROR_ON_UNCONFIGURED_KERNEL(k)
Definition: Validate.h:941
#define ARM_COMPUTE_ERROR_ON_DATA_TYPE_CHANNEL_NOT_IN(t, c,...)
Definition: Validate.h:790
uint8_t pixel_bilinear_c1_clamp(const T *first_pixel_ptr, size_t stride, size_t width, size_t height, float x, float y)
Return the pixel at (x,y) using bilinear interpolation by clamping when out of borders.
Definition: ScaleHelpers.h:235
static constexpr size_t DimY
Alias for dimension 1 also known as Y dimension.
Definition: Window.h:45
ScaleKernelInfo info(interpolation_policy, default_border_mode, PixelValue(), sampling_policy, false)
Information about executing thread and CPU.
Definition: CPPTypes.h:235
unsigned int num_elems_processed_per_iteration
void execute_window_loop(const Window &w, L &&lambda_function, Ts &&... iterators)
Iterate through the passed window, automatically adjusting the iterators and calling the lambda_funct...
Definition: Helpers.inl:77
virtual const Strides & strides_in_bytes() const =0
The strides in bytes for accessing each dimension of the tensor.
Container for valid region of a window.
Definition: Types.h:188
Iterator updated by execute_window_loop for each window element.
Definition: Helpers.h:46
Describe a multidimensional execution window.
Definition: Window.h:39
#define ARM_COMPUTE_ERROR_ON_INVALID_SUBWINDOW(f, s)
Definition: Validate.h:205
__kernel void remap_bilinear(__global uchar *in_ptr, uint in_stride_x, uint in_step_x, uint in_stride_y, uint in_step_y, uint in_offset_first_element_in_bytes, __global uchar *out_ptr, uint out_stride_x, uint out_step_x, uint out_stride_y, uint out_step_y, uint out_offset_first_element_in_bytes, __global uchar *mapx_ptr, uint mapx_stride_x, uint mapx_step_x, uint mapx_stride_y, uint mapx_step_y, uint mapx_offset_first_element_in_bytes, __global uchar *mapy_ptr, uint mapy_stride_x, uint mapy_step_x, uint mapy_stride_y, uint mapy_step_y, uint mapy_offset_first_element_in_bytes, const float width, const float height)
Performs a remapping of an input image to an output given two remapping image using bilinear as inter...
Definition: remap.cl:113