cuda做卷积计算初稿

#include "cuda_runtime.h"
#include "device_launch_parameters.h"
#include <iostream>
using namespace std;
//返回thread和block
int getThreadNum()
{
	cudaDeviceProp prop;//cudaDeviceProp的一个对象
	int count = 0;//GPU的个数
	cudaGetDeviceCount(&count);
	std::cout << "gpu 的个数：" << count << '\n';

	cudaGetDeviceProperties(&prop, 0);//第二参数为那个gpu
	cout << "最大线程数：" << prop.maxThreadsPerBlock << endl;
	cout << "最大网格类型：" << prop.maxGridSize[0] << '\t' << prop.maxGridSize[1] << '\t' << prop.maxGridSize[2] << endl;
	return prop.maxThreadsPerBlock;
}
__global__ void conv(float *imgGpu, float*kernelGpu, float*resultGpu, int width, int height, int kernelSize)
{
	int id = threadIdx.x + blockIdx.x*blockDim.x;
	if (id >= width * height)
	{
		return;
	}
	int row = id / width;//获取img 的行和列
	int clo = id / height;
	//每一个线程处理一次卷积计算
	
	for (int i = 0; i < kernelSize; ++i)
	{
		for (int j = 0; j < kernelSize; ++j)
		{
			float imgValue = 0;//记录结果
			//imgValue += kernelGpu[i*kernelSize + j] * imgGpu[id];
			int curRow = row - kernelSize / 2 + i;
			int curClo = clo - kernelSize / 2 + j;
			if (curRow < 0 || curClo < 0||curRow>=height||curClo>=width)
			{}
			else
			{
				//imgValue += kernelGpu[i*kernelSize + j] * imgGpu[(curRow + i - 1)*width + curClo + j - 1];
				imgValue = imgGpu[curRow*width + curClo];
			}
			resultGpu[id] += kernelGpu[i*kernelSize + j] * imgValue;
		}
	}
}
int main()
{
	//定义一个1080p照片
	const int width = 1920;
	const int height = 1080;
	float *img = new float[width*height];
	//赋值
	for (int row=0; row < height; ++row)
	{
		for (int col = 0; col < width; ++col)
		{
			img[col + row * width] = (col + row) % 255;
		}
	}
	//声明卷积核大小,大小为3*3
	const int kernelSize = 3;
	float*kernel = new float[kernelSize*kernelSize];
	//卷积核赋值
	//第一种方法
	//for (int i = 0; i < kernelSize; ++i)
	//{
	//	for (int j = 0; j < kernelSize; ++j)
	//	{
	//		kernel[j + i * kernelSize] = i - 1;
	//	}
	//}
	//第二种
	for (int i = 0; i < kernelSize*kernelSize; ++i)
	{
		kernel[i] = i % kernelSize - 1;
	}
	//输出img的左上角
	for (int row=0; row < 10; ++row)
	{
		for (int col = 0; col < 10; ++col)
		{
			std::cout << img[col + row * width] << '\t';
		}
		std::cout << '\n';
	}
	for (int i = 0; i < kernelSize*kernelSize; ++i)
	{
		std::cout << kernel[i] << '\t';
	}

	float *imgGpu = 0;//将host值复制到device上面
	float *kernelGpu = 0;//将kernel也复制到device上
	float *resultGpu = 0;//卷积结果
	cudaMalloc(&imgGpu, height*width * sizeof(float));
	cudaMalloc(&kernel, kernelSize*kernelSize * sizeof(float));
	//这个地方捕捉错误，明天改
	cudaMemcpy(imgGpu, img, width*height*sizeof(float), cudaMemcpyHostToDevice);
	cudaMemcpy(kernelGpu, kernel, kernelSize*kernelSize*sizeof(float), cudaMemcpyHostToDevice);
	const int threadNum = getThreadNum();
	const int blockNum = (width*height + threadNum - 1) / threadNum;//这里block使用一维
	conv << <blockNum, threadNum >> > (imgGpu, kernelGpu, resultGpu, width, height, kernelSize);
	float *showImg = new float[height*width];
	cudaMemcpy(showImg, resultGpu, width*height * sizeof(float), cudaMemcpyDeviceToHost);
	return 0;
}