From f47bfb97fa8bc928a7860b84b7b227f716f65e58 Mon Sep 17 00:00:00 2001 From: Sanghoon Lee Date: Tue, 23 Jan 2018 15:16:47 +0000 Subject: COMPMID-594: Implement reference and CL/NEON validation for LocallyConnected Change-Id: I01e7abcf3f1b19458128e277044af850ad9fa224 Reviewed-on: https://eu-gerrit-1.euhpc.arm.com/118610 Reviewed-by: Anthony Barbier Tested-by: Jenkins --- src/core/NEON/kernels/NELocallyConnectedMatrixMultiplyKernel.cpp | 6 +++--- src/runtime/CL/functions/CLLocallyConnectedLayer.cpp | 7 +++++-- 2 files changed, 8 insertions(+), 5 deletions(-) (limited to 'src') diff --git a/src/core/NEON/kernels/NELocallyConnectedMatrixMultiplyKernel.cpp b/src/core/NEON/kernels/NELocallyConnectedMatrixMultiplyKernel.cpp index 58da0402bc..35beb82689 100644 --- a/src/core/NEON/kernels/NELocallyConnectedMatrixMultiplyKernel.cpp +++ b/src/core/NEON/kernels/NELocallyConnectedMatrixMultiplyKernel.cpp @@ -1,5 +1,5 @@ /* - * Copyright (c) 2017, 2018 ARM Limited. + * Copyright (c) 2017-2018 ARM Limited. * * SPDX-License-Identifier: MIT * @@ -186,7 +186,7 @@ void vector_matrix_multiply_f32(const ITensor *input0, const ITensor *input1, IT 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)); + win_a.set(Window::DimX, Window::Dimension(0, 0, 0)); Iterator ina(input0, win_a); Iterator out(output, win_out); @@ -234,7 +234,7 @@ void vector_matrix_multiply_f32(const ITensor *input0, const ITensor *input1, IT asm volatile("PLD [%0, #128*1]" ::"r"(reinterpret_cast(matrix_b + 2 * in_b_stride))); asm volatile("PLD [%0, #128*1]" ::"r"(reinterpret_cast(matrix_b + 3 * in_b_stride))); asm volatile("PLD [%0, #128*1]" ::"r"(reinterpret_cast(matrix_b + 4 * in_b_stride))); -#endif /* __arm __ */ +#endif /* __arm__ */ acc0 = vmlaq_lane_f32(acc0, b00, a0l, 0); acc1 = vmlaq_lane_f32(acc1, b01, a0l, 0); diff --git a/src/runtime/CL/functions/CLLocallyConnectedLayer.cpp b/src/runtime/CL/functions/CLLocallyConnectedLayer.cpp index 9120aadf17..d284949323 100644 --- a/src/runtime/CL/functions/CLLocallyConnectedLayer.cpp +++ b/src/runtime/CL/functions/CLLocallyConnectedLayer.cpp @@ -67,10 +67,13 @@ void CLLocallyConnectedLayer::configure(const ICLTensor *input, const ICLTensor std::tie(stride_x, stride_y) = conv_info.stride(); std::tie(pad_x, pad_y) = conv_info.pad(); + 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), weights->info()->dimension(0), weights->info()->dimension(1), + std::tie(conv_w, conv_h) = scaled_dimensions(input->info()->dimension(0), input->info()->dimension(1), kernel_width, kernel_height, conv_info); ARM_COMPUTE_ERROR_ON_MSG((output->info()->dimension(0) != conv_w) || (output->info()->dimension(1) != conv_h), "Output shape does not match the expected one"); @@ -106,7 +109,7 @@ void CLLocallyConnectedLayer::configure(const ICLTensor *input, const ICLTensor _memory_group.manage(&_gemm_output); // Configure kernels - _input_im2col_kernel.configure(input, &_input_im2col_reshaped, Size2D(conv_w, conv_h), conv_info, _has_bias); + _input_im2col_kernel.configure(input, &_input_im2col_reshaped, Size2D(kernel_width, kernel_height), conv_info, _has_bias); _weights_reshape_kernel.configure(weights, biases, &_weights_reshaped); _mm_kernel.configure(&_input_im2col_reshaped, &_weights_reshaped, &_gemm_output); _output_col2im_kernel.configure(&_gemm_output, output, std::make_pair(conv_w, conv_h)); -- cgit v1.2.1