cuda 内核启动

C++ 使用 __global__ 声明说明符定义内核，并使用新的 <<<...>>> 执行配置语法指定内核调用的 CUDA 线程数（请参阅 C++ 语言扩展）。 每个执行内核的线程都有一个唯一的线程 ID，可以通过内置变量在内核中访问。

示例代码使用内置变量 threadIdx 将两个大小为 N 的向量 A 和 B 相加，并将结果存储到向量 C 中：

#include <cuda_runtime_api.h>
#include <iostream>

#define RANDOM(x) (rand() % x)
#define MAX 10

// single block multiple threads
__global__ void vector_add_gpu_2(int *d_a, int *d_b, int *d_c, int n){
	int tid = threadIdx.x;
	const int t_n = blockDim.x;
	while(tid < n){
		d_c[tid] = d_a[tid] + d_b[tid];
		tid+=t_n;
	}
}

int main(){

	/***向量相加的实现***/
	int n = 5;
	int *a = (int *)malloc(sizeof(int)*n);
	int *b = (int *)malloc(sizeof(int)*n);
	int *c = (int *)malloc(sizeof(int)*n);

	for (size_t i = 0; i<n; i++){
		a[i] = RANDOM(MAX);
		b[i] = RANDOM(MAX);
		std::cout << a[i] << "   " << b[i] << std::endl;
	}

    cudaError_t  cudaStatus;

	// GPU memory allocate
	int *d_a, *d_b, *d_c;
	cudaMalloc((void **)&d_a, sizeof(int)*n);
	cudaMalloc((void **)&d_b, sizeof(int)*n);
	cudaMalloc((void **)&d_c, sizeof(int)*n);

	// data a and b copy to GPU
	cudaStatus = cudaMemcpy(d_a, a, sizeof(int)*n, cudaMemcpyHostToDevice);
	if (cudaStatus != cudaSuccess) {
        std::cout << ("Memory copy failed! error code: %s", cudaGetErrorString(cudaStatus)) << std::endl;
    }
	cudaStatus = cudaMemcpy(d_b, b, sizeof(int)*n, cudaMemcpyHostToDevice);
	if (cudaStatus != cudaSuccess) {
        std::cout << ("Memory copy failed! error code: %s", cudaGetErrorString(cudaStatus)) << std::endl;
    }

	vector_add_gpu_2<<<1, 3>>>(d_a, d_b, d_c, n);

	// result copy back to CPU
	cudaMemcpy(c, d_c, sizeof(int)*n, cudaMemcpyDeviceToHost);
	std::cout << "the result of add is: " << std::endl;

	for (size_t i = 0; i<n; i++){
		std::cout << c[i] << " ";
	}

	std::cout << std::endl;
	// GPU memory free
	cudaFree(d_a);
	cudaFree(d_b);
	cudaFree(d_c);
	free(a);
	free(b);
	free(c);

	return 0;
}

思考：

1 VecAdd 被执行几次？

2 怎么被调用的？

3 有多少个线程在执行？线程间数据相互影响吗？