Lines Matching defs:ptr

262 inline __device__ char __ldg(const char *ptr) { return __nvvm_ldg_c(ptr); }
263 inline __device__ short __ldg(const short *ptr) { return __nvvm_ldg_s(ptr); }
264 inline __device__ int __ldg(const int *ptr) { return __nvvm_ldg_i(ptr); }
265 inline __device__ long __ldg(const long *ptr) { return __nvvm_ldg_l(ptr); }
266 inline __device__ long long __ldg(const long long *ptr) {
267 return __nvvm_ldg_ll(ptr);
269 inline __device__ unsigned char __ldg(const unsigned char *ptr) {
270 return __nvvm_ldg_uc(ptr);
272 inline __device__ signed char __ldg(const signed char *ptr) {
273 return __nvvm_ldg_uc((const unsigned char *)ptr);
275 inline __device__ unsigned short __ldg(const unsigned short *ptr) {
276 return __nvvm_ldg_us(ptr);
278 inline __device__ unsigned int __ldg(const unsigned int *ptr) {
279 return __nvvm_ldg_ui(ptr);
281 inline __device__ unsigned long __ldg(const unsigned long *ptr) {
282 return __nvvm_ldg_ul(ptr);
284 inline __device__ unsigned long long __ldg(const unsigned long long *ptr) {
285 return __nvvm_ldg_ull(ptr);
287 inline __device__ float __ldg(const float *ptr) { return __nvvm_ldg_f(ptr); }
288 inline __device__ double __ldg(const double *ptr) { return __nvvm_ldg_d(ptr); }
290 inline __device__ char2 __ldg(const char2 *ptr) {
292 // We can assume that ptr is aligned at least to char2's alignment, but the
293 // load will assume that ptr is aligned to char2's alignment. This is only
295 c2 rv = __nvvm_ldg_c2(reinterpret_cast<const c2 *>(ptr));
301 inline __device__ char4 __ldg(const char4 *ptr) {
303 c4 rv = __nvvm_ldg_c4(reinterpret_cast<const c4 *>(ptr));
311 inline __device__ short2 __ldg(const short2 *ptr) {
313 s2 rv = __nvvm_ldg_s2(reinterpret_cast<const s2 *>(ptr));
319 inline __device__ short4 __ldg(const short4 *ptr) {
321 s4 rv = __nvvm_ldg_s4(reinterpret_cast<const s4 *>(ptr));
329 inline __device__ int2 __ldg(const int2 *ptr) {
331 i2 rv = __nvvm_ldg_i2(reinterpret_cast<const i2 *>(ptr));
337 inline __device__ int4 __ldg(const int4 *ptr) {
339 i4 rv = __nvvm_ldg_i4(reinterpret_cast<const i4 *>(ptr));
347 inline __device__ longlong2 __ldg(const longlong2 *ptr) {
349 ll2 rv = __nvvm_ldg_ll2(reinterpret_cast<const ll2 *>(ptr));
356 inline __device__ uchar2 __ldg(const uchar2 *ptr) {
358 uc2 rv = __nvvm_ldg_uc2(reinterpret_cast<const uc2 *>(ptr));
364 inline __device__ uchar4 __ldg(const uchar4 *ptr) {
366 uc4 rv = __nvvm_ldg_uc4(reinterpret_cast<const uc4 *>(ptr));
374 inline __device__ ushort2 __ldg(const ushort2 *ptr) {
376 us2 rv = __nvvm_ldg_us2(reinterpret_cast<const us2 *>(ptr));
382 inline __device__ ushort4 __ldg(const ushort4 *ptr) {
384 us4 rv = __nvvm_ldg_us4(reinterpret_cast<const us4 *>(ptr));
392 inline __device__ uint2 __ldg(const uint2 *ptr) {
394 ui2 rv = __nvvm_ldg_ui2(reinterpret_cast<const ui2 *>(ptr));
400 inline __device__ uint4 __ldg(const uint4 *ptr) {
402 ui4 rv = __nvvm_ldg_ui4(reinterpret_cast<const ui4 *>(ptr));
410 inline __device__ ulonglong2 __ldg(const ulonglong2 *ptr) {
412 ull2 rv = __nvvm_ldg_ull2(reinterpret_cast<const ull2 *>(ptr));
419 inline __device__ float2 __ldg(const float2 *ptr) {
421 f2 rv = __nvvm_ldg_f2(reinterpret_cast<const f2 *>(ptr));
427 inline __device__ float4 __ldg(const float4 *ptr) {
429 f4 rv = __nvvm_ldg_f4(reinterpret_cast<const f4 *>(ptr));
437 inline __device__ double2 __ldg(const double2 *ptr) {
439 d2 rv = __nvvm_ldg_d2(reinterpret_cast<const d2 *>(ptr));
579 __device__ inline unsigned __isCtaShared(const void *ptr) {
580 return __isShared(ptr);