Cache Hints for Memory Operations#

These functions provide explicit control over caching behavior for memory operations. They generate PTX instructions with cache policy hints that can optimize specific memory access patterns. All functions support arrays or pointers with all bitwidths of signed/unsigned integer and floating-point types.

See also

Cache Operators in the PTX ISA documentation.

numba.cuda.ldca(array, i)#: Load element i from array with cache-all policy (ld.global.ca). This is the default caching behavior.

numba.cuda.ldcg(array, i)#: Load element i from array with cache-global policy (ld.global.cg). Useful for data shared across thread blocks.

numba.cuda.ldcs(array, i)#: Load element i from array with cache-streaming policy (ld.global.cs). Optimized for streaming data accessed once.

numba.cuda.ldlu(array, i)#: Load element i from array with last-use policy (ld.global.lu). Indicates data is unlikely to be reused.

numba.cuda.ldcv(array, i)#: Load element i from array with cache-volatile policy (ld.global.cv). Used for volatile data that may change externally.

numba.cuda.stcg(array, i, value)#: Store value to array[i] with cache-global policy (st.global.cg). Useful for data shared across thread blocks.

numba.cuda.stcs(array, i, value)#: Store value to array[i] with cache-streaming policy (st.global.cs). Optimized for streaming writes.

numba.cuda.stwb(array, i, value)#: Store value to array[i] with write-back policy (st.global.wb). This is the default caching behavior.

numba.cuda.stwt(array, i, value)#: Store value to array[i] with write-through policy (st.global.wt). Writes through cache hierarchy to memory.