Skip to content

Commit 6f25f5d

Browse files
xinjin01xinjin01
authored
Fix samples memory leak and cleanup issues (#137)
* Fix samples memory leak and cleanup issues Fix memory leaks and cleanup logic in the following sample programs: - samples/core/binaries - samples/core/blur - samples/core/reduce - samples/core/saxpy - samples/core/multi-device - samples/extensions/khr/externalmemory Also fix minor logic and error-handling inconsistencies such as: - Ensuring proper release of cl_event objects - Correct ordering of Vulkan cleanup calls - Improved version string checks and CLI option allocation Signed-off-by: Xin Jin <xin.jin@arm.com> * Fix use-after-free on dev_version in version logic Fix a use-after-free warning reported by GCC: error: pointer 'dev_version' may be used after 'free' else if (opencl_version_contains(dev_version, "2.")) Previously, dev_version was freed immediately after checking for OpenCL 1.0/1.1, but later reused for determining whether to add -cl-std compiler options. This triggered -Werror=use-after-free under GCC with strict diagnostics. This patch: - Defers free(dev_version) to a new 'ver:' cleanup label to avoid premature free. - Replaces raw malloc/free with MEM_CHECK for safety and consistency. - Adds snprintf-based -cl-std option handling with bounds checking. - Moves compiler_options logic earlier to avoid stale pointer usage. - Removes old compiler_options block near clBuildProgram. The refactoring ensures correctness, eliminates UB, and prepares for cleaner version-based behavior in the future. Signed-off-by: Xin Jin <xin.jin@arm.com> * Replace label endl with end To address review comment Signed-off-by: Xin Jin <xin.jin@arm.com> * Remove redundant NULL check before free(dev_version) Calling free() on a NULL pointer is well-defined and has no effect. This simplifies the code by removing an unnecessary conditional. Signed-off-by: Xin Jin <xin.jin@arm.com> --------- Signed-off-by: Xin Jin <xin.jin@arm.com>
1 parent 8cd6194 commit 6f25f5d

7 files changed

Lines changed: 143 additions & 81 deletions

File tree

‎samples/core/binaries/main.c‎

Lines changed: 5 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -258,21 +258,21 @@ int main(int argc, char *argv[])
258258
error, vec);
259259

260260
/// Run kernel
261-
cl_event pass;
262261
OCLERROR_RET(clSetKernelArg(Collatz, 0, sizeof(cl_mem), &buf), error, buff);
263262

264263
GET_CURRENT_TIMER(start_time)
264+
cl_event pass;
265265
OCLERROR_RET(clEnqueueNDRangeKernel(queue, Collatz, 1, &start, &length,
266266
NULL, 0, NULL, &pass),
267267
error, buff);
268-
OCLERROR_RET(clWaitForEvents(1, &pass), error, buff);
268+
OCLERROR_RET(clWaitForEvents(1, &pass), error, ev);
269269
GET_CURRENT_TIMER(end_time)
270270

271271
if (diag_opts.verbose) print_timings(start_time, end_time, &pass, 1);
272272

273273
OCLERROR_RET(clEnqueueReadBuffer(queue, buf, CL_BLOCKING, 0,
274274
sizeof(cl_int) * length, v, 0, NULL, NULL),
275-
error, buff);
275+
error, ev);
276276

277277
/// Show results
278278
int max_steps = 0;
@@ -298,6 +298,8 @@ int main(int argc, char *argv[])
298298
length, start + 1, max_steps, max_ind);
299299

300300
/// Cleanup
301+
ev:
302+
OCLERROR_RET(clReleaseEvent(pass), end_error, buff);
301303
buff:
302304
OCLERROR_RET(clReleaseMemObject(buf), end_error, vec);
303305
vec:

‎samples/core/blur/main.c‎

Lines changed: 55 additions & 34 deletions
Original file line numberDiff line numberDiff line change
@@ -359,19 +359,20 @@ cl_int single_pass_box_blur(state *const s, cl_int size)
359359
OCLERROR_RET(clEnqueueNDRangeKernel(s->queue, blur, 2, origin, image_size,
360360
NULL, 0, NULL, &pass),
361361
error, blr);
362-
OCLERROR_RET(clWaitForEvents(1, &pass), error, blr);
362+
OCLERROR_RET(clWaitForEvents(1, &pass), error, ev);
363363
GET_CURRENT_TIMER(end)
364364

365365
OCLERROR_RET(clEnqueueReadImage(s->queue, s->output_image_buf, CL_BLOCKING,
366366
origin, image_size, 0, 0,
367367
s->output_image.pixels, 0, NULL, NULL),
368-
error, blr);
368+
error, ev);
369369

370370
if (s->verbose) print_timings(start, end, &pass, 1);
371371

372372
// write output file
373-
OCLERROR_RET(finalize_blur(s), error, blr);
374-
373+
OCLERROR_RET(finalize_blur(s), error, ev);
374+
ev:
375+
clReleaseEvent(pass);
375376
blr:
376377
clReleaseKernel(blur);
377378
end:
@@ -417,20 +418,23 @@ cl_int dual_pass_box_blur(state *const s, cl_int size)
417418
error, blr2);
418419
OCLERROR_RET(clEnqueueNDRangeKernel(s->queue, blur2, 2, origin, image_size,
419420
NULL, 0, NULL, pass + 1),
420-
error, blr2);
421-
OCLERROR_RET(clWaitForEvents(2, pass), error, blr2);
421+
error, ev1);
422+
OCLERROR_RET(clWaitForEvents(2, pass), error, ev2);
422423
GET_CURRENT_TIMER(end)
423424

424425
OCLERROR_RET(clEnqueueReadImage(s->queue, s->output_image_buf, CL_BLOCKING,
425426
origin, image_size, 0, 0,
426427
s->output_image.pixels, 0, NULL, NULL),
427-
error, blr2);
428+
error, ev2);
428429

429430
if (s->verbose) print_timings(start, end, pass, 2);
430431

431432
// write output file
432-
OCLERROR_RET(finalize_blur(s), error, blr2);
433-
433+
OCLERROR_RET(finalize_blur(s), error, ev2);
434+
ev2:
435+
OCLERROR_RET(clReleaseEvent(pass[1]), end_error, ev1);
436+
ev1:
437+
OCLERROR_RET(clReleaseEvent(pass[0]), end_error, blr2);
434438
blr2:
435439
OCLERROR_RET(clReleaseKernel(blur2), end_error, blr1);
436440
blr1:
@@ -531,20 +535,23 @@ cl_int dual_pass_local_memory_exchange_box_blur(state *const s, cl_int size)
531535
size_t wgss[3] = { 1, wgs2, 1 };
532536
OCLERROR_RET(clEnqueueNDRangeKernel(s->queue, blur2, 2, origin, work_size2,
533537
wgss, 0, NULL, pass + 1),
534-
error, blr2);
535-
OCLERROR_RET(clWaitForEvents(2, pass), error, blr2);
538+
error, ev1);
539+
OCLERROR_RET(clWaitForEvents(2, pass), error, ev2);
536540
GET_CURRENT_TIMER(end)
537541

538542
OCLERROR_RET(clEnqueueReadImage(s->queue, s->output_image_buf, CL_BLOCKING,
539543
origin, image_size, 0, 0,
540544
s->output_image.pixels, 0, NULL, NULL),
541-
error, blr2);
545+
error, ev2);
542546

543547
if (s->verbose) print_timings(start, end, pass, 2);
544548

545549
// write output file
546-
OCLERROR_RET(finalize_blur(s), error, blr2);
547-
550+
OCLERROR_RET(finalize_blur(s), error, ev2);
551+
ev2:
552+
OCLERROR_RET(clReleaseEvent(pass[1]), end_error, ev1);
553+
ev1:
554+
OCLERROR_RET(clReleaseEvent(pass[0]), end_error, blr2);
548555
blr2:
549556
OCLERROR_RET(clReleaseKernel(blur2), end_error, blr1);
550557
blr1:
@@ -620,21 +627,25 @@ cl_int dual_pass_subgroup_exchange_box_blur(state *const s, cl_int size)
620627
size_t wgss[3] = { 1, wgs2, 1 };
621628
OCLERROR_RET(clEnqueueNDRangeKernel(s->queue, blur2, 2, origin, work_size2,
622629
wgss, 0, NULL, pass + 1),
623-
error, blr2);
624-
OCLERROR_RET(clWaitForEvents(2, pass), error, blr2);
630+
error, ev1);
631+
OCLERROR_RET(clWaitForEvents(2, pass), error, ev2);
625632
GET_CURRENT_TIMER(end)
626633

627634
OCLERROR_RET(clEnqueueReadImage(s->queue, s->output_image_buf, CL_BLOCKING,
628635
origin, image_size, 0, 0,
629636
s->output_image.pixels, 0, NULL, NULL),
630-
error, blr2);
637+
error, ev2);
631638

632639
if (s->verbose) print_timings(start, end, pass, 2);
633640

634641
// write output file
635-
OCLERROR_RET(finalize_blur(s), error, blr2);
642+
OCLERROR_RET(finalize_blur(s), error, ev2);
636643

637644
// cleanup for error handling
645+
ev2:
646+
OCLERROR_RET(clReleaseEvent(pass[1]), end_error, ev1);
647+
ev1:
648+
OCLERROR_RET(clReleaseEvent(pass[0]), end_error, blr2);
638649
blr2:
639650
OCLERROR_RET(clReleaseKernel(blur2), end_error, blr1);
640651
blr1:
@@ -685,20 +696,23 @@ cl_int dual_pass_kernel_blur(state *const s, cl_int size, cl_mem kern)
685696
error, blr2);
686697
OCLERROR_RET(clEnqueueNDRangeKernel(s->queue, blur2, 2, origin, image_size,
687698
NULL, 0, NULL, pass + 1),
688-
error, blr2);
689-
OCLERROR_RET(clWaitForEvents(2, pass), error, blr2);
699+
error, ev1);
700+
OCLERROR_RET(clWaitForEvents(2, pass), error, ev2);
690701
GET_CURRENT_TIMER(end)
691702

692703
OCLERROR_RET(clEnqueueReadImage(s->queue, s->output_image_buf, CL_BLOCKING,
693704
origin, image_size, 0, 0,
694705
s->output_image.pixels, 0, NULL, NULL),
695-
error, blr2);
706+
error, ev2);
696707

697708
if (s->verbose) print_timings(start, end, pass, 2);
698709

699710
// write output file
700-
OCLERROR_RET(finalize_blur(s), error, blr2);
701-
711+
OCLERROR_RET(finalize_blur(s), error, ev2);
712+
ev2:
713+
OCLERROR_RET(clReleaseEvent(pass[1]), end_error, ev1);
714+
ev1:
715+
OCLERROR_RET(clReleaseEvent(pass[0]), end_error, blr2);
702716
blr2:
703717
OCLERROR_RET(clReleaseKernel(blur2), end_error, blr1);
704718
blr1:
@@ -801,20 +815,23 @@ cl_int dual_pass_local_memory_exchange_kernel_blur(state *const s, cl_int size,
801815
size_t wgss[3] = { 1, wgs2, 1 };
802816
OCLERROR_RET(clEnqueueNDRangeKernel(s->queue, blur2, 2, origin, work_size2,
803817
wgss, 0, NULL, pass + 1),
804-
error, blr2);
805-
OCLERROR_RET(clWaitForEvents(2, pass), error, blr2);
818+
error, ev1);
819+
OCLERROR_RET(clWaitForEvents(2, pass), error, ev2);
806820
GET_CURRENT_TIMER(end)
807821

808822
OCLERROR_RET(clEnqueueReadImage(s->queue, s->output_image_buf, CL_BLOCKING,
809823
origin, image_size, 0, 0,
810824
s->output_image.pixels, 0, NULL, NULL),
811-
error, blr2);
825+
error, ev2);
812826

813827
if (s->verbose) print_timings(start, end, pass, 2);
814828

815829
// write output file
816-
OCLERROR_RET(finalize_blur(s), error, blr2);
817-
830+
OCLERROR_RET(finalize_blur(s), error, ev2);
831+
ev2:
832+
OCLERROR_RET(clReleaseEvent(pass[1]), end_error, ev1);
833+
ev1:
834+
OCLERROR_RET(clReleaseEvent(pass[0]), end_error, blr2);
818835
blr2:
819836
OCLERROR_RET(clReleaseKernel(blur2), end_error, blr1);
820837
blr1:
@@ -894,21 +911,25 @@ cl_int dual_pass_subgroup_exchange_kernel_blur(state *const s, cl_int size,
894911
size_t wgss[3] = { 1, wgs2, 1 };
895912
OCLERROR_RET(clEnqueueNDRangeKernel(s->queue, blur2, 2, origin, work_size2,
896913
wgss, 0, NULL, pass + 1),
897-
error, blr2);
898-
OCLERROR_RET(clWaitForEvents(2, pass), error, blr2);
914+
error, ev1);
915+
OCLERROR_RET(clWaitForEvents(2, pass), error, ev2);
899916
GET_CURRENT_TIMER(end)
900917

901918
OCLERROR_RET(clEnqueueReadImage(s->queue, s->output_image_buf, CL_BLOCKING,
902919
origin, image_size, 0, 0,
903920
s->output_image.pixels, 0, NULL, NULL),
904-
error, blr2);
921+
error, ev2);
905922

906923
if (s->verbose) print_timings(start, end, pass, 2);
907924

908925
// write output file
909-
OCLERROR_RET(finalize_blur(s), error, blr2);
926+
OCLERROR_RET(finalize_blur(s), error, ev2);
910927

911928
// cleanup for error handling
929+
ev2:
930+
OCLERROR_RET(clReleaseEvent(pass[1]), end_error, ev1);
931+
ev1:
932+
OCLERROR_RET(clReleaseEvent(pass[0]), end_error, blr2);
912933
blr2:
913934
OCLERROR_RET(clReleaseKernel(blur2), end_error, blr1);
914935
blr1:
@@ -1073,11 +1094,11 @@ int main(int argc, char *argv[])
10731094
dev_version_size, dev_version, NULL),
10741095
error, dev);
10751096
char compiler_options[1024] = "";
1076-
if (opencl_version_contains(dev_version, "3."))
1097+
if (opencl_version_contains(dev_version, "OpenCL C 3."))
10771098
{
10781099
strcat(compiler_options, "-cl-std=CL3.0 ");
10791100
}
1080-
else if (opencl_version_contains(dev_version, "2."))
1101+
else if (opencl_version_contains(dev_version, "OpenCL C 2."))
10811102
{
10821103
strcat(compiler_options, "-cl-std=CL2.0 ");
10831104
}

‎samples/core/multi-device/main.c‎

Lines changed: 49 additions & 23 deletions
Original file line numberDiff line numberDiff line change
@@ -208,12 +208,6 @@ int main(int argc, char* argv[])
208208
dev_opts.triplet.dev_type, &error),
209209
error, end);
210210

211-
// Query OpenCL version supported by device.
212-
char dev_version[64];
213-
OCLERROR_RET(clGetDeviceInfo(dev, CL_DEVICE_VERSION, sizeof(dev_version),
214-
&dev_version, NULL),
215-
error, end);
216-
217211
if (!diag_opts.quiet)
218212
{
219213
cl_util_print_device_info(dev);
@@ -225,6 +219,21 @@ int main(int argc, char* argv[])
225219
fflush(stdout);
226220
}
227221

222+
// Query OpenCL version supported by device.
223+
size_t dev_version_size;
224+
225+
OCLERROR_RET(
226+
clGetDeviceInfo(dev, CL_DEVICE_VERSION, 0, NULL, &dev_version_size),
227+
error, end);
228+
229+
char compiler_options[1023] = "";
230+
char* dev_version = NULL;
231+
MEM_CHECK(dev_version = (char*)malloc(dev_version_size), error, end);
232+
233+
OCLERROR_RET(clGetDeviceInfo(dev, CL_DEVICE_VERSION, dev_version_size,
234+
dev_version, NULL),
235+
error, ver);
236+
228237
if (opencl_version_contains(dev_version, "1.0")
229238
|| opencl_version_contains(dev_version, "1.1"))
230239
{
@@ -233,15 +242,44 @@ int main(int argc, char* argv[])
233242
"1.2 feature, but the device chosen only supports OpenCL %s. "
234243
"Please try with a different OpenCL device instead.\n",
235244
dev_version);
236-
exit(EXIT_SUCCESS);
245+
error = CL_SUCCESS;
246+
goto ver;
247+
}
248+
else
249+
{
250+
// If no -cl-std option is specified then the highest 1.x version
251+
// supported by each device is used to compile the program. Therefore,
252+
// it's only necessary to add the -cl-std option for 2.0 and 3.0 OpenCL
253+
// versions.
254+
255+
int written = 0;
256+
if (opencl_version_contains(dev_version, "3."))
257+
{
258+
written = snprintf(compiler_options, sizeof(compiler_options),
259+
"-cl-std=CL3.0 ");
260+
}
261+
else if (opencl_version_contains(dev_version, "2."))
262+
{
263+
written = snprintf(compiler_options, sizeof(compiler_options),
264+
"-cl-std=CL2.0 ");
265+
}
266+
267+
if (written < 0 || written >= (int)sizeof(compiler_options))
268+
{
269+
fprintf(
270+
stderr,
271+
"Error: compiler_options buffer overflow or encoding error.\n");
272+
free(dev_version);
273+
exit(EXIT_FAILURE);
274+
}
237275
}
238276

239277
// Check if device supports fission.
240278
cl_device_partition_property* dev_props = NULL;
241279
size_t props_size = 0;
242280
OCLERROR_RET(clGetDeviceInfo(dev, CL_DEVICE_PARTITION_PROPERTIES, 0, NULL,
243281
&props_size),
244-
error, end);
282+
error, ver);
245283
if (props_size == 0)
246284
{
247285
fprintf(stdout,
@@ -254,7 +292,7 @@ int main(int argc, char* argv[])
254292

255293
// Check if the "partition equally" type is supported.
256294
MEM_CHECK(dev_props = (cl_device_partition_property*)malloc(props_size),
257-
error, end);
295+
error, ver);
258296
OCLERROR_RET(clGetDeviceInfo(dev, CL_DEVICE_PARTITION_PROPERTIES,
259297
props_size, dev_props, NULL),
260298
error, props);
@@ -328,20 +366,6 @@ int main(int argc, char* argv[])
328366
context, 1, (const char**)&kernel, &program_size, &error),
329367
error, ker);
330368

331-
// If no -cl-std option is specified then the highest 1.x version
332-
// supported by each device is used to compile the program. Therefore,
333-
// it's only necessary to add the -cl-std option for 2.0 and 3.0 OpenCL
334-
// versions.
335-
char compiler_options[1023] = "";
336-
if (opencl_version_contains(dev_version, "3."))
337-
{
338-
strcat(compiler_options, "-cl-std=CL3.0 ");
339-
}
340-
else if (opencl_version_contains(dev_version, "2."))
341-
{
342-
strcat(compiler_options, "-cl-std=CL2.0 ");
343-
}
344-
345369
OCLERROR_RET(
346370
clBuildProgram(program, 2, subdevices, compiler_options, NULL, NULL),
347371
error, prg);
@@ -701,6 +725,8 @@ int main(int argc, char* argv[])
701725
free(subdevices);
702726
props:
703727
free(dev_props);
728+
ver:
729+
free(dev_version);
704730
end:
705731
if (error) cl_util_print_error(error);
706732
return error;

0 commit comments

Comments
 (0)