CUDA 内核在 2 个不同的 GPU 上抛出不同的结果(GeForce 8600M GT vs Quadro FX 770M)

Question

我一直在开发 AES CUDA 应用程序，我有一个在 GPU 上执行 ECB 加密的内核。为了确保算法的逻辑在并行运行时不被修改，我发送了一个由 NIST 提供的已知输入测试 vector ，然后从主机代码将输出与 NIST 提供的带有断言的已知测试 vector 输出进行比较。 我已经在我的 NVIDIA GPU 上运行了这个测试，它是一个 8600M GT。这是在 Windows 7 下运行，驱动程序版本为 3.0。在这种情况下，一切正常，断言成功。

现在，当应用程序在 Quadro FX 770M 上运行时。启动相同的应用程序，发送相同的测试 vector ，但获得的结果不正确，断言失败!!。这在具有相同驱动程序版本的 Linux 上运行 内核由 256 个线程执行。在内核中，为了跳过算法，使用了 256 个元素的预计算查找表。这些表最初加载到全局内存中，启动内核的 256 个线程中的 1 个线程协作加载查找表的 1 个元素，并将该元素移动到共享内存中的新查找表中，从而减少访问延迟。

最初，我考虑过由于 GPU 之间的时钟速度差异而导致的同步问题。因此可能是线程正在使用仍未加载到共享内存中的值或以某种方式仍未处理的值，从而使输出 搞砸了，最后弄错了。

在这里声明了已知的测试 vector ，所以基本上它们被发送到负责设置内核的 AES_set_encrption

void test_vectors ()
{ 

  unsigned char testPlainText[]  = {0x6b, 0xc1, 0xbe, 0xe2, 0x2e, 0x40, 0x9f, 0x96, 0xe9, 0x3d, 0x7e, 0x11, 0x73, 0x93, 0x17, 0x2a}; 
     unsigned char testKeyText[] =  {0x60, 0x3d, 0xeb, 0x10, 0x15, 0xca, 0x71, 0xbe, 0x2b, 0x73, 0xae, 0xf0, 0x85, 0x7d, 0x77,0x1f, 0x35, 0x2c, 0x07, 0x3b, 0x61, 0x08, 0xd7, 0x2d, 0x98, 0x10, 0xa3, 0x09, 0x14, 0xdf, 0xf4};
     unsigned char testCipherText[] = {0xf3, 0xee, 0xd1, 0xbd, 0xb5, 0xd2, 0xa0, 0x3c, 0x06, 0x4b, 0x5a, 0x7e, 0x3d, 0xb1, 0x81, 0xf8};

 unsigned char out[16] = {0x0};
     //AES Encryption
AES_set_encrption( testPlainText, out, 16, (u32*)testKeyText);

 //Display encrypted data
 printf("\n  GPU Encryption: "); 
 for (int i = 0; i < AES_BLOCK_SIZE; i++)
         printf("%x", out[i]);

 //Assert that the encrypted output is the same as the NIST testCipherText vector 
 assert (memcmp (out, testCipherText, 16) == 0);
}

在这里setup函数负责分配内存，调用内核并将结果返回给hos。请注意，我在发送回主机之前进行了同步，因此此时一切都应该完成，这让我认为问题出在内核中。

__host__ double AES_set_encrption (... *input_data,...*output_data, .. input_length, ... ckey )

 //Allocate memory in the device and copy the input buffer from the host to the GPU
  CUDA_SAFE_CALL( cudaMalloc( (void **) &d_input_data,input_length ) ); 
  CUDA_SAFE_CALL( cudaMemcpy( (void*)d_input_data, (void*)input_data, input_length, cudaMemcpyHostToDevice ) ); 

     dim3 dimGrid(1);
     dim3 dimBlock(THREAD_X,THREAD_Y); // THREAD_X = 4 & THREAD_Y = 64
  AES_encrypt<<<dimGrid,dimBlock>>>(d_input_data);

     cudaThreadSynchronize();

     //Copy the data processed by the GPU back to the host 
  cudaMemcpy(output_data, d_input_data, input_length, cudaMemcpyDeviceToHost);

  //Free CUDA resources
  CUDA_SAFE_CALL( cudaFree(d_input_data) );
}

最后在内核中我计算了一组 AES 轮次。因为我当时认为同步问题出在内核中，所以我设置了 __syncthreads();在每一轮或计算操作之后，以确保所有线程同时移动，因此不会评估未计算的值。但这仍然没有解决问题。

这是我使用 8600M GT GPU 时的输出，效果很好:

AES 256 位 key

NIST 测试 vector :

纯文本:6bc1bee22e409f96e93d7e117393172a

key :603deb1015ca71be2b73aef0857d7781

密文:f3eed1bdb5d2a03c64b5a7e3db181f8

GPU 加密:f3eed1bdb5d2a03c64b5a7e3db181f8

测试状态:通过

这是我使用 Quadro FX 770M 失败的情况!!

AES 256 位 key NIST 测试 vector :

纯文本:6bc1bee22e409f96e93d7e117393172a

key :603deb1015ca71be2b73aef0857d7781

密文:f3eed1bdb5d2a03c64b5a7e3db181f8

GPU 加密:c837204eb4c1063ed79c77946893b0

通用断言 memcmp (out, testCipherText, 16) == 0 已引发错误

测试状态:失败

为什么 2 个 GPU 即使处理相同的内核也会计算出不同的结果？？？ 我将不胜感激你们中的任何人可以给我的任何提示或故障排除或任何步骤以解决此问题

提前致谢!

Answer 1

免责声明:我对 AES 加密一无所知。

你使用 double 吗？您可能知道，但可以肯定的是 - 我相信您使用的两张卡都是计算能力 1.1，不支持 double 。也许卡或平台以不同的方式转换为单精度......？有人知道吗？说实话，IEEE 浮点偏差有明确规定，所以我会感到惊讶。