3 #include "hip/hip_runtime.h"
7 hipError_t error = cmd; \
8 if (error != hipSuccess) { \
9 fprintf(stderr, "error: '%s'(%d) at %s:%d\n", hipGetErrorString(error), error, \
10 __FILE__, __LINE__); \
15 // Defining number of elements in Array
18 __device__
int add_one(int no
) {
22 __device__
int multiply_two(int no
) {
27 __device__
int multiply_three(int no
) {
33 __device__
int gpu_operations(int a
,int b
) {
35 c
= multiply_two(a
) + multiply_three(b
);
40 __global__
void gpu_kernel_operations(int *d_a
, int *d_b
, int *d_c
) {
41 // Getting block index of current kernel
42 int tid
= blockIdx
.x
; // handle the data at this index
44 d_c
[tid
] = gpu_operations(d_a
[tid
],d_b
[tid
]);
48 // Defining host arrays
49 int h_a
[N
], h_b
[N
], h_c
[N
];
50 // Defining device pointers
52 // allocate the memory
53 CHECK(hipMalloc((void**)&d_a
, N
* sizeof(int)));
54 CHECK(hipMalloc((void**)&d_b
, N
* sizeof(int)));
55 CHECK(hipMalloc((void**)&d_c
, N
* sizeof(int)));
57 // Initializing Arrays
58 for (int i
= 0; i
< N
; i
++) {
63 // Copy input arrays from host to device memory
64 CHECK(hipMemcpy(d_a
, h_a
, N
* sizeof(int), hipMemcpyHostToDevice
));
65 CHECK(hipMemcpy(d_b
, h_b
, N
* sizeof(int), hipMemcpyHostToDevice
));
67 // Calling kernels with N blocks and one thread per block, passing
68 // device pointers as parameters
69 hipLaunchKernelGGL(gpu_kernel_operations
, dim3(N
), dim3(1 ), 0, 0, d_a
, d_b
, d_c
);
71 // Copy result back to host memory from device memory
72 CHECK(hipMemcpy(h_c
, d_c
, N
* sizeof(int), hipMemcpyDeviceToHost
));
74 // Printing result on console
75 for (int i
= 0; i
< N
; i
++) {
76 if (h_c
[i
] != h_a
[i
]*2 + h_b
[i
]*3 + 1) {
77 fprintf(stderr
, "ERROR: wrong result of %d element is %d*2 + %d*3 + 1 = %d\n",
78 i
, h_a
[i
], h_b
[i
],h_c
[i
]);
This page took 0.033057 seconds and 4 git commands to generate.