Tag: cuda

如何在CUDA / cublas中转置矩阵？

假设我在GPU上有一个尺寸为A*B的矩阵，其中B （列数）是假设C样式的前导维度。 在CUDA（或Cublas）中是否有任何方法将此矩阵转换为FORTRAN样式，其中A （行数）成为主要维度？ 如果它可以在host->device传输期间进行转置，同时保持原始数据不变，那就更好了。

CUDA矢量类型的效率（float2，float3，float4）

我试图从CUDA示例中了解particles_kernel.cu的integrate_functor ： struct integrate_functor { float deltaTime; //constructor for functor //… template __device__ void operator()(Tuple t) { volatile float4 posData = thrust::get(t); volatile float4 velData = thrust::get(t); float3 pos = make_float3(posData.x, posData.y, posData.z); float3 vel = make_float3(velData.x, velData.y, velData.z); // update position and velocity // … // store new position and velocity thrust::get(t) = make_float4(pos, […]

CUDA：平铺矩阵 – 矩阵乘法，共享内存和矩阵大小，是块大小的非倍数

我正在尝试熟悉CUDA编程，并且有一段非常有趣的时间。 我目前正在研究这个 pdf，它处理矩阵乘法，有和没有共享内存。 这两个版本的完整代码都可以在这里找到。 该代码几乎与CUDA矩阵乘法样本中的代码完全相同。 虽然非共享内存版本具有以任何矩阵大小运行的能力，但无论块大小如何，共享内存版本必须与块大小的倍数（我设置为4，默认最初为16）的矩阵一起使用。 pdf末尾提出的问题之一是更改它，以便共享内存版本也可以使用块大小的非倍数。 我认为这将是一个简单的索引检查，就像在非共享版本中一样： int row = blockIdx.y * blockDim.y + threadIdx.y; int col = blockIdx.x * blockDim.x + threadIdx.x; if(row > A.height || col > B.width) return; 但这不起作用。 这是完整的代码，减去主要的方法（有点乱，对不起），我已经有所修改了： void MatMul(const Matrix A, const Matrix B, Matrix C) { // Load A and B to device memory Matrix d_A; d_A.width […]

如何将CUDA代码分成多个文件

我正在尝试将一个CUDA程序分成两个独立的.cu文件，以便更接近于在C ++中编写一个真正的应用程序。 我有一个简单的小程序： 在主机和设备上分配内存。 将主机arrays初始化为一系列数字。 将主机arrays复制到设备arrays使用设备内核查找arrays中所有元素的平方将设备arrays复制回主机arrays打印结果 如果我把它全部放在一个.cu文件中并运行它，这很有效。 当我将它分成两个单独的文件时，我开始得到链接错误。 像我最近的所有问题一样，我知道这很小，但它是什么？ KernelSupport.cu #ifndef _KERNEL_SUPPORT_ #define _KERNEL_SUPPORT_ #include #include int main( int argc, char** argv) { int* hostArray; int* deviceArray; const int arrayLength = 16; const unsigned int memSize = sizeof(int) * arrayLength; hostArray = (int*)malloc(memSize); cudaMalloc((void**) &deviceArray, memSize); std::cout << "Before device\n"; for(int i=0;i<arrayLength;i++) { hostArray[i] = […]

Cuda共享内存数组变量

我试图为矩阵乘法声明一个变量，如下所示： __shared__ float As[BLOCK_SIZE][BLOCK_SIZE]; 我试图让它用户可以输入矩阵的大小来计算，但这意味着改变BLOCK_SIZE。 我更改了它但是我收到了编译器错误：“错误：常量值未知”。 我调查了它，它与这个post类似。 所以我尝试过： __shared__ int buf []; 但后来我得到：“错误：不允许不完整的类型” 谢谢，Dan更新代码（几乎遵循本指南并盯着cuda指南）：通过询问用户矩阵的大小来传递块大小。 他们进入x和y。 块大小仅为x，现在它必须接受与x和y相同的大小。 __global__ void matrixMul( float* C, float* A, float* B, int wA, int wB,size_t block_size) { // Block index int bx = blockIdx.x; int by = blockIdx.y; // Thread index int tx = threadIdx.x; int ty = threadIdx.y; // Index […]

CUDA动态并行MakeFile

这是我的第一个使用动态并行的程序，我无法编译代码。 我需要能够为我在大学的研究项目运行这个项目，我们将非常感谢任何帮助： 我收到以下错误： /cm/shared/apps/cuda50/toolkit/5.0.35/bin/nvcc -m64 -dc -gencode arch=compute_35,code=sm_35 -rdc=true -dlink -po maxrregcount=16 -I/cm/shared/apps/cuda50/toolkit/5.0.35 -I. -I.. -I../../common/inc -o BlackScholes.o -c BlackScholes.cu g++ -m64 -I/cm/shared/apps/cuda50/toolkit/5.0.35 -I. -I.. -I../../common/inc -o BlackScholes_gold.o -c BlackScholes_gold.cpp g++ -m64 -o BlackScholes BlackScholes.o BlackScholes_gold.o -L/cm/shared/apps/cuda50/toolkit/5.0.35/lib64 -lcudart -lcudadevrt BlackScholes.o: In function `__sti____cudaRegisterAll_47_tmpxft_000059cb_00000000_6_BlackScholes_cpp1_ii_c58990ec()’: tmpxft_000059cb_00000000-3_BlackScholes.cudafe1.cpp:(.text+0x1354): undefined reference to `__cudaRegisterLinkedBinary_47_tmpxft_000059cb_00000000_6_BlackScholes_cpp1_ii_c58990ec’ collect2: ld returned 1 exit status make: […]

基于汉明重量的索引

假设我们有一个整数的bitsize n=4; 我所描述的问题是如何根据汉明权重及其知道bitsize值将数字索引到数组位置。 例如，一个包含16个用于bitsize 4的元素的数组将会/看起来像这样： |0|1|2|4|8|3|5|6|9|10|12|7|11|13|14|15| 元素按其汉明重量（必要）分组，并根据大小（不必要）排序。 只要你可以采取例如3（0011）做一些操作并取回索引5,5（0101） – > 6等，就不需要排序。 将出现n位的所有组合，并且不会重复。 例如， 3 bitsize将具有数组： |0|1|2|4|3|5|6|7| 我最好有一个没有循环的解决方案。 或任何讨论simillar解决方案的论文。 或者最后只是抛出任何关于如何做到这一点的想法。

在CUDA中使用常量

哪个是在CUDA中使用常量的最佳方法？ 一种方法是在常量内存中定义常量，如： // CUDA global constants __constant__ int M; int main(void) { … cudaMemcpyToSymbol(“M”, &M, sizeof(M)); … } 一种替代方法是使用C预处理器： #define M = … 我认为使用C预处理器定义常量要快得多。 那么在CUDA设备上使用常量内存的好处是什么？

使用CUDA添加大整数

我一直在GPU上开发一种加密算法，目前坚持使用算法来执行大整数加法。 大整数以通常的方式表示为一堆32位字。 例如，我们可以使用一个线程来添加两个32位字。 为简单起见，假设要添加的数字具有相同的长度和每个块的线程数==字数。 然后： __global__ void add_kernel(int *C, const int *A, const int *B) { int x = A[threadIdx.x]; int y = B[threadIdx.x]; int z = x + y; int carry = (z < x); /** do carry propagation in parallel somehow ? */ ………… z = z + newcarry; // update the resulting […]

在CUDA内核操作中添加Atomic的一些问题

我的kernel.cu类有问题 调用nvcc -v kernel.cu -o kernel.o我收到此错误： kernel.cu(17): error: identifier “atomicAdd” is undefined 我的代码： #include “dot.h” #include #include “device_functions.h” //might call atomicAdd __global__ void dot (int *a, int *b, int *c){ __shared__ int temp[THREADS_PER_BLOCK]; int index = threadIdx.x + blockIdx.x * blockDim.x; temp[threadIdx.x] = a[index] * b[index]; __syncthreads(); if( 0 == threadIdx.x ){ int sum = […]