diff options
author | Georgios Pinitas <georgios.pinitas@arm.com> | 2020-12-01 17:41:34 +0000 |
---|---|---|
committer | Georgios Pinitas <georgios.pinitas@arm.com> | 2020-12-02 15:21:11 +0000 |
commit | 96b16b65dd96351b8af1b2a785856ce13cc8ba84 (patch) | |
tree | d05ae4c07753e9e15d9861eb36df783dbe7a00a0 /src | |
parent | d308df3186b4f6057f94b45b7bed7935c618ea80 (diff) | |
download | ComputeLibrary-96b16b65dd96351b8af1b2a785856ce13cc8ba84.tar.gz |
Remove support for (NE/CL)LocallyConnectedLayer
Remove out-of-date and unmaintained LocallyConnectedLayer for both NEON
and OpenCL.
Resolves: COMPMID-3924
Signed-off-by: Georgios Pinitas <georgios.pinitas@arm.com>
Change-Id: Ia61398ed8cfa3876f41c1b342c4a80d1cca0ca83
Reviewed-on: https://review.mlplatform.org/c/ml/ComputeLibrary/+/4634
Reviewed-by: Michele Di Giorgio <michele.digiorgio@arm.com>
Tested-by: Arm Jenkins <bsgcomp@arm.com>
Comments-Addressed: Arm Jenkins <bsgcomp@arm.com>
Diffstat (limited to 'src')
-rw-r--r-- | src/core/CL/CLKernels.h | 1 | ||||
-rw-r--r-- | src/core/CL/cl_kernels/gemm.cl | 70 | ||||
-rw-r--r-- | src/core/CL/kernels/CLLocallyConnectedMatrixMultiplyKernel.cpp | 145 | ||||
-rw-r--r-- | src/core/CL/kernels/CLLocallyConnectedMatrixMultiplyKernel.h | 85 | ||||
-rw-r--r-- | src/core/NEON/NEKernels.h | 1 | ||||
-rw-r--r-- | src/core/NEON/kernels/NELocallyConnectedMatrixMultiplyKernel.cpp | 391 | ||||
-rw-r--r-- | src/core/NEON/kernels/NELocallyConnectedMatrixMultiplyKernel.h | 79 | ||||
-rw-r--r-- | src/runtime/CL/functions/CLLocallyConnectedLayer.cpp | 226 | ||||
-rw-r--r-- | src/runtime/NEON/functions/NELocallyConnectedLayer.cpp | 203 |
9 files changed, 1 insertions, 1200 deletions
diff --git a/src/core/CL/CLKernels.h b/src/core/CL/CLKernels.h index b335372fa9..eea90eb599 100644 --- a/src/core/CL/CLKernels.h +++ b/src/core/CL/CLKernels.h @@ -103,7 +103,6 @@ #include "src/core/CL/kernels/CLIntegralImageKernel.h" #include "src/core/CL/kernels/CLL2NormalizeLayerKernel.h" #include "src/core/CL/kernels/CLLKTrackerKernel.h" -#include "src/core/CL/kernels/CLLocallyConnectedMatrixMultiplyKernel.h" #include "src/core/CL/kernels/CLMagnitudePhaseKernel.h" #include "src/core/CL/kernels/CLMaxUnpoolingLayerKernel.h" #include "src/core/CL/kernels/CLMeanStdDevKernel.h" diff --git a/src/core/CL/cl_kernels/gemm.cl b/src/core/CL/cl_kernels/gemm.cl index b6afb85aa4..6883aafee5 100644 --- a/src/core/CL/cl_kernels/gemm.cl +++ b/src/core/CL/cl_kernels/gemm.cl @@ -4379,72 +4379,4 @@ __kernel void gemm_ma_f16(TENSOR3D_DECLARATION(src), vstore8(out, 0, (__global half *)dst.ptr); } #endif // defined(ARM_COMPUTE_OPENCL_FP16_ENABLED) -#endif // defined(BETA) - -#if defined(WIDTH_VECTOR_A) -/** This OpenCL kernel computes the vector by matrix multiplication between each row of A (src0) and matrix B (src1) used for locally connected layer - * - * @note The width of A need to be passed at compile time using -DWIDTH_VECTOR_A - * - * @note The input A and matrix B must not be reshaped - * - * @param[in] src0_ptr Pointer to the source matrix. Supported data types: F32 - * @param[in] src0_stride_x Stride of the source matrix in X dimension (in bytes) - * @param[in] src0_step_x src_stride_x * number of elements along X processed per workitem(in bytes) - * @param[in] src0_stride_y Stride of the source matrix in Y dimension (in bytes) - * @param[in] src0_step_y src_stride_y * number of elements along Y processed per workitem(in bytes) - * @param[in] src0_offset_first_element_in_bytes The offset of the first element in the source matrix - * @param[in] src1_ptr Pointer to the source matrix. Supported data types: same as @p src0_ptr - * @param[in] src1_stride_x Stride of the source matrix in X dimension (in bytes) - * @param[in] src1_step_x src_stride_x * number of elements along X processed per workitem(in bytes) - * @param[in] src1_stride_y Stride of the source matrix in Y dimension (in bytes) - * @param[in] src1_step_y src_stride_y * number of elements along Y processed per workitem(in bytes) - * @param[in] src1_stride_z Stride of the source matrix in Z dimension (in bytes) - * @param[in] src1_step_z src_stride_z * number of elements along Z processed per workitem(in bytes) - * @param[in] src1_offset_first_element_in_bytes The offset of the first element in the source matrix - * @param[out] dst_ptr Pointer to the destination matrix Supported data types: same as @p src0_ptr - * @param[in] dst_stride_x Stride of the destination matrix in X dimension (in bytes) - * @param[in] dst_step_x dst_gx_stride_x * number of elements along X processed per workitem(in bytes) - * @param[in] dst_stride_y Stride of the destination matrix in Y dimension (in bytes) - * @param[in] dst_step_y dst_gx_stride_y * number of elements along Y processed per workitem(in bytes) - * @param[in] dst_offset_first_element_in_bytes The offset of the first element in the destination matrix - */ -__kernel void gemm_lc_vm_f32(IMAGE_DECLARATION(src0), - TENSOR3D_DECLARATION(src1), - IMAGE_DECLARATION(dst)) -{ - int idx = get_global_id(0) * 4; - int idy = get_global_id(1); - - // Compute the address for the vector A and matrix B - int2 src_addr = ((int2)(src0_offset_first_element_in_bytes + src0_stride_y * idy, src1_offset_first_element_in_bytes + src1_stride_z * idy)); - src_addr.s1 += idx * sizeof(float); - - int end_row_vec_a = src_addr.s0 + (WIDTH_VECTOR_A * sizeof(float)); - - float4 acc = 0.0f; - - for(; src_addr.s0 <= (end_row_vec_a - 2 * (int)sizeof(float)); src_addr += (int2)(2 * sizeof(float), 2 * src1_stride_y)) - { - float2 a0 = vload2(0, (__global float *)(src0_ptr + src_addr.s0)); - float4 b0 = vload4(0, (__global float *)(src1_ptr + src_addr.s1)); - float4 b1 = vload4(0, (__global float *)(src1_ptr + src_addr.s1 + src1_stride_y)); - - acc += b0 * (float4)a0.s0; - acc += b1 * (float4)a0.s1; - } - - for(; src_addr.s0 < end_row_vec_a; src_addr += (int2)(sizeof(float), src1_stride_y)) - { - float a0 = *((__global float *)(src0_ptr + src_addr.s0)); - float4 b0 = vload4(0, (__global float *)(src1_ptr + src_addr.s1)); - - acc += b0 * (float4)a0; - } - - // Compute destination address - Image dst = CONVERT_TO_IMAGE_STRUCT(dst); - - vstore4(acc, 0, (__global float *)(offset(&dst, 0, 0))); -} -#endif // defined(WIDTH_VECTOR_A) +#endif // defined(BETA)
\ No newline at end of file diff --git a/src/core/CL/kernels/CLLocallyConnectedMatrixMultiplyKernel.cpp b/src/core/CL/kernels/CLLocallyConnectedMatrixMultiplyKernel.cpp deleted file mode 100644 index 49e04c32c2..0000000000 --- a/src/core/CL/kernels/CLLocallyConnectedMatrixMultiplyKernel.cpp +++ /dev/null @@ -1,145 +0,0 @@ -/* - * Copyright (c) 2017-2020 Arm Limited. - * - * SPDX-License-Identifier: MIT - * - * Permission is hereby granted, free of charge, to any person obtaining a copy - * of this software and associated documentation files (the "Software"), to - * deal in the Software without restriction, including without limitation the - * rights to use, copy, modify, merge, publish, distribute, sublicense, and/or - * sell copies of the Software, and to permit persons to whom the Software is - * furnished to do so, subject to the following conditions: - * - * The above copyright notice and this permission notice shall be included in all - * copies or substantial portions of the Software. - * - * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR - * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, - * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE - * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER - * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, - * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE - * SOFTWARE. - */ -#include "src/core/CL/kernels/CLLocallyConnectedMatrixMultiplyKernel.h" - -#include "arm_compute/core/CL/CLHelpers.h" -#include "arm_compute/core/CL/ICLTensor.h" -#include "arm_compute/core/Helpers.h" -#include "arm_compute/core/Utils.h" -#include "src/core/AccessWindowStatic.h" -#include "src/core/CL/CLValidate.h" -#include "src/core/helpers/WindowHelpers.h" - -namespace arm_compute -{ -CLLocallyConnectedMatrixMultiplyKernel::CLLocallyConnectedMatrixMultiplyKernel() - : _input0(nullptr), _input1(nullptr), _output(nullptr) -{ -} - -namespace -{ -Status validate_arguments(const ITensorInfo *input0, const ITensorInfo *input1, const ITensorInfo *output) -{ - ARM_COMPUTE_RETURN_ERROR_ON_NULLPTR(input0, input1, output); - ARM_COMPUTE_RETURN_ERROR_ON_F16_UNSUPPORTED(input0); - ARM_COMPUTE_RETURN_ERROR_ON_DATA_TYPE_CHANNEL_NOT_IN(input0, 1, DataType::F32); - ARM_COMPUTE_RETURN_ERROR_ON_MISMATCHING_DATA_TYPES(input0, input1, output); - ARM_COMPUTE_RETURN_ERROR_ON(input0->dimension(0) != input1->dimension(1)); - - return Status{}; -} - -std::tuple<Status, Window> validate_and_configure_window(ITensorInfo *input0, ITensorInfo *input1, ITensorInfo *output) -{ - const unsigned int num_elems_processed_per_iteration_x = max_cl_vector_width / data_size_from_type(input0->data_type()); - - Window win = calculate_max_window(*output, Steps(num_elems_processed_per_iteration_x)); - - AccessWindowHorizontal input0_access(input0, 0, num_elems_processed_per_iteration_x); - AccessWindowHorizontal input1_access(input1, 0, num_elems_processed_per_iteration_x); - AccessWindowHorizontal output_access(output, 0, num_elems_processed_per_iteration_x); - - bool window_changed = update_window_and_padding(win, input0_access, input1_access, output_access); - - output_access.set_valid_region(win, ValidRegion(Coordinates(), output->tensor_shape())); - - Status err = (window_changed) ? ARM_COMPUTE_CREATE_ERROR(ErrorCode::RUNTIME_ERROR, "Insufficient Padding!") : Status{}; - - return std::make_tuple(err, win); -} -} // namespace - -void CLLocallyConnectedMatrixMultiplyKernel::configure(const ICLTensor *input0, const ICLTensor *input1, ICLTensor *output) -{ - configure(CLKernelLibrary::get().get_compile_context(), input0, input1, output); -} - -void CLLocallyConnectedMatrixMultiplyKernel::configure(const CLCompileContext &compile_context, const ICLTensor *input0, const ICLTensor *input1, ICLTensor *output) -{ - ARM_COMPUTE_ERROR_ON_NULLPTR(input0, input1, output); - ARM_COMPUTE_ERROR_THROW_ON(validate_arguments(input0->info(), input1->info(), output->info())); - - _input0 = input0; - _input1 = input1; - _output = output; - - cl::NDRange lws_hint; - if(output->info()->dimension(1) == 196) - { - lws_hint = cl::NDRange(1, 7); - } - else - { - lws_hint = cl::NDRange(8, 8); - } - - std::ostringstream mm_arguments; - std::set<std::string> build_opts; - - mm_arguments << "-DWIDTH_VECTOR_A=" << input0->info()->dimension(0) << " "; - build_opts.emplace(mm_arguments.str()); - - // Create kernel - std::string data_type_name = lower_string(string_from_data_type(input0->info()->data_type())); - _kernel = create_kernel(compile_context, ("gemm_lc_vm_" + data_type_name), build_opts); - - // Configure kernel window - auto win_config = validate_and_configure_window(input0->info(), input1->info(), output->info()); - - ARM_COMPUTE_ERROR_THROW_ON(std::get<0>(win_config)); - - ICLKernel::configure_internal(std::get<1>(win_config), lws_hint); -} - -Status CLLocallyConnectedMatrixMultiplyKernel::validate(const ITensorInfo *input0, const ITensorInfo *input1, const ITensorInfo *output) -{ - ARM_COMPUTE_RETURN_ON_ERROR(validate_arguments(input0, input1, output)); - ARM_COMPUTE_RETURN_ON_ERROR(std::get<0>(validate_and_configure_window(input0->clone().get(), input1->clone().get(), output->clone().get()))); - - return Status{}; -} - -void CLLocallyConnectedMatrixMultiplyKernel::run(const Window &window, cl::CommandQueue &queue) -{ - ARM_COMPUTE_ERROR_ON_UNCONFIGURED_KERNEL(this); - ARM_COMPUTE_ERROR_ON_INVALID_SUBWINDOW(ICLKernel::window(), window); - - Window slice = window.first_slice_window_2D(); - - Window matrix_b_window; - matrix_b_window.use_tensor_dimensions(_input1->info()->tensor_shape()); - Window slice_matrix_b = matrix_b_window.first_slice_window_3D(); - - do - { - unsigned int idx = 0; - add_2D_tensor_argument(idx, _input0, slice); - add_3D_tensor_argument(idx, _input1, slice_matrix_b); - add_2D_tensor_argument(idx, _output, slice); - enqueue(queue, *this, slice, lws_hint()); - } - while(window.slide_window_slice_2D(slice)); -} -} // namespace arm_compute diff --git a/src/core/CL/kernels/CLLocallyConnectedMatrixMultiplyKernel.h b/src/core/CL/kernels/CLLocallyConnectedMatrixMultiplyKernel.h deleted file mode 100644 index 5d0a22afa5..0000000000 --- a/src/core/CL/kernels/CLLocallyConnectedMatrixMultiplyKernel.h +++ /dev/null @@ -1,85 +0,0 @@ -/* - * Copyright (c) 2017-2020 Arm Limited. - * - * SPDX-License-Identifier: MIT - * - * Permission is hereby granted, free of charge, to any person obtaining a copy - * of this software and associated documentation files (the "Software"), to - * deal in the Software without restriction, including without limitation the - * rights to use, copy, modify, merge, publish, distribute, sublicense, and/or - * sell copies of the Software, and to permit persons to whom the Software is - * furnished to do so, subject to the following conditions: - * - * The above copyright notice and this permission notice shall be included in all - * copies or substantial portions of the Software. - * - * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR - * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, - * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE - * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER - * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, - * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE - * SOFTWARE. - */ -#ifndef ARM_COMPUTE_CLLOCALLYCONNECTEDMATRIXMULTIPLYKERNEL_H -#define ARM_COMPUTE_CLLOCALLYCONNECTEDMATRIXMULTIPLYKERNEL_H - -#include "src/core/CL/ICLKernel.h" - -namespace arm_compute -{ -class ICLTensor; - -/** OpenCL kernel to multiply each row of first tensor with low 2 dimensions of second tensor. - * - * @attention The second input tensor must have at least 2 dimensions (matrix) - * - */ -class CLLocallyConnectedMatrixMultiplyKernel : public ICLKernel -{ -public: - /** Default constructor */ - CLLocallyConnectedMatrixMultiplyKernel(); - /** Prevent instances of this class from being copied (As this class contains pointers) */ - CLLocallyConnectedMatrixMultiplyKernel(const CLLocallyConnectedMatrixMultiplyKernel &) = delete; - /** Prevent instances of this class from being copied (As this class contains pointers) */ - CLLocallyConnectedMatrixMultiplyKernel &operator=(const CLLocallyConnectedMatrixMultiplyKernel &) = delete; - /** Allow instances of this class to be moved */ - CLLocallyConnectedMatrixMultiplyKernel(CLLocallyConnectedMatrixMultiplyKernel &&) = default; - /** Allow instances of this class to be moved */ - CLLocallyConnectedMatrixMultiplyKernel &operator=(CLLocallyConnectedMatrixMultiplyKernel &&) = default; - /** Initialise the kernel's input, output and alpha - * - * @param[in] input0 First input tensor. Data types supported: F32 - * @param[in] input1 Second input tensor. Data type supported: same as @p input0 - * @param[out] output Output tensor to store the result. Data type supported: same as @p input0 - */ - void configure(const ICLTensor *input0, const ICLTensor *input1, ICLTensor *output); - /** Initialise the kernel's input, output and alpha - * - * @param[in] compile_context The compile context to be used. - * @param[in] input0 First input tensor. Data types supported: F32 - * @param[in] input1 Second input tensor. Data type supported: same as @p input0 - * @param[out] output Output tensor to store the result. Data type supported: same as @p input0 - */ - void configure(const CLCompileContext &compile_context, const ICLTensor *input0, const ICLTensor *input1, ICLTensor *output); - /** Static function to check if given info will lead to a valid configuration of @ref CLLocallyConnectedMatrixMultiplyKernel - * - * @param[in] input0 First input tensor info. Data types supported: F32 - * @param[in] input1 Second input tensor info. Data type supported: same as @p input0 - * @param[in] output Output tensor info. Data type supported: same as @p input0 - * - * @return a status - */ - static Status validate(const ITensorInfo *input0, const ITensorInfo *input1, const ITensorInfo *output); - - // Inherited methods overridden: - void run(const Window &window, cl::CommandQueue &queue) override; - -private: - const ICLTensor *_input0; - const ICLTensor *_input1; - ICLTensor *_output; -}; -} // namespace arm_compute -#endif /* ARM_COMPUTE_CLLOCALLYCONNECTEDMATRIXMULTIPLYKERNEL_H */ diff --git a/src/core/NEON/NEKernels.h b/src/core/NEON/NEKernels.h index 88fb8d4023..091130c23a 100644 --- a/src/core/NEON/NEKernels.h +++ b/src/core/NEON/NEKernels.h @@ -99,7 +99,6 @@ #include "src/core/NEON/kernels/NEIntegralImageKernel.h" #include "src/core/NEON/kernels/NEL2NormalizeLayerKernel.h" #include "src/core/NEON/kernels/NELKTrackerKernel.h" -#include "src/core/NEON/kernels/NELocallyConnectedMatrixMultiplyKernel.h" #include "src/core/NEON/kernels/NELogicalKernel.h" #include "src/core/NEON/kernels/NEMagnitudePhaseKernel.h" #include "src/core/NEON/kernels/NEMaxUnpoolingLayerKernel.h" diff --git a/src/core/NEON/kernels/NELocallyConnectedMatrixMultiplyKernel.cpp b/src/core/NEON/kernels/NELocallyConnectedMatrixMultiplyKernel.cpp deleted file mode 100644 index f11694dee4..0000000000 --- a/src/core/NEON/kernels/NELocallyConnectedMatrixMultiplyKernel.cpp +++ /dev/null @@ -1,391 +0,0 @@ -/* - * Copyright (c) 2017-2020 Arm Limited. - * - * SPDX-License-Identifier: MIT - * - * Permission is hereby granted, free of charge, to any person obtaining a copy - * of this software and associated documentation files (the "Software"), to - * deal in the Software without restriction, including without limitation the - * rights to use, copy, modify, merge, publish, distribute, sublicense, and/or - * sell copies of the Software, and to permit persons to whom the Software is - * furnished to do so, subject to the following conditions: - * - * The above copyright notice and this permission notice shall be included in all - * copies or substantial portions of the Software. - * - * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR - * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, - * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE - * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER - * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, - * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE - * SOFTWARE. - */ -#include "src/core/NEON/kernels/NELocallyConnectedMatrixMultiplyKernel.h" - -#include "arm_compute/core/Error.h" -#include "arm_compute/core/Helpers.h" -#include "arm_compute/core/IAccessWindow.h" -#include "arm_compute/core/ITensor.h" -#include "arm_compute/core/TensorInfo.h" -#include "arm_compute/core/Types.h" -#include "arm_compute/core/Utils.h" -#include "arm_compute/core/Validate.h" -#include "arm_compute/core/Window.h" -#include "src/core/CPP/Validate.h" -#include "src/core/NEON/NEFixedPoint.h" -#include "src/core/helpers/AutoConfiguration.h" -#include "src/core/helpers/WindowHelpers.h" - -#include <arm_neon.h> -#include <cstddef> -#include <cstdint> -#include <tuple> - -namespace arm_compute -{ -class Coordinates; - -namespace -{ -void vector_matrix_multiply_f16(const ITensor *input0, const ITensor *input1, ITensor *output, const Window &window, const ThreadInfo &info) -{ -#ifdef __ARM_FEATURE_FP16_VECTOR_ARITHMETIC - const auto width_matrix_b = static_cast<int>(output->info()->dimension(0)); - const auto in_b_stride = static_cast<int>(input1->info()->strides_in_bytes()[1] / data_size_from_type(input1->info()->data_type())); - const auto num_elems_vec_a = static_cast<int>(input0->info()->dimension(0)); - - // The implementation computes 16 elements per iteration - const int window_start_x = 16 * info.thread_id; - const int window_step_x = 16 * info.num_threads; - // Make sure (window_end_x - window_start_x) is a multiple of window_step_x - const int window_end_x = ceil_to_multiple(width_matrix_b - window_start_x, window_step_x) + window_start_x; - - Window win_out(window); - win_out.set(Window::DimX, Window::Dimension(window_start_x, window_end_x, window_step_x)); - - Window win_a(window); - win_a.set(Window::DimX, Window::Dimension(0, 1, 1)); - - Iterator ina(input0, win_a); - Iterator out(output, win_out); - - execute_window_loop(win_out, [&](const Coordinates & id) - { - if(id.x() > width_matrix_b) - { - return; - } - - float16x8_t acc0 = vdupq_n_f16(0.f); - float16x8_t acc1 = vdupq_n_f16(0.f); - float16x8_t acc2 = vdupq_n_f16(0.f); - float16x8_t acc3 = vdupq_n_f16(0.f); - - auto vec_a = reinterpret_cast<const float16_t *>(ina.ptr()); - auto matrix_b = reinterpret_cast<const float16_t *>(input1->ptr_to_element(Coordinates(id[0], 0, id[1]))); - - const float16_t *vec_a_end_addr = vec_a + num_elems_vec_a; - - for(; vec_a <= (vec_a_end_addr - 4);) - { - const float16x4_t a0l = vld1_f16(vec_a); - - float16x8_t b00 = vld1q_f16(matrix_b); - float16x8_t b01 = vld1q_f16(matrix_b + 8 + 0 * in_b_stride); - float16x8_t b02 = vld1q_f16(matrix_b + 16 + 0 * in_b_stride); - float16x8_t b03 = vld1q_f16(matrix_b + 24 + 0 * in_b_stride); - - float16x8_t b10 = vld1q_f16(matrix_b + 0 + 1 * in_b_stride); - float16x8_t b11 = vld1q_f16(matrix_b + 8 + 1 * in_b_stride); - float16x8_t b12 = vld1q_f16(matrix_b + 16 + 1 * in_b_stride); - float16x8_t b13 = vld1q_f16(matrix_b + 24 + 1 * in_b_stride); - - acc0 = vaddq_f16(acc0, vmulq_lane_f16(b00, a0l, 0)); - acc1 = vaddq_f16(acc1, vmulq_lane_f16(b01, a0l, 0)); - acc2 = vaddq_f16(acc2, vmulq_lane_f16(b02, a0l, 0)); - acc3 = vaddq_f16(acc3, vmulq_lane_f16(b03, a0l, 0)); - acc0 = vaddq_f16(acc0, vmulq_lane_f16(b10, a0l, 1)); - acc1 = vaddq_f16(acc1, vmulq_lane_f16(b11, a0l, 1)); - acc2 = vaddq_f16(acc2, vmulq_lane_f16(b12, a0l, 1)); - acc3 = vaddq_f16(acc3, vmulq_lane_f16(b13, a0l, 1)); - - matrix_b += 2 * in_b_stride; - - b00 = vld1q_f16(matrix_b); - b01 = vld1q_f16(matrix_b + 8 + 0 * in_b_stride); - b02 = vld1q_f16(matrix_b + 16 + 0 * in_b_stride); - b03 = vld1q_f16(matrix_b + 24 + 0 * in_b_stride); - b10 = vld1q_f16(matrix_b + 0 + 1 * in_b_stride); - b11 = vld1q_f16(matrix_b + 8 + 1 * in_b_stride); - b12 = vld1q_f16(matrix_b + 16 + 1 * in_b_stride); - b13 = vld1q_f16(matrix_b + 24 + 1 * in_b_stride); - - acc0 = vaddq_f16(acc0, vmulq_lane_f16(b00, a0l, 2)); - acc1 = vaddq_f16(acc1, vmulq_lane_f16(b01, a0l, 2)); - acc2 = vaddq_f16(acc2, vmulq_lane_f16(b02, a0l, 2)); - acc3 = vaddq_f16(acc3, vmulq_lane_f16(b03, a0l, 2)); - acc0 = vaddq_f16(acc0, vmulq_lane_f16(b10, a0l, 3)); - acc1 = vaddq_f16(acc1, vmulq_lane_f16(b11, a0l, 3)); - acc2 = vaddq_f16(acc2, vmulq_lane_f16(b12, a0l, 3)); - acc3 = vaddq_f16(acc3, vmulq_lane_f16(b13, a0l, 3)); - - vec_a += 4; - matrix_b += 2 * in_b_stride; - } - - for(; vec_a < vec_a_end_addr;) - { - const float16_t a0 = *vec_a; - const float16x8_t b00 = vld1q_f16(matrix_b); - const float16x8_t b01 = vld1q_f16(matrix_b + 8 + 0 * in_b_stride); - const float16x8_t b02 = vld1q_f16(matrix_b + 16 + 0 * in_b_stride); - const float16x8_t b03 = vld1q_f16(matrix_b + 24 + 0 * in_b_stride); - - acc0 = vaddq_f16(acc0, vmulq_n_f16(b00, a0)); - acc1 = vaddq_f16(acc1, vmulq_n_f16(b01, a0)); - acc2 = vaddq_f16(acc2, vmulq_n_f16(b02, a0)); - acc3 = vaddq_f16(acc3, vmulq_n_f16(b03, a0)); - - vec_a += 1; - matrix_b += in_b_stride; - } - - const auto vec_out = reinterpret_cast<float16_t *>(out.ptr()); - - vst1q_f16(vec_out + 0, acc0); - vst1q_f16(vec_out + 8, acc1); - vst1q_f16(vec_out + 16, acc2); - vst1q_f16(vec_out + 24, acc3); - }, - ina, out); -#else /* __ARM_FEATURE_FP16_VECTOR_ARITHMETIC */ - ARM_COMPUTE_UNUSED(input0); - ARM_COMPUTE_UNUSED(input1); - ARM_COMPUTE_UNUSED(output); - ARM_COMPUTE_UNUSED(window); - ARM_COMPUTE_UNUSED(info); - ARM_COMPUTE_ERROR("Not supported, recompile with -march=armv8.2-a+fp16+simd."); -#endif /* __ARM_FEATURE_FP16_VECTOR_ARITHMETIC */ -} - -void vector_matrix_multiply_f32(const ITensor *input0, const ITensor *input1, ITensor *output, const Window &window, const ThreadInfo &info) -{ - const auto width_matrix_b = static_cast<int>(output->info()->dimension(0)); - const auto in_b_stride = static_cast<int>(input1->info()->strides_in_bytes()[1] / data_size_from_type(input1->info()->data_type())); - const auto num_elems_vec_a = static_cast<int>(input0->info()->dimension(0)); - - // The implementation computes 16 elements per iteration - const int window_start_x = 16 * info.thread_id; - const int window_step_x = 16 * info.num_threads; - // Make sure (window_end_x - window_start_x) is a multiple of window_step_x - const int window_end_x = ceil_to_multiple(width_matrix_b - window_start_x, window_step_x) + window_start_x; - - Window win_out(window); - win_out.set(Window::DimX, Window::Dimension(window_start_x, window_end_x, window_step_x)); - - Window win_a(window); - win_a.set(Window::DimX, Window::Dimension(0, 0, 0)); - - Iterator ina(input0, win_a); - Iterator out(output, win_out); - - execute_window_loop(win_out, [&](const Coordinates & id) - { - if(id.x() > width_matrix_b) - { - return; - } - - float32x4_t acc0 = vdupq_n_f32(0.f); - float32x4_t acc1 = vdupq_n_f32(0.f); - float32x4_t acc2 = vdupq_n_f32(0.f); - float32x4_t acc3 = vdupq_n_f32(0.f); - - auto vec_a = reinterpret_cast<const float *>(ina.ptr()); - auto matrix_b = reinterpret_cast<const float *>(input1->ptr_to_element(Coordinates(id[0], 0, id[1]))); - -#if __arm__ - asm volatile("PLD [%0, #128*4]" ::"r"(reinterpret_cast<const uint8_t *>(vec_a))); - asm volatile("PLD [%0, #128*4]" ::"r"(reinterpret_cast<const uint8_t *>(matrix_b))); - asm volatile("PLD [%0, #128*4]" ::"r"(reinterpret_cast<const uint8_t *>(matrix_b + in_b_stride))); -#endif /* __arm__ */ - - const float *vec_a_end_addr = vec_a + num_elems_vec_a; - - for(; vec_a <= (vec_a_end_addr - 4);) - { - float32x2_t a0l = vld1_f32(vec_a); - - float32x4_t b00 = vld1q_f32(matrix_b + 0 + 0 * in_b_stride); - float32x4_t b01 = vld1q_f32(matrix_b + 4 + 0 * in_b_stride); - float32x4_t b02 = vld1q_f32(matrix_b + 8 + 0 * in_b_stride); - float32x4_t b03 = vld1q_f32(matrix_b + 12 + 0 * in_b_stride); - - float32x4_t b10 = vld1q_f32(matrix_b + 0 + 1 * in_b_stride); - float32x4_t b11 = vld1q_f32(matrix_b + 4 + 1 * in_b_stride); - float32x4_t b12 = vld1q_f32(matrix_b + 8 + 1 * in_b_stride); - float32x4_t b13 = vld1q_f32(matrix_b + 12 + 1 * in_b_stride); - -#if __arm__ - asm volatile("PLD [%0, #128*4]" ::"r"(reinterpret_cast<const uint8_t *>(vec_a))); - asm volatile("PLD [%0, #128*1]" ::"r"(reinterpret_cast<const uint8_t *>(matrix_b + 1 * in_b_stride))); - asm volatile("PLD [%0, #128*1]" ::"r"(reinterpret_cast<const uint8_t *>(matrix_b + 2 * in_b_stride))); - asm volatile("PLD [%0, #128*1]" ::"r"(reinterpret_cast<const uint8_t *>(matrix_b + 3 * in_b_stride))); - asm volatile("PLD [%0, #128*1]" ::"r"(reinterpret_cast<const uint8_t *>(matrix_b + 4 * in_b_stride))); -#endif /* __arm__ */ - - acc0 = vmlaq_lane_f32(acc0, b00, a0l, 0); - acc1 = vmlaq_lane_f32(acc1, b01, a0l, 0); - acc2 = vmlaq_lane_f32(acc2, b02, a0l, 0); - acc3 = vmlaq_lane_f32(acc3, b03, a0l, 0); - - acc0 = vmlaq_lane_f32(acc0, b10, a0l, 1); - acc1 = vmlaq_lane_f32(acc1, b11, a0l, 1); - acc2 = vmlaq_lane_f32(acc2, b12, a0l, 1); - acc3 = vmlaq_lane_f32(acc3, b13, a0l, 1); - - vec_a += 2; - matrix_b += 2 * in_b_stride; - - a0l = vld1_f32(vec_a); - - b00 = vld1q_f32(matrix_b + 0 + 0 * in_b_stride); - b01 = vld1q_f32(matrix_b + 4 + 0 * in_b_stride); - b02 = vld1q_f32(matrix_b + 8 + 0 * in_b_stride); - b03 = vld1q_f32(matrix_b + 12 + 0 * in_b_stride); - - b10 = vld1q_f32(matrix_b + 0 + 1 * in_b_stride); - b11 = vld1q_f32(matrix_b + 4 + 1 * in_b_stride); - b12 = vld1q_f32(matrix_b + 8 + 1 * in_b_stride); - b13 = vld1q_f32(matrix_b + 12 + 1 * in_b_stride); - - acc0 = vmlaq_lane_f32(acc0, b00, a0l, 0); - acc1 = vmlaq_lane_f32(acc1, b01, a0l, 0); - acc2 = vmlaq_lane_f32(acc2, b02, a0l, 0); - acc3 = vmlaq_lane_f32(acc3, b03, a0l, 0); - - acc0 = vmlaq_lane_f32(acc0, b10, a0l, 1); - acc1 = vmlaq_lane_f32(acc1, b11, a0l, 1); - acc2 = vmlaq_lane_f32(acc2, b12, a0l, 1); - acc3 = vmlaq_lane_f32(acc3, b13, a0l, 1); - - vec_a += 2; - matrix_b += 2 * in_b_stride; - } - - for(; vec_a < vec_a_end_addr;) - { - const float a0 = *vec_a; - - const float32x4_t b00 = vld1q_f32(matrix_b + 0 + 0 * in_b_stride); - const float32x4_t b01 = vld1q_f32(matrix_b + 4 + 0 * in_b_stride); - const float32x4_t b02 = vld1q_f32(matrix_b + 8 + 0 * in_b_stride); - const float32x4_t b03 = vld1q_f32(matrix_b + 12 + 0 * in_b_stride); - - acc0 = vmlaq_n_f32(acc0, b00, a0); - acc1 = vmlaq_n_f32(acc1, b01, a0); - acc2 = vmlaq_n_f32(acc2, b02, a0); - acc3 = vmlaq_n_f32(acc3, b03, a0); - - vec_a += 1; - matrix_b += in_b_stride; - } - - const auto vec_out = reinterpret_cast<float *>(out.ptr()); - - vst1q_f32(vec_out + 0, acc0); - vst1q_f32(vec_out + 4, acc1); - vst1q_f32(vec_out + 8, acc2); - vst1q_f32(vec_out + 12, acc3); - }, - ina, out); -} - -Status validate_arguments(const ITensorInfo *input0, const ITensorInfo *input1, const ITensorInfo *output) -{ - ARM_COMPUTE_RETURN_ERROR_ON_CPU_F16_UNSUPPORTED(input0); - ARM_COMPUTE_RETURN_ERROR_ON_DATA_TYPE_CHANNEL_NOT_IN(input0, 1, DataType::F16, DataType::F32); - ARM_COMPUTE_RETURN_ERROR_ON_DATA_TYPE_CHANNEL_NOT_IN(input1, 1, DataType::F16, DataType::F32); - ARM_COMPUTE_RETURN_ERROR_ON_DATA_TYPE_CHANNEL_NOT_IN(output, 1, DataType::F16, DataType::F32); - ARM_COMPUTE_RETURN_ERROR_ON_MISMATCHING_DATA_TYPES(input0, input1, output); - ARM_COMPUTE_RETURN_ERROR_ON(input0->dimension(0) != input1->dimension(1)); - - return Status{}; -} - -std::tuple<Status, Window> validate_and_configure_window(ITensorInfo *input0, ITensorInfo *input1, ITensorInfo *output) -{ - constexpr unsigned int num_elems_processed_per_iteration_x = 16; - - Window win = calculate_max_window(*output, Steps(num_elems_processed_per_iteration_x)); - - AccessWindowHorizontal input0_access(input0, 0, num_elems_processed_per_iteration_x); - AccessWindowHorizontal input1_access(input1, 0, num_elems_processed_per_iteration_x); - AccessWindowHorizontal output_access(output, 0, num_elems_processed_per_iteration_x); - - bool window_changed = update_window_and_padding(win, input0_access, input1_access, output_access); - - output_access.set_valid_region(win, ValidRegion(Coordinates(), output->tensor_shape())); - - Status err = (window_changed) ? ARM_COMPUTE_CREATE_ERROR(ErrorCode::RUNTIME_ERROR, "Insufficient Padding!") : Status{}; - - return std::make_tuple(err, win); -} -} // namespace - -NELocallyConnectedMatrixMultiplyKernel::NELocallyConnectedMatrixMultiplyKernel() - : _input0(nullptr), _input1(nullptr), _output(nullptr) -{ -} - -void NELocallyConnectedMatrixMultiplyKernel::configure(const ITensor *input0, const ITensor *input1, ITensor *output) -{ - ARM_COMPUTE_ERROR_ON_NULLPTR(input0, input1, output); - ARM_COMPUTE_ERROR_THROW_ON(validate_arguments(input0->info(), input1->info(), output->info())); - - _input0 = input0; - _input1 = input1; - _output = output; - - // Configure kernel window - auto win_config = validate_and_configure_window(input0->info(), input1->info(), output->info()); - - ARM_COMPUTE_ERROR_THROW_ON(std::get<0>(win_config)); - - INEKernel::configure(std::get<1>(win_config)); -} - -Status NELocallyConnectedMatrixMultiplyKernel::validate(const ITensorInfo *input0, const ITensorInfo *input1, const ITensorInfo *output) -{ - ARM_COMPUTE_RETURN_ON_ERROR(validate_arguments(input0, input1, output)); - ARM_COMPUTE_RETURN_ON_ERROR(std::get<0>(validate_and_configure_window(input0->clone().get(), input1->clone().get(), output->clone().get()))); - - return Status{}; -} - -void NELocallyConnectedMatrixMultiplyKernel::run(const Window &window, const ThreadInfo &info) -{ - ARM_COMPUTE_ERROR_ON_UNCONFIGURED_KERNEL(this); - ARM_COMPUTE_ERROR_ON_INVALID_SUBWINDOW(INEKernel::window(), window); - - switch(_input0->info()->data_type()) - { - case DataType::F16: - { - vector_matrix_multiply_f16(_input0, _input1, _output, window, info); - break; - } - case DataType::F32: - { - vector_matrix_multiply_f32(_input0, _input1, _output, window, info); - break; - } - default: - { - ARM_COMPUTE_ERROR("Data type not supported"); - break; - } - } -} -} // namespace arm_compute diff --git a/src/core/NEON/kernels/NELocallyConnectedMatrixMultiplyKernel.h b/src/core/NEON/kernels/NELocallyConnectedMatrixMultiplyKernel.h deleted file mode 100644 index 72093b4bb7..0000000000 --- a/src/core/NEON/kernels/NELocallyConnectedMatrixMultiplyKernel.h +++ /dev/null @@ -1,79 +0,0 @@ -/* - * Copyright (c) 2017-2020 Arm Limited. - * - * SPDX-License-Identifier: MIT - * - * Permission is hereby granted, free of charge, to any person obtaining a copy - * of this software and associated documentation files (the "Software"), to - * deal in the Software without restriction, including without limitation the - * rights to use, copy, modify, merge, publish, distribute, sublicense, and/or - * sell copies of the Software, and to permit persons to whom the Software is - * furnished to do so, subject to the following conditions: - * - * The above copyright notice and this permission notice shall be included in all - * copies or substantial portions of the Software. - * - * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR - * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, - * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE - * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER - * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, - * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE - * SOFTWARE. - */ -#ifndef ARM_COMPUTE_NELOCALLYCONNECTEDMATRIXMULTIPLYKERNEL_H -#define ARM_COMPUTE_NELOCALLYCONNECTEDMATRIXMULTIPLYKERNEL_H - -#include "src/core/NEON/INEKernel.h" - -namespace arm_compute -{ -class ITensor; - -/** NEON kernel to multiply each row of first tensor with low 2 dimensions of second tensor. */ -class NELocallyConnectedMatrixMultiplyKernel : public INEKernel -{ -public: - const char *name() const override - { - return "NELocallyConnectedMatrixMultiplyKernel"; - } - /** Default constructor */ - NELocallyConnectedMatrixMultiplyKernel(); - /** Prevent instances of this class from being copied (As this class contains pointers) */ - NELocallyConnectedMatrixMultiplyKernel(const NELocallyConnectedMatrixMultiplyKernel &) = delete; - /** Prevent instances of this class from being copied (As this class contains pointers) */ - NELocallyConnectedMatrixMultiplyKernel &operator=(const NELocallyConnectedMatrixMultiplyKernel &) = delete; - /** Allow instances of this class to be moved */ - NELocallyConnectedMatrixMultiplyKernel(NELocallyConnectedMatrixMultiplyKernel &&) = default; - /** Allow instances of this class to be moved */ - NELocallyConnectedMatrixMultiplyKernel &operator=(NELocallyConnectedMatrixMultiplyKernel &&) = default; - /** Default destructor */ - ~NELocallyConnectedMatrixMultiplyKernel() = default; - /** Initialise the kernel's input and output - * - * @param[in] input0 First input tensor. Data types supported: F16, F32 - * @param[in] input1 Second input tensor containing the Matrix B. Data type supported: same as @p input0 - * @param[out] output Output tensor to store the result of matrix multiplication. Data type supported: same as @p input0 - */ - void configure(const ITensor *input0, const ITensor *input1, ITensor *output); - /** Static function to check if given info will lead to a valid configuration of @ref NELocallyConnectedMatrixMultiplyKernel - * - * @param[in] input0 First input tensor info. Data types supported: F16, F32 - * @param[in] input1 Second input tensor info. Data type supported: same as @p input0 - * @param[in] output Output tensor info. Data type supported: same as @p input0 - * - * @return a status - */ - static Status validate(const ITensorInfo *input0, const ITensorInfo *input1, const ITensorInfo *output); - - // Inherited methods overridden: - void run(const Window &window, const ThreadInfo &info) override; - -private: - const ITensor *_input0; - const ITensor *_input1; - ITensor *_output; -}; -} // namespace arm_compute -#endif /* ARM_COMPUTE_NELOCALLYCONNECTEDMATRIXMULTIPLYKERNEL_H */ diff --git a/src/runtime/CL/functions/CLLocallyConnectedLayer.cpp b/src/runtime/CL/functions/CLLocallyConnectedLayer.cpp deleted file mode 100644 index 3adae07095..0000000000 --- a/src/runtime/CL/functions/CLLocallyConnectedLayer.cpp +++ /dev/null @@ -1,226 +0,0 @@ -/* - * Copyright (c) 2017-2020 Arm Limited. - * - * SPDX-License-Identifier: MIT - * - * Permission is hereby granted, free of charge, to any person obtaining a copy - * of this software and associated documentation files (the "Software"), to - * deal in the Software without restriction, including without limitation the - * rights to use, copy, modify, merge, publish, distribute, sublicense, and/or - * sell copies of the Software, and to permit persons to whom the Software is - * furnished to do so, subject to the following conditions: - * - * The above copyright notice and this permission notice shall be included in all - * copies or substantial portions of the Software. - * - * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR - * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, - * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE - * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER - * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, - * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE - * SOFTWARE. - */ -#include "arm_compute/runtime/CL/functions/CLLocallyConnectedLayer.h" - -#include "arm_compute/core/PixelValue.h" -#include "arm_compute/core/Utils.h" -#include "arm_compute/core/Validate.h" -#include "arm_compute/runtime/CL/CLScheduler.h" -#include "src/core/CL/kernels/CLCol2ImKernel.h" -#include "src/core/CL/kernels/CLIm2ColKernel.h" -#include "src/core/CL/kernels/CLLocallyConnectedMatrixMultiplyKernel.h" -#include "src/core/CL/kernels/CLWeightsReshapeKernel.h" - -#include <cmath> -#include <tuple> - -using namespace arm_compute; - -namespace -{ -void calculate_shapes(const ITensorInfo *input, const ITensorInfo *weights, const ITensorInfo *biases, const ITensorInfo *output, const PadStrideInfo &conv_info, - TensorShape &shape_wr, TensorShape &shape_im2col, TensorShape &shape_gemm) -{ - ARM_COMPUTE_UNUSED(output); - - const unsigned int kernel_width = weights->dimension(0); - const unsigned int kernel_height = weights->dimension(1); - - bool has_bias = (biases != nullptr); - - // Get convolved dimensions - unsigned int conv_w = 0; - unsigned int conv_h = 0; - std::tie(conv_w, conv_h) = scaled_dimensions(input->dimension(0), - input->dimension(1), - kernel_width, - kernel_height, - conv_info); - - const size_t mat_weights_cols = weights->dimension(3); - const size_t mat_weights_rows = weights->dimension(0) * weights->dimension(1) * weights->dimension(2) + ((has_bias) ? 1 : 0); - const size_t mat_weights_num = weights->dimension(4); - - shape_wr = TensorShape(mat_weights_cols, mat_weights_rows, mat_weights_num); - - const size_t mat_input_cols = mat_weights_rows; - const size_t mat_input_rows = conv_w * conv_h; - - shape_im2col = input->tensor_shape(); - if(shape_im2col.num_dimensions() >= 3) - { - shape_im2col.remove_dimension(2); - } - shape_im2col.set(0, mat_input_cols); - shape_im2col.set(1, mat_input_rows); - - shape_gemm = shape_im2col; - shape_gemm.set(0, mat_weights_cols); - shape_gemm.set(1, mat_input_rows); -} -} // namespace - -CLLocallyConnectedLayer::CLLocallyConnectedLayer(std::shared_ptr<IMemoryManager> memory_manager) - : _memory_group(std::move(memory_manager)), - _input_im2col_kernel(std::make_unique<CLIm2ColKernel>()), - _weights_reshape_kernel(std::make_unique<CLWeightsReshapeKernel>()), - _mm_kernel(std::make_unique<CLLocallyConnectedMatrixMultiplyKernel>()), - _output_col2im_kernel(std::make_unique<CLCol2ImKernel>()), - _input_im2col_reshaped(), - _weights_reshaped(), - _gemm_output(), - _is_prepared(false), - _original_weights(nullptr) -{ -} - -Status CLLocallyConnectedLayer::validate(const ITensorInfo *input, const ITensorInfo *weights, const ITensorInfo *biases, const ITensorInfo *output, const PadStrideInfo &conv_info) -{ - ARM_COMPUTE_RETURN_ERROR_ON_NULLPTR(input, weights, output); - ARM_COMPUTE_RETURN_ERROR_ON(weights->dimension(2) != input->dimension(2)); - ARM_COMPUTE_RETURN_ERROR_ON(!conv_info.padding_is_symmetric()); - - bool has_bias = (biases != nullptr); - - if(has_bias) - { - ARM_COMPUTE_RETURN_ERROR_ON(biases->dimension(0) != weights->dimension(3)); - ARM_COMPUTE_RETURN_ERROR_ON(biases->num_dimensions() > 2); - } - - const unsigned int kernel_width = weights->dimension(0); - const unsigned int kernel_height = weights->dimension(1); - - // Get convolved dimensions - unsigned int conv_w = 0; - unsigned int conv_h = 0; - std::tie(conv_w, conv_h) = scaled_dimensions(input->dimension(0), input->dimension(1), kernel_width, kernel_height, - conv_info); - - ARM_COMPUTE_RETURN_ERROR_ON_MSG((output->dimension(0) != conv_w) || (output->dimension(1) != conv_h), "Output shape does not match the expected one"); - ARM_COMPUTE_RETURN_ERROR_ON_MSG(weights->dimension(4) != (conv_w * conv_h), "Weights shape does not match the expected one"); - - // Calculate intermediate buffer shapes - TensorShape shape_wr; - TensorShape shape_im2col; - TensorShape shape_gemm; - calculate_shapes(input, weights, biases, output, conv_info, shape_wr, shape_im2col, shape_gemm); - - TensorInfo weights_reshaped_info(shape_wr, 1, weights->data_type()); - TensorInfo input_im2col_reshaped_info(shape_im2col, 1, input->data_type()); - TensorInfo gemm_output_info(shape_gemm, 1, input->data_type()); - - ARM_COMPUTE_RETURN_ON_ERROR(CLIm2ColKernel::validate(input, &input_im2col_reshaped_info, Size2D(kernel_width, kernel_height), conv_info, has_bias)); - ARM_COMPUTE_RETURN_ON_ERROR(CLWeightsReshapeKernel::validate(weights, biases, &weights_reshaped_info)); - ARM_COMPUTE_RETURN_ON_ERROR(CLLocallyConnectedMatrixMultiplyKernel::validate(&input_im2col_reshaped_info, &weights_reshaped_info, &gemm_output_info)); - ARM_COMPUTE_RETURN_ON_ERROR(CLCol2ImKernel::validate(&gemm_output_info, output, Size2D(conv_w, conv_h))); - - return Status{}; -} - -#pragma GCC diagnostic push -#pragma GCC diagnostic ignored "-Wdeprecated-declarations" -void CLLocallyConnectedLayer::configure(const ICLTensor *input, const ICLTensor *weights, const ICLTensor *biases, ICLTensor *output, const PadStrideInfo &conv_info) -{ - configure(CLKernelLibrary::get().get_compile_context(), input, weights, biases, output, conv_info); -} -#pragma GCC diagnostic pop - -void CLLocallyConnectedLayer::configure(const CLCompileContext &compile_context, const ICLTensor *input, const ICLTensor *weights, const ICLTensor *biases, ICLTensor *output, - const PadStrideInfo &conv_info) -{ - ARM_COMPUTE_ERROR_ON_NULLPTR(input, weights, output); - ARM_COMPUTE_ERROR_THROW_ON(CLLocallyConnectedLayer::validate(input->info(), weights->info(), biases == nullptr ? nullptr : biases->info(), output->info(), conv_info)); - - bool _has_bias = (biases != nullptr); - _original_weights = weights; - _is_prepared = false; - - const unsigned int kernel_width = weights->info()->dimension(0); - const unsigned int kernel_height = weights->info()->dimension(1); - - // Get convolved dimensions - unsigned int conv_w = 0; - unsigned int conv_h = 0; - std::tie(conv_w, conv_h) = scaled_dimensions(input->info()->dimension(0), input->info()->dimension(1), kernel_width, kernel_height, - conv_info); - - // Calculate intermediate buffer shapes - TensorShape shape_wr; - TensorShape shape_im2col; - TensorShape shape_gemm; - calculate_shapes(input->info(), weights->info(), biases == nullptr ? nullptr : biases->info(), output->info(), conv_info, shape_wr, shape_im2col, shape_gemm); - - _weights_reshaped.allocator()->init(TensorInfo(shape_wr, 1, weights->info()->data_type())); - _input_im2col_reshaped.allocator()->init(TensorInfo(shape_im2col, 1, input->info()->data_type())); - _gemm_output.allocator()->init(TensorInfo(shape_gemm, 1, input->info()->data_type())); - - // Manage intermediate buffers - _memory_group.manage(&_input_im2col_reshaped); - _memory_group.manage(&_gemm_output); - - // Configure kernels - _input_im2col_kernel->configure(compile_context, input, &_input_im2col_reshaped, Size2D(kernel_width, kernel_height), conv_info, _has_bias); - _weights_reshape_kernel->configure(compile_context, weights, biases, &_weights_reshaped); - _mm_kernel->configure(compile_context, &_input_im2col_reshaped, &_weights_reshaped, &_gemm_output); - _output_col2im_kernel->configure(compile_context, &_gemm_output, output, Size2D(conv_w, conv_h)); - - // Allocate intermediate tensors - _input_im2col_reshaped.allocator()->allocate(); - _gemm_output.allocator()->allocate(); - - CLScheduler::get().tune_kernel_static(*_input_im2col_kernel); -} - -void CLLocallyConnectedLayer::run() -{ - prepare(); - - MemoryGroupResourceScope scope_mg(_memory_group); - - // Run input reshaping - CLScheduler::get().enqueue(*_input_im2col_kernel); - - // Runs vector matrix multiply on reshaped matrices - CLScheduler::get().enqueue(*_mm_kernel); - - // Reshape output matrix - CLScheduler::get().enqueue(*_output_col2im_kernel.get(), false); -} - -void CLLocallyConnectedLayer::prepare() -{ - if(!_is_prepared) - { - ARM_COMPUTE_ERROR_ON(!_original_weights->is_used()); - - // Run weights reshaping and mark original weights tensor as unused - _weights_reshaped.allocator()->allocate(); - CLScheduler::get().enqueue(*_weights_reshape_kernel); - _original_weights->mark_as_unused(); - - CLScheduler::get().queue().finish(); - _is_prepared = true; - } -} diff --git a/src/runtime/NEON/functions/NELocallyConnectedLayer.cpp b/src/runtime/NEON/functions/NELocallyConnectedLayer.cpp deleted file mode 100644 index c1164c3bee..0000000000 --- a/src/runtime/NEON/functions/NELocallyConnectedLayer.cpp +++ /dev/null @@ -1,203 +0,0 @@ -/* - * Copyright (c) 2017-2020 Arm Limited. - * - * SPDX-License-Identifier: MIT - * - * Permission is hereby granted, free of charge, to any person obtaining a copy - * of this software and associated documentation files (the "Software"), to - * deal in the Software without restriction, including without limitation the - * rights to use, copy, modify, merge, publish, distribute, sublicense, and/or - * sell copies of the Software, and to permit persons to whom the Software is - * furnished to do so, subject to the following conditions: - * - * The above copyright notice and this permission notice shall be included in all - * copies or substantial portions of the Software. - * - * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR - * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, - * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE - * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER - * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, - * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE - * SOFTWARE. - */ -#include "arm_compute/runtime/NEON/functions/NELocallyConnectedLayer.h" - -#include "arm_compute/core/PixelValue.h" -#include "arm_compute/core/Utils.h" -#include "arm_compute/core/Validate.h" -#include "arm_compute/runtime/NEON/NEScheduler.h" -#include "src/core/NEON/kernels/NEIm2ColKernel.h" -#include "src/core/NEON/kernels/NELocallyConnectedMatrixMultiplyKernel.h" -#include "src/core/NEON/kernels/NEWeightsReshapeKernel.h" - -#include <cmath> -#include <tuple> - -namespace arm_compute -{ -namespace -{ -void calculate_shapes(const ITensorInfo *input, const ITensorInfo *weights, const ITensorInfo *biases, const ITensorInfo *output, const PadStrideInfo &conv_info, - TensorShape &shape_wr, TensorShape &shape_im2col, TensorShape &shape_gemm) -{ - ARM_COMPUTE_UNUSED(output); - - const unsigned int kernel_width = weights->dimension(0); - const unsigned int kernel_height = weights->dimension(1); - - bool has_bias = (biases != nullptr); - - // Get convolved dimensions - unsigned int conv_w = 0; - unsigned int conv_h = 0; - std::tie(conv_w, conv_h) = scaled_dimensions(input->dimension(0), input->dimension(1), kernel_width, kernel_height, - conv_info); - - const size_t mat_weights_cols = weights->dimension(3); - const size_t mat_weights_rows = weights->dimension(0) * weights->dimension(1) * weights->dimension(2) + ((has_bias) ? 1 : 0); - const size_t mat_weights_num = weights->dimension(4); - - shape_wr = TensorShape(mat_weights_cols, mat_weights_rows, mat_weights_num); - - const size_t mat_input_cols = mat_weights_rows; - const size_t mat_input_rows = conv_w * conv_h; - - shape_im2col = input->tensor_shape(); - shape_im2col.set(0, mat_input_cols); - shape_im2col.set(1, mat_input_rows); - shape_im2col.set(2, 1); - - shape_gemm = shape_im2col; - shape_gemm.set(0, mat_weights_cols); - shape_gemm.set(1, mat_input_rows); -} -} // namespace -NELocallyConnectedLayer::~NELocallyConnectedLayer() = default; - -NELocallyConnectedLayer::NELocallyConnectedLayer(std::shared_ptr<IMemoryManager> memory_manager) - : _memory_group(std::move(memory_manager)), _input_im2col(), _weights_reshape_kernel(), _mm_kernel(), _output_col2im(), _input_im2col_reshaped(), _weights_reshaped(), _gemm_output(), - _is_prepared(false), _original_weights(nullptr) -{ -} - -Status NELocallyConnectedLayer::validate(const ITensorInfo *input, const ITensorInfo *weights, const ITensorInfo *biases, const ITensorInfo *output, const PadStrideInfo &conv_info) -{ - ARM_COMPUTE_RETURN_ERROR_ON_NULLPTR(input, weights, output); - ARM_COMPUTE_RETURN_ERROR_ON(weights->dimension(2) != input->dimension(2)); - ARM_COMPUTE_RETURN_ERROR_ON(!conv_info.padding_is_symmetric()); - - bool has_bias = (biases != nullptr); - - if(has_bias) - { - ARM_COMPUTE_RETURN_ERROR_ON(biases->dimension(0) != weights->dimension(3)); - ARM_COMPUTE_RETURN_ERROR_ON(biases->num_dimensions() > 2); - } - - const unsigned int kernel_width = weights->dimension(0); - const unsigned int kernel_height = weights->dimension(1); - - // Get convolved dimensions - unsigned int conv_w = 0; - unsigned int conv_h = 0; - std::tie(conv_w, conv_h) = scaled_dimensions(input->dimension(0), input->dimension(1), kernel_width, kernel_height, - conv_info); - - ARM_COMPUTE_RETURN_ERROR_ON_MSG((output->dimension(0) != conv_w) || (output->dimension(1) != conv_h), "Output shape does not match the expected one"); - ARM_COMPUTE_RETURN_ERROR_ON_MSG(weights->dimension(4) != (conv_w * conv_h), "Weights shape does not match the expected one"); - - // Calculate intermediate buffer shapes - TensorShape shape_wr; - TensorShape shape_im2col; - TensorShape shape_gemm; - calculate_shapes(input, weights, biases, output, conv_info, shape_wr, shape_im2col, shape_gemm); - - TensorInfo weights_reshaped_info(shape_wr, 1, weights->data_type()); - TensorInfo input_im2col_reshaped_info(shape_im2col, 1, input->data_type()); - TensorInfo gemm_output_info(shape_gemm, 1, input->data_type()); - - ARM_COMPUTE_RETURN_ON_ERROR(NEIm2Col::validate(input, &input_im2col_reshaped_info, Size2D(kernel_width, kernel_height), conv_info, has_bias)); - ARM_COMPUTE_RETURN_ON_ERROR(NEWeightsReshapeKernel::validate(weights, biases, &weights_reshaped_info)); - ARM_COMPUTE_RETURN_ON_ERROR(NELocallyConnectedMatrixMultiplyKernel::validate(&input_im2col_reshaped_info, &weights_reshaped_info, &gemm_output_info)); - ARM_COMPUTE_RETURN_ON_ERROR(NECol2Im::validate(&gemm_output_info, output, Size2D(conv_w, conv_h))); - - return Status{}; -} - -void NELocallyConnectedLayer::configure(const ITensor *input, const ITensor *weights, const ITensor *biases, ITensor *output, const PadStrideInfo &conv_info) -{ - ARM_COMPUTE_ERROR_ON_NULLPTR(input, weights, output); - ARM_COMPUTE_ERROR_THROW_ON(NELocallyConnectedLayer::validate(input->info(), weights->info(), biases == nullptr ? nullptr : biases->info(), output->info(), conv_info)); - - bool _has_bias = (biases != nullptr); - _is_prepared = false; - _original_weights = weights; - - const unsigned int kernel_width = weights->info()->dimension(0); - const unsigned int kernel_height = weights->info()->dimension(1); - - // Get convolved dimensions - unsigned int conv_w = 0; - unsigned int conv_h = 0; - std::tie(conv_w, conv_h) = scaled_dimensions(input->info()->dimension(0), input->info()->dimension(1), kernel_width, kernel_height, - conv_info); - - // Calculate intermediate buffer shapes - TensorShape shape_wr; - TensorShape shape_im2col; - TensorShape shape_gemm; - calculate_shapes(input->info(), weights->info(), biases == nullptr ? nullptr : biases->info(), output->info(), conv_info, shape_wr, shape_im2col, shape_gemm); - - _weights_reshaped.allocator()->init(TensorInfo(shape_wr, 1, weights->info()->data_type())); - _input_im2col_reshaped.allocator()->init(TensorInfo(shape_im2col, 1, input->info()->data_type())); - _gemm_output.allocator()->init(TensorInfo(shape_gemm, 1, input->info()->data_type())); - - // Manage intermediate buffers - _memory_group.manage(&_input_im2col_reshaped); - _memory_group.manage(&_gemm_output); - - // Configure kernels - _input_im2col.configure(input, &_input_im2col_reshaped, Size2D(kernel_width, kernel_height), conv_info, _has_bias); - _weights_reshape_kernel = std::make_unique<NEWeightsReshapeKernel>(); - _weights_reshape_kernel->configure(weights, biases, &_weights_reshaped); - _mm_kernel = std::make_unique<NELocallyConnectedMatrixMultiplyKernel>(); - _mm_kernel->configure(&_input_im2col_reshaped, &_weights_reshaped, &_gemm_output); - _output_col2im.configure(&_gemm_output, output, Size2D(conv_w, conv_h)); - - // Allocate intermediate tensors - _input_im2col_reshaped.allocator()->allocate(); - _gemm_output.allocator()->allocate(); -} - -void NELocallyConnectedLayer::run() -{ - prepare(); - - MemoryGroupResourceScope scope_mg(_memory_group); - - // Run input reshaping - _input_im2col.run(); - - // Runs GEMM on reshaped matrices - NEScheduler::get().schedule(_mm_kernel.get(), Window::DimX); - - // Reshape output matrix - _output_col2im.run(); -} - -void NELocallyConnectedLayer::prepare() -{ - if(!_is_prepared) - { - ARM_COMPUTE_ERROR_ON(!_original_weights->is_used()); - - // Run weights reshaping and mark original weights tensor as unused - _weights_reshaped.allocator()->allocate(); - NEScheduler::get().schedule(_weights_reshape_kernel.get(), 3); - _original_weights->mark_as_unused(); - - _is_prepared = true; - } -} -} // namespace arm_compute |