すでに同じ名前のスレッドが 10 個もあるフォーラム スレッドを作成するのは少し気の毒ですが、それらすべてを確認した後、ほとんどのガイドと一緒に、まだ問題を理解できません。
char 配列 [40090][11] があり、その要素の 2 つの可能な組み合わせごとにカスタム操作を行いたい (11 バイトの束全体を要素と見なします)。これは一種の mmatrix 乗算であり、行列は 1 列 1 行であることを理解しています。
SDKマニュアルに従って、出力要素ごとに1つのスレッドを持つことを考えています。40090=19*2110 なので、以下を使用しています:
dim3 threadsperblock(19,19);
dim3 blocksingrid(2110,2110);
xkernel<<<blocksingrid, threadsperblock>>>(dev_b2);
質問 1 : これでよろしいですか?
さて、私はSDKの手動の例に忠実に従っていると思います（共有メモリを使用しているものではありません）。ただし、データに対して必要な操作の一部を敢えて行うと、非常に役に立たないエラー 30 が返されます。不明なエラーです。では、質問 2 : 何が間違っているのでしょうか。注: カーネルがどこにも何も保存しないことは無視してください。
#include "cuda_runtime.h"
#include "device_launch_parameters.h"
#include <cstdlib>
#include <iostream>
#include <fstream>
#include <iomanip>
#include <ctime>
#include <stdio.h>
using namespace std;
cudaError_t cudafunct(void);
__global__ void xkernel(char * dev_b2);
__device__ unsigned char typecheck(unsigned char type1,unsigned char type2);
#define b2c 40090
unsigned char block2[b2c][11];//
//unsigned int i,b1,b2,counter=0;//Block(2),Piece,Rotation,Type(of block2),InterconnectinTriangle
//unsigned char *block4,type=0;
ofstream ofile;
int main()
{
ifstream block2file("2.blk",ios::binary);
block2file.read((char*)(&block2),b2c*11);
block2file.close();
//block4=new unsigned char[200000000];//200MB will do, better than doing constant reallocs
cudaError_t cudaStatus = cudafunct();
if (cudaStatus != cudaSuccess) {
fprintf(stderr, "cudafunct failed!");
system("PAUSE");
return 1;
}
/*
// cudaDeviceReset must be called before exiting in order for profiling and
// tracing tools such as Nsight and Visual Profiler to show complete traces.
cudaStatus = cudaDeviceReset();
if (cudaStatus != cudaSuccess) {
fprintf(stderr, "cudaDeviceReset failed!");
return 1;
}*/
cout<<"Sequence end. Saving to file...\n";
//ofile.open("blk4.et2",ios::binary);
//ofile.write((char*)block4,17*counter);
//ofile.close();
int t=clock();
//cout<<"\nFound a total of "<<counter<<" block4s.\nTime elapsed: "<<t<<" clocks / "<<(double)t/(double)CLOCKS_PER_SEC<<" seconds\n";
system("PAUSE");
}
// Helper function for using CUDA to add vectors in parallel.
cudaError_t cudafunct(void)
{
char *dev_b2 = 0;
cudaError_t cudaStatus;
cudaStatus = cudaMalloc((void**)&dev_b2, sizeof(block2));
if (cudaStatus != cudaSuccess) {
fprintf(stderr, "cudaMalloc failed!");
goto Error;
}
cudaStatus = cudaMemcpy(dev_b2, block2, sizeof(block2), cudaMemcpyHostToDevice);
if (cudaStatus != cudaSuccess) {
fprintf(stderr, "cudaMemcpy failed!");
goto Error;
}
dim3 threadsperblock(19,19);
dim3 blocksingrid(2110,2110);
xkernel<<<blocksingrid, threadsperblock>>>(dev_b2);
// cudaDeviceSynchronize waits for the kernel to finish, and returns
// any errors encountered during the launch.
cudaStatus = cudaDeviceSynchronize();
if (cudaStatus != cudaSuccess) {
fprintf(stderr, "cudaDeviceSynchronize returned error code %d after launching xkernel!\n", cudaStatus);
goto Error;
}
/*
// Copy output vector from GPU buffer to host memory.
cudaStatus = cudaMemcpy(c, dev_c, size * sizeof(int), cudaMemcpyDeviceToHost);
if (cudaStatus != cudaSuccess) {
fprintf(stderr, "cudaMemcpy failed!");
goto Error;
}*/
Error:
cudaFree(dev_b2);
return cudaStatus;
}
__global__ void xkernel(char *dev_b2)
{
int i = blockIdx.x * blockDim.x + threadIdx.x;
int j = blockIdx.y * blockDim.y + threadIdx.y;
/*for(int k=0;k<11;k++)
{
lb2[0][k]=dev_b2[i*b2c+k];
lb2[1][k]=dev_b2[j*b2c+k];
}*/
int b00;
b00=dev_b2[i*b2c];
//int type=typecheck(dev_b2[i*b2c+4],dev_b2[j*b2c+4]);
//if(!j && !(i % 100))cout<<setw(6)<<i<<" / "<<jc<<" ("<<setw(10)<<(float)100*i/jc<<" % )"<<endl;
/*if(
(dev_b2[i*b2c+7]!=dev_b2[j*b2c+9])||//SW~NW
(dev_b2[i*b2c+6]!=dev_b2[j*b2c+10])//SE~NE
) return;
if( (type=typecheck(dev_b2[i*b2c+4],dev_b2[j*b2c+4]) ) ==255) return;*/
/*if(
(dev_b2[i*b2c+0]==dev_b2[j*b2c+0])||//1st=3rd
(dev_b2[i*b2c+0]==dev_b2[j*b2c+2])||//1st=4th
(dev_b2[i*b2c+2]==dev_b2[j*b2c+0])||//2nd=3rd
(dev_b2[i*b2c+2]==dev_b2[j*b2c+2])//2nd=4th
) return;*/
/*
*(block4+counter*17+0)=b2[i][0];//1st piece
*(block4+counter*17+1)=b2[i][1];//1st rotation
*(block4+counter*17+2)=b2[i][2];//2nd piece
*(block4+counter*17+3)=b2[i][3];//2nd rotation
*(block4+counter*17+4)=b2[j][0];//3rd piece
*(block4+counter*17+5)=b2[j][1];//3rd rotation
*(block4+counter*17+6)=b2[j][2];//4th piece
*(block4+counter*17+7)=b2[j][3];//4th rotation
*(block4+counter*17+8)=type;
*(block4+counter*17+9)=b2[i][5];//Right frame colours, down->up
*(block4+counter*17+10)=b2[j][5];
*(block4+counter*17+11)=b2[j][6];//Up frame colours, right->left
*(block4+counter*17+12)=b2[j][7];
*(block4+counter*17+13)=b2[j][8];//Left frame colours, up->down
*(block4+counter*17+14)=b2[i][8];
*(block4+counter*17+15)=b2[i][9];//Down frame colours, left->right
*(block4+counter++*17+16)=b2[i][10];*/
}
__device__ unsigned char typecheck(unsigned char type1,unsigned char type2)
{//Warning! Previous error! First partenthesis is t*2* = upper piece!
if( (type1==4) && (type2==0) ) return 0;
if( (type1==6) && (type2==1) ) return 1;
if( (type1==2) && (type2==6) ) return 2;
if( (type1==3) && (type2==4) ) return 3;
if( (type1==4) && (type2==4) ) return 4;
if( (type1==8) && (type2==5) ) return 5;
if( (type1==6) && (type2==6) ) return 6;
if( (type1==7) && (type2==8) ) return 7;
if( (type1==8) && (type2==8) ) return 8;
if( (type1==9) && (type2==8) ) return 9;
if( (type1==10) && (type2==8) ) return 10;
if( (type1==8) && (type2==11) ) return 11;
if( (type1==8) && (type2==12) ) return 12;
if( (type1==8) && (type2==13) ) return 13;
return 255;
}