Commit | Line | Data |
---|---|---|
ec980634 | 1 | /* Copyright (C) 2019-2020 Free Software Foundation, Inc. |
2 | Copyright (C) 2019-2020 Advanced Micro Devices, Inc. All rights reserved. | |
3 | ||
4 | This file is part of GDB. | |
5 | ||
6 | This program is free software; you can redistribute it and/or modify | |
7 | it under the terms of the GNU General Public License as published by | |
8 | the Free Software Foundation; either version 3 of the License, or | |
9 | (at your option) any later version. | |
10 | ||
11 | This program is distributed in the hope that it will be useful, | |
12 | but WITHOUT ANY WARRANTY; without even the implied warranty of | |
13 | MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the | |
14 | GNU General Public License for more details. | |
15 | ||
16 | You should have received a copy of the GNU General Public License | |
17 | along with this program. If not, see <http://www.gnu.org/licenses/>. */ | |
18 | ||
820c00e6 QS |
19 | #include <stdio.h> |
20 | #include <iostream> | |
21 | #include "hip/hip_runtime.h" | |
22 | ||
23 | #define CHECK(cmd) \ | |
24 | { \ | |
25 | hipError_t error = cmd; \ | |
26 | if (error != hipSuccess) { \ | |
27 | fprintf(stderr, "error: '%s'(%d) at %s:%d\n", hipGetErrorString(error), error, \ | |
28 | __FILE__, __LINE__); \ | |
29 | exit(EXIT_FAILURE); \ | |
30 | } \ | |
31 | } | |
32 | ||
33 | // Defining number of elements in Array | |
34 | #define N 5 | |
35 | ||
36 | __device__ int add_one(int no) { | |
37 | return no + 1; | |
38 | } | |
39 | ||
40 | __device__ int multiply_two(int no) { | |
41 | no = no*2; | |
42 | return no; | |
43 | } | |
44 | ||
45 | __device__ int multiply_three(int no) { | |
46 | no = no*3; | |
47 | return no; | |
48 | } | |
49 | ||
50 | //c = a*2 + b*3 + 1; | |
51 | __device__ int gpu_operations(int a,int b) { | |
52 | int c; | |
53 | c = multiply_two(a) + multiply_three(b); | |
54 | c = add_one(c); | |
55 | return c; | |
56 | } | |
57 | ||
58 | __global__ void gpu_kernel_operations(int *d_a, int *d_b, int *d_c) { | |
59 | // Getting block index of current kernel | |
60 | int tid = blockIdx.x; // handle the data at this index | |
61 | if (tid < N) | |
62 | d_c[tid] = gpu_operations(d_a[tid],d_b[tid]); | |
63 | } | |
64 | ||
65 | int main(void) { | |
66 | // Defining host arrays | |
67 | int h_a[N], h_b[N], h_c[N]; | |
68 | // Defining device pointers | |
69 | int *d_a, *d_b, *d_c; | |
70 | // allocate the memory | |
71 | CHECK(hipMalloc((void**)&d_a, N * sizeof(int))); | |
72 | CHECK(hipMalloc((void**)&d_b, N * sizeof(int))); | |
73 | CHECK(hipMalloc((void**)&d_c, N * sizeof(int))); | |
74 | ||
75 | // Initializing Arrays | |
76 | for (int i = 0; i < N; i++) { | |
77 | h_a[i] = i; | |
78 | h_b[i] = i + 1; | |
79 | } | |
80 | ||
81 | // Copy input arrays from host to device memory | |
82 | CHECK(hipMemcpy(d_a, h_a, N * sizeof(int), hipMemcpyHostToDevice)); | |
83 | CHECK(hipMemcpy(d_b, h_b, N * sizeof(int), hipMemcpyHostToDevice)); | |
84 | ||
85 | // Calling kernels with N blocks and one thread per block, passing | |
86 | // device pointers as parameters | |
87 | hipLaunchKernelGGL(gpu_kernel_operations, dim3(N), dim3(1 ), 0, 0, d_a, d_b, d_c); | |
88 | ||
89 | // Copy result back to host memory from device memory | |
90 | CHECK(hipMemcpy(h_c, d_c, N * sizeof(int), hipMemcpyDeviceToHost)); | |
91 | ||
92 | // Printing result on console | |
93 | for (int i = 0; i < N; i++) { | |
94 | if (h_c[i] != h_a[i]*2 + h_b[i]*3 + 1) { | |
95 | fprintf(stderr, "ERROR: wrong result of %d element is %d*2 + %d*3 + 1 = %d\n", | |
96 | i, h_a[i], h_b[i],h_c[i]); | |
97 | exit(EXIT_FAILURE); | |
98 | } | |
99 | } | |
100 | // Free up memory | |
101 | CHECK(hipFree(d_a)); | |
102 | CHECK(hipFree(d_b)); | |
103 | CHECK(hipFree(d_c)); | |
104 | return 0; | |
105 | } | |
106 |