gpu erreur：遇到非法内存访问
gpu erreur: an illegal memory access was encountered
非法内存访问错误，一切正常，直到我将其添加到内核中
dd->lg[index1] = lg1;
dd->hist[index0] = hist0;
dd->hist[index1] = hist1;
在代码中我使用数组结构 (SoA)
struct info
{
int hist[N];
double lg[N];
};
//info is *dd
完整代码在这里
#include "common.h"
#include <cuda_runtime.h>
#include <stdio.h>
#include <stdlib.h>
#include <cuda.h>
#include <curand_kernel.h>
#include <iostream>
using namespace std;
#define gpuErrchk(ans) { gpuAssert((ans), __FILE__, __LINE__); }
inline void gpuAssert(cudaError_t code, const char *file, int line, bool abort=true)
{
if (code != cudaSuccess)
{
fprintf(stderr,"GPUassert: %s %s %d\n", cudaGetErrorString(code), file, line);
if (abort) exit(code);
}
}
#define L 64
#define Nthread 256
#define Nblock 768*1024
#define frames Nthread*Nblock/(L*L)
#define N Nthread*Nblock
__device__ unsigned int nx = L*L,l = L-1,Msweepthr=1,J = 1;
struct Lattice
{
int x[N];
};
struct info
{
int hist[N];
double lg[N];
};
__global__ void setup_kernel(curandState *state)
{
int id = threadIdx.x + blockIdx.x * blockDim.x;
/* Each thread gets same seed, a different sequence
number, no offset */
//curand_init (unsigned long long seed, unsigned long long sequence, unsigned long long offset, curandState_t *state)
curand_init(1, id, 0, &state[id]);
}
__global__ void initiallattice(Lattice *data,int n)
{
unsigned int i = blockIdx.x * blockDim.x + threadIdx.x;
int sr1 = 1;
unsigned int m,a,c,r;
//generator pseudo- number for start config of lattice
m = 75;
a = 1103515245;
c = 74;
r = (a * i + c) % m;
data->x[i] = (((r)&sr1) == sr1 ? (1) : (-1));
}
__device__ int cell_energy(int s,int s1,int s2)
{
return (-J * ((s) * (s1+s2)));
}
__global__ void wanglanday(Lattice *pp, curandState *state,info *dd){
unsigned int frame_no = blockIdx.x * blockDim.x + threadIdx.x;
unsigned int ix,right,down;
int Rn_i,dE,e_old,e_new,lg0,lg1,p,hist0,hist1;
float Rn_c;
//Copy state to local memory for efficiency
curandState localState = state[frame_no];
int E = 0;
for(unsigned int mm= 0 ; mm<nx; mm++){
ix = frame_no*nx+(mm);
right = (frame_no*nx+(((((mm + 1) % L) == 1) ? (mm+l) : ((mm - 1 + nx) % nx))));
down = (frame_no*nx+((mm + L + nx) % nx));
E += cell_energy(pp->x[ix],pp->x[right],pp->x[down]);
}
//printf("frame_no = %d , energy = %d \n",frame_no,E);
for(unsigned int nn = 0 ; nn < Msweepthr ; nn++){
Rn_i = curand(&localState)%nx;
Rn_c = curand_uniform(&localState);
ix = frame_no*nx+(Rn_i);
right = (frame_no*nx+(((((ix + 1) % L) == 1) ? (ix+l) : ((ix - 1 + nx) % nx))));
down = (frame_no*nx+((ix + L + nx) % nx));
int newpp = pp->x[ix]*(-1);
e_old = cell_energy(pp->x[ix],pp->x[right],pp->x[down]);
e_new = cell_energy(newpp,pp->x[right],pp->x[down]);
// de = (e_new - e_old + (Lattice->x[n]*h)) * 2.0;
dE = (e_new - e_old)*2;
int index0= nx*2+E,index1= nx*2+E+dE;
lg0=dd->lg[index0];
lg1=dd->lg[index1];
hist0=dd->hist[index0];
hist1=dd->hist[index1];
p = lg0-lg1;
if ( expf(p) >= 1 || expf(p) > Rn_c){
lg1 += log(2.71828);
hist1 += 1;
E +=dE;
}else
{
lg0 += log(2.71828);
hist0 += 1;
}
dd->lg[index0] = lg0;
dd->lg[index1] = lg1;
dd->hist[index0] = hist0;
dd->hist[index1] = hist1;
}
//state[blockIdx.x] = localState;
}
int main(int argc, char **argv)
{
clock_t tStart = clock();
// set up device
int dev = 0;
cudaDeviceProp deviceProp;
gpuErrchk(cudaGetDeviceProperties(&deviceProp, dev));
printf("%s test struct wNG ", argv[0]);
printf("device %d: %s \n", dev, deviceProp.name);
gpuErrchk(cudaSetDevice(dev));
curandState *devStates;
// allocate host memory
size_t nBytes = sizeof(Lattice);
gpuErrchk(cudaMalloc((void **)&devStates, frames * sizeof(curandState))); // execution configuration
/* Setup prng states */
setup_kernel<<<frames,1>>>(devStates);
gpuErrchk( cudaPeekAtLastError() );
gpuErrchk( cudaDeviceSynchronize() );
// allocate device memory
Lattice *uzli;
gpuErrchk(cudaMalloc((Lattice**)&uzli, nBytes));
info *datas;
gpuErrchk(cudaMalloc((Lattice**)&datas, sizeof(info)));
initiallattice<<<Nblock,Nthread>>>(uzli, N);
gpuErrchk( cudaPeekAtLastError() );
gpuErrchk( cudaDeviceSynchronize() );
dim3 thr(256,1);
dim3 grid(frames/256,1);
wanglanday<<<grid,thr>>>(uzli,devStates,datas);//i am here
gpuErrchk( cudaPeekAtLastError() );
gpuErrchk( cudaDeviceSynchronize() );
printf("\n\t runing time: : %.2fs \v \n", (double)(clock() - tStart)/CLOCKS_PER_SEC);
// reset device
CHECK(cudaDeviceReset());
return EXIT_SUCCESS;
}
这个错误出现了很多次，我通过改变获取东西或写东西的方式处理了很多次，我寻找解释，但 none 得到了帮助。
代码的目的需要内核内部的循环。
我注意到你没有问一个问题，这是习惯上的。所以我会 imagine/pretend 你的问题是这样的：
Why does this code generate this "illegal memory access" error?
按照配方 here，compute-sanitizer 工具告诉我您的代码在这一行进行非法访问：
e_old = cell_energy(pp->x[ix],pp->x[right],pp->x[down]);
现在，在您的主要代码中，我们可以看到 x 数组的大小由 N 给出。因此，如果您的代码在此行进行非法访问，则 ix 或 right 或 down 必须在 0..N-1
范围之外
我们可以将上面的代码行替换为：
if ((ix < N) && (right < N) && (down < N))
e_old = cell_energy(pp->x[ix],pp->x[right],pp->x[down]);
else {
printf("frame_no: %u, N: %u, ix: %u, right: %u, down: %u\n", frame_no, N, ix, right, down); return;}
当我重新编译 运行 你的代码时，我得到以下输出：
$ ./t1982
./t1982 test struct wNG device 0: Tesla V100-PCIE-32GB
frame_no: 36826, N: 201326592, ix: 150842048, right: 301681407, down: 150842112
frame_no: 36797, N: 201326592, ix: 150723712, right: 301444287, down: 150723776
frame_no: 36739, N: 201326592, ix: 150485888, right: 300968895, down: 150485952
frame_no: 36756, N: 201326592, ix: 150556160, right: 301108799, down: 150556224
frame_no: 35221, N: 201326592, ix: 144267072, right: 288532351, down: 144267136
frame_no: 35223, N: 201326592, ix: 144275712, right: 288549183, down: 144275776
frame_no: 35303, N: 201326592, ix: 144603200, right: 289204351, down: 144603264
frame_no: 35160, N: 201326592, ix: 144017408, right: 288032831, down: 144017472
frame_no: 32212, N: 201326592, ix: 131943936, right: 263884351, down: 131944000
frame_no: 32135, N: 201326592, ix: 131627200, right: 263252223, down: 131627264
frame_no: 32149, N: 201326592, ix: 131685824, right: 263368191, down: 131685888
frame_no: 32226, N: 201326592, ix: 131999872, right: 263997631, down: 131999936
frame_no: 32187, N: 201326592, ix: 131838720, right: 263676735, down: 131838784
frame_no: 32103, N: 201326592, ix: 131494208, right: 262988159, down: 131494272
frame_no: 32034, N: 201326592, ix: 131213824, right: 262425151, down: 131213888
frame_no: 32060, N: 201326592, ix: 131319296, right: 262637119, down: 131319360
frame_no: 30487, N: 201326592, ix: 124875008, right: 249749823, down: 124875072
frame_no: 30518, N: 201326592, ix: 125002048, right: 250003839, down: 125002112
frame_no: 31956, N: 201326592, ix: 130894592, right: 261786431, down: 130894656
frame_no: 31966, N: 201326592, ix: 130935808, right: 261868607, down: 130935872
frame_no: 31921, N: 201326592, ix: 130750208, right: 261498687, down: 130750272
frame_no: 31897, N: 201326592, ix: 130652224, right: 261302399, down: 130652288
frame_no: 31902, N: 201326592, ix: 130674304, right: 261344959, down: 130674368
frame_no: 31799, N: 201326592, ix: 130250240, right: 260499007, down: 130250304
frame_no: 33534, N: 201326592, ix: 137357696, right: 274713023, down: 137357760
frame_no: 40929, N: 201326592, ix: 167645376, right: 335290623, down: 167645440
frame_no: 33356, N: 201326592, ix: 136628736, right: 273254975, down: 136628800
frame_no: 40736, N: 201326592, ix: 166856448, right: 333711167, down: 166856512
frame_no: 40884, N: 201326592, ix: 167461760, right: 334922687, down: 167461824
frame_no: 40917, N: 201326592, ix: 167598848, right: 335194943, down: 167598912
frame_no: 30357, N: 201326592, ix: 124345088, right: 248687423, down: 124345152
frame_no: 40586, N: 201326592, ix: 166240256, right: 332480575, down: 166240320
frame_no: 30296, N: 201326592, ix: 124096256, right: 248188735, down: 124096320
frame_no: 33592, N: 201326592, ix: 137595648, right: 275188543, down: 137595712
frame_no: 30223, N: 201326592, ix: 123794304, right: 247587775, down: 123794368
frame_no: 33705, N: 201326592, ix: 138057664, right: 276113407, down: 138057728
frame_no: 40549, N: 201326592, ix: 166089408, right: 332178175, down: 166089472
frame_no: 40552, N: 201326592, ix: 166104960, right: 332206015, down: 166105024
frame_no: 38355, N: 201326592, ix: 157104000, right: 314206143, down: 157104064
frame_no: 38259, N: 201326592, ix: 156710656, right: 313419583, down: 156710720
frame_no: 37980, N: 201326592, ix: 155568768, right: 311134911, down: 155568832
frame_no: 38391, N: 201326592, ix: 157250176, right: 314499775, down: 157250240
frame_no: 38393, N: 201326592, ix: 157258688, right: 314516479, down: 157258752
frame_no: 38395, N: 201326592, ix: 157268608, right: 314534591, down: 157268672
frame_no: 38061, N: 201326592, ix: 155900160, right: 311798079, down: 155900224
frame_no: 38101, N: 201326592, ix: 156065280, right: 312127039, down: 156065344
frame_no: 38106, N: 201326592, ix: 156085760, right: 312167999, down: 156085824
frame_no: 38120, N: 201326592, ix: 156141952, right: 312281535, down: 156142016
frame_no: 38133, N: 201326592, ix: 156195328, right: 312388159, down: 156195392
frame_no: 37995, N: 201326592, ix: 155628992, right: 311256575, down: 155629056
frame_no: 38007, N: 201326592, ix: 155677120, right: 311353855, down: 155677184
frame_no: 38014, N: 201326592, ix: 155708544, right: 311413951, down: 155708608
frame_no: 36455, N: 201326592, ix: 149321728, right: 298641471, down: 149321792
frame_no: 35067, N: 201326592, ix: 143637184, right: 287271679, down: 143637248
frame_no: 34995, N: 201326592, ix: 143339968, right: 286679551, down: 143340032
frame_no: 34895, N: 201326592, ix: 142932992, right: 285862975, down: 142933056
frame_no: 36589, N: 201326592, ix: 149868608, right: 299737215, down: 149868672
frame_no: 36559, N: 201326592, ix: 149748480, right: 299494207, down: 149748544
frame_no: 36391, N: 201326592, ix: 149060608, right: 298118207, down: 149060672
frame_no: 36396, N: 201326592, ix: 149079296, right: 298157375, down: 149079360
frame_no: 34875, N: 201326592, ix: 142850304, right: 285698367, down: 142850368
frame_no: 36506, N: 201326592, ix: 149532160, right: 299060799, down: 149532224
frame_no: 34914, N: 201326592, ix: 143009664, right: 286017471, down: 143009728
frame_no: 34919, N: 201326592, ix: 143031232, right: 286059519, down: 143031296
frame_no: 34934, N: 201326592, ix: 143092800, right: 286182527, down: 143092864
frame_no: 28837, N: 201326592, ix: 118118144, right: 236234559, down: 118118208
frame_no: 28844, N: 201326592, ix: 118145856, right: 236290943, down: 118145920
frame_no: 39556, N: 201326592, ix: 162023616, right: 324045055, down: 162023680
frame_no: 39424, N: 201326592, ix: 161481216, right: 322961983, down: 161481280
frame_no: 28747, N: 201326592, ix: 117749184, right: 235496959, down: 117749248
frame_no: 33258, N: 201326592, ix: 136225984, right: 272450815, down: 136226048
frame_no: 30897, N: 201326592, ix: 126556800, right: 253110975, down: 126556864
frame_no: 33170, N: 201326592, ix: 135866112, right: 271730495, down: 135866176
frame_no: 33033, N: 201326592, ix: 135305088, right: 270608319, down: 135305152
frame_no: 30874, N: 201326592, ix: 126462208, right: 252922175, down: 126462272
frame_no: 30878, N: 201326592, ix: 126479616, right: 252955967, down: 126479680
frame_no: 34091, N: 201326592, ix: 139638528, right: 279275327, down: 139638592
frame_no: 40216, N: 201326592, ix: 164728128, right: 329452927, down: 164728192
frame_no: 30759, N: 201326592, ix: 125991936, right: 251980863, down: 125992000
frame_no: 30760, N: 201326592, ix: 125994304, right: 251987327, down: 125994368
frame_no: 30844, N: 201326592, ix: 126340608, right: 252677695, down: 126340672
frame_no: 39474, N: 201326592, ix: 161685504, right: 323371071, down: 161685568
frame_no: 30747, N: 201326592, ix: 125941824, right: 251881599, down: 125941888
frame_no: 31142, N: 201326592, ix: 127561600, right: 255119295, down: 127561664
frame_no: 31218, N: 201326592, ix: 127871808, right: 255740799, down: 127871872
frame_no: 31224, N: 201326592, ix: 127894208, right: 255787775, down: 127894272
frame_no: 38785, N: 201326592, ix: 158864256, right: 317727679, down: 158864320
frame_no: 38815, N: 201326592, ix: 158986240, right: 317972543, down: 158986304
frame_no: 31010, N: 201326592, ix: 127018240, right: 254035263, down: 127018304
frame_no: 38898, N: 201326592, ix: 159328384, right: 318654655, down: 159328448
frame_no: 38729, N: 201326592, ix: 158637568, right: 317271615, down: 158637632
frame_no: 40391, N: 201326592, ix: 165444928, right: 330886527, down: 165444992
frame_no: 40308, N: 201326592, ix: 165105088, right: 330206719, down: 165105152
frame_no: 40309, N: 201326592, ix: 165106816, right: 330212543, down: 165106880
frame_no: 40321, N: 201326592, ix: 165158016, right: 330312895, down: 165158080
frame_no: 40249, N: 201326592, ix: 164862592, right: 329722559, down: 164862656
frame_no: 37721, N: 201326592, ix: 154507712, right: 309012991, down: 154507776
frame_no: 37776, N: 201326592, ix: 154732480, right: 309463039, down: 154732544
frame_no: 32887, N: 201326592, ix: 134706176, right: 269411391, down: 134706240
frame_no: 29486, N: 201326592, ix: 120777152, right: 241551871, down: 120777216
frame_no: 38864, N: 201326592, ix: 159187008, right: 318374015, down: 159187072
frame_no: 37675, N: 201326592, ix: 154318528, right: 308635391, down: 154318592
frame_no: 31315, N: 201326592, ix: 128266560, right: 256532863, down: 128266624
frame_no: 31325, N: 201326592, ix: 128310080, right: 256617343, down: 128310144
frame_no: 32797, N: 201326592, ix: 134336768, right: 268673343, down: 134336832
frame_no: 34532, N: 201326592, ix: 141446720, right: 282889855, down: 141446784
frame_no: 29658, N: 201326592, ix: 121479424, right: 242958655, down: 121479488
frame_no: 29177, N: 201326592, ix: 119511040, right: 239020095, down: 119511104
frame_no: 29541, N: 201326592, ix: 121003520, right: 242003519, down: 121003584
frame_no: 29544, N: 201326592, ix: 121015552, right: 242027839, down: 121015616
frame_no: 29358, N: 201326592, ix: 120251840, right: 240502271, down: 120251904
frame_no: 34326, N: 201326592, ix: 140602240, right: 281201599, down: 140602304
frame_no: 29587, N: 201326592, ix: 121188992, right: 242377407, down: 121189056
frame_no: 29306, N: 201326592, ix: 120040256, right: 240077695, down: 120040320
frame_no: 32909, N: 201326592, ix: 134795648, right: 269590975, down: 134795712
frame_no: 29512, N: 201326592, ix: 120883136, right: 241764351, down: 120883200
frame_no: 28979, N: 201326592, ix: 118701184, right: 237399231, down: 118701248
frame_no: 35712, N: 201326592, ix: 146278912, right: 292555327, down: 146278976
frame_no: 29065, N: 201326592, ix: 119053632, right: 238103935, down: 119053696
frame_no: 39997, N: 201326592, ix: 163830080, right: 327657855, down: 163830144
frame_no: 32802, N: 201326592, ix: 134360960, right: 268718015, down: 134361024
frame_no: 29223, N: 201326592, ix: 119698944, right: 239396415, down: 119699008
frame_no: 32858, N: 201326592, ix: 134589632, right: 269176063, down: 134589696
frame_no: 35621, N: 201326592, ix: 145905536, right: 291809215, down: 145905600
frame_no: 32358, N: 201326592, ix: 132539584, right: 265078015, down: 132539648
frame_no: 31646, N: 201326592, ix: 129625600, right: 259247679, down: 129625664
frame_no: 29409, N: 201326592, ix: 120461696, right: 240921023, down: 120461760
frame_no: 29437, N: 201326592, ix: 120575424, right: 241149439, down: 120575488
frame_no: 30056, N: 201326592, ix: 123112000, right: 246221439, down: 123112064
frame_no: 31544, N: 201326592, ix: 129206080, right: 258410367, down: 129206144
frame_no: 29327, N: 201326592, ix: 120125376, right: 240248831, down: 120125440
frame_no: 29330, N: 201326592, ix: 120136448, right: 240272191, down: 120136512
frame_no: 32321, N: 201326592, ix: 132389568, right: 264776447, down: 132389632
frame_no: 32345, N: 201326592, ix: 132486208, right: 264971391, down: 132486272
frame_no: 32348, N: 201326592, ix: 132500864, right: 264998335, down: 132500928
frame_no: 31726, N: 201326592, ix: 129952512, right: 259902271, down: 129952576
frame_no: 32265, N: 201326592, ix: 132161344, right: 264318847, down: 132161408
frame_no: 31660, N: 201326592, ix: 129679808, right: 259359231, down: 129679872
frame_no: 32474, N: 201326592, ix: 133014016, right: 266027583, down: 133014080
frame_no: 29788, N: 201326592, ix: 122015424, right: 244027135, down: 122015488
frame_no: 29898, N: 201326592, ix: 122463296, right: 244925567, down: 122463360
frame_no: 29904, N: 201326592, ix: 122488448, right: 244975295, down: 122488512
frame_no: 32737, N: 201326592, ix: 134091456, right: 268182271, down: 134091520
frame_no: 32750, N: 201326592, ix: 134147328, right: 268291391, down: 134147392
frame_no: 32763, N: 201326592, ix: 134199488, right: 268396799, down: 134199552
frame_no: 40111, N: 201326592, ix: 164298240, right: 328592959, down: 164298304
frame_no: 40126, N: 201326592, ix: 164359936, right: 328716095, down: 164360000
frame_no: 32668, N: 201326592, ix: 133808384, right: 267616575, down: 133808448
frame_no: 40005, N: 201326592, ix: 163863360, right: 327723903, down: 163863424
frame_no: 40154, N: 201326592, ix: 164471808, right: 328942655, down: 164471872
frame_no: 32587, N: 201326592, ix: 133479808, right: 266956223, down: 133479872
frame_no: 34021, N: 201326592, ix: 139350848, right: 278700927, down: 139350912
frame_no: 39948, N: 201326592, ix: 163628672, right: 327255743, down: 163628736
frame_no: 33938, N: 201326592, ix: 139011008, right: 278021119, down: 139011072
frame_no: 29847, N: 201326592, ix: 122253312, right: 244506687, down: 122253376
frame_no: 34673, N: 201326592, ix: 142022016, right: 284042687, down: 142022080
frame_no: 40042, N: 201326592, ix: 164012800, right: 328024895, down: 164012864
frame_no: 40173, N: 201326592, ix: 164552064, right: 329100735, down: 164552128
frame_no: 35478, N: 201326592, ix: 145320960, right: 290638911, down: 145321024
frame_no: 33819, N: 201326592, ix: 138523840, right: 277046527, down: 138523904
frame_no: 34776, N: 201326592, ix: 142446208, right: 284888767, down: 142446272
frame_no: 34707, N: 201326592, ix: 142160000, right: 284319935, down: 142160064
frame_no: 34566, N: 201326592, ix: 141583808, right: 283166207, down: 141583872
frame_no: 32681, N: 201326592, ix: 133862080, right: 267723519, down: 133862144
frame_no: 34626, N: 201326592, ix: 141830400, right: 283658559, down: 141830464
frame_no: 35513, N: 201326592, ix: 145462080, right: 290923391, down: 145462144
frame_no: 39854, N: 201326592, ix: 163245056, right: 326487103, down: 163245120
frame_no: 39867, N: 201326592, ix: 163297536, right: 326592831, down: 163297600
frame_no: 39871, N: 201326592, ix: 163315392, right: 326627071, down: 163315456
frame_no: 36305, N: 201326592, ix: 148706112, right: 297411455, down: 148706176
frame_no: 35850, N: 201326592, ix: 146842560, right: 293684223, down: 146842624
frame_no: 35377, N: 201326592, ix: 144907520, right: 289811775, down: 144907584
frame_no: 35384, N: 201326592, ix: 144936128, right: 289869055, down: 144936192
frame_no: 36066, N: 201326592, ix: 147730176, right: 295456575, down: 147730240
frame_no: 37114, N: 201326592, ix: 152020608, right: 304039615, down: 152020672
frame_no: 36919, N: 201326592, ix: 151221184, right: 302441471, down: 151221248
frame_no: 36925, N: 201326592, ix: 151244864, right: 302489727, down: 151244928
frame_no: 38506, N: 201326592, ix: 157721728, right: 315442367, down: 157721792
frame_no: 38972, N: 201326592, ix: 159632512, right: 319261887, down: 159632576
frame_no: 39145, N: 201326592, ix: 160340800, right: 320678783, down: 160340864
frame_no: 39166, N: 201326592, ix: 160427072, right: 320851071, down: 160427136
frame_no: 39033, N: 201326592, ix: 159880640, right: 319759871, down: 159880704
frame_no: 37143, N: 201326592, ix: 152141056, right: 304278847, down: 152141120
frame_no: 37144, N: 201326592, ix: 152143104, right: 304284991, down: 152143168
frame_no: 36022, N: 201326592, ix: 147546304, right: 295092479, down: 147546368
frame_no: 34792, N: 201326592, ix: 142510976, right: 285019071, down: 142511040
frame_no: 34793, N: 201326592, ix: 142512576, right: 285024767, down: 142512640
frame_no: 34798, N: 201326592, ix: 142535360, right: 285068031, down: 142535424
frame_no: 36174, N: 201326592, ix: 148170560, right: 296339327, down: 148170624
frame_no: 36348, N: 201326592, ix: 148883008, right: 297764479, down: 148883072
frame_no: 36197, N: 201326592, ix: 148263680, right: 296526655, down: 148263744
frame_no: 38538, N: 201326592, ix: 157855488, right: 315707199, down: 157855552
frame_no: 38646, N: 201326592, ix: 158297856, right: 316591935, down: 158297920
frame_no: 38439, N: 201326592, ix: 157446528, right: 314892735, down: 157446592
frame_no: 39309, N: 201326592, ix: 161010944, right: 322020671, down: 161011008
frame_no: 39310, N: 201326592, ix: 161014464, right: 322028287, down: 161014528
frame_no: 36253, N: 201326592, ix: 148496320, right: 296988671, down: 148492288
frame_no: 38603, N: 201326592, ix: 158120704, right: 316238655, down: 158120768
frame_no: 35921, N: 201326592, ix: 147134464, right: 294266943, down: 147134528
frame_no: 36881, N: 201326592, ix: 151067904, right: 302132543, down: 151067968
frame_no: 36894, N: 201326592, ix: 151120960, right: 302238847, down: 151121024
frame_no: 37059, N: 201326592, ix: 151795840, right: 303589567, down: 151795904
frame_no: 37337, N: 201326592, ix: 152933824, right: 305866239, down: 152933888
frame_no: 37344, N: 201326592, ix: 152964608, right: 305925695, down: 152964672
frame_no: 37347, N: 201326592, ix: 152974912, right: 305948287, down: 152974976
<truncated output>
runing time: : 0.38s
所以在我看来很明显，在上述大多数情况下，您对 right 的计算超出了 N-1 的允许范围。
此处给出了 right 的计算结果：
right = (frame_no*nx+(((((mm + 1) % L) == 1) ? (mm+l) : ((mm - 1 + nx) % nx))));
我不知道如何为您解决这个问题。您可以类似地在该语句之后使用 printf 来突出显示计算的索引非法的情况：
if (right >= N) printf("....", ....);
顺便说一句，我注意到在对上述代码进行更改后，运行在 compute-sanitizer 下修改后的代码不会产生任何错误。因此，我怀疑当前代码中与您报告的错误相关的大部分问题都与 right 的非法 calculation/usage 有关。
非法内存访问错误，一切正常，直到我将其添加到内核中
dd->lg[index1] = lg1;
dd->hist[index0] = hist0;
dd->hist[index1] = hist1;
在代码中我使用数组结构 (SoA)
struct info
{
int hist[N];
double lg[N];
};
//info is *dd
完整代码在这里
#include "common.h"
#include <cuda_runtime.h>
#include <stdio.h>
#include <stdlib.h>
#include <cuda.h>
#include <curand_kernel.h>
#include <iostream>
using namespace std;
#define gpuErrchk(ans) { gpuAssert((ans), __FILE__, __LINE__); }
inline void gpuAssert(cudaError_t code, const char *file, int line, bool abort=true)
{
if (code != cudaSuccess)
{
fprintf(stderr,"GPUassert: %s %s %d\n", cudaGetErrorString(code), file, line);
if (abort) exit(code);
}
}
#define L 64
#define Nthread 256
#define Nblock 768*1024
#define frames Nthread*Nblock/(L*L)
#define N Nthread*Nblock
__device__ unsigned int nx = L*L,l = L-1,Msweepthr=1,J = 1;
struct Lattice
{
int x[N];
};
struct info
{
int hist[N];
double lg[N];
};
__global__ void setup_kernel(curandState *state)
{
int id = threadIdx.x + blockIdx.x * blockDim.x;
/* Each thread gets same seed, a different sequence
number, no offset */
//curand_init (unsigned long long seed, unsigned long long sequence, unsigned long long offset, curandState_t *state)
curand_init(1, id, 0, &state[id]);
}
__global__ void initiallattice(Lattice *data,int n)
{
unsigned int i = blockIdx.x * blockDim.x + threadIdx.x;
int sr1 = 1;
unsigned int m,a,c,r;
//generator pseudo- number for start config of lattice
m = 75;
a = 1103515245;
c = 74;
r = (a * i + c) % m;
data->x[i] = (((r)&sr1) == sr1 ? (1) : (-1));
}
__device__ int cell_energy(int s,int s1,int s2)
{
return (-J * ((s) * (s1+s2)));
}
__global__ void wanglanday(Lattice *pp, curandState *state,info *dd){
unsigned int frame_no = blockIdx.x * blockDim.x + threadIdx.x;
unsigned int ix,right,down;
int Rn_i,dE,e_old,e_new,lg0,lg1,p,hist0,hist1;
float Rn_c;
//Copy state to local memory for efficiency
curandState localState = state[frame_no];
int E = 0;
for(unsigned int mm= 0 ; mm<nx; mm++){
ix = frame_no*nx+(mm);
right = (frame_no*nx+(((((mm + 1) % L) == 1) ? (mm+l) : ((mm - 1 + nx) % nx))));
down = (frame_no*nx+((mm + L + nx) % nx));
E += cell_energy(pp->x[ix],pp->x[right],pp->x[down]);
}
//printf("frame_no = %d , energy = %d \n",frame_no,E);
for(unsigned int nn = 0 ; nn < Msweepthr ; nn++){
Rn_i = curand(&localState)%nx;
Rn_c = curand_uniform(&localState);
ix = frame_no*nx+(Rn_i);
right = (frame_no*nx+(((((ix + 1) % L) == 1) ? (ix+l) : ((ix - 1 + nx) % nx))));
down = (frame_no*nx+((ix + L + nx) % nx));
int newpp = pp->x[ix]*(-1);
e_old = cell_energy(pp->x[ix],pp->x[right],pp->x[down]);
e_new = cell_energy(newpp,pp->x[right],pp->x[down]);
// de = (e_new - e_old + (Lattice->x[n]*h)) * 2.0;
dE = (e_new - e_old)*2;
int index0= nx*2+E,index1= nx*2+E+dE;
lg0=dd->lg[index0];
lg1=dd->lg[index1];
hist0=dd->hist[index0];
hist1=dd->hist[index1];
p = lg0-lg1;
if ( expf(p) >= 1 || expf(p) > Rn_c){
lg1 += log(2.71828);
hist1 += 1;
E +=dE;
}else
{
lg0 += log(2.71828);
hist0 += 1;
}
dd->lg[index0] = lg0;
dd->lg[index1] = lg1;
dd->hist[index0] = hist0;
dd->hist[index1] = hist1;
}
//state[blockIdx.x] = localState;
}
int main(int argc, char **argv)
{
clock_t tStart = clock();
// set up device
int dev = 0;
cudaDeviceProp deviceProp;
gpuErrchk(cudaGetDeviceProperties(&deviceProp, dev));
printf("%s test struct wNG ", argv[0]);
printf("device %d: %s \n", dev, deviceProp.name);
gpuErrchk(cudaSetDevice(dev));
curandState *devStates;
// allocate host memory
size_t nBytes = sizeof(Lattice);
gpuErrchk(cudaMalloc((void **)&devStates, frames * sizeof(curandState))); // execution configuration
/* Setup prng states */
setup_kernel<<<frames,1>>>(devStates);
gpuErrchk( cudaPeekAtLastError() );
gpuErrchk( cudaDeviceSynchronize() );
// allocate device memory
Lattice *uzli;
gpuErrchk(cudaMalloc((Lattice**)&uzli, nBytes));
info *datas;
gpuErrchk(cudaMalloc((Lattice**)&datas, sizeof(info)));
initiallattice<<<Nblock,Nthread>>>(uzli, N);
gpuErrchk( cudaPeekAtLastError() );
gpuErrchk( cudaDeviceSynchronize() );
dim3 thr(256,1);
dim3 grid(frames/256,1);
wanglanday<<<grid,thr>>>(uzli,devStates,datas);//i am here
gpuErrchk( cudaPeekAtLastError() );
gpuErrchk( cudaDeviceSynchronize() );
printf("\n\t runing time: : %.2fs \v \n", (double)(clock() - tStart)/CLOCKS_PER_SEC);
// reset device
CHECK(cudaDeviceReset());
return EXIT_SUCCESS;
}
这个错误出现了很多次，我通过改变获取东西或写东西的方式处理了很多次，我寻找解释，但 none 得到了帮助。 代码的目的需要内核内部的循环。
我注意到你没有问一个问题，这是习惯上的。所以我会 imagine/pretend 你的问题是这样的：
Why does this code generate this "illegal memory access" error?
按照配方 here，compute-sanitizer 工具告诉我您的代码在这一行进行非法访问：
e_old = cell_energy(pp->x[ix],pp->x[right],pp->x[down]);
现在，在您的主要代码中，我们可以看到 x 数组的大小由 N 给出。因此，如果您的代码在此行进行非法访问，则 ix 或 right 或 down 必须在 0..N-1
我们可以将上面的代码行替换为：
if ((ix < N) && (right < N) && (down < N))
e_old = cell_energy(pp->x[ix],pp->x[right],pp->x[down]);
else {
printf("frame_no: %u, N: %u, ix: %u, right: %u, down: %u\n", frame_no, N, ix, right, down); return;}
当我重新编译 运行 你的代码时，我得到以下输出：
$ ./t1982
./t1982 test struct wNG device 0: Tesla V100-PCIE-32GB
frame_no: 36826, N: 201326592, ix: 150842048, right: 301681407, down: 150842112
frame_no: 36797, N: 201326592, ix: 150723712, right: 301444287, down: 150723776
frame_no: 36739, N: 201326592, ix: 150485888, right: 300968895, down: 150485952
frame_no: 36756, N: 201326592, ix: 150556160, right: 301108799, down: 150556224
frame_no: 35221, N: 201326592, ix: 144267072, right: 288532351, down: 144267136
frame_no: 35223, N: 201326592, ix: 144275712, right: 288549183, down: 144275776
frame_no: 35303, N: 201326592, ix: 144603200, right: 289204351, down: 144603264
frame_no: 35160, N: 201326592, ix: 144017408, right: 288032831, down: 144017472
frame_no: 32212, N: 201326592, ix: 131943936, right: 263884351, down: 131944000
frame_no: 32135, N: 201326592, ix: 131627200, right: 263252223, down: 131627264
frame_no: 32149, N: 201326592, ix: 131685824, right: 263368191, down: 131685888
frame_no: 32226, N: 201326592, ix: 131999872, right: 263997631, down: 131999936
frame_no: 32187, N: 201326592, ix: 131838720, right: 263676735, down: 131838784
frame_no: 32103, N: 201326592, ix: 131494208, right: 262988159, down: 131494272
frame_no: 32034, N: 201326592, ix: 131213824, right: 262425151, down: 131213888
frame_no: 32060, N: 201326592, ix: 131319296, right: 262637119, down: 131319360
frame_no: 30487, N: 201326592, ix: 124875008, right: 249749823, down: 124875072
frame_no: 30518, N: 201326592, ix: 125002048, right: 250003839, down: 125002112
frame_no: 31956, N: 201326592, ix: 130894592, right: 261786431, down: 130894656
frame_no: 31966, N: 201326592, ix: 130935808, right: 261868607, down: 130935872
frame_no: 31921, N: 201326592, ix: 130750208, right: 261498687, down: 130750272
frame_no: 31897, N: 201326592, ix: 130652224, right: 261302399, down: 130652288
frame_no: 31902, N: 201326592, ix: 130674304, right: 261344959, down: 130674368
frame_no: 31799, N: 201326592, ix: 130250240, right: 260499007, down: 130250304
frame_no: 33534, N: 201326592, ix: 137357696, right: 274713023, down: 137357760
frame_no: 40929, N: 201326592, ix: 167645376, right: 335290623, down: 167645440
frame_no: 33356, N: 201326592, ix: 136628736, right: 273254975, down: 136628800
frame_no: 40736, N: 201326592, ix: 166856448, right: 333711167, down: 166856512
frame_no: 40884, N: 201326592, ix: 167461760, right: 334922687, down: 167461824
frame_no: 40917, N: 201326592, ix: 167598848, right: 335194943, down: 167598912
frame_no: 30357, N: 201326592, ix: 124345088, right: 248687423, down: 124345152
frame_no: 40586, N: 201326592, ix: 166240256, right: 332480575, down: 166240320
frame_no: 30296, N: 201326592, ix: 124096256, right: 248188735, down: 124096320
frame_no: 33592, N: 201326592, ix: 137595648, right: 275188543, down: 137595712
frame_no: 30223, N: 201326592, ix: 123794304, right: 247587775, down: 123794368
frame_no: 33705, N: 201326592, ix: 138057664, right: 276113407, down: 138057728
frame_no: 40549, N: 201326592, ix: 166089408, right: 332178175, down: 166089472
frame_no: 40552, N: 201326592, ix: 166104960, right: 332206015, down: 166105024
frame_no: 38355, N: 201326592, ix: 157104000, right: 314206143, down: 157104064
frame_no: 38259, N: 201326592, ix: 156710656, right: 313419583, down: 156710720
frame_no: 37980, N: 201326592, ix: 155568768, right: 311134911, down: 155568832
frame_no: 38391, N: 201326592, ix: 157250176, right: 314499775, down: 157250240
frame_no: 38393, N: 201326592, ix: 157258688, right: 314516479, down: 157258752
frame_no: 38395, N: 201326592, ix: 157268608, right: 314534591, down: 157268672
frame_no: 38061, N: 201326592, ix: 155900160, right: 311798079, down: 155900224
frame_no: 38101, N: 201326592, ix: 156065280, right: 312127039, down: 156065344
frame_no: 38106, N: 201326592, ix: 156085760, right: 312167999, down: 156085824
frame_no: 38120, N: 201326592, ix: 156141952, right: 312281535, down: 156142016
frame_no: 38133, N: 201326592, ix: 156195328, right: 312388159, down: 156195392
frame_no: 37995, N: 201326592, ix: 155628992, right: 311256575, down: 155629056
frame_no: 38007, N: 201326592, ix: 155677120, right: 311353855, down: 155677184
frame_no: 38014, N: 201326592, ix: 155708544, right: 311413951, down: 155708608
frame_no: 36455, N: 201326592, ix: 149321728, right: 298641471, down: 149321792
frame_no: 35067, N: 201326592, ix: 143637184, right: 287271679, down: 143637248
frame_no: 34995, N: 201326592, ix: 143339968, right: 286679551, down: 143340032
frame_no: 34895, N: 201326592, ix: 142932992, right: 285862975, down: 142933056
frame_no: 36589, N: 201326592, ix: 149868608, right: 299737215, down: 149868672
frame_no: 36559, N: 201326592, ix: 149748480, right: 299494207, down: 149748544
frame_no: 36391, N: 201326592, ix: 149060608, right: 298118207, down: 149060672
frame_no: 36396, N: 201326592, ix: 149079296, right: 298157375, down: 149079360
frame_no: 34875, N: 201326592, ix: 142850304, right: 285698367, down: 142850368
frame_no: 36506, N: 201326592, ix: 149532160, right: 299060799, down: 149532224
frame_no: 34914, N: 201326592, ix: 143009664, right: 286017471, down: 143009728
frame_no: 34919, N: 201326592, ix: 143031232, right: 286059519, down: 143031296
frame_no: 34934, N: 201326592, ix: 143092800, right: 286182527, down: 143092864
frame_no: 28837, N: 201326592, ix: 118118144, right: 236234559, down: 118118208
frame_no: 28844, N: 201326592, ix: 118145856, right: 236290943, down: 118145920
frame_no: 39556, N: 201326592, ix: 162023616, right: 324045055, down: 162023680
frame_no: 39424, N: 201326592, ix: 161481216, right: 322961983, down: 161481280
frame_no: 28747, N: 201326592, ix: 117749184, right: 235496959, down: 117749248
frame_no: 33258, N: 201326592, ix: 136225984, right: 272450815, down: 136226048
frame_no: 30897, N: 201326592, ix: 126556800, right: 253110975, down: 126556864
frame_no: 33170, N: 201326592, ix: 135866112, right: 271730495, down: 135866176
frame_no: 33033, N: 201326592, ix: 135305088, right: 270608319, down: 135305152
frame_no: 30874, N: 201326592, ix: 126462208, right: 252922175, down: 126462272
frame_no: 30878, N: 201326592, ix: 126479616, right: 252955967, down: 126479680
frame_no: 34091, N: 201326592, ix: 139638528, right: 279275327, down: 139638592
frame_no: 40216, N: 201326592, ix: 164728128, right: 329452927, down: 164728192
frame_no: 30759, N: 201326592, ix: 125991936, right: 251980863, down: 125992000
frame_no: 30760, N: 201326592, ix: 125994304, right: 251987327, down: 125994368
frame_no: 30844, N: 201326592, ix: 126340608, right: 252677695, down: 126340672
frame_no: 39474, N: 201326592, ix: 161685504, right: 323371071, down: 161685568
frame_no: 30747, N: 201326592, ix: 125941824, right: 251881599, down: 125941888
frame_no: 31142, N: 201326592, ix: 127561600, right: 255119295, down: 127561664
frame_no: 31218, N: 201326592, ix: 127871808, right: 255740799, down: 127871872
frame_no: 31224, N: 201326592, ix: 127894208, right: 255787775, down: 127894272
frame_no: 38785, N: 201326592, ix: 158864256, right: 317727679, down: 158864320
frame_no: 38815, N: 201326592, ix: 158986240, right: 317972543, down: 158986304
frame_no: 31010, N: 201326592, ix: 127018240, right: 254035263, down: 127018304
frame_no: 38898, N: 201326592, ix: 159328384, right: 318654655, down: 159328448
frame_no: 38729, N: 201326592, ix: 158637568, right: 317271615, down: 158637632
frame_no: 40391, N: 201326592, ix: 165444928, right: 330886527, down: 165444992
frame_no: 40308, N: 201326592, ix: 165105088, right: 330206719, down: 165105152
frame_no: 40309, N: 201326592, ix: 165106816, right: 330212543, down: 165106880
frame_no: 40321, N: 201326592, ix: 165158016, right: 330312895, down: 165158080
frame_no: 40249, N: 201326592, ix: 164862592, right: 329722559, down: 164862656
frame_no: 37721, N: 201326592, ix: 154507712, right: 309012991, down: 154507776
frame_no: 37776, N: 201326592, ix: 154732480, right: 309463039, down: 154732544
frame_no: 32887, N: 201326592, ix: 134706176, right: 269411391, down: 134706240
frame_no: 29486, N: 201326592, ix: 120777152, right: 241551871, down: 120777216
frame_no: 38864, N: 201326592, ix: 159187008, right: 318374015, down: 159187072
frame_no: 37675, N: 201326592, ix: 154318528, right: 308635391, down: 154318592
frame_no: 31315, N: 201326592, ix: 128266560, right: 256532863, down: 128266624
frame_no: 31325, N: 201326592, ix: 128310080, right: 256617343, down: 128310144
frame_no: 32797, N: 201326592, ix: 134336768, right: 268673343, down: 134336832
frame_no: 34532, N: 201326592, ix: 141446720, right: 282889855, down: 141446784
frame_no: 29658, N: 201326592, ix: 121479424, right: 242958655, down: 121479488
frame_no: 29177, N: 201326592, ix: 119511040, right: 239020095, down: 119511104
frame_no: 29541, N: 201326592, ix: 121003520, right: 242003519, down: 121003584
frame_no: 29544, N: 201326592, ix: 121015552, right: 242027839, down: 121015616
frame_no: 29358, N: 201326592, ix: 120251840, right: 240502271, down: 120251904
frame_no: 34326, N: 201326592, ix: 140602240, right: 281201599, down: 140602304
frame_no: 29587, N: 201326592, ix: 121188992, right: 242377407, down: 121189056
frame_no: 29306, N: 201326592, ix: 120040256, right: 240077695, down: 120040320
frame_no: 32909, N: 201326592, ix: 134795648, right: 269590975, down: 134795712
frame_no: 29512, N: 201326592, ix: 120883136, right: 241764351, down: 120883200
frame_no: 28979, N: 201326592, ix: 118701184, right: 237399231, down: 118701248
frame_no: 35712, N: 201326592, ix: 146278912, right: 292555327, down: 146278976
frame_no: 29065, N: 201326592, ix: 119053632, right: 238103935, down: 119053696
frame_no: 39997, N: 201326592, ix: 163830080, right: 327657855, down: 163830144
frame_no: 32802, N: 201326592, ix: 134360960, right: 268718015, down: 134361024
frame_no: 29223, N: 201326592, ix: 119698944, right: 239396415, down: 119699008
frame_no: 32858, N: 201326592, ix: 134589632, right: 269176063, down: 134589696
frame_no: 35621, N: 201326592, ix: 145905536, right: 291809215, down: 145905600
frame_no: 32358, N: 201326592, ix: 132539584, right: 265078015, down: 132539648
frame_no: 31646, N: 201326592, ix: 129625600, right: 259247679, down: 129625664
frame_no: 29409, N: 201326592, ix: 120461696, right: 240921023, down: 120461760
frame_no: 29437, N: 201326592, ix: 120575424, right: 241149439, down: 120575488
frame_no: 30056, N: 201326592, ix: 123112000, right: 246221439, down: 123112064
frame_no: 31544, N: 201326592, ix: 129206080, right: 258410367, down: 129206144
frame_no: 29327, N: 201326592, ix: 120125376, right: 240248831, down: 120125440
frame_no: 29330, N: 201326592, ix: 120136448, right: 240272191, down: 120136512
frame_no: 32321, N: 201326592, ix: 132389568, right: 264776447, down: 132389632
frame_no: 32345, N: 201326592, ix: 132486208, right: 264971391, down: 132486272
frame_no: 32348, N: 201326592, ix: 132500864, right: 264998335, down: 132500928
frame_no: 31726, N: 201326592, ix: 129952512, right: 259902271, down: 129952576
frame_no: 32265, N: 201326592, ix: 132161344, right: 264318847, down: 132161408
frame_no: 31660, N: 201326592, ix: 129679808, right: 259359231, down: 129679872
frame_no: 32474, N: 201326592, ix: 133014016, right: 266027583, down: 133014080
frame_no: 29788, N: 201326592, ix: 122015424, right: 244027135, down: 122015488
frame_no: 29898, N: 201326592, ix: 122463296, right: 244925567, down: 122463360
frame_no: 29904, N: 201326592, ix: 122488448, right: 244975295, down: 122488512
frame_no: 32737, N: 201326592, ix: 134091456, right: 268182271, down: 134091520
frame_no: 32750, N: 201326592, ix: 134147328, right: 268291391, down: 134147392
frame_no: 32763, N: 201326592, ix: 134199488, right: 268396799, down: 134199552
frame_no: 40111, N: 201326592, ix: 164298240, right: 328592959, down: 164298304
frame_no: 40126, N: 201326592, ix: 164359936, right: 328716095, down: 164360000
frame_no: 32668, N: 201326592, ix: 133808384, right: 267616575, down: 133808448
frame_no: 40005, N: 201326592, ix: 163863360, right: 327723903, down: 163863424
frame_no: 40154, N: 201326592, ix: 164471808, right: 328942655, down: 164471872
frame_no: 32587, N: 201326592, ix: 133479808, right: 266956223, down: 133479872
frame_no: 34021, N: 201326592, ix: 139350848, right: 278700927, down: 139350912
frame_no: 39948, N: 201326592, ix: 163628672, right: 327255743, down: 163628736
frame_no: 33938, N: 201326592, ix: 139011008, right: 278021119, down: 139011072
frame_no: 29847, N: 201326592, ix: 122253312, right: 244506687, down: 122253376
frame_no: 34673, N: 201326592, ix: 142022016, right: 284042687, down: 142022080
frame_no: 40042, N: 201326592, ix: 164012800, right: 328024895, down: 164012864
frame_no: 40173, N: 201326592, ix: 164552064, right: 329100735, down: 164552128
frame_no: 35478, N: 201326592, ix: 145320960, right: 290638911, down: 145321024
frame_no: 33819, N: 201326592, ix: 138523840, right: 277046527, down: 138523904
frame_no: 34776, N: 201326592, ix: 142446208, right: 284888767, down: 142446272
frame_no: 34707, N: 201326592, ix: 142160000, right: 284319935, down: 142160064
frame_no: 34566, N: 201326592, ix: 141583808, right: 283166207, down: 141583872
frame_no: 32681, N: 201326592, ix: 133862080, right: 267723519, down: 133862144
frame_no: 34626, N: 201326592, ix: 141830400, right: 283658559, down: 141830464
frame_no: 35513, N: 201326592, ix: 145462080, right: 290923391, down: 145462144
frame_no: 39854, N: 201326592, ix: 163245056, right: 326487103, down: 163245120
frame_no: 39867, N: 201326592, ix: 163297536, right: 326592831, down: 163297600
frame_no: 39871, N: 201326592, ix: 163315392, right: 326627071, down: 163315456
frame_no: 36305, N: 201326592, ix: 148706112, right: 297411455, down: 148706176
frame_no: 35850, N: 201326592, ix: 146842560, right: 293684223, down: 146842624
frame_no: 35377, N: 201326592, ix: 144907520, right: 289811775, down: 144907584
frame_no: 35384, N: 201326592, ix: 144936128, right: 289869055, down: 144936192
frame_no: 36066, N: 201326592, ix: 147730176, right: 295456575, down: 147730240
frame_no: 37114, N: 201326592, ix: 152020608, right: 304039615, down: 152020672
frame_no: 36919, N: 201326592, ix: 151221184, right: 302441471, down: 151221248
frame_no: 36925, N: 201326592, ix: 151244864, right: 302489727, down: 151244928
frame_no: 38506, N: 201326592, ix: 157721728, right: 315442367, down: 157721792
frame_no: 38972, N: 201326592, ix: 159632512, right: 319261887, down: 159632576
frame_no: 39145, N: 201326592, ix: 160340800, right: 320678783, down: 160340864
frame_no: 39166, N: 201326592, ix: 160427072, right: 320851071, down: 160427136
frame_no: 39033, N: 201326592, ix: 159880640, right: 319759871, down: 159880704
frame_no: 37143, N: 201326592, ix: 152141056, right: 304278847, down: 152141120
frame_no: 37144, N: 201326592, ix: 152143104, right: 304284991, down: 152143168
frame_no: 36022, N: 201326592, ix: 147546304, right: 295092479, down: 147546368
frame_no: 34792, N: 201326592, ix: 142510976, right: 285019071, down: 142511040
frame_no: 34793, N: 201326592, ix: 142512576, right: 285024767, down: 142512640
frame_no: 34798, N: 201326592, ix: 142535360, right: 285068031, down: 142535424
frame_no: 36174, N: 201326592, ix: 148170560, right: 296339327, down: 148170624
frame_no: 36348, N: 201326592, ix: 148883008, right: 297764479, down: 148883072
frame_no: 36197, N: 201326592, ix: 148263680, right: 296526655, down: 148263744
frame_no: 38538, N: 201326592, ix: 157855488, right: 315707199, down: 157855552
frame_no: 38646, N: 201326592, ix: 158297856, right: 316591935, down: 158297920
frame_no: 38439, N: 201326592, ix: 157446528, right: 314892735, down: 157446592
frame_no: 39309, N: 201326592, ix: 161010944, right: 322020671, down: 161011008
frame_no: 39310, N: 201326592, ix: 161014464, right: 322028287, down: 161014528
frame_no: 36253, N: 201326592, ix: 148496320, right: 296988671, down: 148492288
frame_no: 38603, N: 201326592, ix: 158120704, right: 316238655, down: 158120768
frame_no: 35921, N: 201326592, ix: 147134464, right: 294266943, down: 147134528
frame_no: 36881, N: 201326592, ix: 151067904, right: 302132543, down: 151067968
frame_no: 36894, N: 201326592, ix: 151120960, right: 302238847, down: 151121024
frame_no: 37059, N: 201326592, ix: 151795840, right: 303589567, down: 151795904
frame_no: 37337, N: 201326592, ix: 152933824, right: 305866239, down: 152933888
frame_no: 37344, N: 201326592, ix: 152964608, right: 305925695, down: 152964672
frame_no: 37347, N: 201326592, ix: 152974912, right: 305948287, down: 152974976
<truncated output>
runing time: : 0.38s
所以在我看来很明显，在上述大多数情况下，您对 right 的计算超出了 N-1 的允许范围。
此处给出了 right 的计算结果：
right = (frame_no*nx+(((((mm + 1) % L) == 1) ? (mm+l) : ((mm - 1 + nx) % nx))));
我不知道如何为您解决这个问题。您可以类似地在该语句之后使用 printf 来突出显示计算的索引非法的情况：
if (right >= N) printf("....", ....);
顺便说一句，我注意到在对上述代码进行更改后，运行在 compute-sanitizer 下修改后的代码不会产生任何错误。因此，我怀疑当前代码中与您报告的错误相关的大部分问题都与 right 的非法 calculation/usage 有关。