如何避免 Cuda 竞争条件，以并行比较哈希图中的值

Question

我有一个问题，在 2D 空间 (x,y) 中我有很多不同大小的点。具有结构 (x,y, size) 的数组表示此空间。我想将此空间下采样到某个特定的网格，其中 x 和 y 舍入到网格中最近的单元格，然后我将最大尺寸保留在单元格中。我的想法是使用哈希映射，其中保留从 2D 索引映射的 1D 索引，保留哈希映射中的最大大小（键：1d 索引，值：最大大小），然后基于此哈希映射创建输出。这是我的内核中的代码，我在其中找到了最大的大小：

  //nx: is the grid size in x
  //dx/dy: is the cell size in x/y
  //points: is the structure keeping the 2d points with the size as (x,y,size)
  //minBounds: has the minimum values of x and y

  int i = blockDim.x * blockIdx.x + threadIdx.x
  std::unordered_map<uint64_t, float> indexSizeMap;

  // Find point index in the grid and flatten it to 1d index
  auto xi = static_cast<uint64_t>(round((points[i].x - minBounds.x) / dx));
  auto yi = static_cast<uint64_t>(round((points[i].y - minBounds.y) / dy));
  uint64_t index1d = to1dIndex(xi, yi, nx + 1); //mapping 2d index to 1d index

  // If the index in the voxel grid is vacant, fill it. Otherwise keep the larger size
  if (indexSizeMap.count(index1d))
  {
     if (indexSizeMap[index1d] < points[i].size)
     {
        indexSizeMap[index1d] = points[i].size;
     }
  }
  else
  {
     indexSizeMap[index1d] = points[i].size;
  }

现在我发现这部分代码可能存在竞争条件，使其具有不确定性，因为可能有两个线程同时填充相同的索引：

  if (indexSizeMap.count(index1d))
  {
     if (indexSizeMap[index1d] < points[i].size)
     {
        indexSizeMap[index1d] = points[i].size;
     }
  }
  else
  {
     indexSizeMap[index1d] = points[i].size;
  }

查看有关竞争条件的文档，我看到了很多处理 +/- 操作的方法，但我没有找到如何处理这种情况。您知道在这种情况下避免竞争条件的方法吗？或者您建议另一种实现吗？

Answer 1

解决方案是另一种实现： 由于在 Cuda 内核中使用哈希映射对我来说效果不佳，因此我最终用与我的点大小相同的数组替换了哈希映射。 我将索引保留在那里。第二个 CUDA 内核（2d）用于删除具有更大尺寸的另一个点的点。所以我最终会得到我需要的点。