【问题标题】:C rotl32 alternativeC rotl32 替代品
【发布时间】:2017-02-11 16:08:24
【问题描述】:

C 语言中rotl32 有替代方案吗?

我发现了这个:Near constant time rotate that does not violate the standards 但仍在尝试获得优化的

我的代码:

k0 = rotl32 ((k3 ^ k2 ^ k ^ k0), 1u)

【问题讨论】:

标签: c opencl


【解决方案1】:

我认为这是最好的便携选择:

uint32_t rotl32(uint32_t var, uint32_t hops)
{
    return (var << hops) | (var >> (32 - hops));
}

【讨论】:

  • 这不适用于hops == 0。无需测试即可修复:return (var &lt;&lt; hops) | (var &gt;&gt; ((32 - hops) &amp; 31));
  • 什么会导致失败?我们向右移动出界,移动从左边输入 0。在 gcc 4.8.5 中工作正常吗?
  • 如果hops == 0var &gt;&gt; (32 - hops) 具有未定义的行为。在 x86 处理器上,它会计算 var &gt;&gt; 0,但在其他处理器上,它可能会计算其他东西,例如 var &gt;&gt; 32,它的计算结果可能为 0,这也可以,但 C 标准不保证这些是唯一可能的行为(很遗憾)。事实上,一切皆有可能,因此必须避免这种极端情况。
【解决方案2】:

您的问题中有 opencl 标记,因此使用内核

__kernel void rotateGpu(__global unsigned int * a,__global unsigned int * b)
{
  int idx = get_global_id(0);
  unsigned int a0=a[idx];
    for(int i=0;i<100;i++)
        a0=rotate(a0,1280u);
  b[idx] = rotate(a0,1280u);

}   

rotate 在 R7-240 GPU 上的性能根据基准:

3200 万个 32b 无符号整数的元素数组,例如a0,内核执行需要 16ms,其中每个线程执行 100 次(10 ms 1 次)旋转 1280u 步长(因此延迟与步长无关)) .它的 200 多个 Gflops(但在整数上)达到了 gpu 的 %40 理论最大值。整数可能比浮点数更快(我想它们需要在移位后进行归一化)。

例子:

__kernel void rotateGpu(__global unsigned int * a,__global unsigned int * b)
{
  int idx = get_global_id(0);
  unsigned int a0=a[idx];

  b[idx] = rotate(a0,2u);

}   

输入:

        buf[0] = 80;
        buf[1] = 12;
        buf[2] = 14;
        buf[3] = 5 ;
        buf[4] = 70;

输出:

320 
48 
56 
20 
280

【讨论】:

  • 它的执行速度是否比 opencl 旋转功能快?
  • 但是我这里写的是opencl rotate函数,只是测试了一下。也许是编译器对其进行优化的能力。也许 nvidia 有不同的时间复杂度或延迟
  • 你能提供一个使用例子吗?
  • 添加了一个向左旋转 2 步长的示例
  • 没看懂,能不能像我上面的代码一样简单点?
【解决方案3】:

dromtrund 发布了一个很好的便携解决方案:

uint32_t rotl32(uint32_t var, uint32_t hops) {
    return (var << hops) | (var >> (32 - hops));
}

不幸的是,这个函数对hops == 0 有未定义的行为。在 x86 处理器上,只有 hops 的低位是重要的。可以通过这种方式强制执行此行为:

uint32_t rotl32(uint32_t var, uint32_t hops) {
    return (var << hops) | (var >> ((32 - hops) & 31));
}

这两个函数都使用 gcc 4.9 及更高版本、clang 3.5 及更高版本和 icc 17 编译为最佳代码,这可以通过 Godbolt's Compiler Explorer 进行验证。

John Regehr 在这个主题上有一个interesting blog article

【讨论】:

    猜你喜欢
    • 2010-12-10
    • 2013-09-02
    • 1970-01-01
    • 1970-01-01
    • 1970-01-01
    • 1970-01-01
    • 1970-01-01
    • 1970-01-01
    • 1970-01-01
    相关资源
    最近更新 更多