assembly - “冗余”将操作移动到 Cuda 中的同一寄存器

Question

我正在查看 CUDA SASS 代码，我注意到对相同寄存器的大量移动操作。前任：

172           MOV R3, R3;
173           MOV R4, R4;
174           MOV R3, R3;
175           MOV R4, R4;
176           MOV R4, R4;
177           MOV R3, R3;
178           MOV R4, R4;

我只是好奇，这些移动操作的目的是什么？是为了时机，他们的行为就像'nop'还是不太明显？

注意：这些不包含在 PTX 代码中，仅包含在 SASS 中。I 和假设在操作之间使用，而不是在 PTX 操作期间使用。虽然，就上下文而言，PTX 是：

85  .reg .u32 a, b; 
86  bfind.s64 a, %rd37; 

实际c++代码是上面 ptx 的内联汇编。

更新：sm_52在Visual Studio 2013compute_52中x64使用 Cuda Toolkit 7.5 编译。设备：GTX 970（麦克斯韦 GP​​U）。

从 切换Debug到Release消除了这种低效率。