HIP: Heterogenous-computing Interface for Portability
Loading...
Searching...
No Matches
hip_ldg.h
1/*
2Copyright (c) 2015 - 2021 Advanced Micro Devices, Inc. All rights reserved.
3
4Permission is hereby granted, free of charge, to any person obtaining a copy
5of this software and associated documentation files (the "Software"), to deal
6in the Software without restriction, including without limitation the rights
7to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
8copies of the Software, and to permit persons to whom the Software is
9furnished to do so, subject to the following conditions:
10
11The above copyright notice and this permission notice shall be included in
12all copies or substantial portions of the Software.
13
14THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
15IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
16FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
17AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
18LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
19OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN
20THE SOFTWARE.
21*/
22
23#ifndef HIP_INCLUDE_HIP_AMD_DETAIL_HIP_LDG_H
24#define HIP_INCLUDE_HIP_AMD_DETAIL_HIP_LDG_H
25
26#if __HIP_CLANG_ONLY__
27#include "amd_hip_vector_types.h"
28#include "host_defines.h"
29
30__device__ inline static char __ldg(const char* ptr) { return *ptr; }
31
32__device__ inline static char2 __ldg(const char2* ptr) { return *ptr; }
33
34__device__ inline static char4 __ldg(const char4* ptr) { return *ptr; }
35
36__device__ inline static signed char __ldg(const signed char* ptr) { return ptr[0]; }
37
38__device__ inline static unsigned char __ldg(const unsigned char* ptr) { return ptr[0]; }
39
40
41__device__ inline static short __ldg(const short* ptr) { return ptr[0]; }
42
43__device__ inline static short2 __ldg(const short2* ptr) { return ptr[0]; }
44
45__device__ inline static short4 __ldg(const short4* ptr) { return ptr[0]; }
46
47__device__ inline static unsigned short __ldg(const unsigned short* ptr) { return ptr[0]; }
48
49
50__device__ inline static int __ldg(const int* ptr) { return ptr[0]; }
51
52__device__ inline static int2 __ldg(const int2* ptr) { return ptr[0]; }
53
54__device__ inline static int4 __ldg(const int4* ptr) { return ptr[0]; }
55
56__device__ inline static unsigned int __ldg(const unsigned int* ptr) { return ptr[0]; }
57
58
59__device__ inline static long __ldg(const long* ptr) { return ptr[0]; }
60
61__device__ inline static unsigned long __ldg(const unsigned long* ptr) { return ptr[0]; }
62
63
64__device__ inline static long long __ldg(const long long* ptr) { return ptr[0]; }
65
66__device__ inline static longlong2 __ldg(const longlong2* ptr) { return ptr[0]; }
67
68__device__ inline static unsigned long long __ldg(const unsigned long long* ptr) { return ptr[0]; }
69
70
71__device__ inline static uchar2 __ldg(const uchar2* ptr) { return ptr[0]; }
72
73__device__ inline static uchar4 __ldg(const uchar4* ptr) { return ptr[0]; }
74
75
76__device__ inline static ushort2 __ldg(const ushort2* ptr) { return ptr[0]; }
77
78
79__device__ inline static uint2 __ldg(const uint2* ptr) { return ptr[0]; }
80
81__device__ inline static uint4 __ldg(const uint4* ptr) { return ptr[0]; }
82
83
84__device__ inline static ulonglong2 __ldg(const ulonglong2* ptr) { return ptr[0]; }
85
86
87__device__ inline static float __ldg(const float* ptr) { return ptr[0]; }
88
89__device__ inline static float2 __ldg(const float2* ptr) { return ptr[0]; }
90
91__device__ inline static float4 __ldg(const float4* ptr) { return ptr[0]; }
92
93
94__device__ inline static double __ldg(const double* ptr) { return ptr[0]; }
95
96__device__ inline static double2 __ldg(const double2* ptr) { return ptr[0]; }
97
98#endif // __HIP_CLANG_ONLY__
99
100#endif // HIP_LDG_H
Definition amd_hip_vector_types.h:1616
Definition amd_hip_vector_types.h:1623
Definition amd_hip_vector_types.h:1653
Definition amd_hip_vector_types.h:1660
Definition amd_hip_vector_types.h:1690
Definition amd_hip_vector_types.h:1697
Definition amd_hip_vector_types.h:1727
Definition amd_hip_vector_types.h:1764
Definition amd_hip_vector_types.h:1771
Definition amd_hip_vector_types.h:1801
Definition amd_hip_vector_types.h:1808
Definition amd_hip_vector_types.h:1912
Definition amd_hip_vector_types.h:1949
Definition amd_hip_vector_types.h:1986
Definition amd_hip_vector_types.h:1993
Definition amd_hip_vector_types.h:2023