C++调用cuda实现gpu并行运算百万数据截取和插值操作

2022年11月9日 21:01 ry 71

最近3个星期都没有更新文章了,花全部心思为了完成上面布置的任务了,需求是对指纹数据进行插值和截取操作,实现算法优化,之前的代码是直接暴力循环的,对于扫描指纹和出图整个过程花费时间25秒左右,和目前最快的速度13秒左右相差太大,对于指纹识别商用肯定不切实际的,因此打算开始优化,首先开始测试每个步骤所花费的时间,经过师兄的测试,发现插值和截取所花费的时间占大头,因此导师把这个任务交给了我,我瑟瑟发抖,毕竟对底层c++加速了解好少,谁叫是研究僧呢,肯定啥都要研究,话不多说,开干,我开始首先配置好cuda环境,这个好弄,我之前用python的tensorflow框架就搞好了,然后就是vscode配置cuda文件代码运行环境了,因为.cu后缀文件支持c++代码的编写了,可以直接调用cuda命令运行,那就好办了,先了解下cuda中gpu并行的知识,一般我们代码没特别函数或者声明都是默认再cpu运行的,在cuda中,有个__global__函数,也叫核函数,是在gpu上线程中并行执行的函数,注意的是这个函数没有返回值,一般类型都设置为void,先看下以下代码

#include<stdio.h>
#include<stdlib.h>
#include<cuda_runtime_api.h>
#include<iostream>
__global__ void kennel(int arr[3][3],int n)
{
    int i = threadIdx.x+blockDim.x*blockIdx.x;
    int j = threadIdx.y+blockDim.y*blockIdx.y;
   
    for(i;i<3;i+=gridDim.x * blockDim.x)
    {
        for(j;j<3;j+=gridDim.y * blockDim.y)
        {
            printf("i:->%d,j:->%d...%d\n",i,j,arr[i][j]);
            
            
        }
    }
}
int main()
{
   int  N = 9;
   float aaa = 3/10;
   int datas[3][3] = {{1,2,3},{4,5,6},{7,8,9}};
   size_t size = N * sizeof(int);
    int (*arr)[3];//定义指针指向n行3列的数组
    cudaMallocManaged(&arr,N*sizeof(int));
    cudaMemcpy(arr,datas,size,cudaMemcpyHostToDevice);
    kennel<<<3,3>>>(arr,N);
    

    cudaDeviceSynchronize();
    return 0;
}

看上去很复杂,其实就是对3行3列的二维数组进行3*3线程并行输出,使用命令运行下,如图所示

PS C:\Users\14499\Desktop\vscodeProject> nvcc .\demo3.cu -o .\demo3
demo3.cu
./demo3.cu : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(825) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(1760) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(2622) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(3461) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(4404) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(5302) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(6213) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(7094) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(7893) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt/device_functions.h : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt/device_functions.h(773) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt/device_functions.h(1629) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失        
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\device_double_functions.h : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失       
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\sm_20_intrinsics.h : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\sm_20_intrinsics.h(926) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
./demo3.cu(10): warning: expression has no effect

./demo3.cu(12): warning: expression has no effect

./demo3.cu : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(825) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(1760) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(2622) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(3461) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(4404) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(5302) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(6213) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(7094) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(7893) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt/device_functions.h : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt/device_functions.h(773) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt/device_functions.h(1629) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失        
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\device_double_functions.h : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失       
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\sm_20_intrinsics.h : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\sm_20_intrinsics.h(926) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
./demo3.cu(10): warning: expression has no effect

./demo3.cu(12): warning: expression has no effect

c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(825) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(1760) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(2622) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(3461) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(4404) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(5302) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(6213) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(7094) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
c:\program files\nvidia gpu computing toolkit\cuda\v10.0\include\crt\math_functions.h(7893) : warning C4819: 该文件包含不能在当前代码页(936)中表示的字符。请将该文件保存为 Unicode 格式以防止数据丢失
   正在创建库 ./demo3.lib 和对象 ./demo3.exp
PS C:\Users\14499\Desktop\vscodeProject> .\demo3
i:->0,j:->0...1
i:->1,j:->0...4
i:->2,j:->0...7
i:->0,j:->1...2
i:->1,j:->1...5
i:->2,j:->1...8
i:->0,j:->2...3
i:->1,j:->2...6
i:->2,j:->2...9

警告可以忽略,这里注意调用核函数要用<<<线程块数,每个线程块对应的线程数>>>(参数一,参数二)表示,除此之外要严格区分cpu上的变量和gpu中的变量,gpu和cpu中的变量不可直接互相调用,对于gpu,必须动态分配gpu内存来存储变量,然后将cpu中的变量拷贝到gpu中,这样就可以实现cpu和gpu之间变量的传递了。还有线程块*线程数就是总线程数了,这里代码中的gridDim.x为3,blockDim.x也为3.总线程就是9个。了解了这个知识后,就开始实现截取和插值的算法优化了,直接上代码,

#include<stdio.h>
#include<fstream>
#include<cuda_runtime_api.h>
#include<iostream>
#include <string>
#include <cstdlib>
#include<cstring>

using namespace std;

int datas[1024000];
int newData[500*1500];
float binData[4096];
//插值
float allendDatas[500*4096];
//读取bin文件
void readBin()
{
  //打开标定文件
      std::ifstream inF("3001800cali.bin", std::ios::binary);
    inF.read((char*)binData, sizeof(float) * (4096));
    inF.close();
    for(int i=0;i<4096;i++)
    {
      printf("%f\n",binData[i]); 
    }
}

//读取txt数据
void readData(int N)
{
  	ifstream infile("100.txt",ios::in);
	  
    int i=0;
 
    
    char data[10];
    while(infile.getline(data, 10))
    {
      int a = atoi(data);
      datas[i] = a;
      printf("this is %d nums:%d\n",i,a);
      
      i++;
    }
}


// CUDA 截取核函数,
__global__ void Speedjiequ(int *p,int *news,int start,int end) {
  int i;
  int j;
  int new_index;
  int index;
  int idx;
  int idy;
for(idy=blockIdx.y * blockDim.y + threadIdx.y;idy<500;idy += blockDim.y * gridDim.y)
{


    for(idx=blockIdx.x *blockDim.x + threadIdx.x;idx<2048;idx+= blockDim.x * gridDim.x)
    {
      
        
          
          if(idx>=start && idx<end)
          {
            index = idy*2048+idx;
            new_index = index+idy*(end-start)-idy*2048-start;
            news[new_index] = p[index-1];
            // printf("new_index:%d\n",new_index);
            // printf("********idy:%d,idx:%d***value is:%d***index:%d\n",idy,idx,p[index-1],index-1);
          }
            
              
         
          
            
        
    }
  
}

  
}
//cuda 插值核函数
__global__ void Speedchazhi(int * pp,float * endDatas,float *BinData)
{
                                        
  int idx;
  int idy;
  float k;
  float b;
 
for(idy=blockIdx.y * blockDim.y + threadIdx.y;idy<500;idy += blockDim.y * gridDim.y)
{


    for(idx=blockIdx.x *blockDim.x + threadIdx.x;idx<4096;idx+= blockDim.x * gridDim.x)
    {
        //binData为一维数组;
      
        
          float bin_index = BinData[idx];
          
          if ((int)bin_index < 1499)
          {
             k = pp[(int)(1500*idy+bin_index+1)]-pp[(int)(1500*idy+bin_index)];
             b = pp[(int)(bin_index+1500*idy)]- k*(int)(bin_index+1500*idy);
            //  printf("k is %f,b is %f\n",k,b);
             endDatas[idy*4096+idx] = k*(bin_index+1500*idy)+b;
          }
          else
          {
            k = pp[(int)(bin_index+1500*idy)]-pp[(int)(bin_index-1+1500*idy)];
            b = pp[(int)(bin_index-1+1500*idy)]-k*(int)(bin_index-1+1500*idy);
            endDatas[idy*4096+idx] = k*(bin_index+1500*idy)+b;
          } 
          
        
        
        
        
        
    }
}
}


int main() {
  //设置显卡设备
  readBin();
    cudaSetDevice(0);

  int N = 1024000;
  int n = 500*1500;
  int *gpuDatas;
  int *newGpuDatas;
  //插值数据
  float *endDatas;
  int *newImgs;
  float *gpuBinData;
  // int *newImgs;
  time_t start,stop,start1,stop1,start2,stop2;
  size_t size = N * sizeof(int);
  size_t size1 = n * sizeof(int);
  size_t size2 = 2*N * sizeof(float);
  size_t size3 = 4096* sizeof(float);
 //截取数据
  cudaMallocManaged(&gpuDatas, size); // 为a分配CPU和GPU空间
  cudaMallocManaged(&newGpuDatas, size1); // 为a分配CPU和GPU空间
  //插值数据
  cudaMallocManaged(&endDatas, size2);
  cudaMallocManaged(&newImgs, size1);
  cudaMallocManaged(&gpuBinData, size3);
  // 将数据从内存拷贝到显存
  start = time(NULL);
  readData(N); // 为数组a赋值
  stop = time(NULL);
  printf("read txt_time is:%d\n",stop-start);
  //截取数据拷贝到显存
  cudaMemcpy(gpuDatas,datas,size,cudaMemcpyHostToDevice);
  cudaMemcpy(newGpuDatas,newData,size1,cudaMemcpyHostToDevice);
  //插值数据拷贝到显存
  cudaMemcpy(endDatas,allendDatas,size2,cudaMemcpyHostToDevice);
  cudaMemcpy(gpuBinData,binData,size3,cudaMemcpyHostToDevice);
  // cudaMemcpy(newImgs,imgs,size1,cudaMemcpyHostToDevice);

  start1 = time(NULL); 
//截取函数
  Speedjiequ<<<1000,1024>>>(gpuDatas,newGpuDatas,300,1800); // 执行核函数
  stop1 = time(NULL);
  printf("gpu_time_is:%d\n",stop1-start1);
  // 将截取数据从显存拷贝到内存并循环输出
  cudaMemcpy(newData,newGpuDatas,n*sizeof(int),cudaMemcpyDeviceToHost);
  // ofstream out("jiequ.txt");
  // for(int i=0;i<500;i++)
  // {
  //   for(int j=0;j<1500;j++)
  //   {
  //     out << newData[i*1500+j] << endl;
  //     printf("alfter jiequ data is:---%d-----\n",newData[i*1500+j]);
   
  //   }
 
  // }
  //插值核函数运行
  cudaMemcpy(newImgs,newData,size1,cudaMemcpyHostToDevice);
  start2 = time(NULL); 
  Speedchazhi<<<2000,1024>>>(newImgs,endDatas,gpuBinData);
  stop2 = time(NULL);
  printf("chazhi_time_is:%d\n",stop1-start1);
   //将插值数据从显存拷贝到内存并打印
  cudaMemcpy(allendDatas,endDatas,size2,cudaMemcpyDeviceToHost);
  ofstream out("chazhi.txt");
  for(int i=0;i<500;i++) 
  {

    for(int j=0;j<4096;j++)
    {
      
      out << allendDatas[i*4096+j] << endl;
      printf("alfter chazhi data is:xxxxxxx%fxxxxxxx\n",allendDatas[i*4096+j]);
  
    }

  }
  out.close();
  
  cudaDeviceSynchronize(); // 同步


//释放所有显存开辟的空间
  cudaFree(gpuDatas); 
  cudaFree(newGpuDatas);
  cudaFree(endDatas);
  cudaFree(gpuBinData);
  cudaDeviceReset();
  printf("gpu_time_is:%d\n",stop1-start1);

  
  return 0;
}

3个星期的结果,我也写吐了,差点放弃了,记录实现后发现对百万数据进行操作,时间基本为0,效率直接提升,原来之前的所用时间为12秒,现在直接节省12秒,只能说gpu并行牛皮,c++牛皮,多谢导师鼓励和给我锻炼的机会,多谢师兄的耐心详细的流程讲解,本次仅为记录,我是菜鸡一枚。

欢迎发表评论~

点击此处登录后即可评论


评论列表
暂时还没有任何评论哦...

赣ICP备2021001574号-1

赣公网安备 36092402000079号