Vector Addition

#include <cuda_runtime.h>

// Kernel is launched as:  vectorAdd<<<ceil(n/256), 256>>>(a, b, c, n)
// gridDim.x  = ceil(n / 256)   (one block per 256 elements)
// blockDim.x = 256
// Thread ID:  int i = blockIdx.x * blockDim.x + threadIdx.x;
__global__ void vectorAdd(const float* a, const float* b, float* c, int n) {
    int tid_in_block = threadIdx.x + 
                       (threadIdx.y * blockDim.x) + 
                       (threadIdx.z * blockDim.x * blockDim.y);

    // 2. Find the 1D block ID within the 3D grid
    int block_id = blockIdx.x + 
                   (blockIdx.y * gridDim.x) + 
                   (blockIdx.z * gridDim.x * gridDim.y);

    // 3. Calculate total threads per block
    int threads_per_block = blockDim.x * blockDim.y * blockDim.z;

    // 4. Combine them for the global 1D thread ID
    int global_tid = tid_in_block + (block_id * threads_per_block);

    if(global_tid < n) {
        c[global_tid] = a[global_tid] + b[global_tid];
    }
}