HIP: Heterogenous-computing Interface for Portability
external
hip-on-vdi
include
hip
nvcc_detail
hip_runtime.h
1
/*
2
Copyright (c) 2015 - present Advanced Micro Devices, Inc. All rights reserved.
3
4
Permission is hereby granted, free of charge, to any person obtaining a copy
5
of this software and associated documentation files (the "Software"), to deal
6
in the Software without restriction, including without limitation the rights
7
to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
8
copies of the Software, and to permit persons to whom the Software is
9
furnished to do so, subject to the following conditions:
10
11
The above copyright notice and this permission notice shall be included in
12
all copies or substantial portions of the Software.
13
14
THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
15
IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
16
FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
17
AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
18
LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
19
OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN
20
THE SOFTWARE.
21
*/
22
23
#ifndef HIP_INCLUDE_HIP_NVCC_DETAIL_HIP_RUNTIME_H
24
#define HIP_INCLUDE_HIP_NVCC_DETAIL_HIP_RUNTIME_H
25
26
#include <cuda_runtime.h>
27
28
#include <
hip/hip_runtime_api.h
>
29
30
#define HIP_KERNEL_NAME(...) __VA_ARGS__
31
32
typedef
int
hipLaunchParm;
33
34
#define hipLaunchKernelGGLInternal(kernelName, numBlocks, numThreads, memPerBlock, streamId, ...) \
35
do { \
36
kernelName<<<numBlocks, numThreads, memPerBlock, streamId>>>(__VA_ARGS__); \
37
} while (0)
38
39
#define hipLaunchKernelGGL(kernelName, ...) hipLaunchKernelGGLInternal((kernelName), __VA_ARGS__)
40
41
#define hipReadModeElementType cudaReadModeElementType
42
43
#ifdef __CUDA_ARCH__
44
45
46
// 32-bit Atomics:
47
#define __HIP_ARCH_HAS_GLOBAL_INT32_ATOMICS__ (__CUDA_ARCH__ >= 110)
48
#define __HIP_ARCH_HAS_GLOBAL_FLOAT_ATOMIC_EXCH__ (__CUDA_ARCH__ >= 110)
49
#define __HIP_ARCH_HAS_SHARED_INT32_ATOMICS__ (__CUDA_ARCH__ >= 120)
50
#define __HIP_ARCH_HAS_SHARED_FLOAT_ATOMIC_EXCH__ (__CUDA_ARCH__ >= 120)
51
#define __HIP_ARCH_HAS_FLOAT_ATOMIC_ADD__ (__CUDA_ARCH__ >= 200)
52
53
// 64-bit Atomics:
54
#define __HIP_ARCH_HAS_GLOBAL_INT64_ATOMICS__ (__CUDA_ARCH__ >= 200)
55
#define __HIP_ARCH_HAS_SHARED_INT64_ATOMICS__ (__CUDA_ARCH__ >= 120)
56
57
// Doubles
58
#define __HIP_ARCH_HAS_DOUBLES__ (__CUDA_ARCH__ >= 120)
59
60
// warp cross-lane operations:
61
#define __HIP_ARCH_HAS_WARP_VOTE__ (__CUDA_ARCH__ >= 120)
62
#define __HIP_ARCH_HAS_WARP_BALLOT__ (__CUDA_ARCH__ >= 200)
63
#define __HIP_ARCH_HAS_WARP_SHUFFLE__ (__CUDA_ARCH__ >= 300)
64
#define __HIP_ARCH_HAS_WARP_FUNNEL_SHIFT__ (__CUDA_ARCH__ >= 350)
65
66
// sync
67
#define __HIP_ARCH_HAS_THREAD_FENCE_SYSTEM__ (__CUDA_ARCH__ >= 200)
68
#define __HIP_ARCH_HAS_SYNC_THREAD_EXT__ (__CUDA_ARCH__ >= 200)
69
70
// misc
71
#define __HIP_ARCH_HAS_SURFACE_FUNCS__ (__CUDA_ARCH__ >= 200)
72
#define __HIP_ARCH_HAS_3DGRID__ (__CUDA_ARCH__ >= 200)
73
#define __HIP_ARCH_HAS_DYNAMIC_PARALLEL__ (__CUDA_ARCH__ >= 350)
74
75
#endif
76
77
#ifdef __CUDACC__
78
79
80
#define hipThreadIdx_x threadIdx.x
81
#define hipThreadIdx_y threadIdx.y
82
#define hipThreadIdx_z threadIdx.z
83
84
#define hipBlockIdx_x blockIdx.x
85
#define hipBlockIdx_y blockIdx.y
86
#define hipBlockIdx_z blockIdx.z
87
88
#define hipBlockDim_x blockDim.x
89
#define hipBlockDim_y blockDim.y
90
#define hipBlockDim_z blockDim.z
91
92
#define hipGridDim_x gridDim.x
93
#define hipGridDim_y gridDim.y
94
#define hipGridDim_z gridDim.z
95
96
#define HIP_SYMBOL(X) &X
97
102
#define HIP_DYNAMIC_SHARED(type, var) extern __shared__ type var[];
103
104
#define HIP_DYNAMIC_SHARED_ATTRIBUTE
105
106
#ifdef __HIP_DEVICE_COMPILE__
107
#define abort_() \
108
{ asm("trap;"); }
109
#undef assert
110
#define assert(COND) \
111
{ \
112
if (!COND) { \
113
abort_(); \
114
} \
115
}
116
#endif
117
118
#define __clock() clock()
119
#define __clock64() clock64()
120
121
#endif
122
123
#endif
hip_runtime_api.h
Defines the API signatures for HIP runtime. This file can be compiled with a standard compiler.
Generated on Tue Jan 19 2021 09:16:12 for HIP: Heterogenous-computing Interface for Portability by
1.8.18