`
暴风雪
  • 浏览: 388750 次
  • 性别: Icon_minigender_2
  • 来自: 杭州
社区版块
存档分类
最新评论

[CUDA]共享内存

    博客分类:
  • cuda
 
阅读更多

    对于GPU上启动的每个线程块上的共享内存,CUDA C编译器都会创建该变量的一个副本。同一线程块的每个线程都共享这块内存,但是线程无法看到也不能修改其他线程块中的共享内存。这样做可以使得一个线程块中的多个线程能够在计算上通信和协作。

    共享内存缓冲区驻留在物理GUP上,因此访问共享内存的延迟远远低于访问普通缓冲区的延迟。

    共享内存的声明方式是在前面加上  __shared__

    为了保持进程同步,可以使用cuda的函数__syncthreads();。这个函数的作用是为了确保线程块的每个线程都执行完__syncthreads();之前的语句后,才会执行下面的语句。

 

    出于易于理解,写了一个简单的程序,大致功能就是对于一列数,每四个数字进行逆转位置

 1 2 3 4 5 6 7 8  ----》 4 3 2 1 8 7 6 5

#include<cuda_runtime.h>
#include<windows.h>
#include<iostream>
using namespace std;
const int nMax = 50;
__global__ void exchangeKernel(float *aaa)
{
	int offset = threadIdx.x + blockDim.x * blockIdx.x;
	int x = threadIdx.x;
	__shared__ float tmp[4];
	int a = offset / 4;
	a = (a + 1) * 4 - (offset - a * 4) - 1; ///a为同一个block对应位置的offset
	tmp[x] = aaa[a];

	__syncthreads();
	aaa[offset] = tmp[x];
}

int main(){
	float a[nMax];
	float *devA;
	for (int i = 0; i < nMax; i++){
		a[i] = i;
	}
	cudaMalloc((void**)&devA, nMax*sizeof(float));

	cudaMemcpy(devA, a, nMax*sizeof(float), cudaMemcpyHostToDevice);

	exchangeKernel << <10, 4 >> >(devA );

	cudaMemcpy(a, devA, nMax*sizeof(float), cudaMemcpyDeviceToHost);


	for (int i = 0; i < 40; i++){
		cout << a[i] << " ";
	}cout << endl;
	cudaFree(devA);
	cin >> a[0];
	return 0;
}

 

1
1
分享到:
评论

相关推荐

    利用CUDA全局内存和共享内存实现图像的腐蚀膨胀

    3. **CUDA共享内存**:共享内存位于每个CUDA线程块内,可被该线程块内的所有线程高速访问。相比于全局内存,共享内存的访问速度快得多,但是容量有限。通过合理利用共享内存,可以减少全局内存的读写,提升性能。 4...

    CUDA——了解和使用共享内存

    ### CUDA——了解和使用共享内存 #### 一、CUDA内存模型概述 CUDA(Compute Unified Device Architecture)是一种由NVIDIA推出的并行计算平台和技术,旨在利用GPU进行通用计算。CUDA提供了丰富的编程模型,允许...

    【CUDA编程】opencv4 + CUDA 并行图像处理:图像均值滤波和图像反色

    二、cuda与OpenCV结合方法 三、代码实例:图像均值滤波和图像反色 3.1 代码 3.2 代码说明 3.3 网格大小与线程块大小的确定 3.3.1 网格与线程块大小的限制 3.3.2 如何确定网格大小与线程块大小? 3.4 并行与串行的...

    CUDA C编程权威指南.pdf

    **5.1 CUDA共享内存概述** - **5.1.1 共享内存** - 共享内存的概念及其在CUDA中的作用。 - 如何在核函数中使用共享内存。 - **5.1.2 共享内存分配** - 分配共享内存的基本方法。 - **5.1.3 共享内存存储体和...

    CUDA实现基于共享内存的位图显示

    CUDA实现基于共享内存的位图显示,线程同步

    基于共享内存的位图

    基于共享内存的位图,重点解释了为什么同步重要 GPU CUDA

    CUDA_Freshman

    CUDA Freshman This project is a set of ...5.1 CUDA共享内存概述 5.2 共享内存的数据布局 5.3 减少全局内存访问 5.4 合并的全局内存访问 5.5 常量内存 5.6 线程束洗牌指令 6.0 流和并发 6.1 流和事件概述 6.2 并发内

    cuda.zip_CUDA ustc_cuda_ustc cuda

    2. **全局内存、共享内存和寄存器**:CUDA中的内存层次结构包括全局内存、共享内存和寄存器。全局内存对所有线程可见,但速度相对较慢;共享内存位于线程块级别,速度较快,适合线程间协作;寄存器是最快的存储,但...

    数组逆序=全局内存版 VS 共享内存版

    相反,共享内存是位于每个CUDA线程块内的局部内存,其存取速度较快,但空间有限,只能被同一线程块内的线程共享。 全局内存版的数组逆序算法通常会将数组的所有元素分发给CUDA线程,每个线程负责一个或多个元素的...

    CUDA中使用共享内存的TILED矩阵乘法。一种高效快.zip

    CUDA中使用共享内存的TILED矩阵乘法。一种高效快.zip

    CUDA.rar_cuda_cuda GPU_cuda cpu_cuda学习_cuda学习资料

    3. **全局内存和共享内存**:全局内存是所有线程都能访问的内存,而共享内存仅对同一线程块内的线程可见。由于共享内存访问速度较快,优化程序时通常会利用它来提升性能。 4. **CUDA核函数**:核函数是运行在GPU上...

    基于CUDA架构的LBM共享内存计算优化.docx

    基于 CUDA 架构的 LBM 共享内存计算优化 在物理学领域中,晶格玻尔兹曼法(LBM)作为一种流体力学常用方法,越来越受到国内外学者的青睐。然而,传统基于 CPU 上的线性运算的运算速度问题也逐渐暴露出来。为了解决...

    CUDA任意维矩阵乘

    6. **优化**: 为了最大化性能,可能需要进行额外的优化,比如限制共享内存的使用、调整动态内存分配,以及根据硬件特性调整线程配置。 7. **释放资源**: 最后,别忘了在完成计算后释放GPU内存。 在压缩包提供的...

    cuda 中值滤波高性能计算MedianFilter.cu

    用cuda实现的3×3中值滤波,排序算法为二分法,利用共享内存,巧妙加速,算法执行效率非常高。 下载后带入数据直接用。

    CUDA by example (中文:GPU高性能编程CUDA实战)代码实例

    主要包括CUDA C/C++、CUDA内核函数、全局内存、共享内存、常量内存、纹理内存等概念。程序员需要了解如何定义和管理这些内存类型,以及如何在CPU和GPU之间传输数据。 2. **CUDA内核**:内核是CUDA程序中执行并行...

    CUDA_CUDA简介_

    3. **Memory Hierarchy**:CUDA提供多种内存类型,包括全局内存、共享内存、常量内存和纹理内存。合理使用这些内存可以显著提高性能,因为全局内存访问速度较慢,而共享内存和常量内存访问速度更快。 4. **...

    cuda安装与使用 cuda博客 cuda入门资料

    主要概念包括线程块(thread block)、网格(grid)、共享内存和全局内存等。 3. **CUDA内存管理**:理解不同类型的内存(全局、共享、常量、纹理、寄存器)至关重要,因为它们在性能上有很大差异。全局内存是所有...

    dot product点积 并行处理集合/共享/分布式内存 CUDA源代码

    详细对于CUDA中内存存储问题的编程,以点积的例子来展现,可以在GPU上运行。

    maopao1.rar_cuda sort_cuda 排序_冒泡排序cuda_基于CUDA的加速

    1. **设备内存管理**:CUDA程序需要在GPU上分配和管理内存,包括全局内存、共享内存、常量内存和纹理内存等。在CUDA冒泡排序中,数据通常存储在全局内存中,线程块内的数据可能利用共享内存进行局部优化。 2. **...

Global site tag (gtag.js) - Google Analytics