0

コードはエラーなしでコンパイルされますが、デバッグ時にアクセス違反のメッセージが表示されます。以下のコードの何が問題になっているのかを誰かが指摘できますか？

私のコードは、実際には、再帰的な非線形方程式である同じ方程式の1000インスタンスの1000回の反復を実行します。目的は、複数の（反復）方程式を並行して実行する機能を評価することです。

#include <stdlib.h>
#include <stdio.h>
#include <string.h>
#include <math.h>
#include <conio.h>
#include <cuda.h>
#include <cutil.h>
#include <time.h> 

#define TOTAL_THREADS 1024
#define THREADS_PER_BLOCK 256
#define TOTAL_BLOCKS 4
#define VALUES_PER_THREAD 1000
#define THETA_VALUES_PER_THREAD 15

__global__ void my_compute(float *y_d, float *theta_d, float *u_d) 
{
   int offset = ((blockIdx.x * blockDim.x) + threadIdx.x) * VALUES_PER_THREAD;
   int theta_offset = ((blockIdx.x * blockDim.x) + threadIdx.x) * THETA_VALUES_PER_THREAD;

   for (int i = 7; i < 1000; i++) {
      y_d[offset + i] = theta_d[theta_offset + 0] * y_d[offset + i - 1] +
                        theta_d[theta_offset + 1] * y_d[offset + i - 3] +
                        theta_d[theta_offset + 2] * u_d[offset + i - 5] * u_d[offset + i - 4] +
                        theta_d[theta_offset + 3] + 
                        theta_d[theta_offset + 4] * u_d[offset + i - 6] +
                        theta_d[theta_offset + 5] * u_d[offset + i - 4] * y_d[offset + i - 6] +
                        theta_d[theta_offset + 6] * u_d[offset + i - 7] +
                        theta_d[theta_offset + 7] * u_d[offset + i - 7] * u_d[offset + i - 6] +
                        theta_d[theta_offset + 8] * y_d[offset + i - 4] +
                        theta_d[theta_offset + 9] * y_d[offset + i - 5] +
                        theta_d[theta_offset + 10] * u_d[offset + i - 4] * y_d[offset + i - 5] +
                        theta_d[theta_offset + 11] * u_d[offset + i - 4] * y_d[offset + i - 2] +
                        theta_d[theta_offset + 12] * u_d[offset + i - 7] * u_d[offset + i - 3] +
                        theta_d[theta_offset + 13] * u_d[offset + i - 5] +
                        theta_d[theta_offset + 14] * u_d[offset + i - 4];
   }
}

int main(void) 
{
   float y[1000000];
   FILE * fpoo;
   FILE * u;
   float theta[15000];
   float u_data[1000000];
   float *y_d;
   float *theta_d;
   float *u_d;
   cudaEvent_t start, stop;
   float time;

   cudaEventCreate(&start);
   cudaEventCreate(&stop);

   //memory allocation
   cudaMalloc((void **) &y_d, 1000000 * sizeof(float));
   cudaMalloc((void **) &theta_d, 15000 * sizeof(float));
   cudaMalloc((void **) &u_d, 1000000 * sizeof(float));
   cudaEventRecord(start, 0);

   // importing data for theta and input of model//
   fpoo = fopen("c:\\Fly_theta.txt", "r");
   u = fopen("c:\\Fly_u.txt", "r");

   for (int i = 0; i < 1000; i++) {
      for (int j = 0; j < 15; j++)
         fscanf(fpoo, "%f\n", &theta[15 * i + j]);
   } 
   for (int i = 0; i < 1000; i++) {
      for (int j = 0; j < 1000; j++)
         fscanf(u, "%f\n", &u_data[1000 * i + j]);
   }
   //initialising past input with the value of zero//
   for (int i = 0; i < 1000; i++) {
      for (int j = 0; j < 8; j++)
         y[8 * i + j] = 0;
   }
   cudaMemcpy(y_d, y, 1000000 * sizeof(float), cudaMemcpyHostToDevice);

   cudaMemcpy(theta_d, theta, 15000 * sizeof(float), cudaMemcpyHostToDevice);
   cudaMemcpy(u_d, u_data, 1000000 * sizeof(float), cudaMemcpyHostToDevice);

   //calling kernel function//
   my_compute <<< 4, 256 >>> (y_d, theta_d, u_d);

   cudaMemcpy(y, y_d, 1000000 * sizeof(float), cudaMemcpyDeviceToHost);

   for (int i = 0; i < 1000; i++) {
      for (int j = 0; j < 1000; j++)
         printf("%f", y[1000 * i + j]);
   } 
   cudaEventRecord(stop, 0);
   cudaEventSynchronize(stop);
   cudaEventElapsedTime(&time, start, stop);
   cudaEventDestroy(start);
   cudaEventDestroy(stop);
   printf("Time to generate:  %3.1f ms \n", time);

   cudaFree(y_d);
   cudaFree(theta_d);
   cudaFree(u_d);

   fclose(u);
   fclose(fpoo);

   //fclose();
   _getche();
   return 0;
}
4

4 に答える 4

1

ユージーンはその場にいます。問題は、theta_dの元の割り当てです。

theta_dのバイトサイズは次のようになります。

 THREADS_PER_BLOCK * TOTAL_BLOCKS * THETA_VALUES_PER_THREAD * ELEMENT_SIZE = 
 256 * 4 * 15 * 4 =  61440

選択したアクセスパターンが配列のすべての要素にアクセスできることを確認します。アプリでは、theta_dは次のように宣言されています：

cudaMalloc((void **) &theta_d, 15000 * sizeof(float));

事実上、theta_dのサイズはわずか60000バイトです。その結果、カーネルはバイト60000を超えるすべてのアクセスでエラーになります。

1つの解決策は、すでに#definedした定数に基づいて割り当てのサイズを設定することです。15000を「THREADS_PER_BLOCK * TOTAL_BLOCKS * THETA_VALUES_PER_THREAD」に置き換えると、割り当てが正しくサイズ設定されます。次の主な機能を試してください。

int main(void) 
{
   float y[1000000];
   FILE * fpoo;
   FILE * u;
   float theta[THREADS_PER_BLOCK * TOTAL_BLOCKS * THETA_VALUES_PER_THREAD];
   float u_data[1000000];
   float *y_d;
   float *theta_d;
   float *u_d;
   cudaEvent_t start, stop;
   float time;

   cudaEventCreate(&start);
   cudaEventCreate(&stop);

   //memory allocation
   cudaMalloc((void **) &y_d, 1000000 * sizeof(float));
   cudaMalloc((void **) &theta_d, THREADS_PER_BLOCK * TOTAL_BLOCKS * THETA_VALUES_PER_THREAD* sizeof(float));
   cudaMalloc((void **) &u_d, 1000000 * sizeof(float));
   cudaEventRecord(start, 0);

   // importing data for theta and input of model//
   fpoo = fopen("c:\\Fly_theta.txt", "r");
   u = fopen("c:\\Fly_u.txt", "r");

   for (int i = 0; i < 1000; i++) {
      for (int j = 0; j < 15; j++)
         fscanf(fpoo, "%f\n", &theta[15 * i + j]);
   } 
   for (int i = 0; i < 1000; i++) {
      for (int j = 0; j < 1000; j++)
         fscanf(u, "%f\n", &u_data[1000 * i + j]);
   }
   //initialising past input with the value of zero//
   for (int i = 0; i < 1000; i++) {
      for (int j = 0; j < 8; j++)
         y[8 * i + j] = 0;
   }
   cudaMemcpy(y_d, y, 1000000 * sizeof(float), cudaMemcpyHostToDevice);

   cudaMemcpy(theta_d, theta, THREADS_PER_BLOCK * TOTAL_BLOCKS * THETA_VALUES_PER_THREAD* sizeof(float), cudaMemcpyHostToDevice);
   cudaMemcpy(u_d, u_data, 1000000 * sizeof(float), cudaMemcpyHostToDevice);

   //calling kernel function//
   my_compute <<< 4, 256 >>> (y_d, theta_d, u_d);

   cudaMemcpy(y, y_d, 1000000 * sizeof(float), cudaMemcpyDeviceToHost);

   for (int i = 0; i < 1000; i++) {
      for (int j = 0; j < 1000; j++)
         printf("%f", y[1000 * i + j]);
   } 
   cudaEventRecord(stop, 0);
   cudaEventSynchronize(stop);
   cudaEventElapsedTime(&time, start, stop);
   cudaEventDestroy(start);
   cudaEventDestroy(stop);
   printf("Time to generate:  %3.1f ms \n", time);

   cudaFree(y_d);
   cudaFree(theta_d);
   cudaFree(u_d);

   fclose(u);
   fclose(fpoo);

   //fclose();
   _getche();
   return 0;
}
于 2012-09-06T02:59:35.853 に答える
1

統合されたcuda-memcheckを備えたToolkit5.0のcuda-gdbでコーディングしてみました。

それが私に示したのは、ブロック（3,0,0）のスレッド（232,0,0）でtheta_offsetが15000であるということです-そしてあなたのコードの私の予想から、それは14999（theta_dサイズは15000なので、使用できる最大インデックスです）

4ブロック*256スレッド*15要素/スレッド=15360であることに注意してください。

于 2012-08-15T16:41:01.047 に答える
1

スタック オーバーフローが発生している可能性があります。

static配列宣言に指定子を追加してみてください。

static float y[1000000];
static float theta[15000];
static float u_data[1000000];
于 2012-08-14T23:21:36.737 に答える
0

コードは確かに正しいです。ファイルからデータを読み取るための「fopenforloops」は、期待どおりにデータを読み取っていませんでした。そこで、以下のループに変更しました。私のコードは完全に実行されます。貢献してくれたすべての人に感謝します。

//シータのデータとモデルの入力をインポートします

fpoo= fopen("c:\\Fly_theta.txt","r");

u= fopen("c:\\Fly_u.txt","r");


        for(int j=0;j<15;j++)
            {
                fscanf(fpoo,"%f\n",&theta[j]);
            }
         for(int i=1;i<1000;i++)
         {
             for(int j=0;j<15;j++)
             {
                theta[15*i+j]=theta[j];
             }
         }
        for(int j=0;j<1000;j++)
            {
                fscanf(u,"%f\n",&u_data[j]);
            }
        for(int i=1;i<1000;i++)
         {
             for(int j=0;j<1000;j++)
             {
                u_data[1000*i+j]=u_data[j];
             }
        }
于 2012-08-18T17:26:27.247 に答える