ARM Compute Library
17.04
|
#include "helpers.h"
Go to the source code of this file.
Macros | |
#define | VATOMIC_INC16(histogram, win_pos) |
Functions | |
__kernel void | hist_local_kernel (__global uchar *input_ptr, uint input_stride_x, uint input_step_x, uint input_stride_y, uint input_step_y, uint input_offset_first_element_in_bytes, __local uint *histogram_local, __global uint *restrict histogram, uint num_bins, uint offset, uint range, uint offrange) |
Calculate the histogram of an 8 bit grayscale image. More... | |
__kernel void | hist_border_kernel (__global uchar *input_ptr, uint input_stride_x, uint input_step_x, uint input_stride_y, uint input_step_y, uint input_offset_first_element_in_bytes, __global uint *restrict histogram, uint num_bins, uint offset, uint range, uint offrange) |
Calculate the histogram of an 8 bit grayscale image's border. More... | |
__kernel void | hist_local_kernel_fixed (__global uchar *input_ptr, uint input_stride_x, uint input_step_x, uint input_stride_y, uint input_step_y, uint input_offset_first_element_in_bytes, __local uint *histogram_local, __global uint *restrict histogram) |
Calculate the histogram of an 8 bit grayscale image with bin size of 256 and window size of 1. More... | |
__kernel void | hist_border_kernel_fixed (__global uchar *input_ptr, uint input_stride_x, uint input_step_x, uint input_stride_y, uint input_step_y, uint input_offset_first_element_in_bytes, __global uint *restrict histogram) |
Calculate the histogram of an 8 bit grayscale image with bin size as 256 and window size as 1. More... | |
#define VATOMIC_INC16 | ( | histogram, | |
win_pos | |||
) |
Definition at line 26 of file histogram.cl.
Referenced by hist_local_kernel().
__kernel void hist_border_kernel | ( | __global uchar * | input_ptr, |
uint | input_stride_x, | ||
uint | input_step_x, | ||
uint | input_stride_y, | ||
uint | input_step_y, | ||
uint | input_offset_first_element_in_bytes, | ||
__global uint *restrict | histogram, | ||
uint | num_bins, | ||
uint | offset, | ||
uint | range, | ||
uint | offrange | ||
) |
Calculate the histogram of an 8 bit grayscale image's border.
Each thread will process one pixel using global atomic. When all work items in a work group are done the resulting local histograms are added to the global histogram using global atomics.
[in] | input_ptr | Pointer to the first source image. Supported data types: U8 |
[in] | input_stride_x | Stride of the first source image in X dimension (in bytes) |
[in] | input_step_x | input_stride_x * number of elements along X processed per workitem(in bytes) |
[in] | input_stride_y | Stride of the first source image in Y dimension (in bytes) |
[in] | input_step_y | input_stride_y * number of elements along Y processed per workitem(in bytes) |
[in] | input_offset_first_element_in_bytes | The offset of the first element in the first source image |
[out] | histogram | The output buffer to hold histogram final result. Supported data types: U32 |
[out] | num_bins | The number of bins |
[out] | offset | The start of values to use (inclusive) |
[out] | range | The range of a bin |
[out] | offrange | The maximum value (exclusive) |
Definition at line 141 of file histogram.cl.
References CONVERT_TO_IMAGE_STRUCT, offset(), and Image::ptr.
__kernel void hist_border_kernel_fixed | ( | __global uchar * | input_ptr, |
uint | input_stride_x, | ||
uint | input_step_x, | ||
uint | input_stride_y, | ||
uint | input_step_y, | ||
uint | input_offset_first_element_in_bytes, | ||
__global uint *restrict | histogram | ||
) |
Calculate the histogram of an 8 bit grayscale image with bin size as 256 and window size as 1.
Each thread will process one pixel using global atomic. When all work items in a work group are done the resulting local histograms are added to the global histogram using global atomics.
[in] | input_ptr | Pointer to the first source image. Supported data types: U8 |
[in] | input_stride_x | Stride of the first source image in X dimension (in bytes) |
[in] | input_step_x | input_stride_x * number of elements along X processed per workitem(in bytes) |
[in] | input_stride_y | Stride of the first source image in Y dimension (in bytes) |
[in] | input_step_y | input_stride_y * number of elements along Y processed per workitem(in bytes) |
[in] | input_offset_first_element_in_bytes | The offset of the first element in the first source image |
[out] | histogram | The output buffer to hold histogram final result. Supported data types: U32 |
Definition at line 238 of file histogram.cl.
References CONVERT_TO_IMAGE_STRUCT, and Image::ptr.
__kernel void hist_local_kernel | ( | __global uchar * | input_ptr, |
uint | input_stride_x, | ||
uint | input_step_x, | ||
uint | input_stride_y, | ||
uint | input_step_y, | ||
uint | input_offset_first_element_in_bytes, | ||
__local uint * | histogram_local, | ||
__global uint *restrict | histogram, | ||
uint | num_bins, | ||
uint | offset, | ||
uint | range, | ||
uint | offrange | ||
) |
Calculate the histogram of an 8 bit grayscale image.
Each thread will process 16 pixels and use one local atomic operation per pixel. When all work items in a work group are done the resulting local histograms are added to the global histogram using global atomics.
[in] | input_ptr | Pointer to the first source image. Supported data types: U8 |
[in] | input_stride_x | Stride of the first source image in X dimension (in bytes) |
[in] | input_step_x | input_stride_x * number of elements along X processed per workitem(in bytes) |
[in] | input_stride_y | Stride of the first source image in Y dimension (in bytes) |
[in] | input_step_y | input_stride_y * number of elements along Y processed per workitem(in bytes) |
[in] | input_offset_first_element_in_bytes | The offset of the first element in the first source image |
[in] | histogram_local | The local buffer to hold histogram result in per workgroup. Supported data types: U32 |
[out] | histogram | The output buffer to hold histogram final result. Supported data types: U32 |
[out] | num_bins | The number of bins |
[out] | offset | The start of values to use (inclusive) |
[out] | range | The range of a bin |
[out] | offrange | The maximum value (exclusive) |
Definition at line 68 of file histogram.cl.
References CONVERT_TO_IMAGE_STRUCT, Image::ptr, and VATOMIC_INC16.
__kernel void hist_local_kernel_fixed | ( | __global uchar * | input_ptr, |
uint | input_stride_x, | ||
uint | input_step_x, | ||
uint | input_stride_y, | ||
uint | input_step_y, | ||
uint | input_offset_first_element_in_bytes, | ||
__local uint * | histogram_local, | ||
__global uint *restrict | histogram | ||
) |
Calculate the histogram of an 8 bit grayscale image with bin size of 256 and window size of 1.
Each thread will process 16 pixels and use one local atomic operation per pixel. When all work items in a work group are done the resulting local histograms are added to the global histogram using global atomics.
[in] | input_ptr | Pointer to the first source image. Supported data types: U8 |
[in] | input_stride_x | Stride of the first source image in X dimension (in bytes) |
[in] | input_step_x | input_stride_x * number of elements along X processed per workitem(in bytes) |
[in] | input_stride_y | Stride of the first source image in Y dimension (in bytes) |
[in] | input_step_y | input_stride_y * number of elements along Y processed per workitem(in bytes) |
[in] | input_offset_first_element_in_bytes | The offset of the first element in the first source image |
[in] | histogram_local | The local buffer to hold histogram result in per workgroup. Supported data types: U32 |
[out] | histogram | The output buffer to hold histogram final result. Supported data types: U32 |
Definition at line 178 of file histogram.cl.
References CONVERT_TO_IMAGE_STRUCT, and Image::ptr.