OpenMP Offloading中map(alloc)的工作机制与使用规范问询

我需要在OpenMP Offloading中创建一个仅用于设备端的临时数组，全程不与宿主端交互，只在设备上初始化、读取。为了测试，我写了一个通过临时数组实现数组复制的示例代码：

#include <cstdio>

int main()
{
    const int count = 10;
    int * src = new int[count];
    int * tmp = new int[count];
    int * dst = new int[count];

    for(int i = 0; i < count; i++) src[i] = i;
    for(int i = 0; i < count; i++) printf(" %3d", src[i]); printf("\n");

    #pragma omp target map(to:src[0:count]) map(from:dst[0:count]) map(alloc:tmp[0:count])
    {
        for(int i = 0; i < count; i++) tmp[i] = src[i];
        for(int i = 0; i < count; i++) dst[i] = tmp[i];
    }

    for(int i = 0; i < count; i++) printf(" %3d", dst[i]); printf("\n");

    delete[] src;
    delete[] tmp;
    delete[] dst;

    return 0;
}

测试结果 #

这段代码在两个平台都能正常运行：

• NVIDIA平台：使用pgc++ -mp=gpu编译
• Intel GPU平台：使用icpx -fiopenmp -fopenmp-targets=spir64编译

但当我尝试不在宿主端分配tmp数组时，出现了兼容性问题：

• 将tmp设为int * tmp = nullptr：NVIDIA平台运行失败，Intel平台仍可运行
• 仅声明int * tmp不初始化：Intel平台运行失败
• 不声明tmp变量：编译失败（此点可理解）

我已通过设置OMP_TARGET_OFFLOAD=MANDATORY确保代码真正卸载到设备执行。

疑惑与核心问题 #

我原本理解map(alloc)会在设备上分配数组供核函数使用，而tmp完全不涉及宿主端操作，为什么宿主端必须分配或初始化该指针？

我想明确以下几点：

1. OpenMP Offloading中map(alloc)的确切要求是什么？
2. map(alloc)的工作机制是怎样的？
3. 如何正确使用map(alloc)实现纯设备端临时数组？

希望能提供示例代码，以及官方文档/标准中的相关说明（我查阅了OpenMP 5.2标准，说明不够详细；现有教程也未深入讲解）。

补充说明 #

我知道可以通过#ifdef处理宿主端tmp数组的分配，保证禁用OpenMP时代码仍可运行；也了解omp_target_alloc()这类手动内存管理方式，但希望通过map(alloc)实现需求。当前使用的编译器版本为pgc++ 22.2-0和icpx 2022.0.0.20211123。

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