동적으로 할당 된 멤버를 포함하는 구조체에 대한 포인터와 openacc를 결합하여 머리를 감싸려고합니다. 아래 코드는 실패합니다.
cuStreamSynchronize에 대한 Thread : 1 호출에 실패하면 오류 700이 반환되었습니다. 커널 실행 중 잘못된 주소
nvc ( "x86-64 Linux의 nvc 20.9-0 LLVM 64 비트 대상 -tp haswell")를 사용하여 컴파일 할 때. 내가 말할 수있는 한 나는 OpenACC '시작하기'가이드에서 제안 된 접근 방식을 따르고있다. 그러나 어떻게 든 아마도 포인터가 장치에 붙어 있지 않습니다 (?). 여기에서 무엇이 잘못되었는지 아는 사람이 있습니까?
#include <stdlib.h>
#include <stdio.h>
typedef struct grid
{
int N;
double *X;
} grid;
void allocate(grid* g, int N)
{
g->N = N;
g->X = (double*) malloc(sizeof(double) * g->N);
#pragma acc enter data create(g[0:1])
#pragma acc enter data create(g->X[0:N])
}
void release(grid* g)
{
#pragma acc exit data delete(g->X[0:g->N])
#pragma acc exit data delete(g[0:1])
free(g->X);
}
void fill(grid * g)
{
int i;
#pragma acc parallel loop
for (i = 0; i < g->N; i++)
{
g->X[i] = 42; // the cuprit, commenting this removes the error too
}
}
int main()
{
grid g;
allocate(&g, 10);
fill(&g);
release(&g);
return 0;
}```
컴파일러 피드백 메시지에서 다음과 같은 내용을 볼 수 있습니다.
fill:
32, Accelerator restriction: size of the GPU copy of g is unknown
Generating Tesla code
32, #pragma acc loop gang, vector(128) /* blockIdx.x threadIdx.x */
32, Generating implicit copyin(g) [if not already present]
37, Generating update self(g->X[:g->N])
문제는 컴파일러가 동적 데이터 멤버를 사용하여 집계 유형을 암시 적으로 복사 할 수 없으므로 g가 이미 장치임을 나타 내기 위해 "present (g)"를 추가해야한다는 것입니다.
또한 장치에서 N 값을 가져오고 종료 데이터 삭제 지시문에 배열 모양을 포함 할 필요가 없도록 복사하고 싶을 것입니다. 예를 들면 :
% cat test.c
#include <stdlib.h>
#include <stdio.h>
typedef struct grid
{
int N;
double *X;
} grid;
void allocate(grid* g, int N)
{
g->N = N;
g->X = (double*) malloc(sizeof(double) * g->N);
#pragma acc enter data copyin(g[0:1])
#pragma acc enter data create(g->X[0:N])
}
void release(grid* g)
{
#pragma acc exit data delete(g->X)
#pragma acc exit data delete(g)
free(g->X);
}
void fill(grid * g)
{
int i;
#pragma acc parallel loop present(g)
for (i = 0; i < g->N; i++)
{
g->X[i] = 42; // the cuprit, commenting this removes the error too
}
#pragma acc update self(g->X[:g->N])
for (i = 0; i < 4; i++)
{
printf("%d : %f \n",i,g->X[i]);
}
}
int main()
{
grid g;
allocate(&g, 10);
fill(&g);
release(&g);
return 0;
}
% nvc -acc test.c -Minfo=accel -V20.9 ; a.out
allocate:
17, Generating enter data copyin(g[:1])
Generating enter data create(g->X[:N])
release:
24, Generating exit data delete(g[:1],g->X[:1])
fill:
32, Generating present(g[:1])
Generating Tesla code
32, #pragma acc loop gang, vector(128) /* blockIdx.x threadIdx.x */
37, Generating update self(g->X[:g->N])
0 : 42.000000
1 : 42.000000
2 : 42.000000
3 : 42.000000
이 기사는 인터넷에서 수집됩니다. 재 인쇄 할 때 출처를 알려주십시오.
침해가 발생한 경우 연락 주시기 바랍니다[email protected] 삭제
몇 마디 만하겠습니다