21.02
|
Interface for the kernel to run a 3x3 depthwise convolution on a tensor. More...
#include <ICLDepthwiseConvolutionLayer3x3Kernel.h>
Public Member Functions | |
ICLDepthwiseConvolutionLayer3x3Kernel () | |
Default constructor. More... | |
ICLDepthwiseConvolutionLayer3x3Kernel (const ICLDepthwiseConvolutionLayer3x3Kernel &)=delete | |
Prevent instances of this class from being copied (As this class contains pointers) More... | |
ICLDepthwiseConvolutionLayer3x3Kernel & | operator= (const ICLDepthwiseConvolutionLayer3x3Kernel &)=delete |
Prevent instances of this class from being copied (As this class contains pointers) More... | |
ICLDepthwiseConvolutionLayer3x3Kernel (ICLDepthwiseConvolutionLayer3x3Kernel &&)=default | |
Default Move Constructor. More... | |
ICLDepthwiseConvolutionLayer3x3Kernel & | operator= (ICLDepthwiseConvolutionLayer3x3Kernel &&)=default |
Default move assignment operator. More... | |
virtual void | configure (const ICLTensor *input, const ICLTensor *weights, const ICLTensor *biases, ICLTensor *output, const PadStrideInfo &conv_info, unsigned int depth_multiplier=1, ActivationLayerInfo act_info=ActivationLayerInfo(), const Size2D &dilation=Size2D(1U, 1U), const ICLTensor *output_multipliers=nullptr, const ICLTensor *output_shifts=nullptr)=0 |
Initialize the function's source, destination, conv and border_size. More... | |
virtual void | configure (const CLCompileContext &compile_context, const ICLTensor *input, const ICLTensor *weights, const ICLTensor *biases, ICLTensor *output, const PadStrideInfo &conv_info, unsigned int depth_multiplier=1, ActivationLayerInfo act_info=ActivationLayerInfo(), const Size2D &dilation=Size2D(1U, 1U), const ICLTensor *output_multipliers=nullptr, const ICLTensor *output_shifts=nullptr)=0 |
Initialize the function's source, destination, conv and border_size. More... | |
Public Member Functions inherited from ICLKernel | |
ICLKernel () | |
Constructor. More... | |
cl::Kernel & | kernel () |
Returns a reference to the OpenCL kernel of this object. More... | |
template<typename T > | |
void | add_1D_array_argument (unsigned int &idx, const ICLArray< T > *array, const Strides &strides, unsigned int num_dimensions, const Window &window) |
Add the passed 1D array's parameters to the object's kernel's arguments starting from the index idx. More... | |
void | add_1D_tensor_argument (unsigned int &idx, const ICLTensor *tensor, const Window &window) |
Add the passed 1D tensor's parameters to the object's kernel's arguments starting from the index idx. More... | |
void | add_1D_tensor_argument_if (bool cond, unsigned int &idx, const ICLTensor *tensor, const Window &window) |
Add the passed 1D tensor's parameters to the object's kernel's arguments starting from the index idx if the condition is true. More... | |
void | add_2D_tensor_argument (unsigned int &idx, const ICLTensor *tensor, const Window &window) |
Add the passed 2D tensor's parameters to the object's kernel's arguments starting from the index idx. More... | |
void | add_2D_tensor_argument_if (bool cond, unsigned int &idx, const ICLTensor *tensor, const Window &window) |
Add the passed 2D tensor's parameters to the object's kernel's arguments starting from the index idx if the condition is true. More... | |
void | add_3D_tensor_argument (unsigned int &idx, const ICLTensor *tensor, const Window &window) |
Add the passed 3D tensor's parameters to the object's kernel's arguments starting from the index idx. More... | |
void | add_4D_tensor_argument (unsigned int &idx, const ICLTensor *tensor, const Window &window) |
Add the passed 4D tensor's parameters to the object's kernel's arguments starting from the index idx. More... | |
virtual void | run (const Window &window, cl::CommandQueue &queue) |
Enqueue the OpenCL kernel to process the given window on the passed OpenCL command queue. More... | |
virtual void | run_op (ITensorPack &tensors, const Window &window, cl::CommandQueue &queue) |
Enqueue the OpenCL kernel to process the given window on the passed OpenCL command queue. More... | |
template<typename T > | |
void | add_argument (unsigned int &idx, T value) |
Add the passed parameters to the object's kernel's arguments starting from the index idx. More... | |
void | set_lws_hint (const cl::NDRange &lws_hint) |
Set the Local-Workgroup-Size hint. More... | |
cl::NDRange | lws_hint () const |
Return the Local-Workgroup-Size hint. More... | |
void | set_wbsm_hint (const cl_int &wbsm_hint) |
Set the workgroup batch size modifier hint. More... | |
cl_int | wbsm_hint () const |
Return the workgroup batch size modifier hint. More... | |
const std::string & | config_id () const |
Get the configuration ID. More... | |
void | set_target (GPUTarget target) |
Set the targeted GPU architecture. More... | |
void | set_target (cl::Device &device) |
Set the targeted GPU architecture according to the CL device. More... | |
GPUTarget | get_target () const |
Get the targeted GPU architecture. More... | |
size_t | get_max_workgroup_size () |
Get the maximum workgroup size for the device the CLKernelLibrary uses. More... | |
template<unsigned int dimension_size> | |
void | add_tensor_argument (unsigned &idx, const ICLTensor *tensor, const Window &window) |
template<typename T , unsigned int dimension_size> | |
void | add_array_argument (unsigned &idx, const ICLArray< T > *array, const Strides &strides, unsigned int num_dimensions, const Window &window) |
Add the passed array's parameters to the object's kernel's arguments starting from the index idx. More... | |
Public Member Functions inherited from IKernel | |
IKernel () | |
Constructor. More... | |
virtual | ~IKernel ()=default |
Destructor. More... | |
virtual bool | is_parallelisable () const |
Indicates whether or not the kernel is parallelisable. More... | |
virtual BorderSize | border_size () const |
The size of the border for that kernel. More... | |
const Window & | window () const |
The maximum window the kernel can be executed on. More... | |
Additional Inherited Members | |
Static Public Member Functions inherited from ICLKernel | |
static constexpr unsigned int | num_arguments_per_1D_array () |
Returns the number of arguments enqueued per 1D array object. More... | |
static constexpr unsigned int | num_arguments_per_1D_tensor () |
Returns the number of arguments enqueued per 1D tensor object. More... | |
static constexpr unsigned int | num_arguments_per_2D_tensor () |
Returns the number of arguments enqueued per 2D tensor object. More... | |
static constexpr unsigned int | num_arguments_per_3D_tensor () |
Returns the number of arguments enqueued per 3D tensor object. More... | |
static constexpr unsigned int | num_arguments_per_4D_tensor () |
Returns the number of arguments enqueued per 4D tensor object. More... | |
static cl::NDRange | gws_from_window (const Window &window) |
Get the global work size given an execution window. More... | |
Interface for the kernel to run a 3x3 depthwise convolution on a tensor.
Definition at line 35 of file ICLDepthwiseConvolutionLayer3x3Kernel.h.
|
inline |
Default constructor.
Definition at line 39 of file ICLDepthwiseConvolutionLayer3x3Kernel.h.
References ICLDepthwiseConvolutionLayer3x3Kernel::configure(), arm_compute::test::validation::conv_info, arm_compute::test::validation::input, ICLDepthwiseConvolutionLayer3x3Kernel::operator=(), and arm_compute::U.
|
delete |
Prevent instances of this class from being copied (As this class contains pointers)
Default Move Constructor.
|
pure virtual |
Initialize the function's source, destination, conv and border_size.
[in] | input | Source tensor. DataType supported: QASYMM8/F16/F32. |
[in] | weights | Weights tensor. A 3D tensor with dimensions [3, 3, IFM]. Data type supported: Same as input , QASYMM8/QSYMM8_PER_CHANNEL when input is QASYMM8. |
[in] | biases | Biases tensor. A 1D tensor with dimensions [IFM]. Must be nullptr if not needed. Data type supported: Same as input , S32 when input is QASYMM8. |
[out] | output | Destination tensor. Data type supported: Same as input . |
[in] | conv_info | Padding and stride information to use for the convolution. |
[in] | depth_multiplier | (Optional) Multiplier to apply to the input's depth in order to retrieve the output's depth. Defaults to 1. |
[in] | act_info | (Optional) Activation layer information in case of a fused activation. Only RELU, BOUNDED_RELU and LU_BOUNDED_RELU are supported for QASYMM8. |
[in] | dilation | (Optional) Dilation, in elements, across x and y. Defaults to (1, 1). |
[in] | output_multipliers | (Optional) Output multipliers tensor for quantized computations. In case of per-channel quantization, the number of multipliers must be equal to the number of filters (IFM). Supported data types: S32 |
[in] | output_shifts | (Optional) Output shifts tensor for quantized computations. In case of per-channel quantization, the number of multipliers must be equal to the number of filters (IFM). Supported data types: S32 |
Implemented in CLDepthwiseConvolutionLayer3x3NHWCKernel, and CLDepthwiseConvolutionLayer3x3NCHWKernel.
Referenced by ICLDepthwiseConvolutionLayer3x3Kernel::ICLDepthwiseConvolutionLayer3x3Kernel().
|
pure virtual |
Initialize the function's source, destination, conv and border_size.
[in] | compile_context | The compile context to be used. |
[in] | input | Source tensor. DataType supported: QASYMM8/F16/F32. |
[in] | weights | Weights tensor. A 3D tensor with dimensions [3, 3, IFM]. Data type supported: Same as input , QASYMM8/QSYMM8_PER_CHANNEL when input is QASYMM8. |
[in] | biases | Biases tensor. A 1D tensor with dimensions [IFM]. Must be nullptr if not needed. Data type supported: Same as input , S32 when input is QASYMM8. |
[out] | output | Destination tensor. Data type supported: Same as input . |
[in] | conv_info | Padding and stride information to use for the convolution. |
[in] | depth_multiplier | (Optional) Multiplier to apply to the input's depth in order to retrieve the output's depth. Defaults to 1. |
[in] | act_info | (Optional) Activation layer information in case of a fused activation. Only RELU, BOUNDED_RELU and LU_BOUNDED_RELU are supported for QASYMM8. |
[in] | dilation | (Optional) Dilation, in elements, across x and y. Defaults to (1, 1). |
[in] | output_multipliers | (Optional) Output multipliers tensor for quantized computations. In case of per-channel quantization, the number of multipliers must be equal to the number of filters (IFM). Supported data types: S32 |
[in] | output_shifts | (Optional) Output shifts tensor for quantized computations. In case of per-channel quantization, the number of multipliers must be equal to the number of filters (IFM). Supported data types: S32 |
Implemented in CLDepthwiseConvolutionLayer3x3NHWCKernel, and CLDepthwiseConvolutionLayer3x3NCHWKernel.
|
delete |
Prevent instances of this class from being copied (As this class contains pointers)
Referenced by ICLDepthwiseConvolutionLayer3x3Kernel::ICLDepthwiseConvolutionLayer3x3Kernel().
|
default |
Default move assignment operator.