HIP - Hello World!

Your first HIP program

HIP is an awesome C++ Runtime API and Kernel Language that allows developers to create portable application. In this lesson we will see how the HIP dialect can be used to add to arrays a[] and b[].

The CPU-only code (gold)

Let's first show you the CPU-only code with no HIP involved or what-sp ever! We call this as the gold code and we use it to match the reults. This code adds the following two arrays a[] and c[] onto the GPU:

// array init
    int a[]{1, 2, 3, 4, 5};
    int b[]{6, 7, 8, 9, 10};

The HIP/GPU code

Now let's modify the code to execute with HIP/ GPU. We aim to perform the addition opertion on the GPU, instead of CPU. The figure below shows how it works.

The idea is simple, we first init the arrays on the host, copy the arrays to the device, note the use of the HIP kernel function hipLaunchKernelGGL() which performs the addition operation on the GPU:

// HIP Kernel
__global__
void kernel(const int* pA, const int* pB, int* pC) {
    const auto gidx = blockIdx.x * blockDim.x + threadIdx.x;

    pC[gidx] = pA[gidx] + pB[gidx];
}

Then returns the result into the host array c[] shows the output.

7 9 11 13 15 

If you don't understand the HIP threads, blocks and grids concepts as yet, don't worry! Well get there shortly.

Source code credit: Github/ROCm-Developer-Tools/HIP-CPU.