local3clc
Address space qualifier.
Overview
OpenCL implements the following disjoint address spaces: __global, __local, __constant, and __private The address space qualifier may be used in variable declarations to specify the region of memory that is used to allocate the object. The C syntax for type qualifiers is extended in OpenCL to include an address space name as a valid type qualifier. If the type of an object is qualified by an address space name, the object is allocated in the specified address name; otherwise, the object is allocated in the generic address space.
The address space names without the __prefix i.e. global, local, constant and private may be substituted for the corresponding address space names with the __prefix.
The generic address space name for arguments to a function in a program, or local variables of a function is __private. All function arguments shall be in the __private address space.
functionQualifiers(3clc) function arguments declared to be a pointer of a type can point to one of the following address spaces only: __global, __local or __constant. A pointer to address space A can only be assigned to a pointer to the same address space A. Casting a pointer to address space A to a pointer to address space B is illegal.
Function arguments of type otherDataTypes(3clc), otherDataTypes(3clc), otherDataTypes(3clc), otherDataTypes(3clc), otherDataTypes(3clc), and otherDataTypes(3clc) refer to image memory objects allocated in the __global address space.
Notes
The __local or local address space name is used to describe variables that need to be allocated in local memory and are shared by all work-items of a work-group. Pointers to the __local address space are allowed as arguments to functions (including __kernel functions). Variables allocated in the __local address space can also be defined inside a __kernel function scope.
Example
There is no generic address space name for program scope variables. All program scope variables must be declared in the __constant address space. For example:
| // declares a pointer p in the __private address space that // points to an int object in address space __global __global int *p; // declares an array of 4 floats in the __private address space. float x[4]; |
There is no address space for function return values. Using an address space qualifier in a function return type declaration will generate a compilation error, unless the return type is declared as a pointer type and the qualifier is used on the points-to address space.
For example:
| __private int f() { ... } // should generate an error __local int *f() { ... } // allowed __local int * __private f() { ... }; // should generate an error. |
| // Examples of variables allocated in the __local address space // inside a __kernel function __kernel void my_func(...) { __local float a; // A single float allocated // in local address space __local float b[10]; // An array of 10 floats // allocated in local address space. if (...) { // example of variable in __local address space but not // declared at __kernel function scope. __local float c; // not allowed. } } |
Variables allocated in the __local address space inside a __kernel function cannot be initialized.
__kernel void my_func(...)
{
local float a = 1; // not allowed __local float b; b = 1; // allowed
}
NOTE: Variables allocated in the __local address space inside a __kernel function are allocated for each work-group executing the kernel.
Specification
OpenCL Specification[1]
Notes
- 1.
-
OpenCL Specification
page 222, section 6.5 - Address Space Qualifiers
Releases
5 releases since Debian 8 (Jessie); the text changed 3 times.