CUDA cudaMemcpy2D的用法

斑竹您好,我想将100100的图像 复制给设备上 但是设备上图大小是102102,周围边框都是0 填充的、cudaMemcpy2D可以实现但是我不知道怎么用。

希望 高手解答下啊 别打击新手的积极性

我的想法是 先在device上建一个102102的0矩阵,再将CPU上的100100的图像复制进去。但是cudaMemcpy2D怎么去用我还是不明白

楼主您好:

直接用cudaMemcpy2D无法完成任务的,只所以这么晚回复你,是因为你上来就用了特别肯定的语气,要求一定必须这样给你实现,这让我很为难,经过考虑,给您两种变通的方式:

(1)cudaMemcpy2D完成复制前100行,每行后面可以指定destination pitch为102 * sizeof(元素类型), 这样可以完成复制前100行,并自动在每行后面填充上两个元素。最后两行则需要您手工复制两行0上去。
您需要考虑手册没有阐明cudaMemcpy2D自动填充的是是否(一般是0),所以请谨慎使用此方式。

而推荐的方式则是:
(2)使用texture(或者surface)访问 + 指定cudaBoundaryModeZero,这样可以有效的自动将您的越界(x < 0 或 y < 0 或 x >= 100 或 y >= 100)的值返回0,相当于自动在周围环绕了一圈0.
这样可以完成您的目标,而且无任何不清晰的地方,我建议您考虑此建议。

当楼主考虑了上文的两个建议后,建议您下次发文不要直接说,我要求必须给我按照某某某方式的答案。这会让论坛很为难的。

感谢您的周末来访。


代为修正个别错字。

好斑竹啊~很耐心~受教了

您客气了,服务您是我们的荣幸。

感谢您的来访。

斑竹您好,非常感谢您的解答。我认真看了您给我的两个建议。可是我还是对第二种方法不是很懂。这是我最近在研究的一个利用CUDA对hog加速的一个程序,其中的关于从CPU端复制数据到GPU上函数(并周边填充hPaddingSizeX、hPaddingSizeY大小的0数据):
host void PadHostImage(uchar4* registeredImage, float4 *paddedRegisteredImage,
int minx, int miny, int maxx, int maxy)
{
hWidthROI = maxx - minx;
hHeightROI = maxy - miny;

int toaddxx = 0, toaddxy = 0, toaddyx = 0, toaddyy = 0;

if (avSizeX) { toaddxx = hWidthROI * marginX / avSizeX; toaddxy = hHeightROI * marginY / avSizeX; }
if (avSizeY) { toaddyx = hWidthROI * marginX / avSizeY; toaddyy = hHeightROI * marginY / avSizeY; }

hPaddingSizeX = max(toaddxx, toaddyx); hPaddingSizeY = max(toaddxy, toaddyy);

hPaddedWidth = hWidthROI + hPaddingSizeX2;
hPaddedHeight = hHeightROI + hPaddingSizeY
2;

cutilSafeCall(cudaMemset(paddedRegisteredImageU4, 0, sizeof(uchar4) * hPaddedWidth * hPaddedHeight));

cutilSafeCall(cudaMemcpy2D(paddedRegisteredImageU4 + hPaddingSizeX + hPaddingSizeY * hPaddedWidth,
hPaddedWidth * sizeof(uchar4), registeredImage + minx + miny * hWidth,
hWidth * sizeof(uchar4), hWidthROI * sizeof(uchar4),
hHeightROI, cudaMemcpyHostToDevice));

Uchar4ToFloat4(paddedRegisteredImageU4, paddedRegisteredImage, hPaddedWidth, hPaddedHeight);
}
可是,我对cudaMemcpy2D()第一个参数和第三个参数的使用很费解。

代码出自[Prisacariu V,2009]fastHOG - a real-time GPU implementation of HOG

楼主您好,这不是您自己的代码吧?又是您抄的?

您这代码通过使用cudaMemset进行预清零(这样末尾两行就不用复制了)。
而您的第一个参数是目标指针,第三个参数是源指针。
分别从paddedRegisteredU4和registeredImage中:
前者越过hPaddingSizeX + hPaddingSizeY * hPaddedWidth个元素,作为起始写入位置;
后者越过minx + miny * hWidth个元素,作为起始读取位置。

而为何具体要越过这些个元素,则可能和您的算法要求有关,这里(您的这段代码)无法看出。

谢谢斑竹,这段代码是 fastHOG-a real-time GPU implementation of HOG这篇paper提供的代码中的一部分~“楼主您好,这不是您自己的代码吧?又是您抄的?”这句话如果去掉的话,我觉得您的回答就非常完美了

LZ您好:

我觉得您引用别人的代码而不给出引用出处,甚至并不表明这是引用他人的代码,这个是十分不严谨的治学态度,而玫瑰斑竹此时对您的评价的每一个字都是合理而中肯的,都是完美的,去掉那句话才会变得不完美。

如果您觉得玫瑰斑竹的这句话不合适,请全部无视玫瑰斑竹的建议,因为它们是一体的。

祝您好运~

斑竹您说的是~~