ONE - On-device Neural Engine
Loading...
Searching...
No Matches
helpers.h File Reference

Go to the source code of this file.

Data Structures

struct  Vector
 
struct  Image
 
struct  Tensor3D
 
struct  Tensor4D
 

Macros

#define GPU_ARCH_MIDGARD   0x100
 
#define GPU_ARCH_BIFROST   0x200
 
#define CONCAT(a, b)   a##b
 
#define EXPAND(x)   x
 
#define CLAMP(x, min_val, max_val)   min(max(x, min_val), max_val)
 
#define VLOAD_STR(size)   vload##size
 
#define VLOAD(size)   VLOAD_STR(size)
 
#define VSTORE_STR(size)   vstore##size
 
#define VSTORE(size)   VSTORE_STR(size)
 
#define float1   float
 
#define half1   half
 
#define char1   char
 
#define uchar1   uchar
 
#define short1   short
 
#define ushort1   ushort
 
#define int1   int
 
#define uint1   uint
 
#define long1   long
 
#define ulong1   ulong
 
#define double1   double
 
#define vload1(OFFSET, PTR)   *(OFFSET + PTR)
 
#define vstore1(DATA, OFFSET, PTR)   *(OFFSET + PTR) = DATA
 
#define convert_float_sat   convert_float
 
#define convert_float1_sat   convert_float
 
#define convert_float2_sat   convert_float2
 
#define convert_float3_sat   convert_float3
 
#define convert_float4_sat   convert_float4
 
#define convert_float8_sat   convert_float8
 
#define convert_float16_sat   convert_float16
 
#define convert_half_sat   convert_float
 
#define convert_half1_sat   convert_half
 
#define convert_half2_sat   convert_half2
 
#define convert_half3_sat   convert_half3
 
#define convert_half4_sat   convert_half4
 
#define convert_half8_sat   convert_half8
 
#define convert_half16_sat   convert_half16
 
#define convert_float1   convert_float
 
#define convert_half1   convert_half
 
#define convert_char1   convert_char
 
#define convert_uchar1   convert_uchar
 
#define convert_short1   convert_short
 
#define convert_ushort1   convert_ushort
 
#define convert_int1   convert_int
 
#define convert_uint1   convert_uint
 
#define convert_long1   convert_long
 
#define convert_ulong1   convert_ulong
 
#define convert_double1   convert_double
 
#define convert_char1_sat   convert_char_sat
 
#define convert_uchar1_sat   convert_uchar_sat
 
#define convert_short1_sat   convert_short_sat
 
#define convert_ushort1_sat   convert_ushort_sat
 
#define convert_int1_sat   convert_int_sat
 
#define convert_uint1_sat   convert_uint_sat
 
#define convert_long1_sat   convert_long_sat
 
#define convert_ulong1_sat   convert_ulong_sat
 
#define convert_double1_sat   convert_double_sat
 
#define VEC_DATA_TYPE_STR(type, size)   type##size
 
#define VEC_DATA_TYPE(type, size)   VEC_DATA_TYPE_STR(type, size)
 
#define CL_VEC_DATA_TYPE_STR(type, size)   type##size
 
#define CL_VEC_DATA_TYPE(type, size)   CL_VEC_DATA_TYPE_STR(type, size)
 
#define CONVERT_STR(x, type)   (convert_##type((x)))
 
#define CONVERT(x, type)   CONVERT_STR(x, type)
 
#define CONVERT_SAT_STR(x, type)   (convert_##type##_sat((x)))
 
#define CONVERT_SAT(x, type)   CONVERT_SAT_STR(x, type)
 
#define CONVERT_SAT_ROUND_STR(x, type, round)   (convert_##type##_sat_##round((x)))
 
#define CONVERT_SAT_ROUND(x, type, round)   CONVERT_SAT_ROUND_STR(x, type, round)
 
#define VECTOR_DECLARATION(name)
 
#define IMAGE_DECLARATION(name)
 
#define TENSOR3D_DECLARATION(name)
 
#define TENSOR4D_DECLARATION(name)
 
#define CONVERT_TO_VECTOR_STRUCT(name)
 
#define CONVERT_TO_VECTOR_STRUCT_NO_STEP(name)    update_vector_workitem_ptr(name##_ptr, name##_offset_first_element_in_bytes, name##_stride_x, 0)
 
#define CONVERT_TO_IMAGE_STRUCT(name)
 
#define CONVERT_TO_IMAGE_STRUCT_NO_STEP(name)
 
#define CONVERT_TENSOR3D_TO_IMAGE_STRUCT(name)
 
#define CONVERT_TENSOR3D_TO_IMAGE_STRUCT_NO_STEP(name)
 
#define CONVERT_TENSOR3D_TO_IMAGE_STRUCT(name)
 
#define CONVERT_TO_TENSOR3D_STRUCT(name)
 
#define CONVERT_TO_TENSOR3D_STRUCT_NO_STEP(name)
 
#define CONVERT_TO_TENSOR4D_STRUCT(name, mod_size)
 
#define CONVERT_TO_TENSOR4D_STRUCT_NO_STEP(name, mod_size)
 
REVn

REVn reverses the given vector whose size is n.

Parameters
[in]xThe vector to be reversed
Returns
The reversed vector
#define REV1(x)   ((x))
 
#define REV2(x)   ((x).s10)
 
#define REV3(x)   ((x).s210)
 
#define REV4(x)   ((x).s3210)
 
#define REV8(x)   ((x).s76543210)
 
#define REV16(x)   ((x).sFEDCBA9876543210)
 
REVERSE

Reverse the given vector.

Parameters
[in]xThe vector to be reversed
[in]sThe size of the vector
Returns
The reversed vector
#define REVERSE_STR(x, s)   REV##s((x))
 
#define REVERSE(x, s)   REVERSE_STR(x, s)
 
ROTs_n

Circular-right-shift (rotate-right) the vector of size s by the amount of n.

Parameters
[in]xThe vector to be shifted
Returns
The shifted vector
#define ROT1_0(x)   ((x))
 
#define ROT2_0(x)   ((x))
 
#define ROT2_1(x)   ((x).s10)
 
#define ROT3_0(x)   ((x))
 
#define ROT3_1(x)   ((x).s201)
 
#define ROT3_2(x)   ((x).s120)
 
#define ROT4_0(x)   ((x))
 
#define ROT4_1(x)   ((x).s3012)
 
#define ROT4_2(x)   ((x).s2301)
 
#define ROT4_3(x)   ((x).s1230)
 
#define ROT8_0(x)   ((x))
 
#define ROT8_1(x)   ((x).s70123456)
 
#define ROT8_2(x)   ((x).s67012345)
 
#define ROT8_3(x)   ((x).s56701234)
 
#define ROT8_4(x)   ((x).s45670123)
 
#define ROT8_5(x)   ((x).s34567012)
 
#define ROT8_6(x)   ((x).s23456701)
 
#define ROT8_7(x)   ((x).s12345670)
 
#define ROT16_0(x)   ((x))
 
#define ROT16_1(x)   ((x).sF0123456789ABCDE)
 
#define ROT16_2(x)   ((x).sEF0123456789ABCD)
 
#define ROT16_3(x)   ((x).sDEF0123456789ABC)
 
#define ROT16_4(x)   ((x).sCDEF0123456789AB)
 
#define ROT16_5(x)   ((x).sBCDEF0123456789A)
 
#define ROT16_6(x)   ((x).sABCDEF0123456789)
 
#define ROT16_7(x)   ((x).s9ABCDEF012345678)
 
#define ROT16_8(x)   ((x).s89ABCDEF01234567)
 
#define ROT16_9(x)   ((x).s789ABCDEF0123456)
 
#define ROT16_10(x)   ((x).s6789ABCDEF012345)
 
#define ROT16_11(x)   ((x).s56789ABCDEF01234)
 
#define ROT16_12(x)   ((x).s456789ABCDEF0123)
 
#define ROT16_13(x)   ((x).s3456789ABCDEF012)
 
#define ROT16_14(x)   ((x).s23456789ABCDEF01)
 
#define ROT16_15(x)   ((x).s123456789ABCDEF0)
 
ROTATE

Circular-right-shift (rotate-right) the given vector by the given amount.

Parameters
[in]xThe vector to be shifted
[in]sThe size of the vector
[in]nThe amount to be shifted
Returns
The shifted vector
#define ROTATE_STR(x, s, n)   ROT##s##_##n(x)
 
#define ROTATE(x, s, n)   ROTATE_STR(x, s, n)
 
V_OFFSn

Creates a vector of size n filled with offset values corresponding to the location of each element.

Parameters
[in]dtThe data type of the output vector
Returns
The vector filled with offset values
#define V_OFFS1(dt)   (dt)(0)
 
#define V_OFFS2(dt)   (dt)(0, 1)
 
#define V_OFFS3(dt)   (dt)(0, 1, 3)
 
#define V_OFFS4(dt)   (dt)(0, 1, 2, 3)
 
#define V_OFFS8(dt)   (dt)(0, 1, 2, 3, 4, 5, 6, 7)
 
#define V_OFFS16(dt)   (dt)(0, 1, 2, 3, 4, 5, 6, 7, 8, 9, 10, 11, 12, 13, 14, 15)
 
VEC_OFFS

Create a vector filled with offset values corresponding to the location of each element.

Parameters
[in]dtThe data type of the output vector
[in]sThe size of the output vector
Returns
The vector filled with offset values
#define VEC_OFFS_STR(dt, s)   V_OFFS##s(dt)
 
#define VEC_OFFS(dt, s)   VEC_OFFS_STR(dt, s)
 

Typedefs

typedef struct Vector Vector
 
typedef struct Image Image
 
typedef struct Tensor3D Tensor3D
 
typedef struct Tensor4D Tensor4D
 

Functions

Vector update_vector_workitem_ptr (__global uchar *ptr, uint offset_first_element_in_bytes, uint stride_x, uint step_x)
 
Image update_image_workitem_ptr (__global uchar *ptr, uint offset_first_element_in_bytes, uint stride_x, uint step_x, uint stride_y, uint step_y)
 
Image update_image_from_tensor3D_workitem_ptr (__global uchar *ptr, uint offset_first_element_in_bytes, uint stride_x, uint step_x, uint stride_y, uint step_y, uint stride_z, uint step_z)
 
Tensor3D update_tensor3D_workitem_ptr (__global uchar *ptr, uint offset_first_element_in_bytes, uint stride_x, uint step_x, uint stride_y, uint step_y, uint stride_z, uint step_z)
 
Tensor4D update_tensor4D_workitem_ptr (__global uchar *ptr, uint offset_first_element_in_bytes, uint stride_x, uint step_x, uint stride_y, uint step_y, uint stride_z, uint step_z, uint stride_w, uint step_w, uint mod_size)
 
__global const uchar * vector_offset (const Vector *vec, int x)
 
__global uchar * offset (const Image *img, int x, int y)
 
__global const uchar * tensor3D_offset (const Tensor3D *tensor, int x, int y, int z)
 
__global const uchar * tensor4D_offset (const Tensor4D *tensor, int x, int y, int z, int w)
 

Macro Definition Documentation

◆ char1

#define char1   char

Definition at line 221 of file helpers.h.

◆ CL_VEC_DATA_TYPE

#define CL_VEC_DATA_TYPE (   type,
  size 
)    CL_VEC_DATA_TYPE_STR(type, size)

Definition at line 278 of file helpers.h.

◆ CL_VEC_DATA_TYPE_STR

#define CL_VEC_DATA_TYPE_STR (   type,
  size 
)    type##size

Definition at line 277 of file helpers.h.

◆ CLAMP

#define CLAMP (   x,
  min_val,
  max_val 
)    min(max(x, min_val), max_val)

Clamp the given value between an upper and lower bound.

Parameters
[in]xThe value to be clamped
[in]min_valThe lower bound
[in]max_valThe upper bound
Returns
The clamped value.

Definition at line 89 of file helpers.h.

◆ CONCAT

#define CONCAT (   a,
 
)    a##b

Concatenate two inputs.

Parameters
[in]aThe first input to be concatenated
[in]bThe second input to be concatenated
Returns
The concatenated output

Definition at line 71 of file helpers.h.

◆ CONVERT

#define CONVERT (   x,
  type 
)    CONVERT_STR(x, type)

Definition at line 281 of file helpers.h.

◆ convert_char1

#define convert_char1   convert_char

Definition at line 254 of file helpers.h.

◆ convert_char1_sat

#define convert_char1_sat   convert_char_sat

Definition at line 264 of file helpers.h.

◆ convert_double1

#define convert_double1   convert_double

Definition at line 262 of file helpers.h.

◆ convert_double1_sat

#define convert_double1_sat   convert_double_sat

Definition at line 272 of file helpers.h.

◆ convert_float1

#define convert_float1   convert_float

Definition at line 252 of file helpers.h.

◆ convert_float16_sat

#define convert_float16_sat   convert_float16

Definition at line 243 of file helpers.h.

◆ convert_float1_sat

#define convert_float1_sat   convert_float

Definition at line 238 of file helpers.h.

◆ convert_float2_sat

#define convert_float2_sat   convert_float2

Definition at line 239 of file helpers.h.

◆ convert_float3_sat

#define convert_float3_sat   convert_float3

Definition at line 240 of file helpers.h.

◆ convert_float4_sat

#define convert_float4_sat   convert_float4

Definition at line 241 of file helpers.h.

◆ convert_float8_sat

#define convert_float8_sat   convert_float8

Definition at line 242 of file helpers.h.

◆ convert_float_sat

#define convert_float_sat   convert_float

Definition at line 237 of file helpers.h.

◆ convert_half1

#define convert_half1   convert_half

Definition at line 253 of file helpers.h.

◆ convert_half16_sat

#define convert_half16_sat   convert_half16

Definition at line 250 of file helpers.h.

◆ convert_half1_sat

#define convert_half1_sat   convert_half

Definition at line 245 of file helpers.h.

◆ convert_half2_sat

#define convert_half2_sat   convert_half2

Definition at line 246 of file helpers.h.

◆ convert_half3_sat

#define convert_half3_sat   convert_half3

Definition at line 247 of file helpers.h.

◆ convert_half4_sat

#define convert_half4_sat   convert_half4

Definition at line 248 of file helpers.h.

◆ convert_half8_sat

#define convert_half8_sat   convert_half8

Definition at line 249 of file helpers.h.

◆ convert_half_sat

#define convert_half_sat   convert_float

Definition at line 244 of file helpers.h.

◆ convert_int1

#define convert_int1   convert_int

Definition at line 258 of file helpers.h.

◆ convert_int1_sat

#define convert_int1_sat   convert_int_sat

Definition at line 268 of file helpers.h.

◆ convert_long1

#define convert_long1   convert_long

Definition at line 260 of file helpers.h.

◆ convert_long1_sat

#define convert_long1_sat   convert_long_sat

Definition at line 270 of file helpers.h.

◆ CONVERT_SAT

#define CONVERT_SAT (   x,
  type 
)    CONVERT_SAT_STR(x, type)

Definition at line 284 of file helpers.h.

◆ CONVERT_SAT_ROUND

#define CONVERT_SAT_ROUND (   x,
  type,
  round 
)    CONVERT_SAT_ROUND_STR(x, type, round)

Definition at line 287 of file helpers.h.

◆ CONVERT_SAT_ROUND_STR

#define CONVERT_SAT_ROUND_STR (   x,
  type,
  round 
)    (convert_##type##_sat_##round((x)))

Definition at line 286 of file helpers.h.

◆ CONVERT_SAT_STR

#define CONVERT_SAT_STR (   x,
  type 
)    (convert_##type##_sat((x)))

Definition at line 283 of file helpers.h.

◆ convert_short1

#define convert_short1   convert_short

Definition at line 256 of file helpers.h.

◆ convert_short1_sat

#define convert_short1_sat   convert_short_sat

Definition at line 266 of file helpers.h.

◆ CONVERT_STR

#define CONVERT_STR (   x,
  type 
)    (convert_##type((x)))

Definition at line 280 of file helpers.h.

◆ CONVERT_TENSOR3D_TO_IMAGE_STRUCT [1/2]

#define CONVERT_TENSOR3D_TO_IMAGE_STRUCT (   name)
Value:
update_image_from_tensor3D_workitem_ptr(name##_ptr, name##_offset_first_element_in_bytes, \
name##_stride_x, name##_step_x, name##_stride_y, \
name##_step_y, name##_stride_z, name##_step_z)
Image update_image_from_tensor3D_workitem_ptr(__global uchar *ptr, uint offset_first_element_in_bytes, uint stride_x, uint step_x, uint stride_y, uint step_y, uint stride_z, uint step_z)
Definition helpers.h:460

Definition at line 322 of file helpers.h.

357{
358 __global uchar *ptr;
359 int offset_first_element_in_bytes;
360 int stride_x;
361} Vector;
362
364typedef struct Image
365{
366 __global uchar *ptr;
368 int stride_x;
369 int stride_y;
370} Image;
371
373typedef struct Tensor3D
374{
375 __global uchar *ptr;
377 int stride_x;
378 int stride_y;
379 int stride_z;
380} Tensor3D;
381
383typedef struct Tensor4D
384{
385 __global uchar *ptr;
387 int stride_x;
388 int stride_y;
389 int stride_z;
390 int stride_w;
391} Tensor4D;
392
404inline Vector update_vector_workitem_ptr(__global uchar *ptr, uint offset_first_element_in_bytes,
405 uint stride_x, uint step_x)
406{
407 Vector vector = {
408 .ptr = ptr,
409 .offset_first_element_in_bytes = offset_first_element_in_bytes,
410 .stride_x = stride_x,
411 };
412 vector.ptr += vector.offset_first_element_in_bytes + get_global_id(0) * step_x;
413 return vector;
414}
415
430inline Image update_image_workitem_ptr(__global uchar *ptr, uint offset_first_element_in_bytes,
431 uint stride_x, uint step_x, uint stride_y, uint step_y)
432{
433 Image img = {.ptr = ptr,
434 .offset_first_element_in_bytes = offset_first_element_in_bytes,
435 .stride_x = stride_x,
436 .stride_y = stride_y};
437 img.ptr +=
438 img.offset_first_element_in_bytes + get_global_id(0) * step_x + get_global_id(1) * step_y;
439 return img;
440}
441
459inline Image update_image_from_tensor3D_workitem_ptr(__global uchar *ptr,
460 uint offset_first_element_in_bytes,
461 uint stride_x, uint step_x, uint stride_y,
462 uint step_y, uint stride_z, uint step_z)
463{
464 Image img = {.ptr = ptr,
465 .offset_first_element_in_bytes = offset_first_element_in_bytes,
466 .stride_x = stride_x,
467 .stride_y = stride_y};
468 img.ptr += img.offset_first_element_in_bytes + get_global_id(0) * step_x +
469 get_global_id(1) * step_y + get_global_id(2) * step_z;
470 return img;
471}
472
490inline Tensor3D update_tensor3D_workitem_ptr(__global uchar *ptr,
491 uint offset_first_element_in_bytes, uint stride_x,
492 uint step_x, uint stride_y, uint step_y, uint stride_z,
493 uint step_z)
494{
495 Tensor3D tensor = {.ptr = ptr,
496 .offset_first_element_in_bytes = offset_first_element_in_bytes,
497 .stride_x = stride_x,
498 .stride_y = stride_y,
499 .stride_z = stride_z};
500 tensor.ptr += tensor.offset_first_element_in_bytes + get_global_id(0) * step_x +
501 get_global_id(1) * step_y + get_global_id(2) * step_z;
502 return tensor;
503}
504
505inline Tensor4D update_tensor4D_workitem_ptr(__global uchar *ptr,
506 uint offset_first_element_in_bytes, uint stride_x,
507 uint step_x, uint stride_y, uint step_y, uint stride_z,
508 uint step_z, uint stride_w, uint step_w, uint mod_size)
509{
510 Tensor4D tensor = {.ptr = ptr,
511 .offset_first_element_in_bytes = offset_first_element_in_bytes,
512 .stride_x = stride_x,
513 .stride_y = stride_y,
514 .stride_z = stride_z,
515 .stride_w = stride_w};
516
517 tensor.ptr += tensor.offset_first_element_in_bytes + get_global_id(0) * step_x +
518 get_global_id(1) * step_y + (get_global_id(2) % mod_size) * step_z +
519 (get_global_id(2) / mod_size) * step_w;
520 return tensor;
521}
522
528inline __global const uchar *vector_offset(const Vector *vec, int x)
529{
530 return vec->ptr + x * vec->stride_x;
531}
532
539inline __global uchar *offset(const Image *img, int x, int y)
540{
541 return img->ptr + x * img->stride_x + y * img->stride_y;
542}
543
551inline __global const uchar *tensor3D_offset(const Tensor3D *tensor, int x, int y, int z)
552{
553 return tensor->ptr + x * tensor->stride_x + y * tensor->stride_y + z * tensor->stride_z;
554}
555
564inline __global const uchar *tensor4D_offset(const Tensor4D *tensor, int x, int y, int z, int w)
565{
566 return tensor->ptr + x * tensor->stride_x + y * tensor->stride_y + z * tensor->stride_z +
567 w * tensor->stride_w;
568}
569
570#endif // _HELPER_H
Image update_image_workitem_ptr(__global uchar *ptr, uint offset_first_element_in_bytes, uint stride_x, uint step_x, uint stride_y, uint step_y)
Definition helpers.h:431
__global const uchar * tensor4D_offset(const Tensor4D *tensor, int x, int y, int z, int w)
Definition helpers.h:565
__global uchar * offset(const Image *img, int x, int y)
Definition helpers.h:540
Tensor4D update_tensor4D_workitem_ptr(__global uchar *ptr, uint offset_first_element_in_bytes, uint stride_x, uint step_x, uint stride_y, uint step_y, uint stride_z, uint step_z, uint stride_w, uint step_w, uint mod_size)
Definition helpers.h:506
Vector update_vector_workitem_ptr(__global uchar *ptr, uint offset_first_element_in_bytes, uint stride_x, uint step_x)
Definition helpers.h:405
Tensor3D update_tensor3D_workitem_ptr(__global uchar *ptr, uint offset_first_element_in_bytes, uint stride_x, uint step_x, uint stride_y, uint step_y, uint stride_z, uint step_z)
Definition helpers.h:491
__global const uchar * tensor3D_offset(const Tensor3D *tensor, int x, int y, int z)
Definition helpers.h:552
__global const uchar * vector_offset(const Vector *vec, int x)
Definition helpers.h:529
int stride_x
Definition helpers.h:369
int offset_first_element_in_bytes
Definition helpers.h:368
__global uchar * ptr
Definition helpers.h:367
int stride_y
Definition helpers.h:370
int stride_y
Definition helpers.h:379
__global uchar * ptr
Definition helpers.h:376
int offset_first_element_in_bytes
Definition helpers.h:377
int stride_z
Definition helpers.h:380
int stride_x
Definition helpers.h:378
int offset_first_element_in_bytes
Definition helpers.h:387
int stride_z
Definition helpers.h:390
int stride_y
Definition helpers.h:389
int stride_w
Definition helpers.h:391
__global uchar * ptr
Definition helpers.h:386
int stride_x
Definition helpers.h:388
int offset_first_element_in_bytes
Definition helpers.h:360
__global uchar * ptr
Definition helpers.h:359
int stride_x
Definition helpers.h:361

◆ CONVERT_TENSOR3D_TO_IMAGE_STRUCT [2/2]

#define CONVERT_TENSOR3D_TO_IMAGE_STRUCT (   name)
Value:
update_image_from_tensor3D_workitem_ptr(name##_ptr, name##_offset_first_element_in_bytes, \
name##_stride_x, name##_step_x, name##_stride_y, \
name##_step_y, name##_stride_z, name##_step_z)

Definition at line 322 of file helpers.h.

◆ CONVERT_TENSOR3D_TO_IMAGE_STRUCT_NO_STEP

#define CONVERT_TENSOR3D_TO_IMAGE_STRUCT_NO_STEP (   name)
Value:
update_image_from_tensor3D_workitem_ptr(name##_ptr, name##_offset_first_element_in_bytes, \
name##_stride_x, 0, name##_stride_y, 0, name##_stride_z, \
name##_step_z)

Definition at line 327 of file helpers.h.

◆ CONVERT_TO_IMAGE_STRUCT

#define CONVERT_TO_IMAGE_STRUCT (   name)
Value:
update_image_workitem_ptr(name##_ptr, name##_offset_first_element_in_bytes, name##_stride_x, \
name##_step_x, name##_stride_y, name##_step_y)

Definition at line 314 of file helpers.h.

◆ CONVERT_TO_IMAGE_STRUCT_NO_STEP

#define CONVERT_TO_IMAGE_STRUCT_NO_STEP (   name)
Value:
update_image_workitem_ptr(name##_ptr, name##_offset_first_element_in_bytes, name##_stride_x, 0, \
name##_stride_y, 0)

Definition at line 318 of file helpers.h.

◆ CONVERT_TO_TENSOR3D_STRUCT

#define CONVERT_TO_TENSOR3D_STRUCT (   name)
Value:
update_tensor3D_workitem_ptr(name##_ptr, name##_offset_first_element_in_bytes, name##_stride_x, \
name##_step_x, name##_stride_y, name##_step_y, name##_stride_z, \
name##_step_z)

Definition at line 337 of file helpers.h.

◆ CONVERT_TO_TENSOR3D_STRUCT_NO_STEP

#define CONVERT_TO_TENSOR3D_STRUCT_NO_STEP (   name)
Value:
update_tensor3D_workitem_ptr(name##_ptr, name##_offset_first_element_in_bytes, name##_stride_x, \
0, name##_stride_y, 0, name##_stride_z, 0)

Definition at line 342 of file helpers.h.

◆ CONVERT_TO_TENSOR4D_STRUCT

#define CONVERT_TO_TENSOR4D_STRUCT (   name,
  mod_size 
)
Value:
update_tensor4D_workitem_ptr(name##_ptr, name##_offset_first_element_in_bytes, name##_stride_x, \
name##_step_x, name##_stride_y, name##_step_y, name##_stride_z, \
name##_step_z, name##_stride_w, name##_step_w, mod_size)

Definition at line 346 of file helpers.h.

◆ CONVERT_TO_TENSOR4D_STRUCT_NO_STEP

#define CONVERT_TO_TENSOR4D_STRUCT_NO_STEP (   name,
  mod_size 
)
Value:
update_tensor4D_workitem_ptr(name##_ptr, name##_offset_first_element_in_bytes, name##_stride_x, \
0, name##_stride_y, 0, name##_stride_z, 0, name##_stride_w, 0, \
mod_size)

Definition at line 351 of file helpers.h.

◆ CONVERT_TO_VECTOR_STRUCT

#define CONVERT_TO_VECTOR_STRUCT (   name)
Value:
update_vector_workitem_ptr(name##_ptr, name##_offset_first_element_in_bytes, name##_stride_x, \
name##_step_x)

Definition at line 307 of file helpers.h.

◆ CONVERT_TO_VECTOR_STRUCT_NO_STEP

#define CONVERT_TO_VECTOR_STRUCT_NO_STEP (   name)     update_vector_workitem_ptr(name##_ptr, name##_offset_first_element_in_bytes, name##_stride_x, 0)

Definition at line 311 of file helpers.h.

◆ convert_uchar1

#define convert_uchar1   convert_uchar

Definition at line 255 of file helpers.h.

◆ convert_uchar1_sat

#define convert_uchar1_sat   convert_uchar_sat

Definition at line 265 of file helpers.h.

◆ convert_uint1

#define convert_uint1   convert_uint

Definition at line 259 of file helpers.h.

◆ convert_uint1_sat

#define convert_uint1_sat   convert_uint_sat

Definition at line 269 of file helpers.h.

◆ convert_ulong1

#define convert_ulong1   convert_ulong

Definition at line 261 of file helpers.h.

◆ convert_ulong1_sat

#define convert_ulong1_sat   convert_ulong_sat

Definition at line 271 of file helpers.h.

◆ convert_ushort1

#define convert_ushort1   convert_ushort

Definition at line 257 of file helpers.h.

◆ convert_ushort1_sat

#define convert_ushort1_sat   convert_ushort_sat

Definition at line 267 of file helpers.h.

◆ double1

#define double1   double

Definition at line 229 of file helpers.h.

◆ EXPAND

#define EXPAND (   x)    x

Expand the given vector

Parameters
[in]xThe vector to be expanded
Returns
The expanded output

Definition at line 79 of file helpers.h.

◆ float1

#define float1   float

Definition at line 219 of file helpers.h.

◆ GPU_ARCH_BIFROST

#define GPU_ARCH_BIFROST   0x200

Definition at line 62 of file helpers.h.

◆ GPU_ARCH_MIDGARD

#define GPU_ARCH_MIDGARD   0x100

Definition at line 61 of file helpers.h.

◆ half1

#define half1   half

Definition at line 220 of file helpers.h.

◆ IMAGE_DECLARATION

#define IMAGE_DECLARATION (   name)
Value:
__global uchar *name##_ptr, uint name##_stride_x, uint name##_step_x, uint name##_stride_y, \
uint name##_step_y, uint name##_offset_first_element_in_bytes

Definition at line 293 of file helpers.h.

◆ int1

#define int1   int

Definition at line 225 of file helpers.h.

◆ long1

#define long1   long

Definition at line 227 of file helpers.h.

◆ REV1

#define REV1 (   x)    ((x))

Definition at line 99 of file helpers.h.

◆ REV16

#define REV16 (   x)    ((x).sFEDCBA9876543210)

Definition at line 104 of file helpers.h.

◆ REV2

#define REV2 (   x)    ((x).s10)

Definition at line 100 of file helpers.h.

◆ REV3

#define REV3 (   x)    ((x).s210)

Definition at line 101 of file helpers.h.

◆ REV4

#define REV4 (   x)    ((x).s3210)

Definition at line 102 of file helpers.h.

◆ REV8

#define REV8 (   x)    ((x).s76543210)

Definition at line 103 of file helpers.h.

◆ REVERSE

#define REVERSE (   x,
 
)    REVERSE_STR(x, s)

Definition at line 117 of file helpers.h.

◆ REVERSE_STR

#define REVERSE_STR (   x,
 
)    REV##s((x))

Definition at line 116 of file helpers.h.

◆ ROT16_0

#define ROT16_0 (   x)    ((x))

Definition at line 151 of file helpers.h.

◆ ROT16_1

#define ROT16_1 (   x)    ((x).sF0123456789ABCDE)

Definition at line 152 of file helpers.h.

◆ ROT16_10

#define ROT16_10 (   x)    ((x).s6789ABCDEF012345)

Definition at line 161 of file helpers.h.

◆ ROT16_11

#define ROT16_11 (   x)    ((x).s56789ABCDEF01234)

Definition at line 162 of file helpers.h.

◆ ROT16_12

#define ROT16_12 (   x)    ((x).s456789ABCDEF0123)

Definition at line 163 of file helpers.h.

◆ ROT16_13

#define ROT16_13 (   x)    ((x).s3456789ABCDEF012)

Definition at line 164 of file helpers.h.

◆ ROT16_14

#define ROT16_14 (   x)    ((x).s23456789ABCDEF01)

Definition at line 165 of file helpers.h.

◆ ROT16_15

#define ROT16_15 (   x)    ((x).s123456789ABCDEF0)

Definition at line 166 of file helpers.h.

◆ ROT16_2

#define ROT16_2 (   x)    ((x).sEF0123456789ABCD)

Definition at line 153 of file helpers.h.

◆ ROT16_3

#define ROT16_3 (   x)    ((x).sDEF0123456789ABC)

Definition at line 154 of file helpers.h.

◆ ROT16_4

#define ROT16_4 (   x)    ((x).sCDEF0123456789AB)

Definition at line 155 of file helpers.h.

◆ ROT16_5

#define ROT16_5 (   x)    ((x).sBCDEF0123456789A)

Definition at line 156 of file helpers.h.

◆ ROT16_6

#define ROT16_6 (   x)    ((x).sABCDEF0123456789)

Definition at line 157 of file helpers.h.

◆ ROT16_7

#define ROT16_7 (   x)    ((x).s9ABCDEF012345678)

Definition at line 158 of file helpers.h.

◆ ROT16_8

#define ROT16_8 (   x)    ((x).s89ABCDEF01234567)

Definition at line 159 of file helpers.h.

◆ ROT16_9

#define ROT16_9 (   x)    ((x).s789ABCDEF0123456)

Definition at line 160 of file helpers.h.

◆ ROT1_0

#define ROT1_0 (   x)    ((x))

Definition at line 128 of file helpers.h.

◆ ROT2_0

#define ROT2_0 (   x)    ((x))

Definition at line 130 of file helpers.h.

◆ ROT2_1

#define ROT2_1 (   x)    ((x).s10)

Definition at line 131 of file helpers.h.

◆ ROT3_0

#define ROT3_0 (   x)    ((x))

Definition at line 133 of file helpers.h.

◆ ROT3_1

#define ROT3_1 (   x)    ((x).s201)

Definition at line 134 of file helpers.h.

◆ ROT3_2

#define ROT3_2 (   x)    ((x).s120)

Definition at line 135 of file helpers.h.

◆ ROT4_0

#define ROT4_0 (   x)    ((x))

Definition at line 137 of file helpers.h.

◆ ROT4_1

#define ROT4_1 (   x)    ((x).s3012)

Definition at line 138 of file helpers.h.

◆ ROT4_2

#define ROT4_2 (   x)    ((x).s2301)

Definition at line 139 of file helpers.h.

◆ ROT4_3

#define ROT4_3 (   x)    ((x).s1230)

Definition at line 140 of file helpers.h.

◆ ROT8_0

#define ROT8_0 (   x)    ((x))

Definition at line 142 of file helpers.h.

◆ ROT8_1

#define ROT8_1 (   x)    ((x).s70123456)

Definition at line 143 of file helpers.h.

◆ ROT8_2

#define ROT8_2 (   x)    ((x).s67012345)

Definition at line 144 of file helpers.h.

◆ ROT8_3

#define ROT8_3 (   x)    ((x).s56701234)

Definition at line 145 of file helpers.h.

◆ ROT8_4

#define ROT8_4 (   x)    ((x).s45670123)

Definition at line 146 of file helpers.h.

◆ ROT8_5

#define ROT8_5 (   x)    ((x).s34567012)

Definition at line 147 of file helpers.h.

◆ ROT8_6

#define ROT8_6 (   x)    ((x).s23456701)

Definition at line 148 of file helpers.h.

◆ ROT8_7

#define ROT8_7 (   x)    ((x).s12345670)

Definition at line 149 of file helpers.h.

◆ ROTATE

#define ROTATE (   x,
  s,
 
)    ROTATE_STR(x, s, n)

Definition at line 180 of file helpers.h.

◆ ROTATE_STR

#define ROTATE_STR (   x,
  s,
 
)    ROT##s##_##n(x)

Definition at line 179 of file helpers.h.

◆ short1

#define short1   short

Definition at line 223 of file helpers.h.

◆ TENSOR3D_DECLARATION

#define TENSOR3D_DECLARATION (   name)
Value:
__global uchar *name##_ptr, uint name##_stride_x, uint name##_step_x, uint name##_stride_y, \
uint name##_step_y, uint name##_stride_z, uint name##_step_z, \
uint name##_offset_first_element_in_bytes

Definition at line 297 of file helpers.h.

◆ TENSOR4D_DECLARATION

#define TENSOR4D_DECLARATION (   name)
Value:
__global uchar *name##_ptr, uint name##_stride_x, uint name##_step_x, uint name##_stride_y, \
uint name##_step_y, uint name##_stride_z, uint name##_step_z, uint name##_stride_w, \
uint name##_step_w, uint name##_offset_first_element_in_bytes

Definition at line 302 of file helpers.h.

◆ uchar1

#define uchar1   uchar

Definition at line 222 of file helpers.h.

◆ uint1

#define uint1   uint

Definition at line 226 of file helpers.h.

◆ ulong1

#define ulong1   ulong

Definition at line 228 of file helpers.h.

◆ ushort1

#define ushort1   ushort

Definition at line 224 of file helpers.h.

◆ V_OFFS1

#define V_OFFS1 (   dt)    (dt)(0)

Definition at line 192 of file helpers.h.

◆ V_OFFS16

#define V_OFFS16 (   dt)    (dt)(0, 1, 2, 3, 4, 5, 6, 7, 8, 9, 10, 11, 12, 13, 14, 15)

Definition at line 197 of file helpers.h.

◆ V_OFFS2

#define V_OFFS2 (   dt)    (dt)(0, 1)

Definition at line 193 of file helpers.h.

◆ V_OFFS3

#define V_OFFS3 (   dt)    (dt)(0, 1, 3)

Definition at line 194 of file helpers.h.

◆ V_OFFS4

#define V_OFFS4 (   dt)    (dt)(0, 1, 2, 3)

Definition at line 195 of file helpers.h.

◆ V_OFFS8

#define V_OFFS8 (   dt)    (dt)(0, 1, 2, 3, 4, 5, 6, 7)

Definition at line 196 of file helpers.h.

◆ VEC_DATA_TYPE

#define VEC_DATA_TYPE (   type,
  size 
)    VEC_DATA_TYPE_STR(type, size)

Definition at line 275 of file helpers.h.

◆ VEC_DATA_TYPE_STR

#define VEC_DATA_TYPE_STR (   type,
  size 
)    type##size

Definition at line 274 of file helpers.h.

◆ VEC_OFFS

#define VEC_OFFS (   dt,
 
)    VEC_OFFS_STR(dt, s)

Definition at line 210 of file helpers.h.

◆ VEC_OFFS_STR

#define VEC_OFFS_STR (   dt,
 
)    V_OFFS##s(dt)

Definition at line 209 of file helpers.h.

◆ VECTOR_DECLARATION

#define VECTOR_DECLARATION (   name)
Value:
__global uchar *name##_ptr, uint name##_stride_x, uint name##_step_x, \
uint name##_offset_first_element_in_bytes

Definition at line 289 of file helpers.h.

◆ VLOAD

#define VLOAD (   size)    VLOAD_STR(size)

Definition at line 214 of file helpers.h.

◆ vload1

#define vload1 (   OFFSET,
  PTR 
)    *(OFFSET + PTR)

Definition at line 231 of file helpers.h.

◆ VLOAD_STR

#define VLOAD_STR (   size)    vload##size

Definition at line 213 of file helpers.h.

◆ VSTORE

#define VSTORE (   size)    VSTORE_STR(size)

Definition at line 217 of file helpers.h.

◆ vstore1

#define vstore1 (   DATA,
  OFFSET,
  PTR 
)    *(OFFSET + PTR) = DATA

Definition at line 232 of file helpers.h.

◆ VSTORE_STR

#define VSTORE_STR (   size)    vstore##size

Definition at line 216 of file helpers.h.

Typedef Documentation

◆ Image

typedef struct Image Image

Structure to hold Image information

◆ Tensor3D

typedef struct Tensor3D Tensor3D

Structure to hold 3D tensor information

◆ Tensor4D

typedef struct Tensor4D Tensor4D

Structure to hold 4D tensor information

◆ Vector

typedef struct Vector Vector

Structure to hold Vector information

Function Documentation

◆ offset()

__global uchar * offset ( const Image img,
int  x,
int  y 
)
inline

Get the pointer position of a Image

Parameters
[in]imgPointer to the starting position of the buffer
[in]xRelative X position
[in]yRelative Y position

Definition at line 540 of file helpers.h.

541{
542 return img->ptr + x * img->stride_x + y * img->stride_y;
543}

References Image::ptr, Image::stride_x, and Image::stride_y.

Referenced by luci_interpreter::StaticMemoryManager::allocate_memory(), ANeuralNetworksExecution_setInputFromMemory(), ANeuralNetworksExecution_setOutputFromMemory(), ANeuralNetworksMemory_createFromFd(), ANeuralNetworksModel_setOperandValueFromMemory(), onert::backend::acl_common::asQuantizationInfo(), Tensor::at(), coco::FeatureLayouts::BCHW::at(), coco::FeatureLayouts::BHWC::at(), coco::FeatureLayouts::BC::at(), mir::Tensor< T >::atOffset(), mir::Tensor< T >::atOffset(), mir::TensorVariant::atOffset(), onert::backend::IPortableTensor::calcOffset(), luci_interpreter::kernels::computePaddingWithOffset(), concatenationPrepare(), onert_micro::train::pal::Conv2DBiasGrad(), onert::backend::cpu::ops::DepthwiseConvolutionLayer::convQ8iHybridPerChannel(), flatbuffers::vector_downward::data_at(), dequantize_qasymm8(), dequantize_qasymm8_signed(), luci_interpreter::kernels::testing::dequantizeTensorData(), enco::CppCode::dump(), flexbuffers::FLATBUFFERS_FINAL_CLASS::Value::ElemWidth(), onert_micro::core::OMTrainingRuntimeModule::evaluateMetric(), onert_micro::core::train::OMTrainingHandler::evaluateMetric(), luci_interpreter::kernels::MirrorPad::execute(), luci_interpreter::execute_kernel_CircleMirrorPad(), luci_interpreter::execute_kernel_CircleSlice(), luci::exportNodes(), luci::CircleReader::file_data(), tflite2circle::CircleModel::finalize(), nnfw::cker::FullyConnectedHybrid(), souschef::ExplicitDataChef< std::string >::generate(), record_minmax::getMovingAverage(), mir::TensorVariant::getOffset(), onert::exec::feature::nchw::Reader< T >::getRef(), onert::exec::feature::nhwc::Reader< T >::getRef(), onert_micro::train::pal::GRUWeightGrads(), onert_micro::core::train::OMTrainingHandler::handleError(), flexbuffers::Indirect(), flexbuffers::Indirect(), onert::loader::BaseLoader< LoaderDomain >::loadOperand(), luci_interpreter::kernels::testing::makeInputTensor(), nnfw::cker::train::MaxPool2D(), nnfw::cker::train::MSEGrad(), record_minmax::DirectoryIterator::next(), nnfw::misc::vector::Object< T >::Object(), nnfw::cker::FusedBatchNorm::operator()(), Eigen::internal::gemm_pack_rhs< Scalar, Index, TensorContractionSubMapper< Scalar, Index, Rhs, TensorEvaluator< const TensorReshapingOp< NewDimension, const TensorImagePatchOp< Rows, Cols, ArgType > >, Device >, nocontract_t, contract_t, packet_size, inner_dim_contiguous, inner_dim_reordered, Alignment >, nr, ColMajor, false, false >::operator()(), Eigen::internal::gemm_pack_rhs< Scalar, Index, TensorContractionSubMapper< Scalar, Index, Rhs, TensorEvaluator< const TensorReshapingOp< NewDimension, const TensorImagePatchOp< Rows, Cols, ArgType > >, Device >, nocontract_t, contract_t, 2, inner_dim_contiguous, inner_dim_reordered, Alignment >, nr, ColMajor, false, false >::operator()(), Eigen::internal::gemm_pack_rhs< Scalar, Index, TensorContractionSubMapper< Scalar, Index, Rhs, TensorEvaluator< const TensorReshapingOp< NewDimension, const TensorImagePatchOp< Rows, Cols, ArgType > >, Device >, nocontract_t, contract_t, 1, inner_dim_contiguous, inner_dim_reordered, Alignment >, nr, ColMajor, false, false >::operator()(), flatbuffers::VectorIterator< T, IT >::operator+(), flatbuffers::VectorIterator< T, IT >::operator+=(), flatbuffers::VectorIterator< T, IT >::operator-(), flatbuffers::VectorIterator< T, IT >::operator-=(), nnfw::cker::PortableAsymmetricQuantizeFloats(), quantize_qasymm8(), tflite::reference_ops::RankOneSelect(), nnfw::cker::RankOneSelect(), nnfw::cker::ReducedOutputOffset(), onert::backend::train::FirstFitPlanner< Index >::release(), onert::backend::basic::FirstFitPlanner::release(), flatbuffers::vector_downward::release_raw(), flatbuffers::FlatBufferBuilder::ReleaseRaw(), flatbuffers::grpc::MessageBuilder::ReleaseRaw(), coco::Arg::reorder(), nnfw::cker::RoPE(), nnfw::cker::functor::FillPhiloxRandomTask< Distribution, true >::Run(), nnfw::cker::functor::FillPhiloxRandomTask< Distribution, false >::Run(), MappedMemory::set(), luci_interpreter::Tensor::set_offset(), ModelArgumentInfo::setFromMemory(), ExecutionBuilder::setInputFromMemory(), ModelBuilder::setOperandValueFromMemory(), ExecutionBuilder::setOutputFromMemory(), and onert_micro::core::OMTrainingRuntimeModule::trainSingleStep().

◆ tensor3D_offset()

__global const uchar * tensor3D_offset ( const Tensor3D tensor,
int  x,
int  y,
int  z 
)
inline

Get the pointer position of a Tensor3D

Parameters
[in]tensorPointer to the starting position of the buffer
[in]xRelative X position
[in]yRelative Y position
[in]zRelative Z position

Definition at line 552 of file helpers.h.

553{
554 return tensor->ptr + x * tensor->stride_x + y * tensor->stride_y + z * tensor->stride_z;
555}

◆ tensor4D_offset()

__global const uchar * tensor4D_offset ( const Tensor4D tensor,
int  x,
int  y,
int  z,
int  w 
)
inline

Get the pointer position of a Tensor4D

Parameters
[in]tensorPointer to the starting position of the buffer
[in]xRelative X position
[in]yRelative Y position
[in]zRelative Z position
[in]wRelative W position

Definition at line 565 of file helpers.h.

566{
567 return tensor->ptr + x * tensor->stride_x + y * tensor->stride_y + z * tensor->stride_z +
568 w * tensor->stride_w;
569}

◆ update_image_from_tensor3D_workitem_ptr()

Image update_image_from_tensor3D_workitem_ptr ( __global uchar *  ptr,
uint  offset_first_element_in_bytes,
uint  stride_x,
uint  step_x,
uint  stride_y,
uint  step_y,
uint  stride_z,
uint  step_z 
)
inline

Wrap 3D tensor information into an image structure, and make the pointer point at this workitem's data.

Parameters
[in]ptrPointer to the starting postion of the buffer
[in]offset_first_element_in_bytesThe offset of the first element in the source image
[in]stride_xStride of the image in X dimension (in bytes)
[in]step_xstride_x * number of elements along X processed per workitem(in bytes)
[in]stride_yStride of the image in Y dimension (in bytes)
[in]step_ystride_y * number of elements along Y processed per workitem(in bytes)
[in]stride_zStride of the image in Z dimension (in bytes)
[in]step_zstride_z * number of elements along Z processed per workitem(in bytes)
Returns
A 3D tensor object

Definition at line 460 of file helpers.h.

464{
465 Image img = {.ptr = ptr,
466 .offset_first_element_in_bytes = offset_first_element_in_bytes,
467 .stride_x = stride_x,
468 .stride_y = stride_y};
469 img.ptr += img.offset_first_element_in_bytes + get_global_id(0) * step_x +
470 get_global_id(1) * step_y + get_global_id(2) * step_z;
471 return img;
472}

References Image::offset_first_element_in_bytes, and Image::ptr.

◆ update_image_workitem_ptr()

Image update_image_workitem_ptr ( __global uchar *  ptr,
uint  offset_first_element_in_bytes,
uint  stride_x,
uint  step_x,
uint  stride_y,
uint  step_y 
)
inline

Wrap image information into an Image structure, and make the pointer point at this workitem's data.

Parameters
[in]ptrPointer to the starting postion of the buffer
[in]offset_first_element_in_bytesThe offset of the first element in the source image
[in]stride_xStride of the image in X dimension (in bytes)
[in]step_xstride_x * number of elements along X processed per workitem(in bytes)
[in]stride_yStride of the image in Y dimension (in bytes)
[in]step_ystride_y * number of elements along Y processed per workitem(in bytes)
Returns
An image object

Definition at line 431 of file helpers.h.

433{
434 Image img = {.ptr = ptr,
435 .offset_first_element_in_bytes = offset_first_element_in_bytes,
436 .stride_x = stride_x,
437 .stride_y = stride_y};
438 img.ptr +=
439 img.offset_first_element_in_bytes + get_global_id(0) * step_x + get_global_id(1) * step_y;
440 return img;
441}

References Image::offset_first_element_in_bytes, and Image::ptr.

◆ update_tensor3D_workitem_ptr()

Tensor3D update_tensor3D_workitem_ptr ( __global uchar *  ptr,
uint  offset_first_element_in_bytes,
uint  stride_x,
uint  step_x,
uint  stride_y,
uint  step_y,
uint  stride_z,
uint  step_z 
)
inline

Wrap 3D tensor information into an tensor structure, and make the pointer point at this workitem's data.

Parameters
[in]ptrPointer to the starting postion of the buffer
[in]offset_first_element_in_bytesThe offset of the first element in the source image
[in]stride_xStride of the image in X dimension (in bytes)
[in]step_xstride_x * number of elements along X processed per workitem(in bytes)
[in]stride_yStride of the image in Y dimension (in bytes)
[in]step_ystride_y * number of elements along Y processed per workitem(in bytes)
[in]stride_zStride of the image in Z dimension (in bytes)
[in]step_zstride_z * number of elements along Z processed per workitem(in bytes)
Returns
A 3D tensor object

Definition at line 491 of file helpers.h.

495{
496 Tensor3D tensor = {.ptr = ptr,
497 .offset_first_element_in_bytes = offset_first_element_in_bytes,
498 .stride_x = stride_x,
499 .stride_y = stride_y,
500 .stride_z = stride_z};
501 tensor.ptr += tensor.offset_first_element_in_bytes + get_global_id(0) * step_x +
502 get_global_id(1) * step_y + get_global_id(2) * step_z;
503 return tensor;
504}

◆ update_tensor4D_workitem_ptr()

Tensor4D update_tensor4D_workitem_ptr ( __global uchar *  ptr,
uint  offset_first_element_in_bytes,
uint  stride_x,
uint  step_x,
uint  stride_y,
uint  step_y,
uint  stride_z,
uint  step_z,
uint  stride_w,
uint  step_w,
uint  mod_size 
)
inline

Definition at line 506 of file helpers.h.

510{
511 Tensor4D tensor = {.ptr = ptr,
512 .offset_first_element_in_bytes = offset_first_element_in_bytes,
513 .stride_x = stride_x,
514 .stride_y = stride_y,
515 .stride_z = stride_z,
516 .stride_w = stride_w};
517
518 tensor.ptr += tensor.offset_first_element_in_bytes + get_global_id(0) * step_x +
519 get_global_id(1) * step_y + (get_global_id(2) % mod_size) * step_z +
520 (get_global_id(2) / mod_size) * step_w;
521 return tensor;
522}

◆ update_vector_workitem_ptr()

Vector update_vector_workitem_ptr ( __global uchar *  ptr,
uint  offset_first_element_in_bytes,
uint  stride_x,
uint  step_x 
)
inline

Wrap vector information into an Vector structure, and make the pointer point at this workitem's data.

Parameters
[in]ptrPointer to the starting postion of the buffer
[in]offset_first_element_in_bytesThe offset of the first element in the source vector
[in]stride_xStride of the vector in X dimension (in bytes)
[in]step_xstride_x * number of elements along X processed per workitem(in bytes)
Returns
An image object

Definition at line 405 of file helpers.h.

407{
408 Vector vector = {
409 .ptr = ptr,
410 .offset_first_element_in_bytes = offset_first_element_in_bytes,
411 .stride_x = stride_x,
412 };
413 vector.ptr += vector.offset_first_element_in_bytes + get_global_id(0) * step_x;
414 return vector;
415}

References Vector::offset_first_element_in_bytes, and Vector::ptr.

◆ vector_offset()

__global const uchar * vector_offset ( const Vector vec,
int  x 
)
inline

Get the pointer position of a Vector

Parameters
[in]vecPointer to the starting position of the buffer
[in]xRelative X position

Definition at line 529 of file helpers.h.

530{
531 return vec->ptr + x * vec->stride_x;
532}

References Vector::ptr, and Vector::stride_x.