asc_stcg
Applicability
Product |
Supported |
|---|---|
Atlas 350 Accelerator Card |
√ |
x |
|
x |
|
x |
|
x |
|
x |
|
x |
Function Usage
Stores specified data to the address of the global memory and caches the data to the L2 cache, but not to the data cache.
Prototype
1 | inline void asc_stcg(long int* address, long int val) |
1 | inline void asc_stcg(unsigned long int* address, unsigned long int val) |
1 | inline void asc_stcg(long long int* address, long long int val) |
1 | inline void asc_stcg(unsigned long long int* address, unsigned long long int val) |
1 | inline void asc_stcg(long2* address, long2 val) |
1 | inline void asc_stcg(ulong2* address, ulong2 val) |
1 | inline void asc_stcg(long4* address, long4 val) |
1 | inline void asc_stcg(ulong4* address, ulong4 val) |
1 | inline void asc_stcg(longlong2* address, longlong2 val) |
1 | inline void asc_stcg(ulonglong2* address, ulonglong2 val) |
1 | inline void asc_stcg(longlong4* address, longlong4 val) |
1 | inline void asc_stcg(ulonglong4* address, ulonglong4 val) |
1 | inline void asc_stcg(signed char* address, signed char val) |
1 | inline void asc_stcg(unsigned char* address, unsigned char val) |
1 | inline void asc_stcg(char2* address, char2 val) |
1 | inline void asc_stcg(uchar2* address, uchar2 val) |
1 | inline void asc_stcg(char4* address, char4 val) |
1 | inline void asc_stcg(uchar4* address, uchar4 val) |
1 | inline void asc_stcg(short* address, short val) |
1 | inline void asc_stcg(unsigned short* address, unsigned short val) |
1 | inline void asc_stcg(short2* address, short2 val) |
1 | inline void asc_stcg(ushort2* address, ushort2 val) |
1 | inline void asc_stcg(short4* address, short4 val) |
1 | inline void asc_stcg(ushort4* address, ushort4 val) |
1 | inline void asc_stcg(int* address, int val) |
1 | inline void asc_stcg(unsigned int* address, unsigned int val) |
1 | inline void asc_stcg(int2* address, int2 val) |
1 | inline void asc_stcg(uint2* address, uint2 val) |
1 | inline void asc_stcg(int4* address, int4 val) |
1 | inline void asc_stcg(uint4* address, uint4 val) |
1 | inline void asc_stcg(float* address, float val) |
1 | inline void asc_stcg(float2* address, float2 val) |
1 | inline void asc_stcg(float4* address, float4 val) |
1 | inline void asc_stcg(bfloat16_t* address, bfloat16_t val) |
1 | inline void asc_stcg(bfloat16x2_t* address, bfloat16x2_t val) |
1 | inline void asc_stcg(half* address, half val) |
1 | inline void asc_stcg(half2* address, half2 val) |
Parameters
Parameter |
Input/Output |
Description |
|---|---|---|
address |
Input |
Address of the Global Memory |
val |
Input |
Source operand. |
Returns
None
Restrictions
For SIMT programming, this API is not supported.
Header Files to Be Included
To use APIs of types other than half, half2, bfloat16_t, and bfloat16x2_t, include the simt_api/device_functions.h header file. To use APIs of the half and half2 types, include the simt_api/asc_fp16.h header file. To use APIs of the bfloat16_t and bfloat16x2_t types, include the simt_api/asc_bf16.h header file.
1 | #include "simt_api/device_functions.h" |
1 | #include "simt_api/asc_fp16.h" |
1 | #include "simt_api/asc_bf16.h" |
Examples
1 2 3 4 5 | __simt_vf__ __launch_bounds__(1024) inline void kernel_asc_stcg(__gm__ float* src, __gm__ float* val) { int idx = threadIdx.x + blockIdx.x * blockDim.x; asc_stcg(src + idx, val[idx]); } |