作者tmbyksdG (雨神妹妹的男朋友)
看板C_and_CPP
标题[问题] CUDA新手上路 平行度及硬体相关问题
时间Sun Nov 27 18:04:32 2016
开发平台(Platform): (Ex: Win10, Linux, ...)
Linux上安装CUDA环境 （CUDA版本为8.0 运算能力为3.7）(Tesla K80)
编译器(Ex: GCC, clang, VC++...)+目标环境(跟开发平台不同的话需列出)
NVCC
额外使用到的函数库(Library Used): (Ex: OpenGL, ...)
问题(Question)：
硬体方面：
1.我执行deviceQuery侦测到2个device(device0:Tesla K80, device1:K80),估狗发现K80
是由两个GK210核心所组成，那我侦测到的device是指有两个K80（4个GK210）的意思吗？
还是侦测到的两个device其实就是GK210?
2.13个SMX,总共有2496个cores,所以我一次可以同时做运算的数量是否为2496个threads?
软体方面：
我写了一个64 * 64的矩阵乘法，我想测试不同的block & thread数量去做运算，哪个执
行时间会比较短，我试了两种block的配置（thread数量刚好为4096，一个thread执行一
个输出矩阵的一个element）。
（1）dim3 dimBlock(32, 32);
dim3 dimGrid(2, 2);
这个配置是以下附的程式码配置，执行结果是正常的。
（2）dim3 dimBlock(4, 128);
dim3 dimGrid(1, 8);
换成这样配置编译之後，结果却只有大约三分之一是正常的值，其余却都是0
，我的threadIdx.x & threadIdx.y都是由0-1023，请问这是发生了什麽错误吗？
另外，我想知道执行运算时thread做了什麽事情，每个thread里面装的是什麽东西呢，是
我输入的资料吗？还是加跟乘这两个指令呢？
喂入的资料(Input)：
预期的正确结果(Expected Output)：
错误结果(Wrong Output)：
程式码(Code)：(请善用置底文网页, 记得排版)
#include <stdio.h>
#include <stdlib.h>
#include <time.h>
#include <cuda_runtime.h>
#include "cuda_error.h"
#define N 64
__global__ void MatrixMulKernel(float* C, float* A, float* B, int n)
{
float Pvalue = 0;
int tx = threadIdx.x + blockIdx.x * blockDim.x;
int ty = threadIdx.y + blockIdx.y * blockDim.y;
for(int k = 0; k < n; k++)
{
float Aelement = A[ty * n + k];
float Belement = B[k * n + tx];
Pvalue += Aelement * Belement;
}
C[ty * n + tx] = Pvalue;
__syncthreads();
}
int main()
{
float A[N][N];
float B[N][N];
float C[N][N];
size_t size = N * N * sizeof(float);
int i, j;
float *gA, *gB, *gC;
for(i = 0; i < N; i++)
{
for(j = 0; j < N; j++)
{
A[i][j] = (float) (1 + ( rand() % 9 ));
B[i][j] = (float) (1 + ( rand() % 9 ));
}
}
/*allocate memory*/
cudaMalloc(&gA, size);
cudaMalloc(&gB, size);
cudaMalloc(&gC, size);
puts(cudaGetErrorString(cudaGetLastError()));
cudaMemcpy(gA, A, size, cudaMemcpyHostToDevice);
cudaMemcpy(gB, B, size, cudaMemcpyHostToDevice);
puts(cudaGetErrorString(cudaGetLastError()));
float etime;
cudaEvent_t start, stop;
cudaEventCreate(&start, cudaEventDefault);
cudaEventCreate(&stop, cudaEventDefault);
dim3 dimBlock( 32, 32 );
dim3 dimGrid( 2, 2 );
cudaEventRecord(start);
/*call GPU kernel*/
MatrixMulKernel<<<dimGrid,dimBlock>>>(gC, gA, gB, N);
cudaEventRecord(stop);
cudaEventSynchronize(stop);
cudaEventElapsedTime(&etime, start, stop);
printf("%f ms\n", etime);
puts(cudaGetErrorString(cudaGetLastError()));
cudaMemcpy(C, gC, size, cudaMemcpyDeviceToHost);
puts(cudaGetErrorString(cudaGetLastError()));
cudaFree(gA);
cudaFree(gB);
cudaFree(gC);
return 0;
}
补充说明(Supplement)：
请版上的各位先进指导一下我，谢谢。另外，手机排版请见谅。
--
※ 发信站: 批踢踢实业坊(ptt.cc), 来自: 59.115.193.128
※ 文章网址: https://webptt.com/cn.aspx?n=bbs/C_and_CPP/M.1480241075.A.D3D.html
1F：→ opl164: 第二个状况的tx,ty 你可以印出来看看跟的一个会不一样吧 11/28 12:03
2F：推 a1u1usul3: 硬体1应该是两个GK210的意思 11/28 14:31
3F：→ a1u1usul3: 硬体2可以说是对的，精确来说是78个wrap吧 11/28 14:32
4F：→ a1u1usul3: 软体问题可以用cuda-memcheck跑看看，是不是access到奇 11/28 14:34
5F：→ a1u1usul3: 怪的地方去了 11/28 14:34
6F：→ arXiv: 第二点错了程式跟着错,请先读文件cuda-c-programming-guide 11/29 00:09