利用显卡完成排序算法（CUDA下的秩排序rankSort）

现在排序算法都是基于CPU的，但是CPU有它的并发局限性

GPU需要处理大量三角形数据和渲染数据，并行计算能力非常强，那么为什么不可以利用GPU的超强并行能力完成排序

（笔者曾经看过FPGA编程里面的排序算法，直接利用大规模并行电路直接完成排序，GPU和FAPG居然有一些相似）

为了追求极致的效率为什么不可以把显卡调用起来取完成排序呢 

想到就去做，下载cuda，可以对cuda完成面向显卡的编程，按教程安装好环境好，发现运行失败，后面发现是显卡算力的问题，查了资料以后解决了

VS2022搭建cuda环境的no kernel image is available for execution on the device问题解决_星空_MAX的博客-CSDN博客

环境为VS2022和cuda11.6

原理

比如给出5个无序数字

4，3，1，2，5

开启五个并行线程

分别统计每个数字比它小的数字有几个

4比较完4，3，1，2，5发现小于等于的数字有4个，那么把新数组的第4个位置放置为4

这是的新数组为：

空 空 空 4 空

3比较完4，3，1，2，5发现比自己小的数字有3个，那么把新数组的第3个位置放置为3

这是的新数组为：

空 空 3 4 空

五个数字分别在五个并行线程中找到了自己的位置：

最后的数组为：

1，2，3，4，5

以上就完成了秩排序

代码

#include <stdio.h>
#include <iostream>
#include <cuda.h>
#include <cuda_runtime.h>
//#include <cuda_runtime_api.h>
#include <device_launch_parameters.h>

#ifndef __CUDACC__
#define __CUDACC__
#endif

#define ARRAY_SIZE 6
#define THREADS_PER_BLOCK 6

__global__ void rankSort(int* d_a, int* d_b)
{
	int count = 0;
	int tid = threadIdx.x;
	int ttid = threadIdx.x + blockIdx.x * blockDim.x;

	int val = d_a[ttid];
	__shared__ int cache[THREADS_PER_BLOCK];
	for (int i = tid; i < ARRAY_SIZE; i += THREADS_PER_BLOCK)
	{
		cache[tid] = d_a[i];

		//等待当前块所有线程完成共享内存数据填充
		__syncthreads();

		//统计共享内存中小于当前值得数量
		for (int j = 0; j < THREADS_PER_BLOCK; j++)
		{
			if (val > cache[j])
			{
				count++;
				//等待所有线程完成统计工作
				__syncthreads();
			}
		}
	}

	d_b[count] = val;
}

int main()
{
	//定义主机(CPU)和设备（GPU）变量
	int h_a[ARRAY_SIZE] = { 5, 9, 3, 4, 1, 8 };
	int h_b[ARRAY_SIZE];
	int* d_a, * d_b;

	//分配设备内存
	cudaMalloc(&d_a, ARRAY_SIZE * sizeof(int));
	cudaMalloc(&d_b, ARRAY_SIZE * sizeof(int));

	//拷贝数据：从CPU到GPU
	cudaMemcpy(d_a, h_a, ARRAY_SIZE * sizeof(int), cudaMemcpyHostToDevice);

	//调用内核函数
	rankSort << <ARRAY_SIZE / THREADS_PER_BLOCK, THREADS_PER_BLOCK >> > (d_a, d_b);  //从这里传入两个数组

	//等待内核函数运行完成
	cudaDeviceSynchronize();

	//拷贝数据：从GPU到CPU
	cudaMemcpy(h_b, d_b, ARRAY_SIZE * sizeof(int), cudaMemcpyDeviceToHost);

	//打印秩排序结果
	printf("result:");
	for (int i = 0; i < ARRAY_SIZE; i++)
	{
		printf("%d ", h_b[i]);
	}

	//释放GPU内存
	cudaFree(d_a);
	cudaFree(d_b);

	system("pause");
	return 0;
}

 可以发现排序成功了