Skip to content

GitLab

Switch branch/tag
Find file Normal viewHistoryPermalink
cuda_commen.h 3.08 KB
Newer Older
ccfd's avatar
first commit  
ccfd committed
1
2
3
4
5
6
7
8
9
10
11
12
13
14
15
16
17
18
19
20
21
22
23
24 
#ifndef _CUDA_HEAD_H
#define _CUDA_HEAD_H

#include "config_parameters.h"
#include "cuda.h"
#include "cuda_runtime.h"
#include "mpi.h"

#ifdef __cplusplus
extern "C"{
#endif

#ifdef DEBUG_MODE
#include "stdio.h"
#define CUDA_CALL(CALL) \
    {\
        cudaError_t error;\
        error = CALL;\
        if(error != cudaSuccess){\
            printf("In file \"%s\" , line %d ( \"%s\" ) , cuda call failed\n",__FILE__ , __LINE__ , __FUNCTION__);\
            printf("Error code : %s  (%s)\n",cudaGetErrorString(error) , cudaGetErrorName(error));\
            MPI_Abort(((MPI_Comm)0x44000000) , 1);\
        }\
    }
ccfd's avatar
...  
ccfd committed
25 
#else
ccfd's avatar
first commit  
ccfd committed
26 
#define CUDA_CALL(CALL) CALL;
ccfd's avatar
...  
ccfd committed
27
28 
#endif
ccfd's avatar
first commit  
ccfd committed
29
30
31
32
33
34
35
36
37
38
39
40
41 
#ifdef DEBUG_MODE
#include "stdio.h"
#define CUDA_LAUNCH(call)\
    {\
        cudaError_t error;\
        call;\
        error = cudaGetLastError();\
        if( error != cudaSuccess){\
            printf("In file \"%s\" , line %d ( \"%s\" ) , cuda call failed\n",__FILE__ , __LINE__ , __FUNCTION__);\
            printf("Error code : %s  (%s)\n",cudaGetErrorString(error) , cudaGetErrorName(error));\
            MPI_Abort(((MPI_Comm)0x44000000) , 1);\
        }\
    }
ccfd's avatar
...  
ccfd committed
42 
#else
ccfd's avatar
first commit  
ccfd committed
43
44 
#define CUDA_LAUNCH(call)\
    call;
ccfd's avatar
...  
ccfd committed
45 
#endif
ccfd's avatar
first commit  
ccfd committed
46
47
48
49
50
51
52
53
54
55
56
57
58
59
60
61
62
63
64
65
66
67
68
69
70
71
72
73
74
75
76
77
78
79
80
81
82
83
84
85
86
87
88
89
90
91
92
93
94
95
96
97
98
99
100
101
102
103
104
105
106
107
108
109
110
111
112
113
114
115
116
117
118
119
120
121
122
123
124
125
126
127
128
129
130
131
132
133
134
135
136 

typedef struct cudaField_ {
    REAL * ptr;
    unsigned int pitch;
    #ifdef __HIPCC__
    unsigned int dummy0;
    #endif
} cudaField;

typedef struct cudaField_int_ {
    int * ptr;
    unsigned int pitch;
    #ifdef __HIPCC__
    unsigned int dummy0;
    #endif
} cudaField_int;

typedef struct cudaSoA_{
    REAL * ptr;
    unsigned int pitch;
    unsigned int length_Y;
    unsigned int length_Z;
    #ifdef __HIPCC__
    unsigned int dummy0;
    #endif
} cudaSoA;


typedef struct cudaFieldPack_{
    REAL * ptr;
} cudaFieldPack;


extern uint32_t BlockDimX;
extern uint32_t BlockDimY;
extern uint32_t BlockDimZ;

extern dim3 BlockDim_X;
extern dim3 BlockDim_Y;
extern dim3 BlockDim_Z;

extern dim3 BlockDim;

extern int MaxThreadsPerBlock;
extern int MaxBlockDimX;
extern int MaxBlockDimY;
extern int MaxBlockDimZ;
extern int MaxGridDimX;
extern int MaxGridDimY;
extern int MaxGridDimZ;


void cuda_commen_init();


typedef struct cudaJobPackage_
{
    dim3 start;
    dim3 end;
    #ifdef __cplusplus
    __host__ __device__ cudaJobPackage_(){};
    __host__ __device__ cudaJobPackage_(const dim3 & start_in , const dim3 & end_in):start(start_in),end(end_in){};
    __host__ __device__ void setup(const dim3 & start_in , const dim3 & end_in){start = start_in ,end = end_in;};
    #endif
} cudaJobPackage;

inline void jobsize(cudaJobPackage *job , dim3 *size){
    size->x = job->end.x - job->start.x;
    size->y = job->end.y - job->start.y;
    size->z = job->end.z - job->start.z;
}

inline void job2size(cudaJobPackage *job , dim3 *size){
    size->x = job->end.x;
    size->y = job->end.y;
    size->z = job->end.z;
}
inline void job2job(cudaJobPackage * job , cudaJobPackage * out){
    out->start.x = job->start.x;
    out->start.y = job->start.y;
    out->start.z = job->start.z;
    
    out->end.x = job->end.x + job->start.x;
    out->end.y = job->end.y + job->start.y;
    out->end.z = job->end.z + job->start.z;
}

#ifdef __cplusplus
}
#endif
#endif