使用OpenACC+OpenCV时pgc++编译报错：duplicate部分存在

错误原因分析 #

报错FATAL ERROR: variable in data clause is partially present on the device: name=duplicate的核心原因是：
cv::Mat是复合对象，包含两部分：小体积的元数据（行数、列数、内存步长等，约96字节），以及指向实际像素数据的指针。OpenACC默认只会自动复制Mat对象本身，不会识别并复制其底层的像素数据缓冲区。当你在GPU并行区域中尝试访问duplicate.at<uchar>(i,j)时，设备端仅存在Mat的元数据，无法找到实际的像素内存，因此触发“部分存在”的错误。

解决步骤 #

1. 显式指定像素数据的传输 #

使用OpenACC的数据子句，直接指定要复制的底层像素数据缓冲区及其大小。注意要用duplicate.step计算每行的实际字节数（而非直接用cols，因为OpenCV会做内存对齐，step可能大于cols*通道数），确保覆盖所有像素内存。

2. 避免在设备端调用Mat成员函数 #

Mat::at()是主机端的成员函数，在GPU设备端调用可能引发未定义行为。建议直接通过指针访问像素数据，同时还能提升性能。

3. 选择合适的数据传输子句 #

• 仅需将主机数据传到设备：用copyin
• 仅需将设备数据传回主机：用copyout
• 双向传输（先传设备，修改后传回主机）：用copy

修改后的核心代码片段 #

// 替换原有的并行循环部分
#pragma acc parallel loop collapse(2) copy(duplicate.data[:vrows * duplicate.step])
for(int i=0;i<vrows;i++){
    // 获取当前行的起始指针
    uchar* row_ptr = duplicate.data + i * duplicate.step;
    for(int j=0;j<vcols;j++){
        // 直接通过指针修改像素值
        row_ptr[j] = 10;
        // 如果需要使用grayScale的数据，需同时复制grayScale的缓冲区：
        // row_ptr[j] = *(grayScale.data + i * grayScale.step + j);
    }
}

如果需要同时使用grayScale的数据，需在数据子句中添加copyin(grayScale.data[:vrows * grayScale.step])，确保GPU能访问到灰度图的像素数据。

可选：提前管理数据生命周期 #

如果需要多次使用Mat的数据，可以用enter data和exit data提前将数据传到设备，避免重复复制：

// 在创建duplicate和grayScale后，提前传入设备
#pragma acc enter data copyin(duplicate.data[:vrows * duplicate.step], grayScale.data[:vrows * grayScale.step])

// 并行区域使用present子句，表明数据已在设备上
#pragma acc parallel loop collapse(2) present(duplicate.data[:vrows * duplicate.step], grayScale.data[:vrows * grayScale.step])
for(int i=0;i<vrows;i++){
    uchar* dup_row = duplicate.data + i * duplicate.step;
    uchar* gray_row = grayScale.data + i * grayScale.step;
    for(int j=0;j<vcols;j++){
        dup_row[j] = gray_row[j];
    }
}

// 程序结束前释放设备上的数据
#pragma acc exit data delete(duplicate.data, grayScale.data)

编译注意事项 #

编译时需指定OpenACC目标设备，并链接OpenCV库：

pgc++ -acc=gpu -Minfo=acc your_code.cpp -o your_exec `pkg-config opencv4 --cflags --libs`

• -acc=gpu：指定编译为GPU目标
• -Minfo=acc：输出OpenACC编译信息，可查看数据传输的细节，帮助调试
• pkg-config opencv4 --cflags --libs：自动获取OpenCV的编译和链接参数（若使用OpenCV3则改为opencv3）

内容的提问来源于stack exchange，提问作者Lance