Cuda C++设计:编译时大小未知的可重用类
Cuda C++ design: reusable class with unknown compile-time size
>我正在寻找一种方便的设计,以便能够在编译时大小未知的设备上使用类。 只需要将此类的一个实例发送到设备,为此应该对 cudaMalloc 和 cudaMemcpy 进行一次调用(理想情况下(。
该类的主机版本如下所示:
Class A {
public:
A(int size) : table(size) {
// some useful initialization of table
}
double get(int i) const {
// return some processed element from table
}
private:
std::vector<int> table;
};
内核:
__global__ void kernel(const A *a){
int idx = threadIdx.x + blockDim.x * blockIdx.x;
a->get(idx); // do something useful with it
}
到目前为止,我设计类的设备版本的方式是这样的:
const int sizeMax = 1000;
Class A {
public:
A(int size) {
// size checking + some useful initialization of table
}
__host__ __device__
double get(int i) const {
//
}
private:
int table[sizeMax];
};
和客户端代码:
A a(128);
A* da;
cudaMalloc((void**)&da, sizeof(A));
cudaMemcpy(da, &a, sizeof(A), cudaMemcpyHostToDevice);
kernel<<<1, 32>>>(da);
cudaDeviceSynchronize();
cudaFree(da);
这是相当丑陋的,因为:
它浪费带宽,因为- 必须使用太大的尺寸Max,以便 为了安全起见
- 该类没有关闭进行修改,sizeMax 的值将 不可避免地需要在某个时候提出
有没有其他方法可以在不对性能产生负面影响的情况下以更清洁的方式实现相同的目标?需要明确的是,我只需要类的设备版本,第一个版本只是等效的非 CUDA 代码来说明表大小应该是动态的事实。
在我的评论中,我说:
- 表的单独主机和设备存储,包含在类中,两者都是动态分配的。2. 在构造函数中动态分配表存储大小,而不是在客户端代码中。这还可能包括在必要时调整大小。3. 类方法的差异化,以使用数据的主机副本或指向数据的设备副本(即指针(,具体取决于方法是在主机还是设备代码中执行 4.一种将数据从主机复制到设备的方法,反之亦然,因为类上下文从主机移动到设备,反之亦然。
以下是我想到的一个例子:
#include <stdio.h>
#include <assert.h>
#include <cuda_runtime_api.h>
#include <iostream>
template <typename T>
class gpuvec{
private:
T *h_vec = NULL;
T *d_vec = NULL;
size_t vsize = 0;
bool iscopy;
public:
__host__ __device__
T * data(){
#ifndef __CUDA_ARCH__
return h_vec;
#else
return d_vec;
#endif
}
__host__ __device__
T& operator[](size_t i) {
assert(i < vsize);
return data()[i];}
void to_device(){
assert(cudaMemcpy(d_vec, h_vec, vsize*sizeof(T), cudaMemcpyHostToDevice) == cudaSuccess);}
void to_host(){
assert(cudaMemcpy(h_vec, d_vec, vsize*sizeof(T), cudaMemcpyDeviceToHost) == cudaSuccess);}
gpuvec(gpuvec &o){
h_vec = o.h_vec;
d_vec = o.d_vec;
vsize = o.vsize;
iscopy = true;}
void copy(gpuvec &o){
free();
iscopy = false;
vsize = o.vsize;
h_vec = (T *)malloc(vsize*sizeof(T));
assert(h_vec != NULL);
assert(cudaMalloc(&d_vec, vsize*sizeof(T)) == cudaSuccess);
memcpy(h_vec, o.h_vec, vsize*sizeof(T));
assert(cudaMemcpy(d_vec, o.d_vec, vsize*sizeof(T), cudaMemcpyDeviceToDevice) == cudaSuccess);}
gpuvec(size_t ds) {
assert(ds > 0);
iscopy = false;
vsize = ds;
h_vec = (T *)malloc(vsize*sizeof(T));
assert(h_vec != NULL);
assert(cudaMalloc(&d_vec, vsize*sizeof(T)) == cudaSuccess);}
gpuvec(){
iscopy = false;
}
~gpuvec(){
if (!iscopy) free();}
void free(){
if (d_vec != NULL) cudaFree(d_vec);
d_vec = NULL;
if (h_vec != NULL) ::free(h_vec);
h_vec = NULL;}
__host__ __device__
size_t size() {
return vsize;}
};
template <typename T>
__global__ void test(gpuvec<T> d){
for (int i = 0; i < d.size(); i++){
d[i] += 1;
}
}
int main(){
size_t ds = 10;
gpuvec<int> A(ds);
A.to_device();
test<<<1,1>>>(A);
A.to_host();
for (size_t i = 0; i < ds; i++)
std::cout << A[i];
std::cout << std::endl;
gpuvec<int> B;
B.copy(A);
A.free();
B.to_device();
test<<<1,1>>>(B);
B.to_host();
for (size_t i = 0; i < ds; i++)
std::cout << B[i];
std::cout << std::endl;
B.free();
}
我相信可以提出很多批评。 这可能不符合关于"矢量语法"应该是什么的任何特定观点。 此外,我确信有些用例它没有涵盖,并且可能包含彻头彻尾的缺陷。 要创建强大的主机/设备矢量实现,可能需要与推力主机和设备矢量一样多的工作和复杂性。 然而,我并不是说推力矢量是这个问题似乎在问什么的直接答案。
基于Robert Crovella的回答,这里有一个简化的(仅限设备,所以忽略第3点和第4点(工作解决方案:
Class A {
public:
A(int size) : table(size) {
// some useful initialization of table
cudaMalloc((void**)&dTable, sizeof(int) * size);
cudaMemcpy(dTable, &table[0], sizeof(int) * size, cudaMemcpyHostToDevice);
}
~A() {
cudaFree(dTable);
}
__device__
double get(int i) const {
// return some processed element of dTable
}
private:
std::vector<int> table;
int *dTable;
};
内核和客户端代码完全相同。
相关文章:
- 继承期间显示未知行为的子类
- 设计一个只能由特定类实例化的类(如果可能的话,通过make_unique)
- 输入中的字符串数未知(以字母表示)
- 具有未知值时的时间复杂性
- 派生类是否可以在抽象工厂设计模式中具有数据成员
- 将寄存器设计成可由C和C++访问的外设的最佳实践
- 询问在设计我的手臂模拟器功能表示格式1
- 链表中写入访问冲突的未知原因
- 正在生成未知类实例
- Bjarne Stroustrup Book - std_lib_facilities.h - 不起作用(未知类型名称)
- 有没有比在库中添加一个并非由所有派生类实现的新虚拟函数更好的设计实践
- 读取文件时运行时的未知行为
- 资源管理设计模式
- 代码在我的计算机上运行良好,但是在将其提交给coursera时遇到未知的信号11问题
- 多态杆件变量 - 类设计
- 初始化多个未知基类
- 使用 make 编译 MPI,几个命名空间错误,例如"错误:未知类型名称'使用'?
- Cuda C++设计:编译时大小未知的可重用类
- C++中的OO设计-用未知类型的子对象装饰父对象
- 如何在c++中从dll调用设计时未知的函数