2017-11-22 18 views
1

1Dデータセットはセグメントに分割され、各ワークアイテムは1つのセグメントを処理します。それはセグメントから要素の数を読みましたか?要素の数はあらかじめ分かっておらず、セグメントごとに異なります。OpenCLの部分和の計算

例えば

:すべてのセグメントが処理された後

+----+----+----+----+----+----+----+----+----+  <-- segments 
    A BCD E FG HIJK L M  N  <-- elements in this segment 

彼らはelements連続に出力メモリを書き込む必要があり、そう

A B C D E F G H I J K L M N 

一方からの元素の絶対出力位置等セグメントは、前のセグメントの要素数に依存します。 Eは、セグメントが1要素(A)を含み、セグメント2が3要素を含むので位置4にある。


OpenCLのカーネルは、共有/ローカルメモリバッファに各セグメントの要素の数を書き込み、この(擬似コード)のように動作

kernel void k(
    constant uchar* input, 
    global int* output, 
    local int* segment_element_counts 
) { 
    int segment = get_local_id(0); 
    int count = count_elements(&input[segment * segment_size]); 

    segment_element_counts[segment] = count; 

    barrier(CLK_LOCAL_MEM_FENCE); 

    ptrdiff_t position = 0; 
    for(int previous_segment = 0; previous_segment < segment; ++previous_segment) 
     position += segment_element_counts[previous_segment]; 

    global int* output_ptr = &output[position]; 
    read_elements(&input[segment * segment_size], output_ptr); 
} 

したがって、各作業項目は、部分和を計算しなければなりませんより大きいidを持つ作業項目がより多くの反復を行うループを使用します。

これを実装するより効率的な方法がありますか(各作業項目は、そのインデックスまでのシーケンスの部分和を計算します)、OpenCL 1.2ではありますか? OpenCL 2はこれにwork_group_scan_inclusive_addを提供するようです。このようなあなたはLOG2にN部分(接頭辞)の合計を行うことができます

答えて

関連する問題