diff options
author | Sang-Hoon Park <sang-hoon.park@arm.com> | 2020-06-24 13:34:04 +0100 |
---|---|---|
committer | Sang-Hoon Park <sang-hoon.park@arm.com> | 2020-06-30 11:03:10 +0000 |
commit | 3687ee1e7719436ff155a35911946b045903e8b6 (patch) | |
tree | 127d4b9da3996f473cef51f01a07fb02b75c0f1b /src/core/NEON/kernels | |
parent | 781cba7f33e056b1ca470ab34eb478177768eaf4 (diff) | |
download | ComputeLibrary-3687ee1e7719436ff155a35911946b045903e8b6.tar.gz |
COMPMID-3539: Ignore align_corners for scaled size of 1
Scale kernels failed to validate when align_corners is true
for scaled output size 1. Change this behavior to ignoring
align_corners value to be aligned with expected behavior of
higher-level frameworks.
Also the minimum output size generated by the fixture for
Scale kernels is changed to 1.
Change-Id: Ib8e479af8bc43de3780005545f0c53fe195dc22e
Signed-off-by: Sang-Hoon Park <sang-hoon.park@arm.com>
Reviewed-on: https://review.mlplatform.org/c/ml/ComputeLibrary/+/3478
Tested-by: Arm Jenkins <bsgcomp@arm.com>
Reviewed-by: Georgios Pinitas <georgios.pinitas@arm.com>
Comments-Addressed: Arm Jenkins <bsgcomp@arm.com>
Diffstat (limited to 'src/core/NEON/kernels')
-rw-r--r-- | src/core/NEON/kernels/NEScaleKernel.cpp | 32 |
1 files changed, 11 insertions, 21 deletions
diff --git a/src/core/NEON/kernels/NEScaleKernel.cpp b/src/core/NEON/kernels/NEScaleKernel.cpp index 857084ef7e..0f329a1c2c 100644 --- a/src/core/NEON/kernels/NEScaleKernel.cpp +++ b/src/core/NEON/kernels/NEScaleKernel.cpp @@ -30,6 +30,8 @@ #include "arm_compute/core/Window.h" #include "arm_compute/core/utils/misc/Utility.h" +#include "src/core/utils/ScaleUtils.h" + #include <arm_neon.h> namespace arm_compute @@ -68,19 +70,7 @@ Status validate_arguments(const ITensorInfo *input, const ITensorInfo *dx, const ARM_COMPUTE_RETURN_ERROR_ON_DATA_TYPE_CHANNEL_NOT_IN(dy, 1, DataType::F32); } - ARM_COMPUTE_RETURN_ERROR_ON(info.align_corners && !is_align_corners_allowed(info.sampling_policy)); - - if(info.align_corners && is_align_corners_allowed(info.sampling_policy)) - { - // For bilinear method with aligned corners, the resize ratio will - // be calculated by (input_size - 1)/(output_size - 1). Belows are - // checking possible overflows. - const auto input_width = input->dimension(width_index); - const auto input_height = input->dimension(height_index); - - ARM_COMPUTE_RETURN_ERROR_ON(input_width == 0 || input_height == 0); - ARM_COMPUTE_RETURN_ERROR_ON((output_width - 1 == 0) || (output_height - 1 == 0)); - } + ARM_COMPUTE_RETURN_ERROR_ON(info.align_corners && !arm_compute::scale_utils::is_align_corners_allowed_sampling_policy(info.sampling_policy)); if(info.interpolation_policy == InterpolationPolicy::AREA) { @@ -378,7 +368,7 @@ void NEScaleKernel::configure(const ITensor *input, const ITensor *dx, const ITe _border_mode = info.border_mode; _constant_border_value = info.constant_border_value; _use_padding = info.use_padding; - _align_corners = info.align_corners; + _align_corners = info.align_corners && arm_compute::scale_utils::is_align_corners_allowed_output_shape(output->info()->tensor_shape(), data_layout); if(info.sampling_policy == SamplingPolicy::CENTER) { @@ -386,8 +376,8 @@ void NEScaleKernel::configure(const ITensor *input, const ITensor *dx, const ITe } // Compute the ratio between source width/height and destination width/height - const auto wr = arm_compute::calculate_resize_ratio(input->info()->dimension(idx_width), output->info()->dimension(idx_width), _align_corners); - const auto hr = arm_compute::calculate_resize_ratio(input->info()->dimension(idx_height), output->info()->dimension(idx_height), _align_corners); + const auto wr = arm_compute::scale_utils::calculate_resize_ratio(input->info()->dimension(idx_width), output->info()->dimension(idx_width), _align_corners); + const auto hr = arm_compute::scale_utils::calculate_resize_ratio(input->info()->dimension(idx_height), output->info()->dimension(idx_height), _align_corners); // Add constant border only on top in case of NHWC layout if(data_layout == DataLayout::NHWC) @@ -437,7 +427,7 @@ void NEScaleKernel::scale_nearest_nchw(const Window &window) const size_t input_stride = _input->info()->strides_in_bytes()[1]; // Compute the ratio between source height and destination height - const auto hr = arm_compute::calculate_resize_ratio(_input->info()->dimension(1), _output->info()->dimension(1), _align_corners); + const auto hr = arm_compute::scale_utils::calculate_resize_ratio(_input->info()->dimension(1), _output->info()->dimension(1), _align_corners); // Don't increment in X and Y direction for the input tensor // A pointer to the start of this plane is needed as base for the precomputed offsets @@ -670,7 +660,7 @@ void NEScaleKernel::scale_bilinear_nchw(const Window &window) ARM_COMPUTE_ERROR_ON_DATA_TYPE_CHANNEL_NOT_IN(_input, 1, DataType::U8, DataType::QASYMM8, DataType::QASYMM8_SIGNED, DataType::S16, DataType::F16, DataType::F32); // Compute the ratio between source height and destination height - const auto hr = arm_compute::calculate_resize_ratio(_input->info()->dimension(1), _output->info()->dimension(1), _align_corners); + const auto hr = arm_compute::scale_utils::calculate_resize_ratio(_input->info()->dimension(1), _output->info()->dimension(1), _align_corners); // Don't increment in X and Y direction for the input tensor // A pointer to the start of this plane is needed as base for the precomputed offsets @@ -971,8 +961,8 @@ void NEScaleKernel::scale_area_nchw(const Window &window) Iterator in(_input, win_in); Iterator out(_output, window); - const auto wr = arm_compute::calculate_resize_ratio(_input->info()->dimension(0), _output->info()->dimension(0), _align_corners); - const auto hr = arm_compute::calculate_resize_ratio(_input->info()->dimension(1), _output->info()->dimension(1), _align_corners); + const auto wr = arm_compute::scale_utils::calculate_resize_ratio(_input->info()->dimension(0), _output->info()->dimension(0), _align_corners); + const auto hr = arm_compute::scale_utils::calculate_resize_ratio(_input->info()->dimension(1), _output->info()->dimension(1), _align_corners); const auto w = _input->info()->dimension(0); const auto h = _input->info()->dimension(1); const size_t in_stride = _input->info()->strides_in_bytes()[1]; @@ -1019,7 +1009,7 @@ void NEScaleKernel::scale_nhwc(const Window &window) const size_t input_stride_c = _input->info()->strides_in_bytes()[idx_channels]; // Compute the ratio between source height and destination height - const auto hr = arm_compute::calculate_resize_ratio(_input->info()->dimension(idx_height), _output->info()->dimension(idx_height), _align_corners); + const auto hr = arm_compute::scale_utils::calculate_resize_ratio(_input->info()->dimension(idx_height), _output->info()->dimension(idx_height), _align_corners); // Don't increment in width/height/channels for the input tensor // A pointer to the start of this plane is needed as base for the precomputed offsets |