2

float[]OpenCL を使用して GPU 上の配列の縮小 (最小値と最大値を見つける) を行っています。

ワークグループごとに、いくつかの要素をメモリからglobalメモリにロードしています。localグローバル サイズがワークグループ サイズの倍数でない場合は、グローバル サイズの倍数になるように、グローバル サイズをパディングします。配列の終わりを過ぎた作業項目は、リダクションのニュートラル要素をlocalメモリに入れます。

しかし、そのニュートラルな要素は何のためにあるべきmax()でしょうか? 最大の機能? OpenCL ドキュメントではMAXFLOAT、HUGE_VALFおよびINFINITYが非常に大きな正の (または符号なし)float値として提供されます。ニュートラル要素を例にするのは理にかなっています-INFINITYか？

現在HUGE_VALF、中立要素としてを使用していますmin()が、ドキュメントではそれHUGE_VALFがエラー値として使用されているとも書かれているため、それは悪い考えかもしれません。

リダクション カーネル (コード):

#define NEUTRAL_ELEMENT HUGE_VALF
#define REDUCTION_OP min

__kernel void reduce(__global float* weights,
                     __local float* weights_cached
                    )
{
  unsigned int id = get_global_id(0);

  // Load data
  if (id < {{ point_count }}) {
    weights_cached[get_local_id(0)] = weights[id];
  } else {
    weights_cached[get_local_id(0)] = NEUTRAL_ELEMENT;
  }

  barrier(CLK_LOCAL_MEM_FENCE);

  // Reduce
  for(unsigned int stride = get_local_size(0) / 2; stride >= 1; stride /= 2) {
    if (get_local_id(0) < stride) {
      weights_cached[get_local_id(0)] = REDUCTION_OP(weights_cached[get_local_id(0)], weights_cached[get_local_id(0) + stride]);
    barrier(CLK_LOCAL_MEM_FENCE);
  }

  // Save
  weights[get_group_id(0)] = weights_cached[0];
}

編集: 私は実際にニュートラル要素としてfmin()and をfmax()一緒に使用することになりました-数値が常に返されるため、これは基本的にOpenCLドキュメントNANに従って動作することが保証されています( は2つの値が指定された場合にのみ返されます)。NANNAN

4

1 に答える 1

2

OpenCL 標準の引用:

HUGE_VALF は +infinity に評価されます。

HUGE_VALFしたがって、 とINFINITY;の使用には (暗黙の意図を除いて) 実質的な違いはありません。minどちらも削減のために正しく機能します。INFINITY明快さという点では、概念的にはエッジケースの返品を意図しているため、私は を少し好みHUGE_VALFますが、これはそうではありません。

同様に、削減-INFINITYのために使用します。max

MAX_FLOAT配列に無限大が含まれている場合、ニュートラル要素として正しく動作しません。

于 2013-01-23T13:28:20.730 に答える