0.内存结构
详细介绍见下图:
1. 内存管理函数
1.1. CPU内存
alloca
alloca是在栈(stack)上申请空间,该变量离开其作用域之后被自动释放，无需手动调用释放函数，不提倡使用
malloc, free
函数原型为: void *malloc(size_t size);
分配所需的内存空间, 并返回一个指向它的指针
realloc
函数原型为: void *realloc(void *mem_address, unsigned int newsize);
更改已经配置的内存空间, 即更改由malloc()函数分配的内存空间的大小, 如果将分配的内存减少，realloc仅仅是改变索引的信息, 如果将分配的内存增大, 分为3种不同的情况.
如果newsize大小为 0，且mem_address指向一个已存在的内存块，则mem_address所指向的内存块会被释放, 相当于free(mem_address).
calloc
函数原型为: void *calloc(size_t n, size_t size)
好处: 不需要人为的计算空间的大小, 并且内部的数据会被初始化为0; 坏处: 运行速度慢.
与malloc, realloc的区别见此处.
memset
函数原型为: void *memset(void *buffer, int c, int count);
memset函数以字节为单位进行赋值，可以方便地清空一个结构类型的变量或数组。
memcpy
函数原型为: void *memcpy (void *dest, const void *src, size_t __n);
从源内存地址的起始位置开始拷贝若干个字节到目标内存地址中，即从源src中拷贝n个字节到目标dest中。
new, delete []
new的底层由malloc, free实现，在自由存储区上分配空间，malloc在堆上分配空间; 对其区分的实际意义并不大, 所以一般认为自由存储区就是堆.
new, delete会自动调用构造函数和析构函数，而malloc, free不会
std::shared_ptr<T>
智能指针, 不需要我们去操心内存释放的问题, 使用RALL机制实现了对内存的自动管理.
1.2. 会出现的问题
1.2.1. 内存泄露(开辟一段内存后, 忘记释放);
1.2.2. 内存碎片(这个是内存管理器层面的, 不好避免);
碎片现象: 在图 1d 的情况下，虽然总共剩余内存为 6，但却满足不了长度大于 4 的内存分配要求.
1.2.3. 野指针(释放完内存后, 忘记把内存地址指向NULL, 或者指针变量没有被初始化).
1.2. GPU显存
cudaMalloc, cuda_free（效率高）
cudaMemset
cudaMemcpy
cudaMallocManaged, cudaDeviceSynchronize, cuda_free(统一内存操作)
统一内存中创建一个托管内存池（CPU上有，GPU上也有），内存池中已分配的空间可以通过相同的指针直接被CPU和GPU访问，底层系统在统一的内存空间中自动的进行设备和主机间的传输。数据传输对应用是透明的，大大简化了代码[2]。
fftw_malloc, fftw_free
2. 代码测试
2.1. 完整代码如下
#include <iostream>
#include <alloca.h>
#include <memory>
#include <string.h>
#include <cuda_runtime.h>
#define COLOR(msg, code) "\033[1;" #code "m" msg "\033[0m"
#define GREEN(msg) COLOR(msg, 32)
const int kNumber = 10;
const int kReallocNumber = 5000;
__global__ void sum_with_self(int *data, int *result) {
int sum = 0;
for (int i = 0; i < kNumber; ++i) {
sum += data[i];
}
*result = sum;
}
int main(int argc, char *argv[]) {
puts(GREEN("--------------normal stack--------------"));
int host_stack_data[kNumber];
for (int i = 0; i < kNumber; ++i) {
host_stack_data[i] = i * 10;
std::cout << &host_stack_data[i] << "->" << host_stack_data[i] << std::endl;
}
puts(GREEN("--------------alloca--------------"));
int *stack_ptr = (int *)alloca(kNumber * sizeof(int)); // 非常不建议使用alloca函数来分配栈上的空间
std::cout << "stack_ptr 栈上分配空间的地址：" << stack_ptr << std::endl;
std::cout << "&stack_ptr 的地址：" << &stack_ptr << std::endl;
memcpy(stack_ptr, host_stack_data, kNumber * sizeof(int));
for (int i = 0; i < kNumber; ++i) {
std::cout << stack_ptr + i << "->" << *(stack_ptr + i) << std::endl; // 如果不进行memcpy，输出可能是0，可能是其他随机值（因为这一部分未分配之前可能存有先前的数据）
}
puts(GREEN("--------------malloc--------------"));
int *heap_ptr = (int *)malloc(kNumber * sizeof(int)); // 不能在堆上分配特别大的空间，那样的地址会显的是在栈上的地址
std::cout << "heap_ptr 指向的地址->" << heap_ptr << std::endl;
std::cout << "heap_ptr 本身在栈上的地址：" << &heap_ptr << std::endl;
memset(heap_ptr, 1, kNumber * sizeof(int)); // 由于memset是按照字节来赋值，所以这里实际是：0b 00000001000000010000000100000001，对应十进制的16843009
for (int i = 0; i < kNumber; ++i) {
std::cout << heap_ptr + i << "->" << *(heap_ptr + i) << std::endl;
}
heap_ptr = (int *)realloc(heap_ptr, kReallocNumber * sizeof(int));
std::cout << "realloc_heap_ptr 指向的地址->" << heap_ptr << std::endl;
free(heap_ptr);
std::cout << "After free, the first element value still is: " << *(heap_ptr) << std::endl; // 为什么还可以解引用*???
heap_ptr = NULL; // 为了避免上一句的野指针, 所以要将其指向NULL
// new的底层由malloc, free实现，在自由存储区上分配空间，malloc在堆上分配空间
// TODO: 一般认为自由存储区就是堆，对其区分的实际意义并不大
// new, delete会自动调用构造函数和析构函数，而malloc, free不会
puts(GREEN("--------------new--------------"));
int *new_ptr = new int[kNumber];
std::cout << "new_ptr 指向的地址->" << new_ptr << std::endl;
std::cout << "new_ptr 本身在栈上的地址：" << &new_ptr << std::endl;
delete [] new_ptr;
new_ptr = NULL;
puts(GREEN("--------------shared_ptr--------------"));
std::shared_ptr<int> shareded_ptr(new int);
std::cout << "shareded_ptr 指向的地址->" << shareded_ptr << std::endl;
std::cout << "shareded_ptr 指向的地址->" << shareded_ptr.get() << std::endl;
std::cout << "shareded_ptr 本身在栈上的地址：" << &shareded_ptr << std::endl;
puts(GREEN("--------------create host data--------------"));
int *host_heap_data = (int*)malloc(kNumber * sizeof(int));
for (int i = 0; i < kNumber; ++i) {
host_heap_data[i] = i;
std::cout << &host_heap_data[i] << "->" << host_heap_data[i] << std::endl;
}
puts(GREEN("--------------cudaMalloc--------------"));
int *gpudata_ptr = NULL, *gpuresult_ptr = NULL;
cudaMalloc((void **)&gpudata_ptr, kNumber * sizeof(int));
cudaMalloc((void **)&gpuresult_ptr, sizeof(int));
std::cout << "分配显存后gpudata_ptr 指向的地址->" << gpudata_ptr << std::endl; // 这里是显存的地址
std::cout << "gpudata_ptr 本身在栈上的地址：" << &gpudata_ptr << std::endl;
#if 1
cudaMemcpy(gpudata_ptr, host_heap_data, kNumber * sizeof(int), cudaMemcpyHostToDevice);
#else
for (int i = 0; i < kNumber; ++i) {
gpudata_ptr[i] = i * 2; // 分配显存后，不能直接赋值，需要借助cudaMemcpy函数；而 cudaMallocManaged分配的显存可以直接赋值。
}
#endif
// for (int i = 0; i < kNumber; ++i) {
// std::cout << gpudata_ptr + i << "->" << *(gpudata_ptr + i) << std::endl;
// }
// TODO: some operation on kernel
sum_with_self<< <1, 1>> >(gpudata_ptr, gpuresult_ptr);
int result;
cudaMemcpy(&result, gpuresult_ptr, sizeof(int), cudaMemcpyDeviceToHost);
std::cout << "The sum of host data is: " << result << std::endl;
cudaFree(gpudata_ptr);
cudaFree(gpuresult_ptr);
free(host_heap_data);
host_heap_data = NULL;
puts(GREEN("--------------cudaMallocManaged--------------"));
int *manageddata_ptr = NULL, *managedresult_ptr = NULL;
cudaMallocManaged((void **)&manageddata_ptr, kNumber * sizeof(int)); // 统一内存管理：并不减少运行时间，只是将内存与设备内存之间的拷贝透明化，使得编程更清晰
cudaMallocManaged((void **)&managedresult_ptr, sizeof(int)); // 统一内存管理：并不减少运行时间，只是将内存与设备内存之间的拷贝透明化，使得编程更清晰
std::cout << "manageddata_ptr 指向的地址->" << manageddata_ptr << std::endl; // 这里是显存的地址
std::cout << "manageddata_ptr 本身在栈上的地址：" << &manageddata_ptr << std::endl;
for (int i = 0; i < kNumber; ++i) {
manageddata_ptr[i] = i;
std::cout << manageddata_ptr + i << "->" << *(manageddata_ptr + i) << std::endl;
}
// TODO: some operation on kernel
sum_with_self<< <1, 1>> >(manageddata_ptr, managedresult_ptr);
cudaDeviceSynchronize();
std::cout << "The sum of managed data is: " << *managedresult_ptr << std::endl; // 可直接解引用
cudaFree(manageddata_ptr);
cudaFree(managedresult_ptr);
puts(GREEN("--------------main function end!--------------"));
return 0;
}
2.2. 程序运行结果
--------------normal stack--------------
0x7fff1158fb00->0
0x7fff1158fb04->10
0x7fff1158fb08->20
0x7fff1158fb0c->30
0x7fff1158fb10->40
0x7fff1158fb14->50
0x7fff1158fb18->60
0x7fff1158fb1c->70
0x7fff1158fb20->80
0x7fff1158fb24->90
--------------alloca--------------
stack_ptr 栈上分配空间的地址：0x7fff1158fa20
&stack_ptr 的地址：0x7fff1158fa90
0x7fff1158fa20->0
0x7fff1158fa24->10
0x7fff1158fa28->20
0x7fff1158fa2c->30
0x7fff1158fa30->40
0x7fff1158fa34->50
0x7fff1158fa38->60
0x7fff1158fa3c->70
0x7fff1158fa40->80
0x7fff1158fa44->90
--------------malloc--------------
heap_ptr 指向的地址->0x1d5d220
heap_ptr 本身在栈上的地址：0x7fff1158fa98
0x1d5d220->16843009
0x1d5d224->16843009
0x1d5d228->16843009
0x1d5d22c->16843009
0x1d5d230->16843009
0x1d5d234->16843009
0x1d5d238->16843009
0x1d5d23c->16843009
0x1d5d240->16843009
0x1d5d244->16843009
realloc_heap_ptr 指向的地址->0x1d5d220
After free, the first element value still is: 16843009
--------------new--------------
new_ptr 指向的地址->0x1d5d220
new_ptr 本身在栈上的地址：0x7fff1158faa0
--------------shared_ptr--------------
shareded_ptr 指向的地址->0x1d5d250
shareded_ptr 指向的地址->0x1d5d250
shareded_ptr 本身在栈上的地址：0x7fff1158faf0
--------------create host data--------------
0x1d5d220->0
0x1d5d224->1
0x1d5d228->2
0x1d5d22c->3
0x1d5d230->4
0x1d5d234->5
0x1d5d238->6
0x1d5d23c->7
0x1d5d240->8
0x1d5d244->9
--------------cudaMalloc--------------
分配显存后gpudata_ptr 指向的地址->0x7f6dcfa00000
gpudata_ptr 本身在栈上的地址：0x7fff1158faa8
The sum of host data is: 45
--------------cudaMallocManaged--------------
manageddata_ptr 指向的地址->0x7f6dc6000000
manageddata_ptr 本身在栈上的地址：0x7fff1158fab8
0x7f6dc6000000->0
0x7f6dc6000004->1
0x7f6dc6000008->2
0x7f6dc600000c->3
0x7f6dc6000010->4
0x7f6dc6000014->5
0x7f6dc6000018->6
0x7f6dc600001c->7
0x7f6dc6000020->8
0x7f6dc6000024->9
The sum of managed data is: 45
--------------main function end!--------------
3. 参考资料
[1]. 统一内存向量加法
[2]. Cuda内存管理
[3]. 什么是代码区、常量区、静态区（全局区）、堆区、栈区？
[4]. Linux虚拟地址空间布局以及进程栈和线程栈总结
[5]. 第3篇-C/C++ 类和内存分配(前)
[6]. 第2篇:C/C++ 内存布局与程序栈
[7]. calloc、malloc、realloc函数的区别及用法
[8]. c++内存泄露内存溢出和野指针
1848
被折叠的 条评论
为什么被折叠?
到【灌水乐园】发言
