-2

N 個のスレッドを (1 つのブロックで) 起動する必要がある

これはコードです。'e' は 1024b の大きな数字です。GPUにコピーして、少しずつ読み取る必要があります。

ホスト コード:

unsigned char *__e;
BIGNUM *e = BN_new();
unsigned char exp[128];

//      e
i = cudaMalloc( (void**)&__e, 128* sizeof(unsigned char) );
if(i != cudaSuccess)
    printf("cudaMalloc __e FAIL! Code: %d\n", i);

BN_bn2bin128B(e, exp);   // copy data in exp

for(i=0; i<128; i++)
    exp[i] = reverse(exp[i]);

i = cudaMemcpy( __e, exp, 128* sizeof(unsigned char), cudaMemcpyHostToDevice);
if(i != cudaSuccess)
    printf("cudaMemcpy __e FAIL! Code: %d\n", i);


unsigned char reverse(unsigned char b) {
 b = (b & 0xF0) >> 4 | (b & 0x0F) << 4;
 b = (b & 0xCC) >> 2 | (b & 0x33) << 2;
 b = (b & 0xAA) >> 1 | (b & 0x55) << 1;
 return b;
}

デバイスコード:

for(int i=0; i<1024; i++)
    if(ISBITSET(__e, i) == 1)
        //do something

ヘッダ：

#define ISBITSET(x,i) ((x[i>>3] & (1<<(i&7)))!=0)

残念ながら、ISBITSET は __e とは異なるものを受け入れないため、__e 自体でさらに値を確認することはできません。

どうすれば解決できますか？それとももっと良い方法がありますか？

4

1 に答える 1

2

GPU は 32 ビット マシンであるため、一度に 8 ではなく 32 ビットずつ 1024 ビットを処理する必要がunsigned charありunsigned intます。

GPU には、一度に 32 ビットを反転するための高速 PTX 命令があるため、GPU に実装することをお勧めします。命令は と呼ばれbrevます。これを使用するには、次のようなインライン PTX を追加します (未テスト)。

asm("brev.b32 %0, %1;" : "=r"(dst_var) : "r"(src_var));

詳細については、NVIDIA のドキュメント「Using Inline PTX Assembly In CUDA」を参照してください。

for(int i=0; i<1024; i++)
    if(ISBITSET(__e, i) == 1)
        //do something

このコードには、パフォーマンス上の問題がある可能性があります。ビットがオンになっている可能性が 50% あると仮定すると、スレッドの半分が待機する必要があり、残りの半分が//do something. 私は回避策を考えることはできません。ループする代わりにスレッドを起動することもできます。

残念ながら、ISBITSET は __e とは異なるものを受け入れないため、__e 自体でさらに値を確認することはできません。

詳しく教えていただけますか？マクロはISBITSET私には問題ないように見え、符号なし文字の任意の配列を処理できるように見えます__e。

于 2012-07-18T23:37:35.587 に答える