git.oblomov.eu Git - clinfo/blob - src/clinfo.c

   1 /* Collect all available information on all available devices
   2  * on all available OpenCL platforms present in the system
   3  */
   4
   5 #include <time.h>
   6 #include <string.h>
   7 #include <dlfcn.h>
   8
   9 #ifndef RTLD_DEFAULT
  10 #define RTLD_DEFAULT ((void*)0)
  11 #endif
  12
  13 /* ISO C forbids assignments between function pointers and void pointers,
  14  * but POSIX allows it. To compile without warnings even in -pedantic mode,
  15  * we use this horrible trick to get a function address from
  16  * clGetExtensionFunctionAddress
  17  */
  18 #define PTR_FUNC_PTR *(void**)&
  19
  20 /* Load STDC format macros (PRI*), or define them
  21  * for those crappy, non-standard compilers
  22  */
  23 #include "fmtmacros.h"
  24
  25 // Support for the horrible MS C compiler
  26 #ifdef _MSC_VER
  27 #include "ms_support.h"
  28 #endif
  29
  30 #include "ext.h"
  31 #include "error.h"
  32 #include "memory.h"
  33 #include "strbuf.h"
  34
  35 #define ARRAY_SIZE(ar) (sizeof(ar)/sizeof(*ar))
  36 #define UNUSED __attribute__((unused))
  37
  38 struct platform_data {
  39         char *pname; /* CL_PLATFORM_NAME */
  40         char *sname; /* CL_PLATFORM_ICD_SUFFIX_KHR or surrogate */
  41         cl_uint ndevs; /* number of devices */
  42         cl_bool has_amd_offline; /* has cl_amd_offline_devices extension */
  43 };
  44
  45 struct platform_info_checks {
  46         int has_khr_icd;
  47         cl_uint plat_version;
  48 };
  49
  50 cl_uint num_platforms;
  51 cl_platform_id *platform;
  52 /* highest version exposed by any platform: if the OpenCL library (the ICD loader)
  53  * has a lower version, problems may arise (such as API calls causing segfaults
  54  * or any other unexpected behavior
  55  */
  56 cl_uint max_plat_version;
  57 /* auto-detected OpenCL version support for the ICD loader */
  58 cl_uint icdl_ocl_version_found = 10;
  59 /* OpenCL version support declared by the ICD loader */
  60 cl_uint icdl_ocl_version;
  61
  62 struct platform_data *pdata;
  63 /* maximum length of a platform's sname */
  64 size_t platform_sname_maxlen;
  65 /* maximum number of devices */
  66 cl_uint maxdevs;
  67 /* line prefix, used to identify the platform/device for each
  68  * device property in RAW output mode */
  69 char *line_pfx;
  70 int line_pfx_len;
  71
  72 cl_uint num_devs_all;
  73
  74 cl_device_id *all_devices;
  75
  76 enum output_modes {
  77         CLINFO_HUMAN = 1, /* more human readable */
  78         CLINFO_RAW = 2, /* property-by-property */
  79         CLINFO_BOTH = CLINFO_HUMAN | CLINFO_RAW
  80 };
  81
  82 enum output_modes output_mode = CLINFO_HUMAN;
  83
  84 /* Specify if we should only be listing the platform and devices;
  85  * can be done in both human and raw mode, and only the platform
  86  * and device names (and number) will be shown
  87  * TODO check if terminal supports UTF-8 and use Unicode line-drawing
  88  * for the tree in list mode
  89  */
  90 cl_bool list_only = CL_FALSE;
  91
  92 static const char unk[] = "Unknown";
  93 static const char none[] = "None";
  94 static const char none_raw[] = "CL_NONE";
  95 static const char na[] = "n/a"; // not available
  96 static const char core[] = "core"; // not available
  97
  98 static const char bytes_str[] = " bytes";
  99 static const char pixels_str[] = " pixels";
 100 static const char images_str[] = " images";
 101
 102 static const char* bool_str[] = { "No", "Yes" };
 103 static const char* bool_raw_str[] = { "CL_FALSE", "CL_TRUE" };
 104
 105 static const char* endian_str[] = { "Big-Endian", "Little-Endian" };
 106
 107 static const cl_device_type devtype[] = { 0,
 108         CL_DEVICE_TYPE_DEFAULT, CL_DEVICE_TYPE_CPU, CL_DEVICE_TYPE_GPU,
 109         CL_DEVICE_TYPE_ACCELERATOR, CL_DEVICE_TYPE_CUSTOM, CL_DEVICE_TYPE_ALL };
 110
 111 const size_t devtype_count = ARRAY_SIZE(devtype);
 112
 113 static const char* device_type_str[] = { unk, "Default", "CPU", "GPU", "Accelerator", "Custom", "All" };
 114 static const char* device_type_raw_str[] = { unk,
 115         "CL_DEVICE_TYPE_DEFAULT", "CL_DEVICE_TYPE_CPU", "CL_DEVICE_TYPE_GPU",
 116         "CL_DEVICE_TYPE_ACCELERATOR", "CL_DEVICE_TYPE_CUSTOM", "CL_DEVICE_TYPE_ALL"
 117 };
 118
 119 static const char* partition_type_str[] = {
 120         "none specified", none, "equally", "by counts", "by affinity domain", "by names (Intel)"
 121 };
 122 static const char* partition_type_raw_str[] = {
 123         "NONE SPECIFIED",
 124         none_raw,
 125         "CL_DEVICE_PARTITION_EQUALLY_EXT",
 126         "CL_DEVICE_PARTITION_BY_COUNTS_EXT",
 127         "CL_DEVICE_PARTITION_BY_AFFINITY_DOMAIN_EXT",
 128         "CL_DEVICE_PARTITION_BY_NAMES_INTEL_EXT"
 129 };
 130
 131 static const char numa[] = "NUMA";
 132 static const char l1cache[] = "L1 cache";
 133 static const char l2cache[] = "L2 cache";
 134 static const char l3cache[] = "L3 cache";
 135 static const char l4cache[] = "L4 cache";
 136
 137 static const char* affinity_domain_str[] = {
 138         numa, l4cache, l3cache, l2cache, l1cache, "next partitionable"
 139 };
 140
 141 static const char* affinity_domain_ext_str[] = {
 142         numa, l4cache, l3cache, l2cache, l1cache, "next fissionable"
 143 };
 144
 145 static const char* affinity_domain_raw_str[] = {
 146         "CL_DEVICE_AFFINITY_DOMAIN_NUMA",
 147         "CL_DEVICE_AFFINITY_DOMAIN_L4_CACHE",
 148         "CL_DEVICE_AFFINITY_DOMAIN_L3_CACHE",
 149         "CL_DEVICE_AFFINITY_DOMAIN_L2_CACHE",
 150         "CL_DEVICE_AFFINITY_DOMAIN_L1_CACHE",
 151         "CL_DEVICE_AFFINITY_DOMAIN_NEXT_PARTITIONABLE"
 152 };
 153
 154 static const char* affinity_domain_raw_ext_str[] = {
 155         "CL_AFFINITY_DOMAIN_NUMA_EXT",
 156         "CL_AFFINITY_DOMAIN_L4_CACHE_EXT",
 157         "CL_AFFINITY_DOMAIN_L3_CACHE_EXT",
 158         "CL_AFFINITY_DOMAIN_L2_CACHE_EXT",
 159         "CL_AFFINITY_DOMAIN_L1_CACHE_EXT",
 160         "CL_AFFINITY_DOMAIN_NEXT_FISSIONABLE_EXT"
 161 };
 162
 163 const size_t affinity_domain_count = ARRAY_SIZE(affinity_domain_str);
 164
 165 static const char* fp_conf_str[] = {
 166         "Denormals", "Infinity and NANs", "Round to nearest", "Round to zero",
 167         "Round to infinity", "IEEE754-2008 fused multiply-add",
 168         "Support is emulated in software",
 169         "Correctly-rounded divide and sqrt operations"
 170 };
 171
 172 static const char* fp_conf_raw_str[] = {
 173         "CL_FP_DENORM",
 174         "CL_FP_INF_NAN",
 175         "CL_FP_ROUND_TO_NEAREST",
 176         "CL_FP_ROUND_TO_ZERO",
 177         "CL_FP_ROUND_TO_INF",
 178         "CL_FP_FMA",
 179         "CL_FP_SOFT_FLOAT",
 180         "CL_FP_CORRECTLY_ROUNDED_DIVIDE_SQRT"
 181 };
 182
 183 const size_t fp_conf_count = ARRAY_SIZE(fp_conf_str);
 184
 185 static const char* svm_cap_str[] = {
 186         "Coarse-grained buffer sharing",
 187         "Fine-grained buffer sharing",
 188         "Fine-grained system sharing",
 189         "Atomics"
 190 };
 191
 192 static const char* svm_cap_raw_str[] = {
 193         "CL_DEVICE_SVM_COARSE_GRAIN_BUFFER",
 194         "CL_DEVICE_SVM_FINE_GRAIN_BUFFER",
 195         "CL_DEVICE_SVM_FINE_GRAIN_SYSTEM",
 196         "CL_DEVICE_SVM_ATOMICS",
 197 };
 198
 199 const size_t svm_cap_count = ARRAY_SIZE(svm_cap_str);
 200
 201 static const char* memsfx[] = {
 202         "B", "KiB", "MiB", "GiB", "TiB"
 203 };
 204
 205 const size_t memsfx_count = ARRAY_SIZE(memsfx);
 206
 207 static const char* lmem_type_str[] = { none, "Local", "Global" };
 208 static const char* lmem_type_raw_str[] = { none_raw, "CL_LOCAL", "CL_GLOBAL" };
 209 static const char* cache_type_str[] = { none, "Read-Only", "Read/Write" };
 210 static const char* cache_type_raw_str[] = { none_raw, "CL_READ_ONLY_CACHE", "CL_READ_WRITE_CACHE" };
 211
 212 static const char* queue_prop_str[] = { "Out-of-order execution", "Profiling" };
 213 static const char* queue_prop_raw_str[] = {
 214         "CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE",
 215         "CL_QUEUE_PROFILING_ENABLE"
 216 };
 217
 218 const size_t queue_prop_count = ARRAY_SIZE(queue_prop_str);
 219
 220 static const char* execap_str[] = { "Run OpenCL kernels", "Run native kernels" };
 221 static const char* execap_raw_str[] = {
 222         "CL_EXEC_KERNEL",
 223         "CL_EXEC_NATIVE_KERNEL"
 224 };
 225
 226 const size_t execap_count = ARRAY_SIZE(execap_str);
 227
 228 static const char* sources[] = {
 229         "#define GWO(type) global type* restrict\n",
 230         "#define GRO(type) global const type* restrict\n",
 231         "#define BODY int i = get_global_id(0); out[i] = in1[i] + in2[i]\n",
 232         "#define _KRN(T, N) void kernel sum##N(GWO(T##N) out, GRO(T##N) in1, GRO(T##N) in2) { BODY; }\n",
 233         "#define KRN(N) _KRN(float, N)\n",
 234         "KRN()\n/* KRN(2)\nKRN(4)\nKRN(8)\nKRN(16) */\n",
 235 };
 236
 237 const char *no_plat(void)
 238 {
 239         return output_mode == CLINFO_HUMAN ?
 240                 "No platform" :
 241                 "CL_INVALID_PLATFORM";
 242 }
 243
 244 const char *no_dev(void)
 245 {
 246         return output_mode == CLINFO_HUMAN ?
 247                 "No devices found in platform" :
 248                 "CL_DEVICE_NOT_FOUND";
 249 }
 250
 251 const char *no_dev_avail(void)
 252 {
 253         return output_mode == CLINFO_HUMAN ?
 254                 "No devices available in platform" :
 255                 "CL_DEVICE_NOT_AVAILABLE";
 256 }
 257
 258
 259 /* preferred workgroup size multiple for each kernel
 260  * have not found a platform where the WG multiple changes,
 261  * but keep this flexible (this can grow up to 5)
 262  */
 263 #define NUM_KERNELS 1
 264 size_t wgm[NUM_KERNELS];
 265
 266 #define INDENT "  "
 267 #define I0_STR "%-48s  "
 268 #define I1_STR "  %-46s  "
 269 #define I2_STR "    %-44s  "
 270
 271 static const char empty_str[] = "";
 272 static const char spc_str[] = " ";
 273 static const char times_str[] = "x";
 274 static const char comma_str[] = ", ";
 275 static const char vbar_str[] = " | ";
 276
 277 int had_error = 0;
 278 const char *cur_sfx = empty_str;
 279
 280 /* parse a CL_DEVICE_VERSION or CL_PLATFORM_VERSION info to determine the OpenCL version.
 281  * Returns an unsigned integer in the form major*10 + minor
 282  */
 283 cl_uint
 284 getOpenCLVersion(const char *version)
 285 {
 286         cl_uint ret = 10;
 287         long parse = 0;
 288         const char *from = version;
 289         char *next = NULL;
 290         parse = strtol(from, &next, 10);
 291
 292         if (next != from) {
 293                 ret = parse*10;
 294                 // skip the dot TODO should we actually check for the dot?
 295                 from = ++next;
 296                 parse = strtol(from, &next, 10);
 297                 if (next != from)
 298                         ret += parse;
 299         }
 300         return ret;
 301 }
 302
 303
 304 /* print strbuf, prefixed by pname, skipping leading whitespace if skip is nonzero,
 305  * affixing cur_sfx */
 306 static inline
 307 void show_strbuf(const char *pname, int skip)
 308 {
 309         printf("%s" I1_STR "%s%s\n",
 310                 line_pfx, pname,
 311                 (skip ? skip_leading_ws(strbuf) : strbuf),
 312                 had_error ? empty_str : cur_sfx);
 313 }
 314
 315 int
 316 platform_info_str(cl_platform_id pid, cl_platform_info param, const char* pname, const struct platform_info_checks * chk UNUSED)
 317 {
 318         error = clGetPlatformInfo(pid, param, 0, NULL, &nusz);
 319         had_error = REPORT_ERROR2("get %s size");
 320         if (!had_error) {
 321                 if (nusz > bufsz) {
 322                         REALLOC(strbuf, nusz, current_param);
 323                         bufsz = nusz;
 324                 }
 325                 error = clGetPlatformInfo(pid, param, bufsz, strbuf, NULL);
 326                 had_error = REPORT_ERROR2("get %s");
 327         }
 328         /* when only listing, do not print anything we're just gathering
 329          * information
 330          */
 331         if (!list_only)
 332                 show_strbuf(pname, 1);
 333         return had_error;
 334 }
 335
 336 int
 337 platform_info_ulong(cl_platform_id pid, cl_platform_info param, const char* pname, const struct platform_info_checks * chk UNUSED)
 338 {
 339         cl_ulong val = 0;
 340
 341         error = clGetPlatformInfo(pid, param, sizeof(val), &val, NULL);
 342         had_error = REPORT_ERROR2("get %s");
 343         /* when only listing, do not print anything we're just gathering
 344          * information
 345          */
 346         if (!list_only) {
 347                 if (had_error)
 348                         show_strbuf(pname, 0);
 349                 else
 350                         printf("%s" I1_STR "%" PRIu64 "%s\n", line_pfx, pname, val, cur_sfx);
 351         }
 352         return had_error;
 353 }
 354
 355 struct platform_info_traits {
 356         cl_platform_info param; // CL_PLATFORM_*
 357         const char *sname; // "CL_PLATFORM_*"
 358         const char *pname; // "Platform *"
 359         const char *sfx; // suffix for the output in non-raw mode
 360         /* pointer to function that shows the parameter */
 361         int (*show_func)(cl_platform_id pid, cl_platform_info param, const char *pname, const struct platform_info_checks *);
 362         /* pointer to function that checks if the parameter should be checked */
 363         int (*check_func)(const struct platform_info_checks *);
 364 };
 365
 366 int khr_icd_p(const struct platform_info_checks *chk)
 367 {
 368         return chk->has_khr_icd;
 369 }
 370
 371 int plat_is_21(const struct platform_info_checks *chk)
 372 {
 373         return !(chk->plat_version < 21);
 374 }
 375
 376 #define PINFO_COND(symbol, name, sfx, typ, funcptr) { symbol, #symbol, "Platform " name, sfx, &platform_info_##typ, &funcptr }
 377 #define PINFO(symbol, name, sfx, typ) { symbol, #symbol, "Platform " name, sfx, &platform_info_##typ, NULL }
 378 struct platform_info_traits pinfo_traits[] = {
 379         PINFO(CL_PLATFORM_NAME, "Name", NULL, str),
 380         PINFO(CL_PLATFORM_VENDOR, "Vendor", NULL, str),
 381         PINFO(CL_PLATFORM_VERSION, "Version", NULL, str),
 382         PINFO(CL_PLATFORM_PROFILE, "Profile", NULL, str),
 383         PINFO(CL_PLATFORM_EXTENSIONS, "Extensions", NULL, str),
 384         PINFO_COND(CL_PLATFORM_HOST_TIMER_RESOLUTION, "Host timer resolution", "ns", ulong, plat_is_21),
 385         PINFO_COND(CL_PLATFORM_ICD_SUFFIX_KHR, "Extensions function suffix", NULL, str, khr_icd_p)
 386 };
 387
 388 /* Print platform info and prepare arrays for device info */
 389 void
 390 printPlatformInfo(cl_uint p)
 391 {
 392         cl_platform_id pid = platform[p];
 393         size_t len = 0;
 394
 395         struct platform_info_checks pinfo_checks = { 0, 10 };
 396
 397         current_function = __func__;
 398
 399         for (current_line = 0; current_line < ARRAY_SIZE(pinfo_traits); ++current_line) {
 400                 const struct platform_info_traits *traits = pinfo_traits + current_line;
 401                 const char *pname = (output_mode == CLINFO_HUMAN ?
 402                         traits->pname : traits->sname);
 403
 404                 current_param = traits->sname;
 405
 406                 if (traits->check_func && !traits->check_func(&pinfo_checks))
 407                         continue;
 408
 409                 cur_sfx = (output_mode == CLINFO_HUMAN && traits->sfx) ? traits->sfx : empty_str;
 410
 411                 had_error = traits->show_func(pid, traits->param,
 412                         pname, &pinfo_checks);
 413
 414                 if (had_error)
 415                         continue;
 416
 417                 /* post-processing */
 418
 419                 switch (traits->param) {
 420                 case CL_PLATFORM_NAME:
 421                         /* Store name for future reference */
 422                         len = strlen(strbuf);
 423                         ALLOC(pdata[p].pname, len+1, "platform name copy");
 424                         /* memcpy instead of strncpy since we already have the len
 425                          * and memcpy is possibly more optimized */
 426                         memcpy(pdata[p].pname, strbuf, len);
 427                         pdata[p].pname[len] = '\0';
 428                         break;
 429                 case CL_PLATFORM_VERSION:
 430                         /* compute numeric value for OpenCL version */
 431                         pinfo_checks.plat_version = getOpenCLVersion(strbuf + 7);
 432                         break;
 433                 case CL_PLATFORM_EXTENSIONS:
 434                         pinfo_checks.has_khr_icd = !!strstr(strbuf, "cl_khr_icd");
 435                         pdata[p].has_amd_offline = !!strstr(strbuf, "cl_amd_offline_devices");
 436                         break;
 437                 case CL_PLATFORM_ICD_SUFFIX_KHR:
 438                         /* Store ICD suffix for future reference */
 439                         len = strlen(strbuf);
 440                         ALLOC(pdata[p].sname, len+1, "platform ICD suffix copy");
 441                         /* memcpy instead of strncpy since we already have the len
 442                          * and memcpy is possibly more optimized */
 443                         memcpy(pdata[p].sname, strbuf, len);
 444                         pdata[p].sname[len] = '\0';
 445                 default:
 446                         /* do nothing */
 447                         break;
 448                 }
 449
 450         }
 451
 452         if (pinfo_checks.plat_version > max_plat_version)
 453                 max_plat_version = pinfo_checks.plat_version;
 454
 455         /* if no CL_PLATFORM_ICD_SUFFIX_KHR, use P### as short/symbolic name */
 456         if (!pdata[p].sname) {
 457 #define SNAME_MAX 32
 458                 ALLOC(pdata[p].sname, SNAME_MAX, "platform symbolic name");
 459                 snprintf(pdata[p].sname, SNAME_MAX, "P%u", p);
 460         }
 461
 462         len = strlen(pdata[p].sname);
 463         if (len > platform_sname_maxlen)
 464                 platform_sname_maxlen = len;
 465
 466         error = clGetDeviceIDs(pid, CL_DEVICE_TYPE_ALL, 0, NULL, &(pdata[p].ndevs));
 467         if (error == CL_DEVICE_NOT_FOUND)
 468                 pdata[p].ndevs = 0;
 469         else
 470                 CHECK_ERROR("number of devices");
 471
 472         num_devs_all += pdata[p].ndevs;
 473
 474         if (pdata[p].ndevs > maxdevs)
 475                 maxdevs = pdata[p].ndevs;
 476 }
 477
 478 int
 479 getWGsizes(cl_platform_id pid, cl_device_id dev)
 480 {
 481         int ret = 0;
 482
 483 #define RR_ERROR(what) do { \
 484         had_error = REPORT_ERROR(what); \
 485         if (had_error) { \
 486                 ret = error; \
 487                 goto out; \
 488         } \
 489 } while(0)
 490
 491
 492         cl_context_properties ctxpft[] = {
 493                 CL_CONTEXT_PLATFORM, (cl_context_properties)pid,
 494                 0, 0 };
 495         cl_uint cursor = 0;
 496         cl_context ctx = NULL;
 497         cl_program prg = NULL;
 498         cl_kernel krn = NULL;
 499
 500         ctx = clCreateContext(ctxpft, 1, &dev, NULL, NULL, &error);
 501         RR_ERROR("create context");
 502         prg = clCreateProgramWithSource(ctx, ARRAY_SIZE(sources), sources, NULL, &error);
 503         RR_ERROR("create program");
 504         error = clBuildProgram(prg, 1, &dev, NULL, NULL, NULL);
 505         had_error = REPORT_ERROR("build program");
 506         if (had_error)
 507                 ret = error;
 508
 509         /* for a program build failure, dump the log to stderr before bailing */
 510         if (error == CL_BUILD_PROGRAM_FAILURE) {
 511                 /* Do not clobber strbuf, shadow it */
 512                 char *strbuf = NULL;
 513                 size_t bufsz = 0, nusz = 0;
 514                 GET_STRING(clGetProgramBuildInfo, CL_PROGRAM_BUILD_LOG, "CL_PROGRAM_BUILD_LOG", prg, dev);
 515                 if (error == CL_SUCCESS) {
 516                         fflush(stdout);
 517                         fflush(stderr);
 518                         fputs("=== CL_PROGRAM_BUILD_LOG ===\n", stderr);
 519                         fputs(strbuf, stderr);
 520                         fflush(stderr);
 521                 }
 522                 free(strbuf);
 523         }
 524         if (had_error)
 525                 goto out;
 526
 527         for (cursor = 0; cursor < NUM_KERNELS; ++cursor) {
 528                 snprintf(strbuf, bufsz, "sum%u", 1<<cursor);
 529                 if (cursor == 0)
 530                         strbuf[3] = 0; // scalar kernel is called 'sum'
 531                 krn = clCreateKernel(prg, strbuf, &error);
 532                 RR_ERROR("create kernel");
 533                 error = clGetKernelWorkGroupInfo(krn, dev, CL_KERNEL_PREFERRED_WORK_GROUP_SIZE_MULTIPLE,
 534                         sizeof(*wgm), wgm + cursor, NULL);
 535                 RR_ERROR("get kernel info");
 536                 clReleaseKernel(krn);
 537                 krn = NULL;
 538         }
 539
 540 out:
 541         if (krn)
 542                 clReleaseKernel(krn);
 543         if (prg)
 544                 clReleaseProgram(prg);
 545         if (ctx)
 546                 clReleaseContext(ctx);
 547         return ret;
 548 }
 549
 550 /*
 551  * Device properties/extensions used in traits checks, and relevant functions
 552  */
 553
 554 struct device_info_checks {
 555         cl_device_type devtype;
 556         cl_device_mem_cache_type cachetype;
 557         cl_device_local_mem_type lmemtype;
 558         cl_bool image_support;
 559         cl_bool compiler_available;
 560         char has_half[12];
 561         char has_double[24];
 562         char has_nv[29];
 563         char has_amd[30];
 564         char has_amd_svm[11];
 565         char has_arm_svm[29];
 566         char has_fission[22];
 567         char has_atomic_counters[26];
 568         char has_image2d_buffer[27];
 569         char has_intel_local_thread[30];
 570         char has_intel_AME[36];
 571         char has_intel_AVC_ME[43];
 572         char has_intel_planar_yuv[20];
 573         char has_intel_required_subgroup_size[32];
 574         char has_altera_dev_temp[29];
 575         char has_spir[12];
 576         char has_qcom_ext_host_ptr[21];
 577         char has_simultaneous_sharing[30];
 578         cl_uint dev_version;
 579 };
 580
 581 #define DEFINE_EXT_CHECK(ext) int dev_has_##ext(const struct device_info_checks *chk) \
 582 { \
 583         return !!(chk->has_##ext[0]); \
 584 }
 585
 586 DEFINE_EXT_CHECK(half)
 587 DEFINE_EXT_CHECK(double)
 588 DEFINE_EXT_CHECK(nv)
 589 DEFINE_EXT_CHECK(amd)
 590 DEFINE_EXT_CHECK(amd_svm)
 591 DEFINE_EXT_CHECK(arm_svm)
 592 DEFINE_EXT_CHECK(fission)
 593 DEFINE_EXT_CHECK(atomic_counters)
 594 DEFINE_EXT_CHECK(image2d_buffer)
 595 DEFINE_EXT_CHECK(intel_local_thread)
 596 DEFINE_EXT_CHECK(intel_AME)
 597 DEFINE_EXT_CHECK(intel_AVC_ME)
 598 DEFINE_EXT_CHECK(intel_planar_yuv)
 599 DEFINE_EXT_CHECK(intel_required_subgroup_size)
 600 DEFINE_EXT_CHECK(altera_dev_temp)
 601 DEFINE_EXT_CHECK(spir)
 602 DEFINE_EXT_CHECK(qcom_ext_host_ptr)
 603 DEFINE_EXT_CHECK(simultaneous_sharing)
 604
 605 /* In the version checks we negate the opposite conditions
 606  * instead of double-negating the actual condition
 607  */
 608
 609 // device supports 1.2
 610 int dev_is_12(const struct device_info_checks *chk)
 611 {
 612         return !(chk->dev_version < 12);
 613 }
 614
 615 // device supports 2.0
 616 int dev_is_20(const struct device_info_checks *chk)
 617 {
 618         return !(chk->dev_version < 20);
 619 }
 620
 621 // device supports 2.1
 622 int dev_is_21(const struct device_info_checks *chk)
 623 {
 624         return !(chk->dev_version < 21);
 625 }
 626
 627 // device does not support 2.0
 628 int dev_not_20(const struct device_info_checks *chk)
 629 {
 630         return !(chk->dev_version >= 20);
 631 }
 632
 633
 634 int dev_is_gpu(const struct device_info_checks *chk)
 635 {
 636         return !!(chk->devtype & CL_DEVICE_TYPE_GPU);
 637 }
 638
 639 int dev_is_gpu_amd(const struct device_info_checks *chk)
 640 {
 641         return dev_is_gpu(chk) && dev_has_amd(chk);
 642 }
 643
 644 int dev_has_svm(const struct device_info_checks *chk)
 645 {
 646         return dev_is_20(chk) || dev_has_amd_svm(chk);
 647 }
 648
 649 int dev_has_partition(const struct device_info_checks *chk)
 650 {
 651         return dev_is_12(chk) || dev_has_fission(chk);
 652 }
 653
 654 int dev_has_cache(const struct device_info_checks *chk)
 655 {
 656         return chk->cachetype != CL_NONE;
 657 }
 658
 659 int dev_has_lmem(const struct device_info_checks *chk)
 660 {
 661         return chk->lmemtype != CL_NONE;
 662 }
 663
 664 int dev_has_images(const struct device_info_checks *chk)
 665 {
 666         return chk->image_support;
 667 }
 668
 669 int dev_has_images_12(const struct device_info_checks *chk)
 670 {
 671         return dev_has_images(chk) && dev_is_12(chk);
 672 }
 673
 674 int dev_has_images_20(const struct device_info_checks *chk)
 675 {
 676         return dev_has_images(chk) && dev_is_20(chk);
 677 }
 678
 679 int dev_has_compiler(const struct device_info_checks *chk)
 680 {
 681         return chk->compiler_available;
 682 }
 683
 684
 685 void identify_device_extensions(const char *extensions, struct device_info_checks *chk)
 686 {
 687 #define _HAS_EXT(ext) (strstr(extensions, ext))
 688 #define HAS_EXT(ext) _HAS_EXT(#ext)
 689 #define CPY_EXT(what, ext) do { \
 690         strncpy(chk->has_##what, has, sizeof(ext)); \
 691         chk->has_##what[sizeof(ext)-1] = '\0'; \
 692 } while (0)
 693 #define CHECK_EXT(what, ext) do { \
 694         has = _HAS_EXT(#ext); \
 695         if (has) CPY_EXT(what, #ext); \
 696 } while(0)
 697
 698         char *has;
 699         CHECK_EXT(half, cl_khr_fp16);
 700         CHECK_EXT(spir, cl_khr_spir);
 701         CHECK_EXT(double, cl_khr_fp64);
 702         if (!dev_has_double(chk))
 703                 CHECK_EXT(double, cl_amd_fp64);
 704         if (!dev_has_double(chk))
 705                 CHECK_EXT(double, cl_APPLE_fp64_basic_ops);
 706         CHECK_EXT(nv, cl_nv_device_attribute_query);
 707         CHECK_EXT(amd, cl_amd_device_attribute_query);
 708         CHECK_EXT(amd_svm, cl_amd_svm);
 709         CHECK_EXT(arm_svm, cl_arm_shared_virtual_memory);
 710         CHECK_EXT(fission, cl_ext_device_fission);
 711         CHECK_EXT(atomic_counters, cl_ext_atomic_counters_64);
 712         if (dev_has_atomic_counters(chk))
 713                 CHECK_EXT(atomic_counters, cl_ext_atomic_counters_32);
 714         CHECK_EXT(image2d_buffer, cl_khr_image2d_from_buffer);
 715         CHECK_EXT(intel_local_thread, cl_intel_exec_by_local_thread);
 716         CHECK_EXT(intel_AME, cl_intel_advanced_motion_estimation);
 717         CHECK_EXT(intel_AVC_ME, cl_intel_device_side_avc_motion_estimation);
 718         CHECK_EXT(intel_planar_yuv, cl_intel_planar_yuv);
 719         CHECK_EXT(intel_required_subgroup_size, cl_intel_required_subgroup_size);
 720         CHECK_EXT(altera_dev_temp, cl_altera_device_temperature);
 721         CHECK_EXT(qcom_ext_host_ptr, cl_qcom_ext_host_ptr);
 722         CHECK_EXT(simultaneous_sharing, cl_intel_simultaneous_sharing);
 723 }
 724
 725
 726
 727 /*
 728  * Device info print functions
 729  */
 730
 731 #define _GET_VAL \
 732         error = clGetDeviceInfo(dev, param, sizeof(val), &val, NULL); \
 733         had_error = REPORT_ERROR2("get %s");
 734
 735 #define _GET_VAL_ARRAY \
 736         error = clGetDeviceInfo(dev, param, 0, NULL, &szval); \
 737         had_error = REPORT_ERROR2("get number of %s"); \
 738         numval = szval/sizeof(val); \
 739         if (!had_error) { \
 740                 REALLOC(val, numval, current_param); \
 741                 error = clGetDeviceInfo(dev, param, szval, val, NULL); \
 742                 had_error = REPORT_ERROR("get %s"); \
 743         }
 744
 745 #define GET_VAL do { \
 746         _GET_VAL \
 747 } while (0)
 748
 749 #define GET_VAL_ARRAY do { \
 750         _GET_VAL_ARRAY \
 751 } while (0)
 752
 753 #define _FMT_VAL(fmt) \
 754         if (had_error) \
 755                 show_strbuf(pname, 0); \
 756         else \
 757                 printf("%s" I1_STR fmt "%s\n", line_pfx, pname, val, cur_sfx);
 758
 759 #define FMT_VAL(fmt) do { \
 760         _FMT_VAL(fmt) \
 761 } while (0)
 762
 763 #define SHOW_VAL(fmt) do { \
 764         _GET_VAL \
 765         _FMT_VAL(fmt) \
 766 } while (0)
 767
 768 #define DEFINE_DEVINFO_SHOW(how, type, fmt) \
 769 int device_info_##how(cl_device_id dev, cl_device_info param, const char *pname, \
 770         const struct device_info_checks *chk UNUSED) \
 771 { \
 772         type val = 0; \
 773         SHOW_VAL(fmt); \
 774         return had_error; \
 775 }
 776
 777 /* Get string-type info without showing it */
 778 int device_info_str_get(cl_device_id dev, cl_device_info param, const char *pname,
 779         const struct device_info_checks *chk UNUSED)
 780 {
 781         current_param = pname;
 782         error = clGetDeviceInfo(dev, param, 0, NULL, &nusz);
 783         had_error = REPORT_ERROR2("get %s size");
 784         if (!had_error) {
 785                 if (nusz > bufsz) {
 786                         REALLOC(strbuf, nusz, current_param);
 787                         bufsz = nusz;
 788                 }
 789                 error = clGetDeviceInfo(dev, param, bufsz, strbuf, NULL);
 790                 had_error = REPORT_ERROR2("get %s");
 791         }
 792         return had_error;
 793 }
 794
 795 int device_info_str(cl_device_id dev, cl_device_info param, const char *pname,
 796         const struct device_info_checks *chk)
 797 {
 798         had_error = device_info_str_get(dev, param, pname, chk);
 799         show_strbuf(pname, 1);
 800         return had_error;
 801 }
 802
 803 DEFINE_DEVINFO_SHOW(int, cl_uint, "%u")
 804 DEFINE_DEVINFO_SHOW(hex, cl_uint, "0x%x")
 805 DEFINE_DEVINFO_SHOW(long, cl_ulong, "%" PRIu64)
 806 DEFINE_DEVINFO_SHOW(sz, size_t, "%" PRIuS)
 807
 808 int device_info_bool(cl_device_id dev, cl_device_info param, const char *pname,
 809         const struct device_info_checks *chk UNUSED)
 810 {
 811         cl_bool val = 0;
 812         const char * const * str = (output_mode == CLINFO_HUMAN ?
 813                 bool_str : bool_raw_str);
 814         GET_VAL;
 815         if (had_error)
 816                 show_strbuf(pname, 0);
 817         else {
 818                 printf("%s" I1_STR "%s%s\n", line_pfx, pname, str[val], cur_sfx);
 819                 /* abuse strbuf to pass the bool value up to the caller,
 820                  * this is used e.g. by CL_DEVICE_IMAGE_SUPPORT
 821                  */
 822                 memcpy(strbuf, &val, sizeof(val));
 823         }
 824         return had_error;
 825 }
 826
 827 int device_info_bits(cl_device_id dev, cl_device_info param, const char *pname,
 828         const struct device_info_checks *chk UNUSED)
 829 {
 830         cl_uint val;
 831         GET_VAL;
 832         if (!had_error)
 833                 sprintf(strbuf, "%u bits (%u bytes)", val, val/8);
 834         show_strbuf(pname, 0);
 835         return had_error;
 836 }
 837
 838
 839 size_t strbuf_mem(cl_ulong val, size_t szval)
 840 {
 841         double dbl = val;
 842         size_t sfx = 0;
 843         while (dbl > 1024 && sfx < memsfx_count) {
 844                 dbl /= 1024;
 845                 ++sfx;
 846         }
 847         return sprintf(strbuf + szval, " (%.4lg%s)",
 848                 dbl, memsfx[sfx]);
 849 }
 850
 851 int device_info_mem(cl_device_id dev, cl_device_info param, const char *pname,
 852         const struct device_info_checks *chk UNUSED)
 853 {
 854         cl_ulong val = 0;
 855         size_t szval = 0;
 856         GET_VAL;
 857         if (!had_error) {
 858                 szval += sprintf(strbuf, "%" PRIu64, val);
 859                 if (output_mode == CLINFO_HUMAN && val > 1024)
 860                         strbuf_mem(val, szval);
 861         }
 862         show_strbuf(pname, 0);
 863         return had_error;
 864 }
 865
 866 int device_info_mem_int(cl_device_id dev, cl_device_info param, const char *pname,
 867         const struct device_info_checks *chk UNUSED)
 868 {
 869         cl_uint val = 0;
 870         size_t szval = 0;
 871         GET_VAL;
 872         if (!had_error) {
 873                 szval += sprintf(strbuf, "%u", val);
 874                 if (output_mode == CLINFO_HUMAN && val > 1024)
 875                         strbuf_mem(val, szval);
 876         }
 877         show_strbuf(pname, 0);
 878         return had_error;
 879 }
 880
 881 int device_info_free_mem_amd(cl_device_id dev, cl_device_info param, const char *pname,
 882         const struct device_info_checks *chk UNUSED)
 883 {
 884         size_t *val = NULL;
 885         size_t szval = 0, numval = 0;
 886         GET_VAL_ARRAY;
 887         if (!had_error) {
 888                 size_t cursor = 0;
 889                 szval = 0;
 890                 for (cursor = 0; cursor < numval; ++cursor) {
 891                         if (szval > 0) {
 892                                 strbuf[szval] = ' ';
 893                                 ++szval;
 894                         }
 895                         szval += sprintf(strbuf + szval, "%" PRIuS, val[cursor]);
 896                         if (output_mode == CLINFO_HUMAN)
 897                                 szval += strbuf_mem(val[cursor]*UINT64_C(1024), szval);
 898                 }
 899         }
 900         show_strbuf(pname, 0);
 901         free(val);
 902         return had_error;
 903 }
 904
 905 int device_info_time_offset(cl_device_id dev, cl_device_info param, const char *pname,
 906         const struct device_info_checks *chk UNUSED)
 907 {
 908         cl_ulong val = 0;
 909         GET_VAL;
 910         if (!had_error) {
 911                 size_t szval = 0;
 912                 time_t time = val/UINT64_C(1000000000);
 913                 szval += snprintf(strbuf, bufsz, "%" PRIu64 "ns (", val);
 914                 szval += bufcpy(szval, ctime(&time));
 915                 /* overwrite ctime's newline with the closing parenthesis */
 916                 if (szval < bufsz)
 917                         strbuf[szval - 1] = ')';
 918         }
 919         show_strbuf(pname, 0);
 920         return had_error;
 921 }
 922
 923 int device_info_szptr(cl_device_id dev, cl_device_info param, const char *pname,
 924         const struct device_info_checks *chk UNUSED)
 925 {
 926         size_t *val = NULL;
 927         size_t szval = 0, numval = 0;
 928         GET_VAL_ARRAY;
 929         if (!had_error) {
 930                 size_t counter = 0;
 931                 set_separator(output_mode == CLINFO_HUMAN ? times_str : spc_str);
 932                 szval = 0;
 933                 for (counter = 0; counter < numval; ++counter) {
 934                         add_separator(&szval);
 935                         szval += snprintf(strbuf + szval, bufsz - szval - 1, "%" PRIuS, val[counter]);
 936                         if (szval >= bufsz) {
 937                                 trunc_strbuf();
 938                                 break;
 939                         }
 940                 }
 941         }
 942         show_strbuf(pname, 0);
 943         free(val);
 944         return had_error;
 945 }
 946
 947 int device_info_wg(cl_device_id dev, cl_device_info param UNUSED, const char *pname,
 948         const struct device_info_checks *chk UNUSED)
 949 {
 950         cl_platform_id val = NULL;
 951         {
 952                 /* shadow */
 953                 cl_device_info param = CL_DEVICE_PLATFORM;
 954                 current_param = "CL_DEVICE_PLATFORM";
 955                 GET_VAL;
 956         }
 957         current_param = pname;
 958         if (!had_error)
 959                 had_error = getWGsizes(val, dev);
 960         if (!had_error) {
 961                 sprintf(strbuf, "%" PRIuS, wgm[0]);
 962         }
 963         show_strbuf(pname, 0);
 964         return had_error;
 965 }
 966
 967 int device_info_img_sz_2d(cl_device_id dev, cl_device_info param, const char *pname,
 968         const struct device_info_checks *chk UNUSED)
 969 {
 970         size_t width = 0, height = 0, val = 0;
 971         GET_VAL; /* HEIGHT */
 972         if (!had_error) {
 973                 height = val;
 974                 param = CL_DEVICE_IMAGE2D_MAX_WIDTH;
 975                 current_param = "CL_DEVICE_IMAGE2D_MAX_WIDTH";
 976                 GET_VAL;
 977                 if (!had_error) {
 978                         width = val;
 979                         sprintf(strbuf, "%" PRIuS "x%" PRIuS, width, height);
 980                 }
 981         }
 982         show_strbuf(pname, 0);
 983         return had_error;
 984 }
 985
 986 int device_info_img_sz_intel_planar_yuv(cl_device_id dev, cl_device_info param, const char *pname,
 987         const struct device_info_checks *chk UNUSED)
 988 {
 989         size_t width = 0, height = 0, val = 0;
 990         GET_VAL; /* HEIGHT */
 991         if (!had_error) {
 992                 height = val;
 993                 param = CL_DEVICE_PLANAR_YUV_MAX_WIDTH_INTEL;
 994                 current_param = "CL_DEVICE_PLANAR_YUV_MAX_WIDTH_INTEL";
 995                 GET_VAL;
 996                 if (!had_error) {
 997                         width = val;
 998                         sprintf(strbuf, "%" PRIuS "x%" PRIuS, width, height);
 999                 }
1000         }
1001         show_strbuf(pname, 0);
1002         return had_error;
1003 }
1004
1005
1006 int device_info_img_sz_3d(cl_device_id dev, cl_device_info param, const char *pname,
1007         const struct device_info_checks *chk UNUSED)
1008 {
1009         size_t width = 0, height = 0, depth = 0, val = 0;
1010         GET_VAL; /* HEIGHT */
1011         if (!had_error) {
1012                 height = val;
1013                 param = CL_DEVICE_IMAGE3D_MAX_WIDTH;
1014                 current_param = "CL_DEVICE_IMAGE3D_MAX_WIDTH";
1015                 GET_VAL;
1016                 if (!had_error) {
1017                         width = val;
1018                         param = CL_DEVICE_IMAGE3D_MAX_DEPTH;
1019                         current_param = "CL_DEVICE_IMAGE3D_MAX_DEPTH";
1020                         GET_VAL;
1021                         if (!had_error) {
1022                                 depth = val;
1023                                 sprintf(strbuf, "%" PRIuS "x%" PRIuS "x%" PRIuS,
1024                                         width, height, depth);
1025                         }
1026                 }
1027         }
1028         show_strbuf(pname, 0);
1029         return had_error;
1030 }
1031
1032
1033 int device_info_devtype(cl_device_id dev, cl_device_info param, const char *pname,
1034         const struct device_info_checks *chk UNUSED)
1035 {
1036         cl_device_type val = 0;
1037         GET_VAL;
1038         if (!had_error) {
1039                 /* iterate over device type strings, appending their textual form
1040                  * to strbuf.
1041                  * TODO: check for extra bits/no bits
1042                  */
1043                 cl_uint i = devtype_count - 1; /* skip CL_DEVICE_TYPE_ALL */
1044                 const char * const *devstr = (output_mode == CLINFO_HUMAN ?
1045                         device_type_str : device_type_raw_str);
1046                 size_t szval = 0;
1047                 strbuf[szval] = '\0';
1048                 set_separator(output_mode == CLINFO_HUMAN ? comma_str : vbar_str);
1049                 for (; i > 0; --i) {
1050                         /* assemble CL_DEVICE_TYPE_* from index i */
1051                         cl_device_type cur = (cl_device_type)(1) << (i-1);
1052                         if (val & cur) {
1053                                 /* match: add separator if not first match */
1054                                 add_separator(&szval);
1055                                 szval += bufcpy(szval, devstr[i]);
1056                         }
1057                 }
1058         }
1059         show_strbuf(pname, 0);
1060         /* we abuse global strbuf to pass the device type over to the caller */
1061         if (!had_error)
1062                 memcpy(strbuf, &val, sizeof(val));
1063         return had_error;
1064 }
1065
1066 int device_info_cachetype(cl_device_id dev, cl_device_info param, const char *pname,
1067         const struct device_info_checks *chk UNUSED)
1068 {
1069         cl_device_mem_cache_type val = 0;
1070         GET_VAL;
1071         if (!had_error) {
1072                 const char * const *ar = (output_mode == CLINFO_HUMAN ?
1073                         cache_type_str : cache_type_raw_str);
1074                 bufcpy(0, ar[val]);
1075         }
1076         show_strbuf(pname, 0);
1077         /* we abuse global strbuf to pass the cache type over to the caller */
1078         if (!had_error)
1079                 memcpy(strbuf, &val, sizeof(val));
1080         return had_error;
1081 }
1082
1083 int device_info_lmemtype(cl_device_id dev, cl_device_info param, const char *pname,
1084         const struct device_info_checks *chk UNUSED)
1085 {
1086         cl_device_local_mem_type val = 0;
1087         GET_VAL;
1088         if (!had_error) {
1089                 const char * const *ar = (output_mode == CLINFO_HUMAN ?
1090                         lmem_type_str : lmem_type_raw_str);
1091                 bufcpy(0, ar[val]);
1092         }
1093         show_strbuf(pname, 0);
1094         /* we abuse global strbuf to pass the lmem type over to the caller */
1095         if (!had_error)
1096                 memcpy(strbuf, &val, sizeof(val));
1097         return had_error;
1098 }
1099
1100 /* stringify a cl_device_topology_amd */
1101 void devtopo_str(const cl_device_topology_amd *devtopo)
1102 {
1103         switch (devtopo->raw.type) {
1104         case 0:
1105                 if (output_mode == CLINFO_HUMAN)
1106                         sprintf(strbuf, "(%s)", na);
1107                 else
1108                         sprintf(strbuf, none_raw);
1109                 break;
1110         case CL_DEVICE_TOPOLOGY_TYPE_PCIE_AMD:
1111                 sprintf(strbuf, "PCI-E, %02x:%02x.%u",
1112                         (cl_uchar)(devtopo->pcie.bus),
1113                         devtopo->pcie.device, devtopo->pcie.function);
1114                 break;
1115         default:
1116                 sprintf(strbuf, "<unknown (%u): %u %u %u %u %u>",
1117                         devtopo->raw.type,
1118                         devtopo->raw.data[0], devtopo->raw.data[1],
1119                         devtopo->raw.data[2],
1120                         devtopo->raw.data[3], devtopo->raw.data[4]);
1121         }
1122 }
1123
1124 int device_info_devtopo_amd(cl_device_id dev, cl_device_info param, const char *pname,
1125         const struct device_info_checks *chk UNUSED)
1126 {
1127         cl_device_topology_amd val;
1128         GET_VAL;
1129         /* TODO how to do this in CLINFO_RAW mode */
1130         if (!had_error) {
1131                 devtopo_str(&val);
1132         }
1133         show_strbuf(pname, 0);
1134         return had_error;
1135 }
1136
1137 /* we assemble a cl_device_topology_amd struct from the NVIDIA info */
1138 int device_info_devtopo_nv(cl_device_id dev, cl_device_info param, const char *pname,
1139         const struct device_info_checks *chk UNUSED)
1140 {
1141         cl_device_topology_amd devtopo;
1142         cl_uint val = 0;
1143
1144         devtopo.raw.type = CL_DEVICE_TOPOLOGY_TYPE_PCIE_AMD;
1145
1146         GET_VAL; /* CL_DEVICE_PCI_BUS_ID_NV */
1147
1148         if (!had_error) {
1149                 devtopo.pcie.bus = val & 0xff;
1150
1151                 param = CL_DEVICE_PCI_SLOT_ID_NV;
1152                 current_param = "CL_DEVICE_PCI_SLOT_ID_NV";
1153
1154                 GET_VAL;
1155
1156                 if (!had_error) {
1157                         devtopo.pcie.device = val >> 3;
1158                         devtopo.pcie.function = val & 7;
1159                         devtopo_str(&devtopo);
1160                 }
1161         }
1162
1163         show_strbuf(pname, 0);
1164         return had_error;
1165 }
1166
1167 /* NVIDIA Compute Capability */
1168 int device_info_cc_nv(cl_device_id dev, cl_device_info param, const char *pname,
1169         const struct device_info_checks *chk UNUSED)
1170 {
1171         cl_uint major = 0, val = 0;
1172         GET_VAL; /* MAJOR */
1173         if (!had_error) {
1174                 major = val;
1175                 param = CL_DEVICE_COMPUTE_CAPABILITY_MINOR_NV;
1176                 current_param = "CL_DEVICE_COMPUTE_CAPABILITY_MINOR_NV";
1177                 GET_VAL;
1178                 if (!had_error)
1179                         snprintf(strbuf, bufsz, "%u.%u", major, val);
1180         }
1181
1182         show_strbuf(pname, 0);
1183         return had_error;
1184 }
1185
1186 /* AMD GFXIP */
1187 int device_info_gfxip_amd(cl_device_id dev, cl_device_info param, const char *pname,
1188         const struct device_info_checks *chk UNUSED)
1189 {
1190         cl_uint major = 0, val = 0;
1191         GET_VAL; /* MAJOR */
1192         if (!had_error) {
1193                 major = val;
1194                 param = CL_DEVICE_GFXIP_MINOR_AMD;
1195                 current_param = "CL_DEVICE_GFXIP_MINOR_AMD";
1196                 GET_VAL;
1197                 if (!had_error)
1198                         snprintf(strbuf, bufsz, "%u.%u", major, val);
1199         }
1200
1201         show_strbuf(pname, 0);
1202         return had_error;
1203 }
1204
1205
1206 /* Device Partition, CLINFO_HUMAN header */
1207 int device_info_partition_header(cl_device_id dev UNUSED, cl_device_info param UNUSED,
1208         const char *pname, const struct device_info_checks *chk)
1209 {
1210         int is_12 = dev_is_12(chk);
1211         int has_fission = dev_has_fission(chk);
1212         size_t szval = snprintf(strbuf, bufsz, "(%s%s%s)",
1213                 (is_12 ? core : empty_str),
1214                 (is_12 && has_fission ? comma_str : empty_str),
1215                 chk->has_fission);
1216         if (szval >= bufsz)
1217                 trunc_strbuf();
1218
1219         show_strbuf(pname, 0);
1220         had_error = CL_SUCCESS;
1221         return had_error;
1222 }
1223
1224 /* Device partition properties */
1225 int device_info_partition_types(cl_device_id dev, cl_device_info param, const char *pname,
1226         const struct device_info_checks *chk UNUSED)
1227 {
1228         size_t numval = 0, szval = 0, cursor = 0, slen = 0;
1229         cl_device_partition_property *val = NULL;
1230         const char * const *ptstr = (output_mode == CLINFO_HUMAN ?
1231                 partition_type_str : partition_type_raw_str);
1232
1233         set_separator(output_mode == CLINFO_HUMAN ? comma_str : vbar_str);
1234
1235         GET_VAL_ARRAY;
1236
1237         szval = 0;
1238         if (!had_error) {
1239                 for (cursor = 0; cursor < numval; ++cursor) {
1240                         int str_idx = -1;
1241
1242                         /* add separator for values past the first */
1243                         add_separator(&szval);
1244
1245                         switch (val[cursor]) {
1246                         case 0: str_idx = 1; break;
1247                         case CL_DEVICE_PARTITION_EQUALLY: str_idx = 2; break;
1248                         case CL_DEVICE_PARTITION_BY_COUNTS: str_idx = 3; break;
1249                         case CL_DEVICE_PARTITION_BY_AFFINITY_DOMAIN: str_idx = 4; break;
1250                         case CL_DEVICE_PARTITION_BY_NAMES_INTEL: str_idx = 5; break;
1251                         default:
1252                                 szval += snprintf(strbuf + szval, bufsz - szval - 1, "by <unknown> (0x%" PRIXPTR ")", val[cursor]);
1253                                 break;
1254                         }
1255                         if (str_idx > 0) {
1256                                 /* string length, minus _EXT */
1257                                 slen = strlen(ptstr[str_idx]);
1258                                 if (output_mode == CLINFO_RAW && str_idx > 1)
1259                                         slen -= 4;
1260                                 szval += bufcpy_len(szval, ptstr[str_idx], slen);
1261                         }
1262                         if (szval >= bufsz) {
1263                                 trunc_strbuf();
1264                                 break;
1265                         }
1266                 }
1267                 if (szval == 0) {
1268                         bufcpy(szval, ptstr[0]);
1269                 } else if (szval < bufsz)
1270                         strbuf[szval] = '\0';
1271         }
1272
1273         show_strbuf(pname, 0);
1274
1275         free(val);
1276         return had_error;
1277 }
1278
1279 int device_info_partition_types_ext(cl_device_id dev, cl_device_info param, const char *pname,
1280         const struct device_info_checks *chk UNUSED)
1281 {
1282         size_t numval = 0, szval = 0, cursor = 0, slen = 0;
1283         cl_device_partition_property_ext *val = NULL;
1284         const char * const *ptstr = (output_mode == CLINFO_HUMAN ?
1285                 partition_type_str : partition_type_raw_str);
1286
1287         set_separator(output_mode == CLINFO_HUMAN ? comma_str : vbar_str);
1288
1289         GET_VAL_ARRAY;
1290
1291         szval = 0;
1292         if (!had_error) {
1293                 for (cursor = 0; cursor < numval; ++cursor) {
1294                         int str_idx = -1;
1295
1296                         /* add separator for values past the first */
1297                         add_separator(&szval);
1298
1299                         switch (val[cursor]) {
1300                         case 0: str_idx = 1; break;
1301                         case CL_DEVICE_PARTITION_EQUALLY_EXT: str_idx = 2; break;
1302                         case CL_DEVICE_PARTITION_BY_COUNTS_EXT: str_idx = 3; break;
1303                         case CL_DEVICE_PARTITION_BY_AFFINITY_DOMAIN_EXT: str_idx = 4; break;
1304                         case CL_DEVICE_PARTITION_BY_NAMES_EXT: str_idx = 5; break;
1305                         default:
1306                                 szval += snprintf(strbuf + szval, bufsz - szval - 1, "by <unknown> (0x%" PRIX64 ")", val[cursor]);
1307                                 break;
1308                         }
1309                         if (str_idx > 0) {
1310                                 /* string length */
1311                                 slen = strlen(ptstr[str_idx]);
1312                                 strncpy(strbuf + szval, ptstr[str_idx], slen);
1313                                 szval += slen;
1314                         }
1315                         if (szval >= bufsz) {
1316                                 trunc_strbuf();
1317                                 break;
1318                         }
1319                 }
1320                 if (szval == 0) {
1321                         slen = strlen(ptstr[0]);
1322                         memcpy(strbuf, ptstr[0], slen);
1323                         szval += slen;
1324                 }
1325                 if (szval < bufsz)
1326                         strbuf[szval] = '\0';
1327         }
1328
1329         show_strbuf(pname, 0);
1330
1331         free(val);
1332         return had_error;
1333 }
1334
1335
1336 /* Device partition affinity domains */
1337 int device_info_partition_affinities(cl_device_id dev, cl_device_info param, const char *pname,
1338         const struct device_info_checks *chk UNUSED)
1339 {
1340         cl_device_affinity_domain val;
1341         GET_VAL;
1342         if (!had_error && val) {
1343                 /* iterate over affinity domain strings appending their textual form
1344                  * to strbuf
1345                  * TODO: check for extra bits/no bits
1346                  */
1347                 size_t szval = 0;
1348                 cl_uint i = 0;
1349                 const char * const *affstr = (output_mode == CLINFO_HUMAN ?
1350                         affinity_domain_str : affinity_domain_raw_str);
1351                 set_separator(output_mode == CLINFO_HUMAN ? comma_str : vbar_str);
1352                 for (i = 0; i < affinity_domain_count; ++i) {
1353                         cl_device_affinity_domain cur = (cl_device_affinity_domain)(1) << i;
1354                         if (val & cur) {
1355                                 /* match: add separator if not first match */
1356                                 add_separator(&szval);
1357                                 szval += bufcpy(szval, affstr[i]);
1358                         }
1359                         if (szval >= bufsz)
1360                                 break;
1361                 }
1362         }
1363         if (val || had_error)
1364                 show_strbuf(pname, 0);
1365         return had_error;
1366 }
1367
1368 int device_info_partition_affinities_ext(cl_device_id dev, cl_device_info param, const char *pname,
1369         const struct device_info_checks *chk UNUSED)
1370 {
1371         size_t numval = 0, szval = 0, cursor = 0, slen = 0;
1372         cl_device_partition_property_ext *val = NULL;
1373         const char * const *ptstr = (output_mode == CLINFO_HUMAN ?
1374                 affinity_domain_ext_str : affinity_domain_raw_ext_str);
1375
1376         set_separator(output_mode == CLINFO_HUMAN ? comma_str : vbar_str);
1377
1378         GET_VAL_ARRAY;
1379
1380         szval = 0;
1381         if (!had_error) {
1382                 for (cursor = 0; cursor < numval; ++cursor) {
1383                         int str_idx = -1;
1384
1385                         /* add separator for values past the first */
1386                         add_separator(&szval);
1387
1388                         switch (val[cursor]) {
1389                         case CL_AFFINITY_DOMAIN_NUMA_EXT: str_idx = 0; break;
1390                         case CL_AFFINITY_DOMAIN_L4_CACHE_EXT: str_idx = 1; break;
1391                         case CL_AFFINITY_DOMAIN_L3_CACHE_EXT: str_idx = 2; break;
1392                         case CL_AFFINITY_DOMAIN_L2_CACHE_EXT: str_idx = 3; break;
1393                         case CL_AFFINITY_DOMAIN_L1_CACHE_EXT: str_idx = 4; break;
1394                         case CL_AFFINITY_DOMAIN_NEXT_FISSIONABLE_EXT: str_idx = 5; break;
1395                         default:
1396                                 szval += snprintf(strbuf + szval, bufsz - szval - 1, "<unknown> (0x%" PRIX64 ")", val[cursor]);
1397                                 break;
1398                         }
1399                         if (str_idx >= 0) {
1400                                 /* string length */
1401                                 const char *str = ptstr[str_idx];
1402                                 slen = strlen(str);
1403                                 strncpy(strbuf + szval, str, slen);
1404                                 szval += slen;
1405                         }
1406                         if (szval >= bufsz) {
1407                                 trunc_strbuf();
1408                                 break;
1409                         }
1410                 }
1411                 strbuf[szval] = '\0';
1412         }
1413
1414         show_strbuf(pname, 0);
1415
1416         free(val);
1417         return had_error;
1418 }
1419
1420 /* Preferred / native vector widths */
1421 int device_info_vecwidth(cl_device_id dev, cl_device_info param, const char *pname,
1422         const struct device_info_checks *chk)
1423 {
1424         cl_uint preferred = 0, val = 0;
1425         GET_VAL;
1426         if (!had_error) {
1427                 preferred = val;
1428
1429                 /* we get called with PREFERRED, NATIVE is at +0x30 offset, except for HALF,
1430                  * which is at +0x08 */
1431                 param += (param == CL_DEVICE_PREFERRED_VECTOR_WIDTH_HALF ? 0x08 : 0x30);
1432                 /* TODO update current_param */
1433                 GET_VAL;
1434
1435                 if (!had_error) {
1436                         size_t szval = 0;
1437                         const char *ext = (param == CL_DEVICE_NATIVE_VECTOR_WIDTH_HALF ?
1438                                 chk->has_half : (param == CL_DEVICE_NATIVE_VECTOR_WIDTH_DOUBLE ?
1439                                 chk->has_double : NULL));
1440                         szval = sprintf(strbuf, "%8u / %-8u", preferred, val);
1441                         if (ext)
1442                                 sprintf(strbuf + szval, " (%s)", *ext ? ext : na);
1443                 }
1444         }
1445         show_strbuf(pname, 0);
1446         return had_error;
1447 }
1448
1449 /* Floating-point configurations */
1450 int device_info_fpconf(cl_device_id dev, cl_device_info param, const char *pname,
1451         const struct device_info_checks *chk)
1452 {
1453         cl_device_fp_config val = 0;
1454         int get_it = (
1455                 (param == CL_DEVICE_SINGLE_FP_CONFIG) ||
1456                 (param == CL_DEVICE_HALF_FP_CONFIG && dev_has_half(chk)) ||
1457                 (param == CL_DEVICE_DOUBLE_FP_CONFIG && dev_has_double(chk)));
1458         if (get_it)
1459                 GET_VAL;
1460         else
1461                 had_error = CL_SUCCESS;
1462
1463         if (!had_error) {
1464                 size_t szval = 0;
1465                 cl_uint i = 0;
1466                 const char * const *fpstr = (output_mode == CLINFO_HUMAN ?
1467                         fp_conf_str : fp_conf_raw_str);
1468                 set_separator(vbar_str);
1469                 if (output_mode == CLINFO_HUMAN) {
1470                         const char *why = na;
1471                         switch (param) {
1472                         case CL_DEVICE_HALF_FP_CONFIG:
1473                                 if (get_it)
1474                                         why = chk->has_half;
1475                                 break;
1476                         case CL_DEVICE_SINGLE_FP_CONFIG:
1477                                 why = core;
1478                                 break;
1479                         case CL_DEVICE_DOUBLE_FP_CONFIG:
1480                                 if (get_it)
1481                                         why = chk->has_double;
1482                                 break;
1483                         default:
1484                                 /* "this can't happen" (unless OpenCL starts supporting _other_ floating-point formats, maybe) */
1485                                 fprintf(stderr, "unsupported floating-point configuration parameter %s\n", pname);
1486
1487                         }
1488                         /* show 'why' it's being shown */
1489                         szval += sprintf(strbuf, "(%s)", why);
1490                 }
1491                 if (get_it) {
1492                         for (i = 0; i < fp_conf_count; ++i) {
1493                                 cl_device_fp_config cur = (cl_device_fp_config)(1) << i;
1494                                 if (output_mode == CLINFO_HUMAN) {
1495                                         szval += sprintf(strbuf + szval, "\n%s" I2_STR "%s",
1496                                                 line_pfx, fpstr[i], bool_str[!!(val & cur)]);
1497                                 } else if (val & cur) {
1498                                         add_separator(&szval);
1499                                         szval += bufcpy(szval, fpstr[i]);
1500                                 }
1501                         }
1502                 }
1503         }
1504
1505         /* only print this for HUMAN output or if we actually got the value */
1506         if (output_mode == CLINFO_HUMAN || get_it)
1507                 show_strbuf(pname, 0);
1508         return had_error;
1509 }
1510
1511 /* Queue properties */
1512 int device_info_qprop(cl_device_id dev, cl_device_info param, const char *pname,
1513         const struct device_info_checks *chk)
1514 {
1515         cl_command_queue_properties val = 0;
1516         GET_VAL;
1517         if (!had_error) {
1518                 size_t szval = 0;
1519                 cl_uint i = 0;
1520                 const char * const *qpstr = (output_mode == CLINFO_HUMAN ?
1521                         queue_prop_str : queue_prop_raw_str);
1522                 set_separator(vbar_str);
1523                 for (i = 0; i < queue_prop_count; ++i) {
1524                         cl_command_queue_properties cur = (cl_command_queue_properties)(1) << i;
1525                         if (output_mode == CLINFO_HUMAN) {
1526                                 szval += sprintf(strbuf + szval, "\n%s" I2_STR "%s",
1527                                         line_pfx, qpstr[i], bool_str[!!(val & cur)]);
1528                         } else if (val & cur) {
1529                                 add_separator(&szval);
1530                                 szval += bufcpy(szval, qpstr[i]);
1531                         }
1532                 }
1533                 if (output_mode == CLINFO_HUMAN && param == CL_DEVICE_QUEUE_PROPERTIES &&
1534                         dev_has_intel_local_thread(chk))
1535                         sprintf(strbuf + szval, "\n%s" I2_STR "%s",
1536                                 line_pfx, "Local thread execution (Intel)", bool_str[CL_TRUE]);
1537         }
1538         show_strbuf(pname, 0);
1539         return had_error;
1540 }
1541
1542 /* Execution capbilities */
1543 int device_info_execap(cl_device_id dev, cl_device_info param, const char *pname,
1544         const struct device_info_checks *chk UNUSED)
1545 {
1546         cl_device_exec_capabilities val = 0;
1547         GET_VAL;
1548         if (!had_error) {
1549                 size_t szval = 0;
1550                 cl_uint i = 0;
1551                 const char * const *qpstr = (output_mode == CLINFO_HUMAN ?
1552                         execap_str : execap_raw_str);
1553                 set_separator(vbar_str);
1554                 for (i = 0; i < execap_count; ++i) {
1555                         cl_device_exec_capabilities cur = (cl_device_exec_capabilities)(1) << i;
1556                         if (output_mode == CLINFO_HUMAN) {
1557                                 szval += sprintf(strbuf + szval, "\n%s" I2_STR "%s",
1558                                         line_pfx, qpstr[i], bool_str[!!(val & cur)]);
1559                         } else if (val & cur) {
1560                                 add_separator(&szval);
1561                                 szval += bufcpy(szval, qpstr[i]);
1562                         }
1563                 }
1564         }
1565         show_strbuf(pname, 0);
1566         return had_error;
1567 }
1568
1569 /* Arch bits and endianness (HUMAN) */
1570 int device_info_arch(cl_device_id dev, cl_device_info param, const char *pname,
1571         const struct device_info_checks *chk UNUSED)
1572 {
1573         cl_uint bits = 0;
1574         {
1575                 cl_uint val = 0;
1576                 GET_VAL;
1577                 if (!had_error)
1578                         bits = val;
1579         }
1580         if (!had_error) {
1581                 cl_bool val = 0;
1582                 param = CL_DEVICE_ENDIAN_LITTLE;
1583                 current_param = "CL_DEVICE_ENDIAN_LITTLE";
1584                 GET_VAL;
1585                 if (!had_error)
1586                         sprintf(strbuf, "%u, %s", bits, endian_str[val]);
1587         }
1588         show_strbuf(pname, 0);
1589         return had_error;
1590 }
1591
1592 /* SVM capabilities */
1593 int device_info_svm_cap(cl_device_id dev, cl_device_info param, const char *pname,
1594         const struct device_info_checks *chk)
1595 {
1596         cl_device_svm_capabilities val = 0;
1597         const int is_20 = dev_is_20(chk);
1598         const int has_amd_svm = (param == CL_DEVICE_SVM_CAPABILITIES && dev_has_amd_svm(chk));
1599
1600         GET_VAL;
1601
1602         if (!had_error) {
1603                 size_t szval = 0;
1604                 cl_uint i = 0;
1605                 const char * const *scstr = (output_mode == CLINFO_HUMAN ?
1606                         svm_cap_str : svm_cap_raw_str);
1607                 set_separator(vbar_str);
1608                 if (output_mode == CLINFO_HUMAN && param == CL_DEVICE_SVM_CAPABILITIES) {
1609                         /* show 'why' it's being shown */
1610                         szval += sprintf(strbuf, "(%s%s%s)",
1611                                 (is_20 ? core : empty_str),
1612                                 (is_20 && has_amd_svm ? comma_str : empty_str),
1613                                 chk->has_amd_svm);
1614                 }
1615                 for (i = 0; i < svm_cap_count; ++i) {
1616                         cl_device_svm_capabilities cur = (cl_device_svm_capabilities)(1) << i;
1617                         if (output_mode == CLINFO_HUMAN) {
1618                                 szval += sprintf(strbuf + szval, "\n%s" I2_STR "%s",
1619                                         line_pfx, scstr[i], bool_str[!!(val & cur)]);
1620                         } else if (val & cur) {
1621                                 add_separator(&szval);
1622                                 szval += bufcpy(szval, scstr[i]);
1623                         }
1624                 }
1625         }
1626
1627         show_strbuf(pname, 0);
1628         return had_error;
1629 }
1630
1631 /*
1632  * Device info traits
1633  */
1634
1635 /* A CL_FALSE param means "just print pname" */
1636
1637 struct device_info_traits {
1638         enum output_modes output_mode;
1639         cl_device_info param; // CL_DEVICE_*
1640         const char *sname; // "CL_DEVICE_*"
1641         const char *pname; // "Device *"
1642         const char *sfx; // suffix for the output in non-raw mode
1643         /* pointer to function that shows the parameter */
1644         int (*show_func)(cl_device_id dev, cl_device_info param, const char *pname, const struct device_info_checks *);
1645         /* pointer to function that checks if the parameter should be checked */
1646         int (*check_func)(const struct device_info_checks *);
1647 };
1648
1649 #define DINFO_SFX(symbol, name, sfx, typ) symbol, #symbol, name, sfx, device_info_##typ
1650 #define DINFO(symbol, name, typ) symbol, #symbol, name, NULL, device_info_##typ
1651
1652 struct device_info_traits dinfo_traits[] = {
1653         { CLINFO_BOTH, DINFO(CL_DEVICE_NAME, "Device Name", str), NULL },
1654         { CLINFO_BOTH, DINFO(CL_DEVICE_VENDOR, "Device Vendor", str), NULL },
1655         { CLINFO_BOTH, DINFO(CL_DEVICE_VENDOR_ID, "Device Vendor ID", hex), NULL },
1656         { CLINFO_BOTH, DINFO(CL_DEVICE_VERSION, "Device Version", str), NULL },
1657         { CLINFO_BOTH, DINFO(CL_DRIVER_VERSION, "Driver Version", str), NULL },
1658         { CLINFO_BOTH, DINFO(CL_DEVICE_OPENCL_C_VERSION, "Device OpenCL C Version", str), NULL },
1659         { CLINFO_BOTH, DINFO(CL_DEVICE_EXTENSIONS, "Device Extensions", str_get), NULL },
1660         { CLINFO_BOTH, DINFO(CL_DEVICE_TYPE, "Device Type", devtype), NULL },
1661
1662         { CLINFO_BOTH, DINFO(CL_DEVICE_AVAILABLE, "Device Available", bool), NULL },
1663
1664         { CLINFO_BOTH, DINFO(CL_DEVICE_PROFILE, "Device Profile", str), NULL },
1665         { CLINFO_BOTH, DINFO(CL_DEVICE_BOARD_NAME_AMD, "Device Board Name (AMD)", str), dev_has_amd },
1666         { CLINFO_BOTH, DINFO(CL_DEVICE_TOPOLOGY_AMD, "Device Topology (AMD)", devtopo_amd), dev_has_amd },
1667
1668         /* Device Topology (NV) is multipart, so different for HUMAN and RAW */
1669         { CLINFO_HUMAN, DINFO(CL_DEVICE_PCI_BUS_ID_NV, "Device Topology (NV)", devtopo_nv), dev_has_nv },
1670         { CLINFO_RAW, DINFO(CL_DEVICE_PCI_BUS_ID_NV, "Device PCI bus (NV)", int), dev_has_nv },
1671         { CLINFO_RAW, DINFO(CL_DEVICE_PCI_SLOT_ID_NV, "Device PCI slot (NV)", int), dev_has_nv },
1672
1673         { CLINFO_BOTH, DINFO(CL_DEVICE_MAX_COMPUTE_UNITS, "Max compute units", int), NULL },
1674         { CLINFO_BOTH, DINFO(CL_DEVICE_SIMD_PER_COMPUTE_UNIT_AMD, "SIMD per compute unit (AMD)", int), dev_is_gpu_amd },
1675         { CLINFO_BOTH, DINFO(CL_DEVICE_SIMD_WIDTH_AMD, "SIMD width (AMD)", int), dev_is_gpu_amd },
1676         { CLINFO_BOTH, DINFO(CL_DEVICE_SIMD_INSTRUCTION_WIDTH_AMD, "SIMD instruction width (AMD)", int), dev_is_gpu_amd },
1677         { CLINFO_BOTH, DINFO_SFX(CL_DEVICE_MAX_CLOCK_FREQUENCY, "Max clock frequency", "MHz", int), NULL },
1678
1679         /* Device Compute Capability (NV) is multipart, so different for HUMAN and RAW */
1680         { CLINFO_HUMAN, DINFO(CL_DEVICE_COMPUTE_CAPABILITY_MAJOR_NV, "Compute Capability (NV)", cc_nv), dev_has_nv },
1681         { CLINFO_RAW, DINFO(CL_DEVICE_COMPUTE_CAPABILITY_MAJOR_NV, INDENT "Compute Capability Major (NV)", int), dev_has_nv },
1682         { CLINFO_RAW, DINFO(CL_DEVICE_COMPUTE_CAPABILITY_MINOR_NV, INDENT "Compute Capability Minor (NV)", int), dev_has_nv },
1683
1684         /* GFXIP (AMD) is multipart, so different for HUMAN and RAW */
1685         /* TODO: find a better human-friendly name than GFXIP; v3 of the cl_amd_device_attribute_query
1686          * extension specification calls it “core engine GFXIP”, which honestly is not better than
1687          * our name choice. */
1688         { CLINFO_HUMAN, DINFO(CL_DEVICE_GFXIP_MAJOR_AMD, "Graphics IP (AMD)", gfxip_amd), dev_is_gpu_amd },
1689         { CLINFO_RAW, DINFO(CL_DEVICE_GFXIP_MAJOR_AMD, INDENT "Graphics IP MAJOR (AMD)", int), dev_is_gpu_amd },
1690         { CLINFO_RAW, DINFO(CL_DEVICE_GFXIP_MINOR_AMD, INDENT "Graphics IP MINOR (AMD)", int), dev_is_gpu_amd },
1691
1692         { CLINFO_BOTH, DINFO_SFX(CL_DEVICE_CORE_TEMPERATURE_ALTERA, "Core Temperature (Altera)", " C", int), dev_has_altera_dev_temp },
1693
1694         /* Device partition support: summary is only presented in HUMAN case */
1695         { CLINFO_HUMAN, DINFO(CL_DEVICE_PARTITION_MAX_SUB_DEVICES, "Device Partition", partition_header), dev_has_partition },
1696         { CLINFO_BOTH, DINFO(CL_DEVICE_PARTITION_MAX_SUB_DEVICES, INDENT "Max number of sub-devices", int), dev_is_12 },
1697         { CLINFO_BOTH, DINFO(CL_DEVICE_PARTITION_PROPERTIES, INDENT "Supported partition types", partition_types), dev_is_12 },
1698         { CLINFO_BOTH, DINFO(CL_DEVICE_PARTITION_AFFINITY_DOMAIN, INDENT "Supported affinity domains", partition_affinities), dev_is_12 },
1699         { CLINFO_BOTH, DINFO(CL_DEVICE_PARTITION_TYPES_EXT, INDENT "Supported partition types (ext)", partition_types_ext), dev_has_fission },
1700         { CLINFO_BOTH, DINFO(CL_DEVICE_AFFINITY_DOMAINS_EXT, INDENT "Supported affinity domains (ext)", partition_affinities_ext), dev_has_fission },
1701
1702         { CLINFO_BOTH, DINFO(CL_DEVICE_MAX_WORK_ITEM_DIMENSIONS, "Max work item dimensions", int), NULL },
1703         { CLINFO_BOTH, DINFO(CL_DEVICE_MAX_WORK_ITEM_SIZES, "Max work item sizes", szptr), NULL },
1704         { CLINFO_BOTH, DINFO(CL_DEVICE_MAX_WORK_GROUP_SIZE, "Max work group size", sz), NULL },
1705
1706         { CLINFO_BOTH, DINFO(CL_DEVICE_COMPILER_AVAILABLE, "Compiler Available", bool), NULL },
1707         { CLINFO_BOTH, DINFO(CL_DEVICE_LINKER_AVAILABLE, "Linker Available", bool), dev_is_12 },
1708         { CLINFO_BOTH, DINFO(CL_KERNEL_PREFERRED_WORK_GROUP_SIZE_MULTIPLE, "Preferred work group size multiple", wg), dev_has_compiler },
1709         { CLINFO_BOTH, DINFO(CL_DEVICE_WARP_SIZE_NV, "Warp size (NV)", int), dev_has_nv },
1710         { CLINFO_BOTH, DINFO(CL_DEVICE_WAVEFRONT_WIDTH_AMD, "Wavefront width (AMD)", int), dev_is_gpu_amd },
1711         { CLINFO_BOTH, DINFO(CL_DEVICE_MAX_NUM_SUB_GROUPS, "Max sub-groups per work group", int), dev_is_21 },
1712         { CLINFO_BOTH, DINFO(CL_DEVICE_SUB_GROUP_SIZES_INTEL, "Sub-group sizes (Intel)", szptr), dev_has_intel_required_subgroup_size },
1713
1714         /* Preferred/native vector widths: header is only presented in HUMAN case, that also pairs
1715          * PREFERRED and NATIVE in a single line */
1716 #define DINFO_VECWIDTH(Type, type) \
1717         { CLINFO_HUMAN, DINFO(CL_DEVICE_PREFERRED_VECTOR_WIDTH_##Type, INDENT #type, vecwidth), NULL }, \
1718         { CLINFO_RAW, DINFO(CL_DEVICE_PREFERRED_VECTOR_WIDTH_##Type, INDENT #type, int), NULL }, \
1719         { CLINFO_RAW, DINFO(CL_DEVICE_NATIVE_VECTOR_WIDTH_##Type, INDENT #type, int), NULL }
1720
1721         { CLINFO_HUMAN, DINFO(CL_FALSE, "Preferred / native vector sizes", str), NULL },
1722         DINFO_VECWIDTH(CHAR, char),
1723         DINFO_VECWIDTH(SHORT, short),
1724         DINFO_VECWIDTH(INT, int),
1725         DINFO_VECWIDTH(LONG, long),
1726         DINFO_VECWIDTH(HALF, half),
1727         DINFO_VECWIDTH(FLOAT, float),
1728         DINFO_VECWIDTH(DOUBLE, double),
1729
1730         /* Floating point configurations */
1731 #define DINFO_FPCONF(Type, type, cond) \
1732         { CLINFO_BOTH, DINFO(CL_DEVICE_##Type##_FP_CONFIG, #type "-precision Floating-point support", fpconf), NULL }
1733
1734         DINFO_FPCONF(HALF, Half, dev_has_half),
1735         DINFO_FPCONF(SINGLE, Single, NULL),
1736         DINFO_FPCONF(DOUBLE, Double, dev_has_double),
1737
1738         /* Address bits and endianness are written together for HUMAN, separate for RAW */
1739         { CLINFO_HUMAN, DINFO(CL_DEVICE_ADDRESS_BITS, "Address bits", arch), NULL },
1740         { CLINFO_RAW, DINFO(CL_DEVICE_ADDRESS_BITS, "Address bits", int), NULL },
1741         { CLINFO_RAW, DINFO(CL_DEVICE_ENDIAN_LITTLE, "Little Endian", bool), NULL },
1742
1743         /* Global memory */
1744         { CLINFO_BOTH, DINFO(CL_DEVICE_GLOBAL_MEM_SIZE, "Global memory size", mem), NULL },
1745         { CLINFO_BOTH, DINFO(CL_DEVICE_GLOBAL_FREE_MEMORY_AMD, "Global free memory (AMD)", free_mem_amd), dev_is_gpu_amd },
1746         { CLINFO_BOTH, DINFO(CL_DEVICE_GLOBAL_MEM_CHANNELS_AMD, "Global memory channels (AMD)", int), dev_is_gpu_amd },
1747         { CLINFO_BOTH, DINFO(CL_DEVICE_GLOBAL_MEM_CHANNEL_BANKS_AMD, "Global memory banks per channel (AMD)", int), dev_is_gpu_amd },
1748         { CLINFO_BOTH, DINFO_SFX(CL_DEVICE_GLOBAL_MEM_CHANNEL_BANK_WIDTH_AMD, "Global memory bank width (AMD)", bytes_str, int), dev_is_gpu_amd },
1749         { CLINFO_BOTH, DINFO(CL_DEVICE_ERROR_CORRECTION_SUPPORT, "Error Correction support", bool), NULL },
1750         { CLINFO_BOTH, DINFO(CL_DEVICE_MAX_MEM_ALLOC_SIZE, "Max memory allocation", mem), NULL },
1751         { CLINFO_BOTH, DINFO(CL_DEVICE_HOST_UNIFIED_MEMORY, "Unified memory for Host and Device", bool), NULL },
1752         { CLINFO_BOTH, DINFO(CL_DEVICE_INTEGRATED_MEMORY_NV, "Integrated memory (NV)", bool), dev_has_nv },
1753
1754         { CLINFO_BOTH, DINFO(CL_DEVICE_SVM_CAPABILITIES, "Shared Virtual Memory (SVM) capabilities", svm_cap), dev_has_svm },
1755         { CLINFO_BOTH, DINFO(CL_DEVICE_SVM_CAPABILITIES_ARM, "Shared Virtual Memory (SVM) capabilities (ARM)", svm_cap), dev_has_arm_svm },
1756
1757         /* Alignment */
1758         { CLINFO_BOTH, DINFO_SFX(CL_DEVICE_MIN_DATA_TYPE_ALIGN_SIZE, "Minimum alignment for any data type", bytes_str, int), NULL },
1759         { CLINFO_HUMAN, DINFO(CL_DEVICE_MEM_BASE_ADDR_ALIGN, "Alignment of base address", bits), NULL },
1760         { CLINFO_RAW, DINFO(CL_DEVICE_MEM_BASE_ADDR_ALIGN, "Alignment of base address", int), NULL },
1761
1762         { CLINFO_BOTH, DINFO_SFX(CL_DEVICE_PAGE_SIZE_QCOM, "Page size (QCOM)", bytes_str, sz), dev_has_qcom_ext_host_ptr },
1763         { CLINFO_BOTH, DINFO_SFX(CL_DEVICE_EXT_MEM_PADDING_IN_BYTES_QCOM, "Externa memory padding (QCOM)", bytes_str, sz), dev_has_qcom_ext_host_ptr },
1764
1765         /* Atomics alignment, with HUMAN-only header */
1766         { CLINFO_HUMAN, DINFO(CL_FALSE, "Preferred alignment for atomics", str), dev_is_20 },
1767         { CLINFO_BOTH, DINFO_SFX(CL_DEVICE_PREFERRED_PLATFORM_ATOMIC_ALIGNMENT, INDENT "SVM", bytes_str, int), dev_is_20 },
1768         { CLINFO_BOTH, DINFO_SFX(CL_DEVICE_PREFERRED_GLOBAL_ATOMIC_ALIGNMENT, INDENT "Global", bytes_str, int), dev_is_20 },
1769         { CLINFO_BOTH, DINFO_SFX(CL_DEVICE_PREFERRED_LOCAL_ATOMIC_ALIGNMENT, INDENT "Local", bytes_str, int), dev_is_20 },
1770
1771         /* Global variables. TODO some 1.2 devices respond to this too */
1772         { CLINFO_BOTH, DINFO(CL_DEVICE_MAX_GLOBAL_VARIABLE_SIZE, "Max size for global variable", mem), dev_is_20 },
1773         { CLINFO_BOTH, DINFO(CL_DEVICE_GLOBAL_VARIABLE_PREFERRED_TOTAL_SIZE, "Preferred total size of global vars", mem), dev_is_20 },
1774
1775         /* Global memory cache */
1776         { CLINFO_BOTH, DINFO(CL_DEVICE_GLOBAL_MEM_CACHE_TYPE, "Global Memory cache type", cachetype), NULL },
1777         { CLINFO_BOTH, DINFO(CL_DEVICE_GLOBAL_MEM_CACHE_SIZE, "Global Memory cache size", sz), dev_has_cache },
1778         { CLINFO_BOTH, DINFO_SFX(CL_DEVICE_GLOBAL_MEM_CACHELINE_SIZE, "Global Memory cache line", " bytes", int), dev_has_cache },
1779
1780         /* Image support */
1781         { CLINFO_BOTH, DINFO(CL_DEVICE_IMAGE_SUPPORT, "Image support", bool), NULL },
1782         { CLINFO_BOTH, DINFO(CL_DEVICE_MAX_SAMPLERS, INDENT "Max number of samplers per kernel", int), dev_has_images },
1783         { CLINFO_BOTH, DINFO_SFX(CL_DEVICE_IMAGE_MAX_BUFFER_SIZE, INDENT "Max size for 1D images from buffer", pixels_str, sz), dev_has_images_12 },
1784         { CLINFO_BOTH, DINFO_SFX(CL_DEVICE_IMAGE_MAX_ARRAY_SIZE, INDENT "Max 1D or 2D image array size", images_str, sz), dev_has_images_12 },
1785         { CLINFO_BOTH, DINFO_SFX(CL_DEVICE_IMAGE_BASE_ADDRESS_ALIGNMENT, INDENT "Base address alignment for 2D image buffers", bytes_str, sz), dev_has_image2d_buffer },
1786         { CLINFO_BOTH, DINFO_SFX(CL_DEVICE_IMAGE_PITCH_ALIGNMENT, INDENT "Pitch alignment for 2D image buffers", bytes_str, sz), dev_has_image2d_buffer },
1787
1788         /* Image dimensions are split for RAW, combined for HUMAN */
1789         { CLINFO_HUMAN, DINFO_SFX(CL_DEVICE_IMAGE2D_MAX_HEIGHT, INDENT "Max 2D image size",  pixels_str, img_sz_2d), dev_has_images },
1790         { CLINFO_RAW, DINFO(CL_DEVICE_IMAGE2D_MAX_HEIGHT, INDENT "Max 2D image height",  sz), dev_has_images },
1791         { CLINFO_RAW, DINFO(CL_DEVICE_IMAGE2D_MAX_WIDTH, INDENT "Max 2D image width",  sz), dev_has_images },
1792         { CLINFO_HUMAN, DINFO_SFX(CL_DEVICE_PLANAR_YUV_MAX_HEIGHT_INTEL, INDENT "Max planar YUV image size",  pixels_str, img_sz_2d), dev_has_intel_planar_yuv },
1793         { CLINFO_RAW, DINFO(CL_DEVICE_PLANAR_YUV_MAX_HEIGHT_INTEL, INDENT "Max planar YUV image height",  sz), dev_has_intel_planar_yuv },
1794         { CLINFO_RAW, DINFO(CL_DEVICE_PLANAR_YUV_MAX_WIDTH_INTEL, INDENT "Max planar YUV image width",  sz), dev_has_intel_planar_yuv },
1795         { CLINFO_HUMAN, DINFO_SFX(CL_DEVICE_IMAGE3D_MAX_HEIGHT, INDENT "Max 3D image size",  pixels_str, img_sz_3d), dev_has_images },
1796         { CLINFO_RAW, DINFO(CL_DEVICE_IMAGE3D_MAX_HEIGHT, INDENT "Max 3D image height",  sz), dev_has_images },
1797         { CLINFO_RAW, DINFO(CL_DEVICE_IMAGE3D_MAX_WIDTH, INDENT "Max 3D image width",  sz), dev_has_images },
1798         { CLINFO_RAW, DINFO(CL_DEVICE_IMAGE3D_MAX_DEPTH, INDENT "Max 3D image depth",  sz), dev_has_images },
1799
1800         { CLINFO_BOTH, DINFO(CL_DEVICE_MAX_READ_IMAGE_ARGS, INDENT "Max number of read image args", int), dev_has_images },
1801         { CLINFO_BOTH, DINFO(CL_DEVICE_MAX_WRITE_IMAGE_ARGS, INDENT "Max number of write image args", int), dev_has_images },
1802         { CLINFO_BOTH, DINFO(CL_DEVICE_MAX_READ_WRITE_IMAGE_ARGS, INDENT "Max number of read/write image args", int), dev_has_images_20 },
1803
1804         /* Pipes */
1805         { CLINFO_BOTH, DINFO(CL_DEVICE_MAX_PIPE_ARGS, "Max number of pipe args", int), dev_is_20 },
1806         { CLINFO_BOTH, DINFO(CL_DEVICE_PIPE_MAX_ACTIVE_RESERVATIONS, "Max active pipe reservations", int), dev_is_20 },
1807         { CLINFO_BOTH, DINFO(CL_DEVICE_PIPE_MAX_PACKET_SIZE, "Max pipe packet size", mem_int), dev_is_20 },
1808
1809         /* Local memory */
1810         { CLINFO_BOTH, DINFO(CL_DEVICE_LOCAL_MEM_TYPE, "Local memory type", lmemtype), NULL },
1811         { CLINFO_BOTH, DINFO(CL_DEVICE_LOCAL_MEM_SIZE, "Local memory size", mem), dev_has_lmem },
1812         { CLINFO_BOTH, DINFO(CL_DEVICE_LOCAL_MEM_SIZE_PER_COMPUTE_UNIT_AMD, "Local memory syze per CU (AMD)", mem), dev_is_gpu_amd },
1813         { CLINFO_BOTH, DINFO(CL_DEVICE_LOCAL_MEM_BANKS_AMD, "Local memory banks (AMD)", int), dev_is_gpu_amd },
1814         { CLINFO_BOTH, DINFO(CL_DEVICE_REGISTERS_PER_BLOCK_NV, "Registers per block (NV)", int), dev_has_nv },
1815
1816         /* Constant memory */
1817         { CLINFO_BOTH, DINFO(CL_DEVICE_MAX_CONSTANT_BUFFER_SIZE, "Max constant buffer size", mem), NULL },
1818         { CLINFO_BOTH, DINFO(CL_DEVICE_MAX_CONSTANT_ARGS, "Max number of constant args", int), NULL },
1819
1820         { CLINFO_BOTH, DINFO(CL_DEVICE_MAX_PARAMETER_SIZE, "Max size of kernel argument", mem), NULL },
1821         { CLINFO_BOTH, DINFO(CL_DEVICE_MAX_ATOMIC_COUNTERS_EXT, "Max number of atomic counters", sz), dev_has_atomic_counters },
1822
1823         /* Queue properties */
1824         { CLINFO_BOTH, DINFO(CL_DEVICE_QUEUE_PROPERTIES, "Queue properties", qprop), dev_not_20 },
1825         { CLINFO_BOTH, DINFO(CL_DEVICE_QUEUE_PROPERTIES, "Queue properties (on host)", qprop), dev_is_20 },
1826         { CLINFO_BOTH, DINFO(CL_DEVICE_QUEUE_ON_DEVICE_PROPERTIES, "Queue properties (on device)", qprop), dev_is_20 },
1827         { CLINFO_BOTH, DINFO(CL_DEVICE_QUEUE_ON_DEVICE_PREFERRED_SIZE, INDENT "Preferred size", mem), dev_is_20 },
1828         { CLINFO_BOTH, DINFO(CL_DEVICE_QUEUE_ON_DEVICE_MAX_SIZE, INDENT "Max size", mem), dev_is_20 },
1829         { CLINFO_BOTH, DINFO(CL_DEVICE_MAX_ON_DEVICE_QUEUES, "Max queues on device", int), dev_is_20 },
1830         { CLINFO_BOTH, DINFO(CL_DEVICE_MAX_ON_DEVICE_EVENTS, "Max events on device", int), dev_is_20 },
1831
1832         /* Interop */
1833         { CLINFO_BOTH, DINFO(CL_DEVICE_PREFERRED_INTEROP_USER_SYNC, "Prefer user sync for interop", bool), dev_is_12 },
1834         { CLINFO_BOTH, DINFO(CL_DEVICE_NUM_SIMULTANEOUS_INTEROPS_INTEL, "Number of simulataneous interops (Intel)", int), dev_has_simultaneous_sharing },
1835         /* TODO: this needs defines for the possible values of the context interops,
1836         { CLINFO_BOTH, DINFO(CL_DEVICE_SIMULTANEOUS_INTEROPS_INTEL, "Simulataneous interops", interop_list), dev_has_simultaneous_sharing },
1837          */
1838
1839         /* Profiling resolution */
1840         { CLINFO_BOTH, DINFO_SFX(CL_DEVICE_PROFILING_TIMER_RESOLUTION, "Profiling timer resolution", "ns", sz), NULL },
1841         { CLINFO_HUMAN, DINFO(CL_DEVICE_PROFILING_TIMER_OFFSET_AMD, "Profiling timer offset since Epoch (AMD)", time_offset), dev_has_amd },
1842         { CLINFO_RAW, DINFO(CL_DEVICE_PROFILING_TIMER_OFFSET_AMD, "Profiling timer offset since Epoch (AMD)", long), dev_has_amd },
1843
1844         /* Kernel execution capabilities */
1845         { CLINFO_BOTH, DINFO(CL_DEVICE_EXECUTION_CAPABILITIES, "Execution capabilities", execap), NULL },
1846         { CLINFO_BOTH, DINFO(CL_DEVICE_SUB_GROUP_INDEPENDENT_FORWARD_PROGRESS, INDENT "Sub-group independent forward progress", bool), dev_is_21 },
1847         { CLINFO_BOTH, DINFO(CL_DEVICE_THREAD_TRACE_SUPPORTED_AMD, INDENT "Thread trace supported (AMD)", bool), dev_is_gpu_amd },
1848         { CLINFO_BOTH, DINFO(CL_DEVICE_KERNEL_EXEC_TIMEOUT_NV, INDENT "Kernel execution timeout (NV)", bool), dev_has_nv },
1849         { CLINFO_BOTH, DINFO(CL_DEVICE_GPU_OVERLAP_NV, "Concurrent copy and kernel execution (NV)", bool), dev_has_nv },
1850         { CLINFO_BOTH, DINFO(CL_DEVICE_ATTRIBUTE_ASYNC_ENGINE_COUNT_NV, INDENT "Number of async copy engines", int), dev_has_nv },
1851         /* TODO FIXME Current drivers don't seem to respond to this, should probably be queried based on driver version,
1852          * or maybe it depends on some other device property?
1853         { CLINFO_BOTH, DINFO(CL_DEVICE_AVAILABLE_ASYNC_QUEUES_AMD, INDENT "Number of async queues (AMD)", int), dev_is_gpu_amd },
1854          */
1855         { CLINFO_BOTH, DINFO(CL_DEVICE_IL_VERSION, INDENT "IL version", str), dev_is_21, },
1856         { CLINFO_BOTH, DINFO(CL_DEVICE_SPIR_VERSIONS, INDENT "SPIR versions", str), dev_has_spir },
1857         { CLINFO_BOTH, DINFO(CL_DEVICE_PRINTF_BUFFER_SIZE, "printf() buffer size", mem), dev_is_12 },
1858         { CLINFO_BOTH, DINFO(CL_DEVICE_BUILT_IN_KERNELS, "Built-in kernels", str), dev_is_12 },
1859         { CLINFO_BOTH, DINFO(CL_DEVICE_ME_VERSION_INTEL, "Motion Estimation accelerator version (Intel)", int), dev_has_intel_AME },
1860         { CLINFO_BOTH, DINFO(CL_DEVICE_AVC_ME_VERSION_INTEL, INDENT "Device-side AVC Motion Estimation version", int), dev_has_intel_AVC_ME },
1861         { CLINFO_BOTH, DINFO(CL_DEVICE_AVC_ME_SUPPORTS_TEXTURE_SAMPLER_USE_INTEL, INDENT INDENT "Supports texture sampler use", bool), dev_has_intel_AVC_ME },
1862         { CLINFO_BOTH, DINFO(CL_DEVICE_AVC_ME_SUPPORTS_PREEMPTION_INTEL, INDENT INDENT "Supports preemption", bool), dev_has_intel_AVC_ME },
1863 };
1864
1865 /* Process all the device info in the traits, except if param_whitelist is not NULL,
1866  * in which case only those in the whitelist will be processed.
1867  * If present, the whitelist should be sorted in the order of appearance of the parameters
1868  * in the traits table, and terminated by the value CL_FALSE
1869  */
1870
1871 void
1872 printDeviceInfo(const cl_device_id *device, cl_uint d,
1873         const cl_device_info *param_whitelist) /* list of device info to process, or NULL */
1874 {
1875         cl_device_id dev = device[d];
1876
1877         char *extensions = NULL;
1878
1879         /* pointer to the traits for CL_DEVICE_EXTENSIONS */
1880         const struct device_info_traits *extensions_traits = NULL;
1881
1882         struct device_info_checks chk;
1883         memset(&chk, 0, sizeof(chk));
1884         chk.dev_version = 10;
1885
1886         current_function = __func__;
1887
1888         for (current_line = 0; current_line < ARRAY_SIZE(dinfo_traits); ++current_line) {
1889
1890                 const struct device_info_traits *traits = dinfo_traits + current_line;
1891                 const char *pname = (output_mode == CLINFO_HUMAN ?
1892                         traits->pname : traits->sname);
1893
1894                 current_param = traits->sname;
1895
1896                 /* Whitelist check: finish if done traversing the list,
1897                  * skip current param if it's not the right one
1898                  */
1899                 if (param_whitelist) {
1900                         if (*param_whitelist == CL_FALSE)
1901                                 break;
1902                         if (traits->param != *param_whitelist)
1903                                 continue;
1904                         ++param_whitelist;
1905                 }
1906
1907                 /* skip if it's not for this output mode */
1908                 if (!(output_mode & traits->output_mode))
1909                         continue;
1910
1911                 if (traits->check_func && !traits->check_func(&chk))
1912                         continue;
1913
1914                 cur_sfx = (output_mode == CLINFO_HUMAN && traits->sfx) ? traits->sfx : empty_str;
1915
1916                 /* Handle headers */
1917                 if (traits->param == CL_FALSE) {
1918                         strbuf[0] = '\0';
1919                         show_strbuf(pname, 0);
1920                         had_error = CL_FALSE;
1921                         continue;
1922                 }
1923
1924                 had_error = traits->show_func(dev, traits->param,
1925                         pname, &chk);
1926
1927                 if (traits->param == CL_DEVICE_EXTENSIONS) {
1928                         /* make a backup of the extensions string, regardless of
1929                          * errors */
1930                         size_t len = strlen(strbuf);
1931                         extensions_traits = traits;
1932                         ALLOC(extensions, len+1, "extensions");
1933                         memcpy(extensions, strbuf, len);
1934                         extensions[len] = '\0';
1935                 }
1936
1937                 if (had_error)
1938                         continue;
1939
1940                 switch (traits->param) {
1941                 case CL_DEVICE_VERSION:
1942                         /* compute numeric value for OpenCL version */
1943                         chk.dev_version = getOpenCLVersion(strbuf + 7);
1944                         break;
1945                 case CL_DEVICE_EXTENSIONS:
1946                         identify_device_extensions(extensions, &chk);
1947                         break;
1948                 case CL_DEVICE_TYPE:
1949                         /* strbuf was abused to give us the dev type */
1950                         memcpy(&(chk.devtype), strbuf, sizeof(chk.devtype));
1951                         break;
1952                 case CL_DEVICE_GLOBAL_MEM_CACHE_TYPE:
1953                         /* strbuf was abused to give us the cache type */
1954                         memcpy(&(chk.cachetype), strbuf, sizeof(chk.cachetype));
1955                         break;
1956                 case CL_DEVICE_LOCAL_MEM_TYPE:
1957                         /* strbuf was abused to give us the lmem type */
1958                         memcpy(&(chk.lmemtype), strbuf, sizeof(chk.lmemtype));
1959                         break;
1960                 case CL_DEVICE_IMAGE_SUPPORT:
1961                         /* strbuf was abused to give us boolean value */
1962                         memcpy(&(chk.image_support), strbuf, sizeof(chk.image_support));
1963                         break;
1964                 case CL_DEVICE_COMPILER_AVAILABLE:
1965                         /* strbuf was abused to give us boolean value */
1966                         memcpy(&(chk.compiler_available), strbuf, sizeof(chk.compiler_available));
1967                         break;
1968                 default:
1969                         /* do nothing */
1970                         break;
1971                 }
1972         }
1973
1974         // and finally the extensions, if we retrieved them
1975         if (extensions)
1976                 printf("%s" I1_STR "%s\n", line_pfx, (output_mode == CLINFO_HUMAN ?
1977                                 extensions_traits->pname :
1978                                 extensions_traits->sname), extensions);
1979         free(extensions);
1980         extensions = NULL;
1981 }
1982
1983 /* list of allowed properties for AMD offline devices */
1984 /* everything else seems to be set to 0, and all the other string properties
1985  * actually segfault the driver */
1986
1987 static const cl_device_info amd_offline_info_whitelist[] = {
1988         CL_DEVICE_NAME,
1989         /* These are present, but all the same, so just skip them:
1990         CL_DEVICE_VENDOR,
1991         CL_DEVICE_VENDOR_ID,
1992         CL_DEVICE_VERSION,
1993         CL_DRIVER_VERSION,
1994         CL_DEVICE_OPENCL_C_VERSION,
1995         */
1996         CL_DEVICE_EXTENSIONS,
1997         CL_DEVICE_TYPE,
1998         CL_DEVICE_MAX_WORK_GROUP_SIZE,
1999         CL_DEVICE_AVAILABLE
2000 };
2001
2002 /* process offline devices from the cl_amd_offline_devices extension */
2003 int processOfflineDevicesAMD(cl_uint p)
2004 {
2005         int ret = 0;
2006
2007         cl_platform_id pid = platform[p];
2008         cl_device_id *device = NULL;
2009         cl_int num_devs, d;
2010
2011         cl_context_properties ctxpft[] = {
2012                 CL_CONTEXT_PLATFORM, (cl_context_properties)pid,
2013                 CL_CONTEXT_OFFLINE_DEVICES_AMD, (cl_context_properties)CL_TRUE,
2014                 0
2015         };
2016
2017         cl_context ctx = NULL;
2018
2019         if (!list_only)
2020                 printf("%s" I0_STR, line_pfx,
2021                         (output_mode == CLINFO_HUMAN ?
2022                          "Number of offline devices (AMD)" : "#OFFDEVICES"));
2023
2024         ctx = clCreateContextFromType(ctxpft, CL_DEVICE_TYPE_ALL, NULL, NULL, &error);
2025         RR_ERROR("create context");
2026
2027         error = clGetContextInfo(ctx, CL_CONTEXT_NUM_DEVICES, sizeof(num_devs), &num_devs, NULL);
2028         RR_ERROR("get num devs");
2029
2030         ALLOC(device, num_devs, "offline devices");
2031
2032         error = clGetContextInfo(ctx, CL_CONTEXT_DEVICES, num_devs*sizeof(*device), device, NULL);
2033         RR_ERROR("get devs");
2034
2035         if (!list_only)
2036                 printf("%d\n", num_devs);
2037
2038         for (d = 0; d < num_devs; ++d) {
2039                 if (list_only) {
2040                         /*
2041                         if (output_mode == CLINFO_HUMAN)
2042                                 puts(" |");
2043                         */
2044                         if (d == num_devs - 1 && output_mode != CLINFO_RAW)
2045                                 line_pfx[1] = '`';
2046                         had_error = device_info_str_get(device[d], CL_DEVICE_NAME, "CL_DEVICE_NAME", NULL);
2047                         printf("%s%u: %s\n", line_pfx, d, strbuf);
2048                 } else {
2049                         if (line_pfx_len > 0) {
2050                                 sprintf(strbuf, "[%s/%u]", pdata[p].sname, -d);
2051                                 sprintf(line_pfx, "%*s", -line_pfx_len, strbuf);
2052                         }
2053                         printDeviceInfo(device, d, amd_offline_info_whitelist);
2054                         if (d < num_devs - 1)
2055                                 puts("");
2056                 }
2057                 fflush(stdout);
2058                 fflush(stderr);
2059         }
2060
2061         had_error = CL_FALSE;
2062 out:
2063         free(device);
2064         if (ctx)
2065                 clReleaseContext(ctx);
2066         return ret;
2067
2068 }
2069
2070 void listPlatformsAndDevices(cl_bool show_offline)
2071 {
2072         cl_uint p, d;
2073         cl_device_id *device;
2074
2075         if (output_mode == CLINFO_RAW)
2076                 sprintf(strbuf, "%u", num_platforms);
2077         else
2078                 sprintf(strbuf, " +-- %sDevice #", (show_offline ? "Offline" : ""));
2079
2080         line_pfx_len = strlen(strbuf) + 1;
2081         REALLOC(line_pfx, line_pfx_len, "line prefix");
2082
2083         for (p = 0, device = all_devices; p < num_platforms; device += pdata[p++].ndevs) {
2084                 printf("%s%u: %s\n",
2085                         (output_mode == CLINFO_HUMAN ? "Platform #" : ""),
2086                         p, pdata[p].pname);
2087                 if (output_mode == CLINFO_RAW)
2088                         sprintf(line_pfx, "%u:", p);
2089                 else
2090                         sprintf(line_pfx, " +-- Device #");
2091
2092                 if (pdata[p].ndevs > 0) {
2093                         error = clGetDeviceIDs(platform[p], CL_DEVICE_TYPE_ALL, pdata[p].ndevs, device, NULL);
2094                         CHECK_ERROR("device IDs");
2095                         for (d = 0; d < pdata[p].ndevs; ++d) {
2096                                 /*
2097                                 if (output_mode == CLINFO_HUMAN)
2098                                         puts(" |");
2099                                 */
2100                                 cl_bool last_device = (d == pdata[p].ndevs - 1 && output_mode != CLINFO_RAW &&
2101                                         (!show_offline || !pdata[p].has_amd_offline));
2102                                 if (last_device)
2103                                         line_pfx[1] = '`';
2104                                 had_error = device_info_str_get(device[d], CL_DEVICE_NAME, "CL_DEVICE_NAME", NULL);
2105                                 printf("%s%u: %s\n", line_pfx, d, strbuf);
2106                                 fflush(stdout);
2107                                 fflush(stderr);
2108                         }
2109                 }
2110
2111                 if (show_offline && pdata[p].has_amd_offline) {
2112                         if (output_mode == CLINFO_RAW)
2113                                 sprintf(line_pfx, "%u*", p);
2114                         else
2115                                 sprintf(line_pfx, " +-- Offline Device #");
2116                         had_error = processOfflineDevicesAMD(p);
2117                         if (had_error)
2118                                 puts(strbuf);
2119                 }
2120         }
2121 }
2122
2123 void showDevices(cl_bool show_offline)
2124 {
2125         cl_uint p, d;
2126         cl_device_id *device;
2127
2128         /* TODO consider enabling this for both output modes */
2129         if (output_mode == CLINFO_RAW) {
2130                 sprintf(strbuf, "%u", maxdevs);
2131                 line_pfx_len = platform_sname_maxlen + strlen(strbuf) + 4;
2132                 REALLOC(line_pfx, line_pfx_len, "line prefix");
2133         }
2134
2135         for (p = 0, device = all_devices; p < num_platforms; device += pdata[p++].ndevs) {
2136                 if (line_pfx_len > 0) {
2137                         sprintf(strbuf, "[%s/*]", pdata[p].sname);
2138                         sprintf(line_pfx, "%*s", -line_pfx_len, strbuf);
2139                 }
2140                 printf("%s" I1_STR "%s\n",
2141                         line_pfx,
2142                         (output_mode == CLINFO_HUMAN ?
2143                          pinfo_traits[0].pname : pinfo_traits[0].sname),
2144                         pdata[p].pname);
2145                 printf("%s" I0_STR "%u\n",
2146                         line_pfx,
2147                         (output_mode == CLINFO_HUMAN ?
2148                          "Number of devices" : "#DEVICES"),
2149                         pdata[p].ndevs);
2150
2151                 if (pdata[p].ndevs > 0) {
2152                         error = clGetDeviceIDs(platform[p], CL_DEVICE_TYPE_ALL, pdata[p].ndevs, device, NULL);
2153                         CHECK_ERROR("device IDs");
2154                 }
2155                 for (d = 0; d < pdata[p].ndevs; ++d) {
2156                         if (line_pfx_len > 0) {
2157                                 sprintf(strbuf, "[%s/%u]", pdata[p].sname, d);
2158                                 sprintf(line_pfx, "%*s", -line_pfx_len, strbuf);
2159                         }
2160                         printDeviceInfo(device, d, NULL);
2161                         if (d < pdata[p].ndevs - 1)
2162                                 puts("");
2163                         fflush(stdout);
2164                         fflush(stderr);
2165                 }
2166                 if (show_offline && pdata[p].has_amd_offline) {
2167                         puts("");
2168                         had_error = processOfflineDevicesAMD(p);
2169                         if (had_error)
2170                                 puts(strbuf);
2171                 }
2172                 puts("");
2173         }
2174 }
2175
2176 /* check the behavior of clGetPlatformInfo() when given a NULL platform ID */
2177 void checkNullGetPlatformName(void)
2178 {
2179         current_param = "CL_PLATFORM_NAME";
2180
2181         error = clGetPlatformInfo(NULL, CL_PLATFORM_NAME, bufsz, strbuf, NULL);
2182         if (error == CL_INVALID_PLATFORM) {
2183                 bufcpy(0, no_plat());
2184         } else {
2185                 current_line = __LINE__+1;
2186                 had_error = REPORT_ERROR2("get %s");
2187         }
2188         printf(I1_STR "%s\n",
2189                 "clGetPlatformInfo(NULL, CL_PLATFORM_NAME, ...)", strbuf);
2190 }
2191
2192 /* check the behavior of clGetDeviceIDs() when given a NULL platform ID;
2193  * return the index of the default platform in our array of platform IDs,
2194  * or num_platforms (which is an invalid platform index) in case of errors
2195  * or no platform or device found.
2196  */
2197 cl_uint checkNullGetDevices(void)
2198 {
2199         cl_uint i = 0; /* generic iterator */
2200         cl_device_id dev = NULL; /* sample device */
2201         cl_platform_id plat = NULL; /* detected platform */
2202
2203         cl_uint found = 0; /* number of platforms found */
2204         cl_uint pidx = num_platforms; /* index of the platform found */
2205         cl_uint numdevs = 0;
2206
2207         current_function = __func__;
2208         current_param = "device IDs";
2209
2210         error = clGetDeviceIDs(NULL, CL_DEVICE_TYPE_ALL, 0, NULL, &numdevs);
2211         /* TODO we should check other CL_DEVICE_TYPE_* combinations, since a smart
2212          * implementation might give you a different default platform for GPUs
2213          * and for CPUs.
2214          * Of course the “no devices” case would then need to be handled differently.
2215          * The logic might be maintained similarly, provided we also gather
2216          * the number of devices of each type for each platform, although it's
2217          * obviously more likely to have multiple platforms with no devices
2218          * of a given type.
2219          */
2220
2221         switch (error) {
2222         case CL_INVALID_PLATFORM:
2223                 bufcpy(0, no_plat());
2224                 break;
2225         case CL_DEVICE_NOT_FOUND:
2226                  /* No devices were found, see if there are platforms with
2227                   * no devices, and if there's only one, assume this is the
2228                   * one being used as default by the ICD loader */
2229                 for (i = 0; i < num_platforms; ++i) {
2230                         if (pdata[i].ndevs == 0) {
2231                                 ++found;
2232                                 if (found > 1)
2233                                         break;
2234                                 else {
2235                                         plat = platform[i];
2236                                         pidx = i;
2237                                 }
2238                         }
2239                 }
2240
2241                 switch (found) {
2242                 case 0:
2243                         bufcpy(0, (output_mode == CLINFO_HUMAN ?
2244                                 "<error: 0 devices, no matching platform!>" :
2245                                 "CL_DEVICE_NOT_FOUND | CL_INVALID_PLATFORM"));
2246                         break;
2247                 case 1:
2248                         bufcpy(0, (output_mode == CLINFO_HUMAN ?
2249                                 pdata[pidx].pname :
2250                                 pdata[pidx].sname));
2251                         break;
2252                 default: /* found > 1 */
2253                         bufcpy(0, (output_mode == CLINFO_HUMAN ?
2254                                 "<error: 0 devices, multiple matching platforms!>" :
2255                                 "CL_DEVICE_NOT_FOUND | ????"));
2256                         break;
2257                 }
2258                 break;
2259         default:
2260                 current_line = __LINE__+1;
2261                 had_error = REPORT_ERROR2("get number of %s");
2262                 if (had_error)
2263                         break;
2264
2265                 /* Determine platform by looking at the CL_DEVICE_PLATFORM of
2266                  * one of the devices */
2267                 error = clGetDeviceIDs(NULL, CL_DEVICE_TYPE_ALL, 1, &dev, NULL);
2268                 current_line = __LINE__+1;
2269                 had_error = REPORT_ERROR2("get %s");
2270                 if (had_error)
2271                         break;
2272
2273                 current_param = "CL_DEVICE_PLATFORM";
2274                 error = clGetDeviceInfo(dev, CL_DEVICE_PLATFORM,
2275                         sizeof(plat), &plat, NULL);
2276                 current_line = __LINE__+1;
2277                 had_error = REPORT_ERROR2("get %s");
2278                 if (had_error)
2279                         break;
2280
2281                 for (i = 0; i < num_platforms; ++i) {
2282                         if (platform[i] == plat) {
2283                                 pidx = i;
2284                                 sprintf(strbuf, "%s [%s]",
2285                                         (output_mode == CLINFO_HUMAN ? "Success" : "CL_SUCCESS"),
2286                                         pdata[i].sname);
2287                                 break;
2288                         }
2289                 }
2290                 if (i == num_platforms) {
2291                         sprintf(strbuf, "<error: platform 0x%p not found>", (void*)plat);
2292                 }
2293         }
2294         printf(I1_STR "%s\n",
2295                 "clGetDeviceIDs(NULL, CL_DEVICE_TYPE_ALL, ...)", strbuf);
2296         return pidx;
2297 }
2298
2299 void checkNullCtx(cl_uint pidx, const cl_device_id *dev, const char *which)
2300 {
2301         cl_context ctx = clCreateContext(NULL, 1, dev, NULL, NULL, &error);
2302
2303         current_function = __func__;
2304         current_param = which;
2305         current_line = __LINE__+2;
2306
2307         had_error = REPORT_ERROR2("create context with device from %s platform");
2308         if (!had_error)
2309                 sprintf(strbuf, "%s [%s]",
2310                         (output_mode == CLINFO_HUMAN ? "Success" : "CL_SUCCESS"),
2311                         pdata[pidx].sname);
2312         if (ctx) {
2313                 clReleaseContext(ctx);
2314                 ctx = NULL;
2315         }
2316 }
2317
2318 /* check behavior of clCreateContextFromType() with NULL cl_context_properties */
2319 void checkNullCtxFromType(void)
2320 {
2321         size_t t; /* type iterator */
2322         size_t i; /* generic iterator */
2323         char def[1024];
2324         cl_context ctx = NULL;
2325
2326         size_t ndevs = 8;
2327         size_t szval = 0;
2328         size_t cursz = ndevs*sizeof(cl_device_id);
2329         cl_platform_id plat = NULL;
2330         cl_device_id *devs = NULL;
2331
2332         const char *platname_prop = (output_mode == CLINFO_HUMAN ?
2333                 pinfo_traits[0].pname :
2334                 pinfo_traits[0].sname);
2335
2336         const char *devname_prop = (output_mode == CLINFO_HUMAN ?
2337                 dinfo_traits[0].pname :
2338                 dinfo_traits[0].sname);
2339
2340         ALLOC(devs, ndevs, "context devices");
2341
2342         current_function = __func__;
2343         for (t = 2; t < devtype_count; ++t) { /* we skip 0 and _DEFAULT */
2344                 current_param = device_type_raw_str[t];
2345
2346                 sprintf(strbuf, "clCreateContextFromType(NULL, %s)", current_param);
2347                 sprintf(def, I1_STR, strbuf);
2348
2349                 current_line = __LINE__+1;
2350                 ctx = clCreateContextFromType(NULL, devtype[t], NULL, NULL, &error);
2351
2352                 switch (error) {
2353                 case CL_INVALID_PLATFORM:
2354                         bufcpy(0, no_plat()); break;
2355                 case CL_DEVICE_NOT_FOUND:
2356                 case CL_INVALID_DEVICE_TYPE: /* e.g. _CUSTOM device on 1.1 platform */
2357                         bufcpy(0, no_dev()); break;
2358                 case CL_DEVICE_NOT_AVAILABLE:
2359                         bufcpy(0, no_dev_avail()); break;
2360                 default:
2361                         had_error = REPORT_ERROR2("create context from type %s");
2362                         if (had_error)
2363                                 break;
2364
2365                         /* get the devices */
2366                         current_param = "CL_CONTEXT_DEVICES";
2367                         current_line = __LINE__+2;
2368
2369                         error = clGetContextInfo(ctx, CL_CONTEXT_DEVICES, 0, NULL, &szval);
2370                         had_error = REPORT_ERROR2("get %s size");
2371                         if (had_error)
2372                                 break;
2373                         if (szval > cursz) {
2374                                 REALLOC(devs, szval, "context devices");
2375                                 cursz = szval;
2376                         }
2377
2378                         current_line = __LINE__+1;
2379                         error = clGetContextInfo(ctx, CL_CONTEXT_DEVICES, cursz, devs, NULL);
2380                         had_error = REPORT_ERROR2("get %s");
2381                         if (had_error)
2382                                 break;
2383                         ndevs = szval/sizeof(cl_device_id);
2384                         if (ndevs < 1) {
2385                                 bufcpy(0, "<error: context created with no devices>");
2386                         }
2387
2388                         /* get the platform from the first device */
2389                         current_param = "CL_DEVICE_PLATFORM";
2390                         current_line = __LINE__+1;
2391                         error = clGetDeviceInfo(*devs, CL_DEVICE_PLATFORM, sizeof(plat), &plat, NULL);
2392                         had_error = REPORT_ERROR2("get %s");
2393                         if (had_error)
2394                                 break;
2395
2396                         szval = 0;
2397                         for (i = 0; i < num_platforms; ++i) {
2398                                 if (platform[i] == plat)
2399                                         break;
2400                         }
2401                         if (i == num_platforms) {
2402                                 sprintf(strbuf, "<error: platform 0x%p not found>", (void*)plat);
2403                                 break;
2404                         } else {
2405                                 szval += sprintf(strbuf, "%s (%" PRIuS ")",
2406                                         (output_mode == CLINFO_HUMAN ? "Success" : "CL_SUCCESS"),
2407                                         ndevs);
2408                                 szval += snprintf(strbuf + szval, bufsz - szval, "\n" I2_STR "%s",
2409                                         platname_prop, pdata[i].pname);
2410                         }
2411                         for (i = 0; i < ndevs; ++i) {
2412                                 size_t szname = 0;
2413                                 /* for each device, show the device name */
2414                                 /* TODO some other unique ID too, e.g. PCI address, if available? */
2415
2416                                 szval += snprintf(strbuf + szval, bufsz - szval, "\n" I2_STR, devname_prop);
2417                                 if (szval >= bufsz) {
2418                                         trunc_strbuf();
2419                                         break;
2420                                 }
2421
2422                                 current_param = "CL_DEVICE_NAME";
2423                                 current_line = __LINE__+1;
2424                                 error = clGetDeviceInfo(devs[i], CL_DEVICE_NAME, bufsz - szval, strbuf + szval, &szname);
2425                                 had_error = REPORT_ERROR2("get %s");
2426                                 if (had_error)
2427                                         break;
2428                                 szval += szname - 1;
2429
2430
2431                         }
2432                         if (i != ndevs)
2433                                 break; /* had an error earlier, bail */
2434
2435                 }
2436
2437                 if (ctx) {
2438                         clReleaseContext(ctx);
2439                         ctx = NULL;
2440                 }
2441                 printf("%s%s\n", def, strbuf);
2442         }
2443         free(devs);
2444 }
2445
2446 /* check the behavior of NULL platform in clGetDeviceIDs (see checkNullGetDevices)
2447  * and in clCreateContext() */
2448 void checkNullBehavior(void)
2449 {
2450         cl_device_id *dev = NULL;
2451         cl_uint p = 0;
2452         cl_uint pidx;
2453
2454         printf("NULL platform behavior\n");
2455
2456         checkNullGetPlatformName();
2457
2458         pidx = checkNullGetDevices();
2459
2460         /* If there's a default platform, and it has devices, try
2461          * creating a context with its first device and see if it works */
2462
2463         if (pidx == num_platforms) {
2464                 bufcpy(0, no_plat());
2465         } else if (pdata[pidx].ndevs == 0) {
2466                 bufcpy(0, no_dev());
2467         } else {
2468                 p = 0;
2469                 dev = all_devices;
2470                 while (p < num_platforms && p != pidx) {
2471                         dev += pdata[p++].ndevs;
2472                 }
2473                 if (p < num_platforms) {
2474                         checkNullCtx(pidx, dev, "default");
2475                 } else {
2476                         /* this shouldn't happen, but still ... */
2477                         bufcpy(0, "<error: overflow in default platform scan>");
2478                 }
2479         }
2480         printf(I1_STR "%s\n", "clCreateContext(NULL, ...) [default]", strbuf);
2481
2482         /* Look for a device from a non-default platform, if there are any */
2483         if (pidx == num_platforms || num_platforms > 1) {
2484                 p = 0;
2485                 dev = all_devices;
2486                 while (p < num_platforms && (p == pidx || pdata[p].ndevs == 0)) {
2487                         dev += pdata[p++].ndevs;
2488                 }
2489                 if (p < num_platforms) {
2490                         checkNullCtx(p, dev, "non-default");
2491                 } else {
2492                         bufcpy(0, "<error: no devices in non-default plaforms>");
2493                 }
2494                 printf(I1_STR "%s\n", "clCreateContext(NULL, ...) [other]", strbuf);
2495         }
2496
2497         checkNullCtxFromType();
2498
2499 }
2500
2501
2502 /* Get properties of the ocl-icd loader, if available */
2503 /* All properties are currently char[] */
2504 typedef enum {
2505         CL_ICDL_OCL_VERSION=1,
2506         CL_ICDL_VERSION=2,
2507         CL_ICDL_NAME=3,
2508         CL_ICDL_VENDOR=4,
2509 } cl_icdl_info;
2510
2511 /* Function pointer to the ICD loader info function */
2512 cl_int (*clGetICDLoaderInfoOCLICD)(cl_icdl_info, size_t, void*, size_t*);
2513
2514 /* We want to auto-detect the OpenCL version supported by the ICD loader.
2515  * To do this, we will progressively find symbols introduced in new APIs,
2516  * until a NULL symbol is found.
2517  */
2518
2519 struct icd_loader_test {
2520         cl_uint version;
2521         const char *symbol;
2522 } icd_loader_tests[] = {
2523         { 11, "clCreateSubBuffer" },
2524         { 12, "clCreateImage" },
2525         { 20, "clSVMAlloc" },
2526         { 21, "clGetHostTimer" },
2527         { 0, NULL }
2528 };
2529
2530 int
2531 icdl_info_str(cl_icdl_info param, const char* pname)
2532 {
2533         error = clGetICDLoaderInfoOCLICD(param, 0, NULL, &nusz);
2534         had_error = REPORT_ERROR2("get %s size");
2535         if (!had_error) {
2536                 if (nusz > bufsz) {
2537                         REALLOC(strbuf, nusz, current_param);
2538                         bufsz = nusz;
2539                 }
2540                 error = clGetICDLoaderInfoOCLICD(param, bufsz, strbuf, NULL);
2541                 had_error = REPORT_ERROR2("get %s");
2542         }
2543         show_strbuf(pname, 1);
2544         return had_error;
2545 }
2546
2547 struct icdl_info_traits {
2548         cl_icdl_info param; // CL_ICDL_*
2549         const char *sname; // "CL_ICDL_*"
2550         const char *pname; // "ICD loader *"
2551 };
2552
2553 static const char * const oclicdl_pfx = "OCLICD";
2554
2555 #define LINFO(symbol, name) { symbol, #symbol, "ICD loader " name }
2556 struct icdl_info_traits linfo_traits[] = {
2557         LINFO(CL_ICDL_NAME, "Name"),
2558         LINFO(CL_ICDL_VENDOR, "Vendor"),
2559         LINFO(CL_ICDL_VERSION, "Version"),
2560         LINFO(CL_ICDL_OCL_VERSION, "Profile")
2561 };
2562
2563 /* GCC < 4.6 does not support the diagnostic push _inside_ the function,
2564  * so we have to put it outside
2565  */
2566 #if defined __GNUC__ && ((__GNUC__*10 + __GNUC_MINOR__) < 46)
2567 #pragma GCC diagnostic ignored "-Wdeprecated-declarations"
2568 #endif
2569
2570 void oclIcdProps(void)
2571 {
2572         /* First of all, we try to auto-detect the supported ICD loader version */
2573         int i = 0;
2574
2575         do {
2576                 struct icd_loader_test check = icd_loader_tests[i];
2577                 if (check.symbol == NULL)
2578                         break;
2579                 if (dlsym(RTLD_DEFAULT, check.symbol) == NULL)
2580                         break;
2581                 icdl_ocl_version_found = check.version;
2582                 ++i;
2583         } while (1);
2584
2585
2586         /* We find the clGetICDLoaderInfoOCLICD extension address, and use it to query
2587          * the ICD loader properties. It should be noted however that
2588          * clGetExtensionFunctionAddress is marked deprecated as of OpenCL 1.2, so
2589          * to use it and compile cleanly we need disable the relevant warning.
2590          * It should be noted that in this specific case we cannot replace the
2591          * call to clGetExtensionFunctionAddress with a call to the superseding function
2592          * clGetExtensionFunctionAddressForPlatform because the extension is in the
2593          * loader itself, not in a specific platform.
2594          */
2595
2596 #ifdef _MSC_VER
2597 #pragma warning(push)
2598 #pragma warning(disable : 4996)
2599 #elif defined __GNUC__ && ((__GNUC__*10 + __GNUC_MINOR__) >= 46)
2600 #pragma GCC diagnostic push
2601 #pragma GCC diagnostic ignored "-Wdeprecated-declarations"
2602 #endif
2603
2604         PTR_FUNC_PTR clGetICDLoaderInfoOCLICD = clGetExtensionFunctionAddress("clGetICDLoaderInfoOCLICD");
2605
2606 #ifdef _MSC_VER
2607 #pragma warning(pop)
2608 #elif defined __GNUC__ && ((__GNUC__*10 + __GNUC_MINOR__) >= 46)
2609 #pragma GCC diagnostic pop
2610 #endif
2611
2612         if (clGetICDLoaderInfoOCLICD != NULL) {
2613                 /* TODO think of a sensible header in CLINFO_RAW */
2614                 if (output_mode != CLINFO_RAW)
2615                         puts("\nICD loader properties");
2616                 current_function = __func__;
2617
2618                 if (output_mode == CLINFO_RAW) {
2619                         line_pfx_len = strlen(oclicdl_pfx) + 5;
2620                         REALLOC(line_pfx, line_pfx_len, "line prefix OCL ICD");
2621                         sprintf(strbuf, "[%s/*]", oclicdl_pfx);
2622                         sprintf(line_pfx, "%*s", -line_pfx_len, strbuf);
2623                 }
2624
2625                 for (current_line = 0; current_line < ARRAY_SIZE(linfo_traits); ++current_line) {
2626                         const struct icdl_info_traits *traits = linfo_traits + current_line;
2627                         current_param = traits->sname;
2628
2629                         had_error = icdl_info_str(traits->param,
2630                                 output_mode == CLINFO_HUMAN ?
2631                                 traits->pname : traits->sname);
2632
2633                         if (!had_error && traits->param == CL_ICDL_OCL_VERSION) {
2634                                 icdl_ocl_version = getOpenCLVersion(strbuf + 7);
2635                         }
2636                 }
2637         }
2638
2639         if (output_mode == CLINFO_HUMAN) {
2640                 if (icdl_ocl_version &&
2641                         icdl_ocl_version != icdl_ocl_version_found) {
2642                         printf( "\tNOTE:\tyour OpenCL library declares to support OpenCL %u.%u,\n"
2643                                 "\t\tbut it seems to support up to OpenCL %u.%u %s.\n",
2644                                 icdl_ocl_version / 10, icdl_ocl_version % 10,
2645                                 icdl_ocl_version_found / 10, icdl_ocl_version_found % 10,
2646                                 icdl_ocl_version_found < icdl_ocl_version  ?
2647                                 "only" : "too");
2648                 }
2649                 if (icdl_ocl_version_found < max_plat_version) {
2650                         printf( "\tNOTE:\tyour OpenCL library only supports OpenCL %u.%u,\n"
2651                                 "\t\tbut some installed platforms support OpenCL %u.%u.\n"
2652                                 "\t\tPrograms using %u.%u features may crash\n"
2653                                 "\t\tor behave unexepectedly\n",
2654                                 icdl_ocl_version_found / 10, icdl_ocl_version_found % 10,
2655                                 max_plat_version / 10, max_plat_version % 10,
2656                                 max_plat_version / 10, max_plat_version % 10);
2657                 }
2658         }
2659 }
2660
2661 #if defined __GNUC__ && ((__GNUC__*10 + __GNUC_MINOR__) < 46)
2662 #pragma GCC diagnostic warning "-Wdeprecated-declarations"
2663 #endif
2664
2665 void version(void)
2666 {
2667         puts("clinfo version 2.1.16.01.12");
2668 }
2669
2670 void usage(void)
2671 {
2672         version();
2673         puts("Display properties of all available OpenCL platforms and devices");
2674         puts("Usage: clinfo [options ...]\n");
2675         puts("Options:");
2676         puts("\t--human\t\thuman-friendly output (default)");
2677         puts("\t--raw\t\traw output");
2678         puts("\t--offline\talso show offline devices");
2679         puts("\t--list, -l\tonly list the platforms and devices by name");
2680         puts("\t-h, -?\t\tshow usage");
2681         puts("\t--version, -v\tshow version\n");
2682         puts("Defaults to raw mode if invoked with");
2683         puts("a name that contains the string \"raw\"");
2684 }
2685
2686 int main(int argc, char *argv[])
2687 {
2688         cl_uint p;
2689         int a = 0;
2690
2691         cl_bool show_offline = CL_FALSE;
2692
2693         /* if there's a 'raw' in the program name, switch to raw output mode */
2694         if (strstr(argv[0], "raw"))
2695                 output_mode = CLINFO_RAW;
2696
2697         /* process command-line arguments */
2698         while (++a < argc) {
2699                 if (!strcmp(argv[a], "--raw"))
2700                         output_mode = CLINFO_RAW;
2701                 else if (!strcmp(argv[a], "--human"))
2702                         output_mode = CLINFO_HUMAN;
2703                 else if (!strcmp(argv[a], "--offline"))
2704                         show_offline = CL_TRUE;
2705                 else if (!strcmp(argv[a], "-l") || !strcmp(argv[a], "--list"))
2706                         list_only = CL_TRUE;
2707                 else if (!strcmp(argv[a], "-?") || !strcmp(argv[a], "-h")) {
2708                         usage();
2709                         return 0;
2710                 } else if (!strcmp(argv[a], "--version") || !strcmp(argv[a], "-v")) {
2711                         version();
2712                         return 0;
2713                 } else {
2714                         fprintf(stderr, "ignoring unknown command-line parameter %s\n", argv[a]);
2715                 }
2716         }
2717
2718
2719         ALLOC(strbuf, 1024, "general string buffer");
2720         bufsz = 1024;
2721
2722         error = clGetPlatformIDs(0, NULL, &num_platforms);
2723         if (error != CL_PLATFORM_NOT_FOUND_KHR)
2724                 CHECK_ERROR("number of platforms");
2725
2726         if (!list_only)
2727                 printf(I0_STR "%u\n",
2728                         (output_mode == CLINFO_HUMAN ?
2729                          "Number of platforms" : "#PLATFORMS"),
2730                         num_platforms);
2731         if (!num_platforms)
2732                 return 0;
2733
2734         ALLOC(platform, num_platforms, "platform IDs");
2735         error = clGetPlatformIDs(num_platforms, platform, NULL);
2736         CHECK_ERROR("platform IDs");
2737
2738         ALLOC(pdata, num_platforms, "platform data");
2739         ALLOC(line_pfx, 1, "line prefix");
2740
2741         for (p = 0; p < num_platforms; ++p) {
2742                 printPlatformInfo(p);
2743                 if (!list_only)
2744                         puts("");
2745         }
2746
2747         if (num_devs_all > 0) {
2748                 ALLOC(all_devices, num_devs_all, "device IDs");
2749         }
2750
2751         if (list_only) {
2752                 listPlatformsAndDevices(show_offline);
2753         } else {
2754                 showDevices(show_offline);
2755                 if (output_mode != CLINFO_RAW)
2756                         checkNullBehavior();
2757                 oclIcdProps();
2758         }
2759
2760         return 0;
2761 }