天道酬勤,学无止境

在 CUDA 中并行化 for 循环(1D Naive Convolution)(Parallelizing a for loop (1D Naive Convolution) in CUDA)

问题

有人可以帮我将嵌套的 for 循环转换为 CUDA 内核吗? 这是我试图转换为 CUDA 内核的函数:

// Convolution on Host
void conv(int* A, int* B, int* out) {

    for (int i = 0; i < N; ++i)
        for (int j = 0; j < N; ++j)
            out[i + j] += A[i] * B[j];
}

我非常努力地并行化这段代码。
这是我的尝试:

__global__ void conv_Kernel(int* A, int* B, int* out) {

    int i = blockIdx.x;
    int j = threadIdx.x;

    __shared__ int temp[N];

    __syncthreads();
    temp[i + j] = A[i] * B[j];
    __syncthreads();

    int sum = 0;
    for (int k = 0; k < N; k++)
        sum += temp[k];
    out[i + j] = sum;
}
回答1

您的 cpu conv函数似乎正在执行此操作(例如, N = 4):

A0B0  A0B1  A0B2  A0B3                   +     ^
      A1B0  A1B1  A1B2  A1B3             +     N
            A2B0  A2B1  A2B2  A2B3       +    rows
                  A3B0  A3B1  A3B2  A3B3 =     v
------------------------------------------
out0  out1  out2  out3  out4  out5  out6
    <-  (2*N)-1 columns ->

您的卷积(对我而言)的区别在于它正在对 2 个等长的信号进行卷积。 由于 GPU 喜欢处理“大”问题,这意味着N应该很大。 但是,您的conv_Kernel实现的一个直接问题是它意味着块维度将用于索引A ,而线程维度将用于索引B 。 但是对于当前的 CUDA GPU,线程维度 ( threadIdx.x ) 被限制为 512 或 1024。 这将使我们只能解决非常小的问题。

你的实现还有各种其他问题。 一个问题是分配的共享内存大小不足以适应i+j范围(可以从 0->2*(N-1))。 这当然很容易解决,但更严重的问题是我没有看到一种方法可以将您的算术映射到与上述所需模式类似的任何东西上。 在考虑了你的内核之后,我放弃了它。

卷积问题有大量与之相关的研究,并且可以通过各种方式针对 GPU 等大规模并行架构进行优化。 因此,我将专注于两个非常简单的实现,它们会根据上图立即提示。

第一个实现只是重新创建上面的图表。 我们将创建一个中间temp数组来存储所有单独的 AxBy 产品,计算这些产品并将其存储在conv_Kernel 。 然后我们将启动第二个内核 ( sum_Kernel ),它只是对temp数组的列进行求和,以生成各种out值。 第一个内核需要N线程,当我们迭代N for 循环迭代时,每行一个,以倾斜的方式连续计算上图中的每一行。 第二个内核需要 (2*N)-1 个线程,每个列/ out值一个。

我的第二个实现 (conv_Kernel2) 不需要temp数组,只需为每一列/ out值分配一个线程,并遍历N行,逐行计算必要的乘积,并将这些乘积“在-苍蝇”。 然后将求和结果直接存储在out数组中。

仅考虑计算,而不是数据移动/初始化所需的时间,在 K20x GPU 上大约N = 512 时,GPU 实现开始比原始单线程 CPU 实现更快,这正是我碰巧使用的。 第二个实现也值得称赞,因为所需的唯一数据移动是 A、B 和结果。 第一个实现还需要分配temp数组并将其初始化为全零。 temp数组的大小与N * N成正比,所以第二种实现还有一个好处,它不需要这个临时存储。

这是一个完整的测试用例,运行和计时您提供的 CPU 实现以及我创建的两个略有不同的 GPU 实现:

$ cat t617.cu
#include <stdio.h>
#include <stdlib.h>
#include <time.h>
#include <sys/time.h>

#define N 4096
#define RG 10
#define USECPSEC 1000000ULL
#define nTPB 256


void conv(int* A, int* B, int* out) {

    for (int i = 0; i < N; ++i)
        for (int j = 0; j < N; ++j)
            out[i + j] += A[i] * B[j];
}

unsigned long long dtime_usec(unsigned long long prev){
  timeval tv1;
  gettimeofday(&tv1,0);
  return ((tv1.tv_sec * USECPSEC)+tv1.tv_usec) - prev;
}


__global__ void conv_Kernel(int* A, int *B, int* temp) {

    int idx = threadIdx.x+blockDim.x*blockIdx.x;
    if (idx < N){
      int my_B = B[idx];
      for (int i = 0; i < N; i++)
        temp[idx + (i*2*N) + i] = my_B * A[i];
      }
}

__global__ void sum_Kernel(int *temp, int *out){
    int idx = threadIdx.x+blockDim.x*blockIdx.x;
    if (idx < (2*N)-1){
      int my_sum = 0;
      for (int i = 0; i < N; i++) my_sum += temp[idx + (i*2*N)];
      out[idx] = my_sum;}
}

__global__ void conv_Kernel2(int *A, int *B, int *out){
    int idx = threadIdx.x+blockDim.x*blockIdx.x;
    if (idx < (2*N)-1){
      int my_sum = 0;
      for (int i = 0; i < N; i++)
        if (((idx < N) && (i <= idx)) || ((idx >= N) && (i > (idx-N)))) my_sum += A[i]*B[idx-i];
      out[idx] = my_sum;
    }
}

int main(){

  int *h_A, *d_A, *h_result, *d_result, *result, *h_B, *d_B, *A, *B, *d_temp;

  B   = (int *)malloc(N*sizeof(int));
  A   = (int *)malloc(N*sizeof(int));
  h_A = (int *)malloc(N*sizeof(int));
  h_B = (int *)malloc(N*sizeof(int));
  h_result = (int *)malloc(2*N*sizeof(int));
  result   = (int *)malloc(2*N*sizeof(int));

  cudaMalloc(&d_B, N*sizeof(int));
  cudaMalloc(&d_A, N*sizeof(int));
  cudaMalloc(&d_result, 2*N*sizeof(int));
  cudaMalloc(&d_temp, 2*N*N*sizeof(int));

  for (int i=0; i < N; i++){
    A[i] = rand()%RG;
    B[i] = rand()%RG;
    h_A[i] = A[i];
    h_B[i] = B[i];}

  for (int i=0; i < 2*N; i++){
    result[i]   = 0;
    h_result[i] = 0;}

  unsigned long long cpu_time = dtime_usec(0);
  conv(A, B, result);
  cpu_time = dtime_usec(cpu_time);

  cudaMemcpy(d_A, h_A, N*sizeof(int), cudaMemcpyHostToDevice);
  cudaMemcpy(d_B, h_B, N*sizeof(int), cudaMemcpyHostToDevice);
  cudaMemset(d_result, 0, 2*N*sizeof(int));
  cudaMemset(d_temp, 0, 2*N*N*sizeof(int));

  unsigned long long gpu_time = dtime_usec(0);
  conv_Kernel<<<(N+nTPB-1)/nTPB,nTPB>>>(d_A, d_B, d_temp);
  sum_Kernel<<<((2*(N-1))+nTPB-1)/nTPB, nTPB>>>(d_temp, d_result);
  cudaDeviceSynchronize();
  gpu_time = dtime_usec(gpu_time);

  cudaMemcpy(h_result, d_result, 2*N*sizeof(int), cudaMemcpyDeviceToHost);
  for (int i = 0; i < 2*N; i++) if (result[i] != h_result[i]) {printf("mismatch at %d, cpu: %d, gpu %d\n", i, result[i], h_result[i]); return 1;}
  printf("Finished.  Results match.  cpu time: %ldus, gpu  time: %ldus\n", cpu_time, gpu_time);


  cudaMemset(d_result, 0, 2*N*sizeof(int)); // just for error checking, the kernel2 require no initialization of the result

  gpu_time = dtime_usec(0);
  conv_Kernel2<<<((2*(N-1))+nTPB-1)/nTPB,nTPB>>>(d_A, d_B, d_result);
  cudaDeviceSynchronize();
  gpu_time = dtime_usec(gpu_time);

  cudaMemcpy(h_result, d_result, 2*N*sizeof(int), cudaMemcpyDeviceToHost);
  for (int i = 0; i < 2*N; i++) if (result[i] != h_result[i]) {printf("mismatch2 at %d, cpu: %d, gpu %d\n", i, result[i], h_result[i]); return 1;}
  printf("Finished.  Results match.  cpu time: %ldus, gpu2 time: %ldus\n", cpu_time, gpu_time);
  return 0;
}
$ nvcc -arch=sm_35 -o t617 t617.cu
$ ./t617
Finished.  Results match.  cpu time: 69059us, gpu  time: 3204us
Finished.  Results match.  cpu time: 69059us, gpu2 time: 1883us
$ nvcc -arch=sm_35 -O3 -o t617 t617.cu
$ ./t617
Finished.  Results match.  cpu time: 13750us, gpu  time: 3214us
Finished.  Results match.  cpu time: 13750us, gpu2 time: 1886us
$

(请注意,即使仅使用 -O3 参数也会对 CPU 代码执行产生显着影响)

正如我所提到的,我认为我的两个示例对于 GPU 代码也非常“幼稚”(例如,都不使用共享内存),但它们可能会给您一些关于如何开始的想法。

为简洁起见,我省略了 CUDA 错误检查。 但是,我建议您在任何时候遇到 CUDA 代码问题时,都应执行适当的 cuda 错误检查。 就您的conv_Kernel ,我相信它会指示一些错误(如果您尝试运行它。)作为快速测试,您始终可以使用cuda-memcheck运行任何 CUDA 代码,以查看是否发生了任何 API 错误。

编辑:我尝试了我的conv_Kernel2一个简单共享内存版本,但它并没有更快。 我相信这样做的原因是这些数据集(在N = 4096 处, AB为 16Kbytes, out大约为 32Kbytes)足够小,可以轻松放入 GPU L2 缓存,而不会出现抖动。

但是,对于较新的体系结构(cc 3.5 和更新版本),如果只读输入数据被正确识别为内核,CUDA 编译器有时可以进行额外的优化。 因此,如果我们将conv_Kernel2定义更改为:

__global__ void conv_Kernel2(const int * __restrict__ A, const int * __restrict__ B, int *out){

然后我目睹了执行时​​间略有改善,就我而言:

$ ./t617
Finished.  Results match.  cpu time: 13792us, gpu  time: 3209us
Finished.  Results match.  cpu time: 13792us, gpu2 time: 1626us
$

我创建了代码的修改版本,它执行以下操作:

  1. N在命令行中指定
  2. 仅包含 cpu conv和 gpu conv_Kernel2
  3. 将数据移入/移出 GPU 的时间成本包含在 GPU 计时测量中
  4. 一个typedef ... mytype; 提供了这样的代码,以便可以轻松地重新编译代码以测试各种数据类型的行为。
  5. 打印出“加速因子”,即cpu时间除以gpu时间。

修改后的代码:

#include <stdio.h>
#include <stdlib.h>
#include <time.h>
#include <sys/time.h>

// RG*RG*MAXN must fit within mytype

#define MAXN 100000
#define RG 10
#define USECPSEC 1000000ULL
#define nTPB 256

typedef double mytype;

void conv(const mytype *A, const mytype *B, mytype* out, int N) {

    for (int i = 0; i < N; ++i)
        for (int j = 0; j < N; ++j)
            out[i + j] += A[i] * B[j];
}

unsigned long long dtime_usec(unsigned long long prev){
  timeval tv1;
  gettimeofday(&tv1,0);
  return ((tv1.tv_sec * USECPSEC)+tv1.tv_usec) - prev;
}



__global__ void conv_Kernel2(const mytype * __restrict__ A, const mytype * __restrict__ B, mytype *out, const int N){
    int idx = threadIdx.x+blockDim.x*blockIdx.x;
    if (idx < (2*N)-1){
      mytype my_sum = 0;
      for (int i = 0; i < N; i++)
        if (((idx < N) && (i <= idx)) || ((idx >= N) && (i > (idx-N)))) my_sum += A[i]*B[idx-i];
      out[idx] = my_sum;
    }
}

int main(int argc, char *argv[]){


  mytype *h_A, *d_A, *h_result, *d_result, *result, *h_B, *d_B, *A, *B;
  if (argc != 2) {printf("must specify N on the command line\n"); return 1;}
  int my_N = atoi(argv[1]);
  if ((my_N < 1) || (my_N > MAXN)) {printf("N out of range\n"); return 1;}
  B   = (mytype *)malloc(my_N*sizeof(mytype));
  A   = (mytype *)malloc(my_N*sizeof(mytype));
  h_A = (mytype *)malloc(my_N*sizeof(mytype));
  h_B = (mytype *)malloc(my_N*sizeof(mytype));
  h_result = (mytype *)malloc(2*my_N*sizeof(mytype));
  result   = (mytype *)malloc(2*my_N*sizeof(mytype));

  cudaMalloc(&d_B, my_N*sizeof(mytype));
  cudaMalloc(&d_A, my_N*sizeof(mytype));
  cudaMalloc(&d_result, 2*my_N*sizeof(mytype));

  for (int i=0; i < my_N; i++){
    A[i] = rand()%RG;
    B[i] = rand()%RG;
    h_A[i] = A[i];
    h_B[i] = B[i];}

  for (int i=0; i < 2*my_N; i++){
    result[i]   = 0;
    h_result[i] = 0;}

  unsigned long long cpu_time = dtime_usec(0);
  conv(A, B, result, my_N);
  cpu_time = dtime_usec(cpu_time);

  cudaMemset(d_result, 0, 2*my_N*sizeof(mytype));

  unsigned long long gpu_time = dtime_usec(0);
  cudaMemcpy(d_A, h_A, my_N*sizeof(mytype), cudaMemcpyHostToDevice);
  cudaMemcpy(d_B, h_B, my_N*sizeof(mytype), cudaMemcpyHostToDevice);
  conv_Kernel2<<<((2*(my_N-1))+nTPB-1)/nTPB,nTPB>>>(d_A, d_B, d_result, my_N);
  cudaDeviceSynchronize();
  cudaMemcpy(h_result, d_result, 2*my_N*sizeof(mytype), cudaMemcpyDeviceToHost);
  gpu_time = dtime_usec(gpu_time);

  for (int i = 0; i < 2*my_N; i++) if (result[i] != h_result[i]) {printf("mismatch2 at %d, cpu: %d, gpu %d\n", i, result[i], h_result[i]); return 1;}
  printf("Finished.  Results match.  cpu time: %ldus, gpu time: %ldus\n", cpu_time, gpu_time);
  printf("cpu/gpu = %f\n", cpu_time/(float)gpu_time);
  return 0;
}

受限制的 HTML

  • 允许的HTML标签:<a href hreflang> <em> <strong> <cite> <blockquote cite> <code> <ul type> <ol start type> <li> <dl> <dt> <dd> <h2 id> <h3 id> <h4 id> <h5 id> <h6 id>
  • 自动断行和分段。
  • 网页和电子邮件地址自动转换为链接。

相关推荐
  • Parallelizing a for loop (1D Naive Convolution) in CUDA
    Can someone please help me convert a nested for loop into a CUDA kernel? Here is the function I am trying to convert into a CUDA kernel: // Convolution on Host void conv(int* A, int* B, int* out) { for (int i = 0; i < N; ++i) for (int j = 0; j < N; ++j) out[i + j] += A[i] * B[j]; } I have tried very hard to parallelize this code. Here is my attempt: __global__ void conv_Kernel(int* A, int* B, int* out) { int i = blockIdx.x; int j = threadIdx.x; __shared__ int temp[N]; __syncthreads(); temp[i + j] = A[i] * B[j]; __syncthreads(); int sum = 0; for (int k = 0; k < N; k++) sum += temp[k]; out[i + j
  • Node.js中的并行任务(Parallelizing tasks in Node.js)
    问题 我有一些我想在JS中完成的任务,这些任务会占用大量资源。 对于这个问题,让我们假设它们是一些繁重的计算,而不是系统访问。 现在,我想同时运行任务A,B和C,并在完成后执行一些功能D。 异步库为此提供了一个很好的框架: async.parallel([A, B, C], D); 如果我正在做的只是计算,那么它将仍然同步运行(除非库将任务本身放在不同的线程上,我希望情况并非如此)。 我如何使它实际上是平行的? 异步代码通常不阻止调用者的操作是什么(使用NodeJS时)? 它开始一个子进程吗? 回答1 我如何使它实际上是平行的? 首先,在单节点应用程序中您实际上不会并行运行。 节点应用程序在单个线程上运行,并且节点的事件循环一次仅处理一个事件。 即使在多核计算机上运行,​​您也不会在节点应用程序中获得并行处理。 也就是说,您可以通过将代码分叉到单独的节点进程中或通过生成子进程来在多核计算机上获得处理并行性。 实际上,这允许您创建节点本身的多个实例,并以不同的方式(例如,stdout,进程派生IPC机制)与这些进程进行通信。 另外,您可以选择将功能(按职责)分离到各自的节点应用程序/服务器中,并通过RPC进行调用。 异步代码通常不阻止调用者的操作是什么(使用NodeJS时)? 它开始一个子进程吗? 它没有开始新的过程。 在下面,当在node.js中使用async.parallel时
  • 使用从bash中的文件读取的数组并行化while循环(Parallelizing a while loop with arrays read from a file in bash)
    问题 我在 Bash 中处理了一个 while 循环,如下所示: while IFS=$'\t' read -r -a line; do myprogram ${line[0]} ${line[1]} ${line[0]}_vs_${line[1]}.result; done < fileinput 它从具有此结构的文件中读取,以供参考: foo bar baz foobar 等等(制表符分隔)。 我想使用 GNU 并行并行化这个循环(因为条目很多并且处理可能很慢),但是示例并不清楚我如何将每一行分配给数组,就像我在这里做的那样。 什么是可能的解决方案(也可以替代 GNU 并行工作)? 回答1 从 https://www.gnu.org/software/parallel/man.html#EXAMPLE:-Use-a-table-as-input: """ table_file.tsv 的内容: foo<TAB>bar baz <TAB> quux 跑步: cmd -o bar -i foo cmd -o quux -i baz 你可以运行: parallel -a table_file.tsv --colsep '\t' cmd -o {2} -i {1} """ 所以在你的情况下,它将是: cat fileinput | parallel --colsep '\t'
  • 在 Python 中并行化四个嵌套循环(Parallelizing four nested loops in Python)
    问题 我有一个相当简单的嵌套 for 循环,它迭代四个数组: for a in a_grid: for b in b_grid: for c in c_grid: for d in d_grid: do_some_stuff(a,b,c,d) # perform calculations and write to file 也许这不是开始在 4D 网格上执行计算的最有效方法。 我知道joblib能够像这样并行化两个嵌套的 for 循环,但是我无法将它推广到四个嵌套循环。 有任何想法吗? 回答1 我通常使用这种形式的代码: #!/usr/bin/env python3 import itertools import multiprocessing #Generate values for each parameter a = range(10) b = range(10) c = range(10) d = range(10) #Generate a list of tuples where each tuple is a combination of parameters. #The list will contain all possible combinations of parameters. paramlist = list(itertools.product(a,b,c
  • 有系统地并行化fortran 2008`do concurrent`,可能与openmp并行(Parallelizing fortran 2008 `do concurrent` systematically, possibly with openmp)
    问题 fortran 2008 do concurrent构造是一个do循环,它告诉编译器没有迭代会影响其他任何迭代。 因此,它可以安全地并行化。 一个有效的例子: program main implicit none integer :: i integer, dimension(10) :: array do concurrent( i= 1: 10) array(i) = i end do end program main 可以以任何顺序进行迭代的地方。 你可以在这里读更多关于它的内容。 据我所知,gfortran不会自动并行化这些do concurrent循环,而我还记得有关此操作的gfortran-diffusion-list邮件(此处)。 只是将它们转换为经典的do循环。 我的问题:您知道一种系统地并行化do concurrent循环的方法吗? 例如,使用系统的openmp语法? 回答1 自动进行操作并非易事。 DO CONCURRENT构造具有forall标头,这意味着它可以接受多个循环,索引变量定义和掩码。 基本上,您需要替换: DO CONCURRENT([<type-spec> :: ]<forall-triplet-spec 1>, <forall-triplet-spec 2>, ...[, <scalar-mask-expression>]) <block
  • 在 .net 4.5 中步进并行化 for 循环(Parallelizing a for loop with stepping in .net 4.5)
    问题 我有一个类似于下面代码的函数。 它的目的是从一组点中一次一个地取一个三角形面,其中每三个点是一个面,然后细分它们,用边长不超过 nodeSize 的较小面的列表替换面。 自然,此功能对于任何逼真的小平面网格来说都是耗时的。 我想重构它以使用一些粗略的并行化。 但是,Parallel.For 似乎没有办法在保留索引号的同时每隔一段时间遍历数组中的索引。 记住循环内的SplitTriangle函数在计算上不利于并行化,我该如何重构这个函数? Protected Shared Function SplitTriangles(Points As IEnumerable(Of Point3D), nodeSize As Single) As List(Of Point3D) Dim resultList As New List(Of Point3D) For i As Integer = 0 To Points.Count - 1 Step 3 resultList.AddRange(SplitTriangle(Points(i), Points(i + 1), Points(i + 2), nodeSize * 4)) Next Return resultList End Function 回答1 您可以使用Enumerable.Range()函数为您生成索引。 我不太熟悉 VB
  • 使用 CUDA 并行化四个或更多嵌套循环(Parallelize four and more nested loops with CUDA)
    问题 我正在研究生成并行 C++ 代码的编译器。 我是 CUDA 编程的新手,但我正在尝试使用 CUDA 并行化 C++ 代码。 目前,如果我有以下顺序 C++ 代码: for(int i = 0; i < a; i++) { for(int j = 0; j < b; j++) { for(int k = 0; k < c; k++) { A[i*y*z + j*z + k*z +l] = 1; } } } 这会产生以下 CUDA 代码: __global__ void kernelExample() { int _cu_x = ((blockIdx.x*blockDim.x)+threadIdx.x); int _cu_y = ((blockIdx.y*blockDim.y)+threadIdx.y); int _cu_z = ((blockIdx.z*blockDim.z)+threadIdx.z); A[_cu_x*y*z + _cu_y*z + _cu_z] = 1; } 所以每个循环嵌套都映射到一维,但是并行化四个或更多嵌套循环的正确方法是什么: for(int i = 0; i < a; i++) { for(int j = 0; j < b; j++) { for(int k = 0; k < c; k++) { for(int l = 0; l < d; l++)
  • 并行化 IO 绑定(网络)ForEach 循环(Parallelizing IO Bound (Network) ForEach Loop)
    问题 根据选择的选项,我有几种不同的方法可以在我的应用程序中将整个目录上传到 Amazon S3。 目前,其中一个选项将并行上传多个目录。 我不确定这是否是个好主意,因为在某些情况下它会加快上传速度,而在其他情况下会减慢上传速度。 当有一堆小目录时,速度似乎加快,但如果批处理中有大目录,它会变慢。 我正在使用下面看到的并行 ForEach 循环,并使用 AWS API 的TransferUtility.UploadDirectoryAsync()方法: Parallel.ForEach(dirs,myParallelOptions, async dir => { await MyUploadMethodAsync(dir) }; TransferUtility.UploadDirectoryAsync()方法在MyUploadMethodAsync() 。 TransferUtility的上传方法都执行单个文件部分的并行上传(如果大小足够大),因此执行目录的并行上传可能有点过头了。 显然,我们仍然受到可用带宽量的限制,所以这可能是一种浪费,我应该只使用带有UploadDirectoryAsync()方法的常规 foreach 循环。 任何人都可以提供一些关于这是否是并行化的坏情况的见解? 回答1 你真的测试过这个吗? 您使用它的方式, Parallel
  • Java:通过多线程并行化快速排序(Java: Parallelizing quick sort via multi-threading)
    问题 我正在试验 Java 中的并行化算法。 我从合并排序开始,并在这个问题中发布了我的尝试。 我修改后的尝试在下面的代码中,我现在尝试并行化快速排序。 我的多线程实现或解决此问题的方法中是否有任何菜鸟错误? 如果不是,我不应该期望双核上的顺序算法和并行算法之间的速度提高超过 32%(请参阅底部的计时)? 这是多线程算法: public class ThreadedQuick extends Thread { final int MAX_THREADS = Runtime.getRuntime().availableProcessors(); CountDownLatch doneSignal; static int num_threads = 1; int[] my_array; int start, end; public ThreadedQuick(CountDownLatch doneSignal, int[] array, int start, int end) { this.my_array = array; this.start = start; this.end = end; this.doneSignal = doneSignal; } public static void reset() { num_threads = 1; } public void run(
  • Clojure中pmap的更好替代品,用于在大数据上并行化价格适中的函数?(Better alternative to pmap in Clojure for parallelizing moderately inexpensive functions over big data?)
    问题 使用clojure,我在一个序列中有大量数据,我想用较少的内核数(4到8)并行处理它。 最简单的方法是使用pmap而不是map ,以在数据序列上映射我的处理函数。 但是在我的情况下,协调开销导致净损失。 我认为原因是pmap假设跨数据映射的功能非常昂贵。 综观PMAP的源代码看来构建future用于依次序列的每个元素,因此每次调用函数上一个单独的线程(循环超过可用核的数量)发生。 这是pmap的相关资料: (defn pmap "Like map, except f is applied in parallel. Semi-lazy in that the parallel computation stays ahead of the consumption, but doesn't realize the entire result unless required. Only useful for computationally intensive functions where the time of f dominates the coordination overhead." ([f coll] (let [n (+ 2 (.. Runtime getRuntime availableProcessors)) rets (map #(future (f %))
  • 如何并行化一个简单的Python循环?(How do I parallelize a simple Python loop?)
    问题 这可能是一个琐碎的问题,但是如何在python中并行化以下循环? # setup output lists output1 = list() output2 = list() output3 = list() for j in range(0, 10): # calc individual parameter value parameter = j * offset # call the calculation out1, out2, out3 = calc_stuff(parameter = parameter) # put results into correct output list output1.append(out1) output2.append(out2) output3.append(out3) 我知道如何在Python中启动单线程,但我不知道如何“收集”结果。 多个过程也可以-在这种情况下最简单的方法。 我目前使用的是Linux,但代码也应同时在Windows和Mac上运行。 并行化此代码的最简单方法是什么? 回答1 由于具有全局解释器锁(GIL),因此在CPython上使用多个线程不会为纯Python代码带来更好的性能。 我建议改用多处理模块: pool = multiprocessing.Pool(4) out1, out2, out3 = zip(
  • 使用 CUDA 卷积多个小矩阵的最佳方法(Best approach for convolution of multiple small matrices using CUDA)
    问题 我需要用小矩阵和内核执行多个卷积,我希望利用 GPU 的许多处理器能让我尽可能快地完成它。 问题如下:我有很多矩阵(~1,000 到 ~10,000)或相对较小的尺寸(~15x15 到 1x1 - 如标量),以及一定数量的卷积掩码(~20 到 1)。 我需要用每个卷积掩码示例对所有矩阵进行卷积: A; %5,000 matrices of size 10x10, A(i) = a 10x10 matrix B; 10 matrices of size 5x5, B(k) = a 5x5 matrix res(j)=conv(A,B(1)); %res(j) is the result of convolving all 5,000 %matrices in A by the j'th kernel B(j) 目标是尽快计算 res(1),...,res(10) 我想听听有关如何实现最有效算法的建议。 基于 FFT 的卷积可能太慢了。 到目前为止,我看到的每个实现都是针对 2d 卷积的,旨在对 2 个大矩阵进行卷积,而我需要对许多小矩阵进行卷积。 我现在对 CUDA 编程知之甚少,但我正在学习中。 我希望自己解决这个问题,但由于时间限制,我不得不在我学习如何在 CUDA 中编码的同时寻求任何有经验的人能给我的建议。 谢谢! ps任何指向适合我目的的实现的指针都非常感谢。
  • Rails 中的并行化方法(Parallelizing methods in Rails)
    问题 我的 Rails Web 应用程序有几十种方法,从调用 API 到处理查询结果。 这些方法具有以下结构: def method_one batch_query_API process_data end .......... def method_nth batch_query_API process_data end def summary method_one ...... method_nth collect_results end 如何在 Rails 中同时运行所有查询方法而不是按顺序运行(当然,无需启动多个工作程序)? 编辑:所有方法都是从单个实例变量调用的。 我认为这限制了使用 Sidekiq 或 Delay 同时提交作业。 回答1 Ruby 拥有出色的 promise gem。 您的示例如下所示: require 'future' def method_one ... def method_nth def summary result1 = future { method_one } ...... resultn = future { method_nth } collect_results result1, ..., resultn end 很简单,不是吗? 但让我们了解更多细节。 这是一个未来的对象: result1 = future { method_one
  • 在 GPU 上运行 FFTW 与使用 CUFFT(running FFTW on GPU vs using CUFFT)
    问题 我有一个基本的 C++ FFTW 实现,如下所示: for (int i = 0; i < N; i++){ // declare pointers and plan fftw_complex *in, *out; fftw_plan p; // allocate in = (fftw_complex*) fftw_malloc(sizeof(fftw_complex) * N); out = (fftw_complex*) fftw_malloc(sizeof(fftw_complex) * N); // initialize "in" ... // create plan p = fftw_plan_dft_1d(N, in, out, FFTW_FORWARD, FFTW_ESTIMATE); // execute plan fftw_execute(p); // clean up fftw_destroy_plan(p); fftw_free(in); fftw_free(out); } 我正在 for 循环中执行 N fft。 我知道我可以使用 FFTW 一次执行多个计划,但是在我的实现中,每个循环的进出都不同。 关键是我在 for 循环中执行整个 FFTW 管道。 我想过渡到使用 CUDA 来加快速度。 我知道 CUDA 有自己的 FFT 库 CUFFT。
  • 计算机视觉中的各种卷积(Convolution in Computer Vision)
    目录 1. 卷积与互相关Cross-correlation2. 深度学习中的卷积(单通道版本,多通道版本)(single channel version, multi-channel version)3. 3D 卷积4. 1×1 卷积5. 卷积算术Convolution Arithmetic6. 转置卷积(去卷积、棋盘效应)Transposed Convolution (Deconvolution, checkerboard artifacts)7. 扩张卷积Dilated Convolution (Atrous Convolution)8. 可分卷积(空间可分卷积,深度可分卷积) (Spatially Separable Convolution, Depthwise Separable Convolution)9. 平展卷积Flattened Convolution10. 分组卷积Grouped Convolution11. 混洗分组卷积Shuffled Grouped Convolution12. 逐点分组卷积Pointwise Grouped Convolution13. 动态卷积Dynamic Convolution==代码实现(Pytorch)==参考文献 1. 卷积与互相关Cross-correlation 在信号处理、图像处理和其它工程/科学领域
  • N-body 算法:为什么这并行更慢?(N-body algorithm: why is this slower in parallel?)
    问题 我整理了一个示例程序,模仿我正在处理的数据结构类型。 即我有n对象,我需要在每个可能的对之间迭代一次并执行(对称)计算。 此操作涉及将数据写入两个对。 在串行中,这将采用这样的循环形式 for(int i = 0; i < N-1; ++i) for(int j = i + 1; j < N; ++j) ... 然而,并没有在互联网上进行太多搜索就找到了“缓存不经意并行实现”,我在下面编写并复制了它。 我在这里链接了一篇详细描述该算法的帖子(使用英特尔 TBB)。 https://software.intel.com/en-us/blogs/2010/07/01/n-bodies-a-parallel-tbb-solution-parallel-code-balanced-recursive-parallelism-with-parallel_invoke 我尝试使用 OpenMP 任务来执行相同的操作,但它的运行速度总是比串行对应的任务慢(只需在没有 -fopenmp 的情况下进行编译)。 我用g++ -Wall -std=c++11 -O3 test.cpp -o test编译它。 使用或不使用-O3都会观察到相同的情况; 串行总是更快。 要添加更多信息,在我的实际应用程序中,通常有几百到几千个元素(下面示例中的变量n )需要以这种成对方式循环多次。 数百万次。
  • 为什么 gcc 自动向量化对大于 3x3 的卷积矩阵不起作用?(Why gcc autovectorization does not work on convolution matrix biger than 3x3?)
    问题 我已经为卷积矩阵实现了以下程序 #include <stdio.h> #include <time.h> #define NUM_LOOP 1000 #define N 128 //input or output dimention 1 #define M N //input or output dimention 2 #define P 5 //convolution matrix dimention 1 if you want a 3x3 convolution matrix it must be 3 #define Q P //convolution matrix dimention 2 #define Csize P*Q #define Cdiv 1 //div for filter #define Coffset 0 //offset //functions void unusual(); //unusual implementation of convolution void naive(); //data unsigned short int input[N][M] __attribute__(( aligned(32))); // input data unsigned short int output[N][M] __attribute__((
  • 并行化 pandas pyodbc SQL 数据库调用(Parallelizing pandas pyodbc SQL database calls)
    问题 我目前正在通过pandas.io.sql.read_sql()命令将数据查询到数据pandas.io.sql.read_sql() 。 我想并行化类似于这些人所倡导的调用:(使用 Python 进行令人尴尬的并行数据库调用(PyData Paris 2015)) 像(非常一般): pools = [ThreadedConnectionPool(1,20,dsn=d) for d in dsns] connections = [pool.getconn() for pool in pools] parallel_connection = ParallelConnection(connections) pandas_cursor = parallel_connection.cursor() pandas_cursor.execute(my_query) 这样的事情可能吗? 回答1 是的,这应该可行,但需要注意的是,您需要在您网站的谈话中更改 parallel_connection.py。 在该代码中有一个fetchall函数,它并行执行每个游标,然后组合结果。 这是你将要改变的核心: 旧代码: def fetchall(self): results = [None] * len(self.cursors) def do_work(index, cursor): results
  • CUDA小内核2D卷积-如何做到(CUDA small kernel 2d convolution - how to do it)
    问题 我已经尝试使用CUDA内核数天,以在500x500图像(但是我也可以改变尺寸)和非常小的2D内核(拉普拉斯2d内核,因此它是3x3内核)之间执行快速2D卷积。以便在所有cuda线程中发挥巨大的优势)。 我创建了一个CPU经典实现(两个循环,就像您想的那样容易),然后开始创建CUDA内核。 经过几次令人失望的尝试以更快的速度进行卷积后,我最终得到了以下代码:http://www.evl.uic.edu/sjames/cs525/final.html(请参阅“共享内存”部分),它基本上允许16x16线程将所需的所有卷积数据分块加载到共享内存中,然后执行卷积。 没什么,CPU仍然快很多。 我没有尝试FFT方法,因为CUDA SDK指出,在较大的内核大小下它是有效的。 无论您是否阅读我写的所有内容,我的问题是: 如何使用CUDA在相对较大的图像和很小的内核(3x3)之间执行快速2D卷积? 回答1 您是对的,因为3x3内核不适合基于FFT的方法。 解决此问题的最佳方法是将内核推送到恒定内存中(或者,如果您使用的是fermi +卡,则不要紧)。 由于您知道内核的大小,因此最快的方法是将输入图像/信号的块读入共享内存,并执行展开的乘法和加法操作。 -- 如果您愿意使用库来执行此操作,则ArrayFire和OpenCV具有高度优化的卷积例程,可以节省大量开发时间。 我对OpenCV不太熟悉
  • CUDA编程笔记
    围绕图灵系显卡 常见术语 Streaming Multiprocessor (SM):GPU中的处理器核心 Graphics Processing Clusters (GPCs) Texture Processing Clusters (TPCs) Raster Operations Units(ROPs):光栅化处理单元。光栅化操作,是发生在模型完全建立,并且完成基本光照及对应纹理之后的操作环节。除了满足二维平面输出对坐标变换的要求之外,Rasterizer(光栅化)最大的意义在是:由于透视固有的视线前后遮蔽问题,建立好的模型存在很多看不到的部分,光栅化过程对Z值得判断,可以将这些看不到的部分剔除掉] 硬件概况 核心代号 产品型号RTX 2080 TiRTX 2080RTX 2070RTX 2060GTX 1660 Ti核心代号TU102TU104TU106TU106TU116 The TU104 and TU106 GPUs utilize the same basic architecture as TU102, scaled down to different degrees for different usage models and market segments. 硬件架构 Graphics Processing Cluster ├── 6 * Texture