cuda数组的拷贝

原文链接

简单描述一下自己要做的事情:(1)CPU三维的vector--->(2)CPU三维数组--->(3)转换到GPU中的三维数组--->(4)转换到CPU中的三维数组,而其中问题主要出在第3、4步。

主要是没有理解一个问题,那就是“cuda的各种拷贝一定要是内存连续的”。而自己在申请三维数组的时候用的是new或者malloc,这种在申请一维数组的时候是连续的,但是在申请多维数组就会出现不连续,因此在这里犯了致命错误。

http://hpcbbs.it168.com/thread-7366-1-1.html这个帖子给了很好的建议,“vector<vector<float> > 并不是二维数组吧,它只是实现了二维数组的操作(比如[][]).内存是不连续的。要用cudaMemcpy还是得定义 float 2darray[N][M] 或者 直接 float *2darray = new float(M*N);”。反正就是这样,纸上得来终觉浅,自己多亲身力为一下。

 1 #include "example1.cuh"
 2 #include "Struct.h"
 3 /************************************************************************/
 4 /* 转换成设备可以识别的                                                 */
 5 /************************************************************************/
 6 void InitCPUData(DataMatrix &datamatrix,std::vector<std::vector<std::vector<float > > > vec3D1,
 7                  std::vector<std::vector<std::vector<float > > > vec3D2,int width,int height,int depth)
 8 {
 9     int i,j,k;
10     for (i=0;i<depth;i++)
11     {
12         for (j=0;j<height;j++)
13         {
14             for (k=0;k<width;k++)
15             {
16                 datamatrix.Mat3D1[i][j][k]=vec3D1[i][j][k];
17                 datamatrix.Mat3D2[i][j][k]=vec3D2[i][j][k];
18             }
19         }
20     }
21 }
22
23 /************************************************************************/
24 /* 分配并且赋值                                                         */
25 /************************************************************************/
26 __host__ void AllocDataAndVal(DataStruct &datastruct,DataMatrix datamatrix,int width,int height,int depth)
27 {
28     //分配内存
29     cudaExtent extent=make_cudaExtent(sizeof(float)*6,7,8);
30     cutilSafeCall(cudaMalloc3D(&(datastruct.Vec3D1),extent));
31     cutilSafeCall(cudaMalloc3D(&(datastruct.Vec3D2),extent));
32     //赋值
33     cudaMemcpy3DParms Parms3D1={0};
34     cudaMemcpy3DParms Parms3D2={0};
35     Parms3D1.dstPtr=datastruct.Vec3D1;
36     Parms3D2.dstPtr=datastruct.Vec3D2;
37     Parms3D1.srcPtr=make_cudaPitchedPtr((void*)datamatrix.Mat3D1,width*sizeof(float),width,height);
38     Parms3D2.srcPtr=make_cudaPitchedPtr((void*)datamatrix.Mat3D2,width*sizeof(float),width,height);
39     Parms3D1.extent=extent;
40     Parms3D2.extent=extent;
41     Parms3D1.kind=cudaMemcpyHostToDevice;
42     Parms3D2.kind=cudaMemcpyHostToDevice;
43     cudaMemcpy3D(&Parms3D1);
44     cudaMemcpy3D(&Parms3D2);
45 }
46
47
48 /************************************************************************/
49 /* 核函数                                                               */
50 /************************************************************************/
51 __global__ void kernel(DataStruct datastruct,int width,int height,int depth) //实现类中两个数组的相加,保持到第一个数组中
52 {
53     char* devPtr1=(char*)datastruct.Vec3D1.ptr; //起始地址
54     char* devPtr2=(char*)datastruct.Vec3D2.ptr;
55     int pitch=datastruct.Vec3D1.pitch; //pitch,相当于宽度
56     int SlicePitch=pitch*height;
57     //用线程
58     int xid=threadIdx.x;
59     int yid=threadIdx.y;
60     int zid=threadIdx.z;
61     if (xid<width&&yid<height&&zid<depth)
62     {
63         ((float*)((char*)(devPtr1+zid*SlicePitch)+yid*pitch))[zid]=((float*)((char*)(devPtr1+zid*SlicePitch)+yid*pitch))[zid]+
64             ((float*)((char*)(devPtr2+zid*SlicePitch)+yid*pitch))[zid];
65     }
66 }
67
68 /************************************************************************/
69 /* 返回到主机上                                                         */
70 /************************************************************************/
71 __host__ void GPU2CPU(DataStruct &datastruct,DataMatrix datamatrix, int width,int height,int depth)
72 {
73     cudaExtent extent=make_cudaExtent(sizeof(float)*6,7,8);
74     cudaMemcpy3DParms Parms3D1={0};
75     cudaMemcpy3DParms Parms3D2={0};
76     Parms3D1.srcPtr=datastruct.Vec3D1;
77     Parms3D2.srcPtr=datastruct.Vec3D2;
78     Parms3D1.dstPtr=make_cudaPitchedPtr((void*)datamatrix.Mat3D1,width*sizeof(float),width,height);
79     Parms3D2.dstPtr=make_cudaPitchedPtr((void*)datamatrix.Mat3D2,width*sizeof(float),width,height);
80     Parms3D1.extent=extent;
81     Parms3D2.extent=extent;
82     Parms3D1.kind=cudaMemcpyDeviceToHost;
83     Parms3D2.kind=cudaMemcpyDeviceToHost;
84     cudaMemcpy3D(&Parms3D1);
85     cudaMemcpy3D(&Parms3D2);
86
87 }

主函数:

 1 // 说明:在cu中host和device的虽然写在一起,但是是分开编译的,这个在一起只是形式上的。如果函数前面有__global__由主机调用设备执行,__device__设备调用设备执行,__host__主机调用主机执行。其分别对应三种形式为核函数、核函数中的函数、一般函数。
 2
 3 #include <iostream>
 4 #include <vector>
 5 #include <algorithm>
 6 #include "example1.cuh"
 7 #include "Struct.h"
 8
 9 int main()
10 {
11     int i,j,k;
12     int width=6;
13     int height=7;
14     int depth=8;
15     std::vector<std::vector<std::vector<float > > > vec3D1(width); //建立6*7*8的三维数组,范文depth-height-width
16     std::vector<std::vector<std::vector<float > > > vec3D2(width);
17
18     vec3D1.resize(depth);
19     vec3D2.resize(depth);
20     for (i=0;i<depth;i++)
21     {
22         vec3D1[i].resize(height);
23         vec3D2[i].resize(height);
24             for (j=0;j<height;j++)
25             {
26                 vec3D1[i][j].resize(width);
27                 vec3D2[i][j].resize(width);
28                 for (k=0;k<width;k++)
29                 {
30                     vec3D1[i][j][k]=i+j+k;
31                     vec3D2[i][j][k]=i*j*k;
32                 }
33             }
34     }
35
36     //////////////////////////////////////////////////////////////////////////
37     //将数据转换成设备可以接受的形式,为赋值做准备,这个是在主机上进行
38     DataMatrix datamatrix;
39     InitCPUData(datamatrix,vec3D1,vec3D2,width,height,depth);
40
41     //////////////////////////////////////////////////////////////////////////
42     //给设备分配内存并且赋值,这个是在设备上进行
43     DataStruct datastruct;
44     AllocDataAndVal(datastruct,datamatrix,width,height,depth);
45
46     //////////////////////////////////////////////////////////////////////////
47     //调用核函数
48     dim3 dimBlock(8,7,6);
49     kernel<<<1,dimBlock>>>(datastruct,width,height,depth);
50
51     //////////////////////////////////////////////////////////////////////////
52     //返回到主机,并显示出来
53     GPU2CPU(datastruct,datamatrix,width,height,depth);
54     for (i=0;i<depth;i++)
55     {
56         for (j=0;j<height;j++)
57         {
58             for (k=0;k<width;k++)
59             {
60                 printf("%f  ",datamatrix.Mat3D1[i][j][k]);
61             }
62             printf("\n");
63         }
64             printf("\n");
65             printf("\n");
66     }
67
68
69     //释放空间
70     cudaFree(&(datastruct.Vec3D1));
71     cudaFree(&(datastruct.Vec3D2));
72
73 }
时间: 2025-01-14 06:38:46

cuda数组的拷贝的相关文章

PHP 数组的拷贝是按值传递 or 按引用传递

在记忆中 PHP 简单变量的拷贝是按值传递,数组和对象的拷贝是按引用传递,即通过引用来实现. 简单变量和对象好理解: <?php // 简单变量的拷贝 $a = 'human'; $b = $a; $b = 'cat'; var_dump($a); // string 'human' (length=5) // 对象的拷贝 class A{} $a = new A(); $b = $a; $b->name = 'jack'; var_dump($a); // object(A)[1] publ

CUDA零内存拷贝 疑问考证

今天思考了一下CUDA零内存拷贝的问题,感觉在即将设计的程序中会派上用场,于是就查了一下相关信息. 以下是一些有帮助的链接: cuda中的零拷贝用法--针对二维指针 cuda中的零拷贝用法--针对一维指针 cuda零拷贝用法-二维结构体指针 浅谈CUDA零拷贝内存 经过调查发现,零拷贝技术适用于集中计算.较少内存拷贝次数的问题.比如向量点积.求和运算等问题. 既然零拷贝技术是在CPU上开辟内存空间,GPU可以直接访问该空间,那么我就产生了一个疑问:"如果CPU上开辟的空间大于GPU的可用空间的时

Java学习笔记(for-each循环与数组的拷贝)

for-each循环 //第一次循环,k=data[0]; //第二次循环,k=data[1]; //... for(int k:data) //data是数组 { xxxxx; } 数组的拷贝 int a;//a是所有者 int[]a=new int[100];//a是管理者 int[]b=a;//b是管理者 a与b共同管理一个数组(改变b(a)[x]的值即改变a(b)[x]的值) 因此把数组a拷贝给数组b的方法如下: int[]a=new int[xx]; int[]b=new int[a.

javascript数组操作(创建、元素删除、数组的拷贝)

这篇文章主要介绍了javascript数组操作,包括创建.元素的访问.元素删除.数组的拷贝等操作,还有其它示例,需要的朋友可以参考下 1.数组的创建 复制代码 代码如下: var arrayObj = new Array(); //创建一个数组var arrayObj = new Array([size]); //创建一个数组并指定长度,注意不是上限,是长度var arrayObj = new Array([element0[, element1[, ...[, elementN]]]]); 创

数组 复制 拷贝 相等判断

js 数组的拷贝(不影响原数组),数组相等判断 数组拷贝 数组属于引用类型:简单的赋值只是添加了一个指向数组的指针:ex: var a = [1,2,3]; var b = a; b.push(2); console.log(a)//[1,2,3,2] 对新数组 b 的操作同样会改变原始数组 a 那么如何实现独立的拷贝?介绍下面两种方法:两种方法性能上相差不大,不同浏览器内核上各有千秋: //方法1 var a = [1,2,3]; var b = a.slice(); a.reverse; c

C++动态申请二维数组与拷贝构造函数

一.C++动态申请二维数组 在C++中不能直接动态申请二维数组,经过一番搜索,发现一种动态申请二维数组较好的方法. 代码如下(MATRIX_TYPE为某一种类型,Lines和Columns): MATRIX_TYPE** elem; //C++二维矩阵动态申请空间 elem = new MATRIX_TYPE*[Lines]; elem[0] = new MATRIX_TYPE[Lines * Columns]; for(int i = 1; i < Lines; i++) elem[i] =

js数组的拷贝赋值复制二三事总结

今天在看React-native性能优化的时候,看到如何避免shouldComponentUpdate的异常数据时,脑内一阵风暴,从而牵连出一连串的问题,于是有了这一篇关于js数组的复制(深浅拷贝)与赋值等为何能产生异常数据的文章. 有什么问题欢迎指正 现在进入正题: 首先异常数据的产生在于我们在复制赋值时,会有或没有改变到本身的值. 一.push与concat push的定义是:像数组末尾添加一个或更多元素,并返回新的长度.该方法会改变数组的长度. concat的定义是:连接两个或更多的数组,

js 数组的拷贝

在js中,数组Array是引用类型,直接将数组赋值给一个变量名,二者所指向的地址是一样的. 所以直接复制数组会产生意想不到的结构. 要想解决拷贝一个数组但是对副本的修改不影响原来的数组,有以下方式: //方法1 var a = [1,2,3]; var b = a.slice(); a.reverse; console.log(a);//[3,2,1] console.log(b);//[1,2,3] //方法2 var c = [4,5,6]; var d = c.concat(); c.re

CUDA数组分配

原问链接 概述:数组分配可以通过cudaMallocArray()和cudaMalloc3DArray() 1.cudaMallocArray() cudaError_t cudaMallocArray ( struct cudaArray ** array, const struct cudaChannelFormatDesc * desc, size_t width, size_t height = 0, unsigned int flags = 0 ) 例:2DArray分配 1 cuda