muRAND 开发者指南

muRAND 是基于 MUSA 开发的高效伪随机数和准随机数生成器，适用于蒙特卡洛模拟、深度学习初始化、随机算法等场景。

概述

什么是 muRAND

muRAND（Pseudo-Random and Quasi-Random Numbers based on MUSA，随机数生成库）提供基于 MUSA 的随机数生成接口，支持在 MTGPU 上生成伪随机数和准随机数，适用于蒙特卡洛模拟、模型初始化和随机采样等场景。

关键特性

• 支持多种伪随机数和准随机数生成器。
• 支持均匀分布、正态分布和泊松分布等常用分布。
• 可与 MUSA stream 配合，实现设备侧随机数生成和异步数据传输。
• 适用于批量随机数生成和高吞吐模拟任务。

支持的生成器类型

生成器	说明	应用场景
MURAND_RNG_PSEUDO_MTGP32	Mersenne Twister for GPU	通用伪随机数
MURAND_RNG_PSEUDO_PHILOX4_32_10	Philox4x32 10 轮	高质量伪随机数
MURAND_RNG_QUASI_SOBOL32	Sobol 准随机数	蒙特卡洛积分

支持的分布

• 均匀分布 (Uniform): murandGenerateUniform()
• 正态分布 (Gaussian/Normal): murandGenerateNormal()
• 泊松分布 (Poisson): murandGeneratePoisson()

---

快速开始

生成均匀分布随机数

#include <musa_runtime.h>
#include <murand.h>
#include <stdio.h>
#include <stdlib.h>

int main() {
    int rand_n = 128;       // 生成 128 个随机数
    int seed = 777;         // 随机种子

    // 创建 stream
    musaStream_t stream;
    musaStreamCreateWithFlags(&stream, musaStreamNonBlocking);

    // 分配设备内存
    float *d_Rand;
    musaMalloc((void **)&d_Rand, rand_n * sizeof(float));

    // 创建随机数生成器
    murandGenerator_t prngGPU;
    murandCreateGenerator(&prngGPU, MURAND_RNG_PSEUDO_MTGP32);
    murandSetStream(prngGPU, stream);
    murandSetPseudoRandomGeneratorSeed(prngGPU, seed);

    // 分配主机 pinned 内存
    float *h_RandGPU;
    musaMallocHost(&h_RandGPU, rand_n * sizeof(float));

    printf("Generating random numbers on GPU...\n");
    
    // 生成均匀分布随机数 [0, 1)
    murandGenerateUniform(prngGPU, (float *)d_Rand, rand_n);

    // 复制结果回主机
    printf("Reading back the results...\n");
    musaMemcpyAsync(h_RandGPU, d_Rand, rand_n * sizeof(float), 
                    musaMemcpyDeviceToHost, stream);
    musaDeviceSynchronize();

    // 打印结果
    for(int i = 0; i < rand_n; i++) {
        printf("r[%d] = %f\n", i, h_RandGPU[i]);
    }

    // 释放资源
    murandDestroyGenerator(prngGPU);
    musaStreamDestroy(stream);
    musaFree(d_Rand);
    musaFreeHost(h_RandGPU);

    return EXIT_SUCCESS;
}

编译和运行

# 使用 mcc 编译
mcc -x musa rand_example.cpp \
    -L/usr/local/musa/lib \
    -I/usr/local/musa/include \
    -lmurand -lmusart \
    -o rand_example

# 运行
./rand_example

---

API 参考

生成器管理

创建生成器

murandStatus_t murandCreateGenerator(murandGenerator_t* generator,
                                      murandRngType_t rng_type);

参数:

参数	说明
generator	输出的生成器 handle
rng_type	随机数生成器类型

支持的 rng_type:

• MURAND_RNG_PSEUDO_MTGP32: Mersenne Twister for GPU
• MURAND_RNG_PSEUDO_PHILOX4_32_10: Philox4x32 10-round

销毁生成器

murandStatus_t murandDestroyGenerator(murandGenerator_t generator);

生成器配置

设置种子

murandStatus_t murandSetPseudoRandomGeneratorSeed(murandGenerator_t generator,
                                                   unsigned long long seed);

设置 Stream

murandStatus_t murandSetStream(murandGenerator_t generator,
                               musaStream_t stream);

设置偏移量

murandStatus_t murandSetGeneratorOffset(murandGenerator_t generator,
                                         unsigned long long offset);

随机数生成

均匀分布

murandStatus_t murandGenerateUniform(murandGenerator_t generator,
                                      float* output_ptr,
                                      size_t n);

生成 n 个均匀分布的浮点随机数，范围 [0, 1)。

murandStatus_t murandGenerateUniformDouble(murandGenerator_t generator,
                                            double* output_ptr,
                                            size_t n);

生成 n 个均匀分布的双精度浮点随机数，范围 [0, 1)。

正态分布

murandStatus_t murandGenerateNormal(murandGenerator_t generator,
                                     float* output_ptr,
                                     size_t n,
                                     float mean,
                                     float stddev);

生成 n 个正态分布随机数，均值为 mean，标准差为 stddev。

泊松分布

murandStatus_t murandGeneratePoisson(murandGenerator_t generator,
                                      unsigned int* output_ptr,
                                      size_t n,
                                      double lambda);

生成 n 个泊松分布随机数，参数为 lambda。

---

使用示例

生成正态分布随机数

#include <musa_runtime.h>
#include <murand.h>
#include <stdio.h>

int main() {
    const int n = 1024;
    const float mean = 0.0f;
    const float stddev = 1.0f;
    
    // 创建生成器
    murandGenerator_t generator;
    murandCreateGenerator(&generator, MURAND_RNG_PSEUDO_MTGP32);
    
    // 设置种子
    murandSetPseudoRandomGeneratorSeed(generator, 12345);
    
    // 分配设备内存
    float* d_output;
    musaMalloc(&d_output, n * sizeof(float));
    
    // 生成正态分布随机数
    murandGenerateNormal(generator, d_output, n, mean, stddev);
    
    // 复制结果回主机进行验证
    float* h_output = (float*)malloc(n * sizeof(float));
    musaMemcpy(h_output, d_output, n * sizeof(float), musaMemcpyDeviceToHost);
    
    printf("Generated %d normal distribution random numbers\n", n);
    printf("Mean = %.2f, StdDev = %.2f\n", mean, stddev);
    printf("First 10 values:\n");
    for (int i = 0; i < 10; i++) {
        printf("  %d: %f\n", i, h_output[i]);
    }
    
    // 清理
    murandDestroyGenerator(generator);
    musaFree(d_output);
    free(h_output);
    
    return 0;
}

批处理随机数生成

#include <musa_runtime.h>
#include <murand.h>
#include <vector>

int main() {
    const int n = 1024;
    const int batch = 32;
    
    // 创建生成器
    murandGenerator_t generator;
    murandCreateGenerator(&generator, MURAND_RNG_PSEUDO_PHILOX4_32_10);
    murandSetPseudoRandomGeneratorSeed(generator, 42);
    
    // 分配设备内存
    std::vector<float*> d_outputs(batch);
    for (int i = 0; i < batch; i++) {
        musaMalloc(&d_outputs[i], n * sizeof(float));
    }
    
    // 使用不同偏移量生成多批随机数
    for (int i = 0; i < batch; i++) {
        murandSetGeneratorOffset(generator, i * n);
        murandGenerateUniform(generator, d_outputs[i], n);
    }
    
    // 清理
    for (int i = 0; i < batch; i++) {
        musaFree(d_outputs[i]);
    }
    murandDestroyGenerator(generator);
    
    return 0;
}

使用 Philox 生成器

#include <musa_runtime.h>
#include <murand.h>
#include <stdio.h>
#include <vector>

int main() {
    // 使用 Philox4x32 10-round 生成器（更高质量）
    murandGenerator_t generator;
    murandCreateGenerator(&generator, MURAND_RNG_PSEUDO_PHILOX4_32_10);
    
    // 设置种子
    murandSetPseudoRandomGeneratorSeed(generator, 0x12345678);
    
    // 分配设备内存
    const int n = 10000;
    float* d_output;
    musaMalloc(&d_output, n * sizeof(float));
    
    // 生成均匀分布随机数
    murandGenerateUniform(generator, d_output, n);
    
    // 复制结果回主机
    std::vector<float> h_output(n);
    musaMemcpy(h_output.data(), d_output, n * sizeof(float), 
               musaMemcpyDeviceToHost);
    
    // 验证：计算平均值（应接近 0.5）
    float sum = 0.0f;
    for (int i = 0; i < n; i++) {
        sum += h_output[i];
    }
    printf("Average: %f (expected: ~0.5)\n", sum / n);
    
    // 清理
    murandDestroyGenerator(generator);
    musaFree(d_output);
    
    return 0;
}

蒙特卡洛积分示例

使用 Sobol 准随机数计算 π 值：

#include <musa_runtime.h>
#include <murand.h>
#include <stdio.h>
#include <math.h>
#include <vector>

// 计算 π 的蒙特卡洛方法
__global__ void estimatePiKernel(float *d_x, float *d_y, int *d_results, int n)
{
    int idx = blockIdx.x * blockDim.x + threadIdx.x;
    if (idx < n)
    {
        // 检查点是否在单位圆内
        if (d_x[idx] * d_x[idx] + d_y[idx] * d_y[idx] <= 1.0f)
        {
            d_results[idx] = 1;
        }
        else
        {
            d_results[idx] = 0;
        }
    }
}

int main()
{
    const int n = 1000000;

    // 创建 Sobol 准随机数生成器
    murandGenerator_t generator;
    murandCreateGenerator(&generator, MURAND_RNG_QUASI_SOBOL32);
    murandSetQuasiRandomGeneratorDimensions(generator, 2);

    // 分配设备内存
    float *d_xy;
    int *d_results;
    musaMalloc(&d_xy, 2 * n * sizeof(float));
    musaMalloc(&d_results, n * sizeof(int));
    float *d_x = d_xy;
    float *d_y = d_xy + n;

    // 生成二维 Sobol 序列
    murandGenerateUniform(generator, d_xy, 2 * n);

    // 启动 kernel 计算圆内点数
    int blockSize = 256;
    int gridSize = (n + blockSize - 1) / blockSize;
    estimatePiKernel<<<gridSize, blockSize>>>(d_x, d_y, d_results, n);

    // 复制结果回主机并求和
    std::vector<int> h_results(n);
    musaMemcpy(h_results.data(), d_results, n * sizeof(int),
               musaMemcpyDeviceToHost);

    int count = 0;
    for (int i = 0; i < n; i++)
    {
        count += h_results[i];
    }

    // 计算 π: π/4 ≈ count/n
    float pi_estimate = 4.0f * count / n;
    printf("Estimated π: %f (actual: %f)\n", pi_estimate, M_PI);
    printf("Error: %f%%\n", fabs(pi_estimate - M_PI) / M_PI * 100);

    // 清理
    murandDestroyGenerator(generator);
    musaFree(d_xy);
    musaFree(d_results);

    return 0;
}

---

最佳实践

种子选择

// ✅ 好的做法：使用时间戳或其他熵源
#include <musa_runtime.h>
#include <chrono>
auto seed = std::chrono::system_clock::now().time_since_epoch().count();
murandSetPseudoRandomGeneratorSeed(generator, seed);

// ✅ 好的做法：固定种子用于可重复实验
murandSetPseudoRandomGeneratorSeed(generator, 42);

多 Stream 并发

// 创建多个生成器用于并发
murandGenerator_t generators[4];
musaStream_t streams[4];

for (int i = 0; i < 4; i++) {
    musaStreamCreate(&streams[i]);
    murandCreateGenerator(&generators[i], MURAND_RNG_PSEUDO_MTGP32);
    murandSetStream(generators[i], streams[i]);
    murandSetPseudoRandomGeneratorSeed(generators[i], 1000 * i);
}

// 并发生成
for (int i = 0; i < 4; i++) {
    murandGenerateUniform(generators[i], d_outputs[i], n);
}

避免重复序列

使用支持 offset 的生成器（例如 Philox）并为每个批次设置不同偏移量：

murandGenerator_t generator;
murandCreateGenerator(&generator, MURAND_RNG_PSEUDO_PHILOX4_32_10);

// 每个 batch 使用不同的偏移量
for (int i = 0; i < batch; i++) {
    murandSetGeneratorOffset(generator, i * n);
    murandGenerateUniform(generator, d_output + i * n, n);
}

---

错误处理

murandStatus_t status = murandGenerateUniform(generator, d_output, n);

if (status != MURAND_STATUS_SUCCESS) {
    fprintf(stderr, "muRAND error: %d\n", status);
    
    switch (status) {
        case MURAND_STATUS_NOT_INITIALIZED:
            fprintf(stderr, "Generator not initialized\n");
            break;
        case MURAND_STATUS_ALLOCATION_FAILED:
            fprintf(stderr, "Memory allocation failed\n");
            break;
        case MURAND_STATUS_OUT_OF_RANGE:
            fprintf(stderr, "Invalid parameter value\n");
            break;
        default:
            fprintf(stderr, "Unknown error\n");
    }
}

---

相关文档

• muRAND Host API 参考
• muRAND Device API 参考