CUDA编程 矩阵相乘kernel函数

#include 
#include "cuda_runtime.h"
#include "device_launch_parameters.h"
using namespace std;

__global__ void MatMul(float* x, int x_h, int x_w, float* w, int w_h, int w_w, float* out)
{
	int row = threadIdx.y + blockIdx.y*blockDim.y;
	int col = threadIdx.x + blockIdx.x*blockDim.x;
	if (row < x_h&&col < w_w)//对row和col的范围进行限制
	{
		float temp = 0.0;
		for (int i = 0; i < x_w; i++)
		{
			temp += x[row*x_w + i] * w[i*w_w + col];
		}
		out[row*w_w + col] = temp;
	}
}
int main()
{
	const int insize = 28*28;
	const int outsize = 10;
	float w[insize*outsize] = {};
	float x[insize] = {};
	float out[outsize] = {};
	for (int i = 0; i < insize*outsize; i++) { w[i] = 1.0; }
	for (int i = 0; i < insize; i++) { x[i] = 1.0; }

	float* d_w = nullptr;
	float* d_x = nullptr;
	float* d_out = nullptr;
	cudaMalloc(&d_w, sizeof(float) * insize*outsize);
	cudaMemcpy(d_w, w, sizeof(float) * insize*outsize, cudaMemcpyHostToDevice);
	cudaMalloc(&d_x, sizeof(float) * insize);
	cudaMemcpy(d_x, x, sizeof(float) * insize, cudaMemcpyHostToDevice);
	cudaMalloc(&d_out, sizeof(float) * outsize);
	cudaMemcpy(d_out, out, sizeof(float) * outsize, cudaMemcpyHostToDevice);

	dim3 blockSize(32, 32);
	dim3 gridSize((insize + blockSize.x - 1) / blockSize.x,
		(insize + blockSize.y - 1) / blockSize.y);
	// 执行kernel
	MatMul << < gridSize, blockSize >> > (d_x,1,insize,d_w,insize,outsize,d_out);
	cudaMemcpy(out, d_out, sizeof(float) * outsize, cudaMemcpyDeviceToHost);
	for (int i = 0; i < outsize; i++)
	{
		cout << out[i] << endl;
	}
	return 0;
}