Khronos

WebCL logo

WebCL Specification

Version 1.0, 14 March 2014

This version:
http://www.khronos.org/registry/webcl/specs/1.0.0/
Latest version:
http://www.khronos.org/registry/webcl/specs/latest/1.0/
Previous version:
None
Editors:
Tomi Aarnio (Nokia Research)
Mikaël Bourges-Sévenier (Motorola Mobility, Inc.)

Copyright © 2014 Khronos Group


Abstract

This specification defines WebCL (Web Computing Language). WebCL is a JavaScript binding to the Khronos OpenCL standard for heterogeneous parallel computing. It enables web applications to harness GPU and multi-core CPU parallel processing from within a Web browser, enabling significant acceleration of computationally intensive applications, such as image and video processing, data visualization, and 3D games.

Status of this document

Public discussion of this specification is welcome on the (archived) WebCL mailing list public_webcl@khronos.org (see instructions).

Contents

Introduction

With the popularity of web-centric platforms, there is a need to leverage high-performance parallel processing to accelerate compute-intensive web applications. Such applications include, for example, data visualization, image and video processing, 3D gaming, computational photography, computer vision, and cryptography. By providing application developers with a standardized JavaScript API and a portable kernel programming language, WebCL enables parallel computing on heterogeneous multicore hardware across a breadth of devices, including mobile, desktop, and server. The WebCL API and kernel language are designed to be substantially similar to those of OpenCL (Open Computing Language), and implementable on top of a conformant OpenCL driver.

Conventions

Basic types

The following basic types are used in this document.

typedef boolean             CLboolean;
typedef long                CLint;      // 32-bit signed integer
typedef long long           CLlong;     // 64-bit signed integer
typedef unsigned long       CLuint;     // 32-bit unsigned integer
typedef unsigned long long  CLulong;    // 64-bit unsigned integer
typedef const unsigned long CLenum;     // Used for enumerated types, such as WebCL.DEVICE_TYPE_GPU

The implementation of any WebCL function that accepts numeric values as input must ensure that no undefined behavior will occur if the input value is outside of the specified numeric range. Where necessary, the implementation may fit the incoming JavaScript number into the valid range using a clamp or modulo operator, or alternatively throw an exception.

Exceptions

WebCL methods throw exceptions instead of returning error codes as return values as in OpenCL. Exceptions that may be thrown by each method are listed in that method's documentation. Additionally, almost all methods may throw `OUT_OF_RESOURCES`, `OUT_OF_HOST_MEMORY`, or the WebCL-specific `WEBCL_IMPLEMENTATION_FAILURE`, so these are not listed separately. Furthermore, calling any function (except `release`) on an object that has been released, will throw INVALID_TYPE, where `TYPE` is `CONTEXT`, `COMMAND_QUEUE`, `MEM_OBJECT`, `SAMPLER`, `PROGRAM`, `KERNEL`, or `EVENT`. These exceptions are also not listed separately in the method descriptions.

exception WebCLException : DOMException {
  DOMString name;              // A string representation of the numeric error code, e.g. "INVALID_VALUE"
  DOMString? message;          // An implementation-specific description of what caused the exception
};

Both the `name` and `message` fields are present in any WebCLException thrown by the implementation, but the message may be `null`.

Resource management

Each dynamically created WebCL object has a `release` method that releases the resources consumed by that object. This does not cause the object to be deleted or garbage collected; it will remain in place, but trying to use it will cause an exception. Trying to release an object that has already been released will be silently ignored.

For convenience, the WebCL and WebCLContext classes contain an additional `releaseAll` function, which releases the context(s), as well as all other WebCL objects created from them. The usage and behavior of the release methods are illustrated in the example below.

  var ctx1 = webcl.createContext(...);
  var ctx2 = webcl.createContext(...);
  var ctx3 = webcl.createContext(...);
  var A = ctx1.createBuffer(...);
  var B = ctx2.createBuffer(...);
  ctx1.release();                    // releases ctx1, but not buffer A
  ctx2.releaseAll();                 // releases ctx2 and buffer B
  A.release();                       // releases buffer A 
  B.release();                       // does nothing: B is already released
  var C = ctx1.createBuffer(...);    // EXCEPTION: ctx1 is no longer valid
  webcl.releaseAll();                // releases ctx3

Applications are strongly recommended to explicitly release all WebCL objects as soon as they are no longer needed, instead of relying on the JavaScript garbage collector. This is necessary because garbage collectors typically do not give any guarantees on when (or indeed if) they will reclaim the objects that are no longer in scope.

When the global `document` object goes out of scope, the WebCL implementation must implicitly call `release` on all remaining WebCL objects.

Concurrency

Depending on the implementation, WebCL operations may be running concurrently with JavaScript. In particular, it may be possible for the application to modify an ArrayBuffer while it's being asynchronously copied to/from a WebCLMemoryObject. To avoid corrupting the contents of either buffer, applications should not modify an ArrayBuffer that has been enqueued for async read/write, until the relevant WebCL command queue has finished.

Callbacks

WebCL allows certain long-running functions to be executed either synchronously or asynchronously. The asynchronous mode is used if a user-defined callback function is given as an argument to such functions. The asynchronous mode is strongly recommended, so as to avoid blocking the JavaScript main thread. If a callback function is associated with a WebCL object that is subsequently released, the callback will no longer be invoked. The signature of the callback function is as follows:

callback WebCLCallback = void (optional WebCLEvent event);

To avoid blocking the main thread, the callback function should complete quickly and not call any potentially long-running functions. Certain WebCL functions are explicitly prohibited, and will cause an `INVALID_OPERATION` exception if invoked from a callback. The disallowed functions include `createContext`, `createCommandQueue`, and the blocking forms of `finish`, `waitForEvents`, `build`, and enqueue{Read, Write}{Buffer, BufferRect, Image}.

Interfaces

WebCL

The WebCL API is accessible through a singleton `webcl` object that resides in the global `window` namespace. The `webcl` object implements the WebCL interface, providing functions for creating computing contexts and querying the available WebCL platforms and extensions.

partial interface Window {
  readonly attribute WebCL webcl;
};
interface WebCL {

  sequence<WebCLPlatform> getPlatforms();

  WebCLContext createContext(optional CLenum deviceType);

  WebCLContext createContext(WebCLPlatform platform, optional CLenum deviceType);

  WebCLContext createContext(WebCLDevice device);

  WebCLContext createContext(sequence<WebCLDevice> devices);

  sequence<DOMString>? getSupportedExtensions();

  CLboolean enableExtension(DOMString extensionName);

  void waitForEvents(sequence<WebCLEvent> eventWaitList,
                     optional WebCLCallback whenFinished);

  void releaseAll();


  /* Error Codes */
  CLint SUCCESS                                   = 0;
  CLint DEVICE_NOT_FOUND                          = -1;
  CLint DEVICE_NOT_AVAILABLE                      = -2;
  CLint COMPILER_NOT_AVAILABLE                    = -3;
  CLint MEM_OBJECT_ALLOCATION_FAILURE             = -4;
  CLint OUT_OF_RESOURCES                          = -5;
  CLint OUT_OF_HOST_MEMORY                        = -6;
  CLint PROFILING_INFO_NOT_AVAILABLE              = -7;
  CLint MEM_COPY_OVERLAP                          = -8;
  CLint IMAGE_FORMAT_MISMATCH                     = -9;
  CLint IMAGE_FORMAT_NOT_SUPPORTED                = -10;
  CLint BUILD_PROGRAM_FAILURE                     = -11;
  CLint MAP_FAILURE                               = -12;
  CLint MISALIGNED_SUB_BUFFER_OFFSET              = -13;
  CLint EXEC_STATUS_ERROR_FOR_EVENTS_IN_WAIT_LIST = -14;
  CLint INVALID_VALUE                             = -30;
  CLint INVALID_DEVICE_TYPE                       = -31;
  CLint INVALID_PLATFORM                          = -32;
  CLint INVALID_DEVICE                            = -33;
  CLint INVALID_CONTEXT                           = -34;
  CLint INVALID_QUEUE_PROPERTIES                  = -35;
  CLint INVALID_COMMAND_QUEUE                     = -36;
  CLint INVALID_HOST_PTR                          = -37;
  CLint INVALID_MEM_OBJECT                        = -38;
  CLint INVALID_IMAGE_FORMAT_DESCRIPTOR           = -39;
  CLint INVALID_IMAGE_SIZE                        = -40;
  CLint INVALID_SAMPLER                           = -41;
  CLint INVALID_BINARY                            = -42;
  CLint INVALID_BUILD_OPTIONS                     = -43;
  CLint INVALID_PROGRAM                           = -44;
  CLint INVALID_PROGRAM_EXECUTABLE                = -45;
  CLint INVALID_KERNEL_NAME                       = -46;
  CLint INVALID_KERNEL_DEFINITION                 = -47;
  CLint INVALID_KERNEL                            = -48;
  CLint INVALID_ARG_INDEX                         = -49;
  CLint INVALID_ARG_VALUE                         = -50;
  CLint INVALID_ARG_SIZE                          = -51;
  CLint INVALID_KERNEL_ARGS                       = -52;
  CLint INVALID_WORK_DIMENSION                    = -53;
  CLint INVALID_WORK_GROUP_SIZE                   = -54;
  CLint INVALID_WORK_ITEM_SIZE                    = -55;
  CLint INVALID_GLOBAL_OFFSET                     = -56;
  CLint INVALID_EVENT_WAIT_LIST                   = -57;
  CLint INVALID_EVENT                             = -58;
  CLint INVALID_OPERATION                         = -59;
  //CLint INVALID_GL_OBJECT                         = -60; // moved to extension
  CLint INVALID_BUFFER_SIZE                       = -61;
  //CLint INVALID_MIP_LEVEL                         = -62; // moved to extension
  CLint INVALID_GLOBAL_WORK_SIZE                  = -63;
  CLint INVALID_PROPERTY                          = -64;

  /* cl_bool */
  CLenum FALSE                                     = 0;
  CLenum TRUE                                      = 1;

  /* cl_platform_info */
  CLenum PLATFORM_PROFILE                          = 0x0900;
  CLenum PLATFORM_VERSION                          = 0x0901;
  CLenum PLATFORM_NAME                             = 0x0902;
  CLenum PLATFORM_VENDOR                           = 0x0903;
  CLenum PLATFORM_EXTENSIONS                       = 0x0904;

  /* cl_device_type - bitfield */
  CLenum DEVICE_TYPE_DEFAULT                       = 0x1;
  CLenum DEVICE_TYPE_CPU                           = 0x2;
  CLenum DEVICE_TYPE_GPU                           = 0x4;
  CLenum DEVICE_TYPE_ACCELERATOR                   = 0x8;
  CLenum DEVICE_TYPE_ALL                           = 0xFFFFFFFF;

  /* cl_device_info */
  CLenum DEVICE_TYPE                               = 0x1000;
  CLenum DEVICE_VENDOR_ID                          = 0x1001;
  CLenum DEVICE_MAX_COMPUTE_UNITS                  = 0x1002;
  CLenum DEVICE_MAX_WORK_ITEM_DIMENSIONS           = 0x1003;
  CLenum DEVICE_MAX_WORK_GROUP_SIZE                = 0x1004;
  CLenum DEVICE_MAX_WORK_ITEM_SIZES                = 0x1005;
  CLenum DEVICE_PREFERRED_VECTOR_WIDTH_CHAR        = 0x1006;
  CLenum DEVICE_PREFERRED_VECTOR_WIDTH_SHORT       = 0x1007;
  CLenum DEVICE_PREFERRED_VECTOR_WIDTH_INT         = 0x1008;
  CLenum DEVICE_PREFERRED_VECTOR_WIDTH_LONG        = 0x1009;
  CLenum DEVICE_PREFERRED_VECTOR_WIDTH_FLOAT       = 0x100A;
  //CLenum DEVICE_PREFERRED_VECTOR_WIDTH_DOUBLE      = 0x100B; // moved to extension
  CLenum DEVICE_MAX_CLOCK_FREQUENCY                = 0x100C;
  CLenum DEVICE_ADDRESS_BITS                       = 0x100D;
  CLenum DEVICE_MAX_READ_IMAGE_ARGS                = 0x100E;
  CLenum DEVICE_MAX_WRITE_IMAGE_ARGS               = 0x100F;
  CLenum DEVICE_MAX_MEM_ALLOC_SIZE                 = 0x1010;
  CLenum DEVICE_IMAGE2D_MAX_WIDTH                  = 0x1011;
  CLenum DEVICE_IMAGE2D_MAX_HEIGHT                 = 0x1012;
  CLenum DEVICE_IMAGE3D_MAX_WIDTH                  = 0x1013;
  CLenum DEVICE_IMAGE3D_MAX_HEIGHT                 = 0x1014;
  CLenum DEVICE_IMAGE3D_MAX_DEPTH                  = 0x1015;
  CLenum DEVICE_IMAGE_SUPPORT                      = 0x1016;
  CLenum DEVICE_MAX_PARAMETER_SIZE                 = 0x1017;
  CLenum DEVICE_MAX_SAMPLERS                       = 0x1018;
  CLenum DEVICE_MEM_BASE_ADDR_ALIGN                = 0x1019;
  //CLenum DEVICE_MIN_DATA_TYPE_ALIGN_SIZE           = 0x101A; // removed; deprecated in OpenCL 1.2
  CLenum DEVICE_SINGLE_FP_CONFIG                   = 0x101B;
  CLenum DEVICE_GLOBAL_MEM_CACHE_TYPE              = 0x101C;
  CLenum DEVICE_GLOBAL_MEM_CACHELINE_SIZE          = 0x101D;
  CLenum DEVICE_GLOBAL_MEM_CACHE_SIZE              = 0x101E;
  CLenum DEVICE_GLOBAL_MEM_SIZE                    = 0x101F;
  CLenum DEVICE_MAX_CONSTANT_BUFFER_SIZE           = 0x1020;
  CLenum DEVICE_MAX_CONSTANT_ARGS                  = 0x1021;
  CLenum DEVICE_LOCAL_MEM_TYPE                     = 0x1022;
  CLenum DEVICE_LOCAL_MEM_SIZE                     = 0x1023;
  CLenum DEVICE_ERROR_CORRECTION_SUPPORT           = 0x1024;
  CLenum DEVICE_PROFILING_TIMER_RESOLUTION         = 0x1025;
  CLenum DEVICE_ENDIAN_LITTLE                      = 0x1026;
  CLenum DEVICE_AVAILABLE                          = 0x1027;
  CLenum DEVICE_COMPILER_AVAILABLE                 = 0x1028;
  CLenum DEVICE_EXECUTION_CAPABILITIES             = 0x1029;
  CLenum DEVICE_QUEUE_PROPERTIES                   = 0x102A;
  CLenum DEVICE_NAME                               = 0x102B;
  CLenum DEVICE_VENDOR                             = 0x102C;
  CLenum DRIVER_VERSION                            = 0x102D;
  CLenum DEVICE_PROFILE                            = 0x102E;
  CLenum DEVICE_VERSION                            = 0x102F;
  CLenum DEVICE_EXTENSIONS                         = 0x1030;
  CLenum DEVICE_PLATFORM                           = 0x1031;
  //CLenum DEVICE_DOUBLE_FP_CONFIG                   = 0x1032; // moved to extension
  //CLenum DEVICE_HALF_FP_CONFIG                     = 0x1033; // moved to extension
  //CLenum DEVICE_PREFERRED_VECTOR_WIDTH_HALF        = 0x1034; // moved to extension
  CLenum DEVICE_HOST_UNIFIED_MEMORY                = 0x1035;
  CLenum DEVICE_NATIVE_VECTOR_WIDTH_CHAR           = 0x1036;
  CLenum DEVICE_NATIVE_VECTOR_WIDTH_SHORT          = 0x1037;
  CLenum DEVICE_NATIVE_VECTOR_WIDTH_INT            = 0x1038;
  CLenum DEVICE_NATIVE_VECTOR_WIDTH_LONG           = 0x1039;
  CLenum DEVICE_NATIVE_VECTOR_WIDTH_FLOAT          = 0x103A;
  //CLenum DEVICE_NATIVE_VECTOR_WIDTH_DOUBLE         = 0x103B; // moved to extension
  //CLenum DEVICE_NATIVE_VECTOR_WIDTH_HALF           = 0x103C; // moved to extension
  CLenum DEVICE_OPENCL_C_VERSION                   = 0x103D;

  /* cl_device_fp_config - bitfield */
  CLenum FP_DENORM                                 = 0x1;
  CLenum FP_INF_NAN                                = 0x2;
  CLenum FP_ROUND_TO_NEAREST                       = 0x4;
  CLenum FP_ROUND_TO_ZERO                          = 0x8;
  CLenum FP_ROUND_TO_INF                           = 0x10;
  CLenum FP_FMA                                    = 0x20;
  CLenum FP_SOFT_FLOAT                             = 0x40;

  /* cl_device_mem_cache_type */
  CLenum NONE                                      = 0x0;
  CLenum READ_ONLY_CACHE                           = 0x1;
  CLenum READ_WRITE_CACHE                          = 0x2;

  /* cl_device_local_mem_type */
  CLenum LOCAL                                     = 0x1;
  CLenum GLOBAL                                    = 0x2;

  /* cl_device_exec_capabilities - bitfield */
  CLenum EXEC_KERNEL                               = 0x1;
  //CLenum EXEC_NATIVE_KERNEL                        = 0x2; // disallowed

  /* cl_command_queue_properties - bitfield */
  CLenum QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE       = 0x1;
  CLenum QUEUE_PROFILING_ENABLE                    = 0x2;

  /* cl_context_info  */
  //CLenum CONTEXT_REFERENCE_COUNT                   = 0x1080; // disallowed
  CLenum CONTEXT_DEVICES                           = 0x1081;
  //CLenum CONTEXT_PROPERTIES                        = 0x1082; // disallowed; no context properties in WebCL
  CLenum CONTEXT_NUM_DEVICES                       = 0x1083;

  /* cl_context_properties */
  //CLenum CONTEXT_PLATFORM                          = 0x1084; // disallowed; no context properties in WebCL

  /* cl_command_queue_info */
  CLenum QUEUE_CONTEXT                             = 0x1090;
  CLenum QUEUE_DEVICE                              = 0x1091;
  //CLenum QUEUE_REFERENCE_COUNT                     = 0x1092; // disallowed
  CLenum QUEUE_PROPERTIES                          = 0x1093;

  /* cl_mem_flags - bitfield */
  CLenum MEM_READ_WRITE                            = 0x1;
  CLenum MEM_WRITE_ONLY                            = 0x2;
  CLenum MEM_READ_ONLY                             = 0x4;

  /* cl_channel_order */
  CLenum R                                         = 0x10B0;
  CLenum A                                         = 0x10B1;
  CLenum RG                                        = 0x10B2;
  CLenum RA                                        = 0x10B3;
  CLenum RGB                                       = 0x10B4;
  CLenum RGBA                                      = 0x10B5;
  CLenum BGRA                                      = 0x10B6;
  CLenum ARGB                                      = 0x10B7;
  CLenum INTENSITY                                 = 0x10B8;
  CLenum LUMINANCE                                 = 0x10B9;
  CLenum Rx                                        = 0x10BA;
  CLenum RGx                                       = 0x10BB;
  CLenum RGBx                                      = 0x10BC;

  /* cl_channel_type */
  CLenum SNORM_INT8                                = 0x10D0;
  CLenum SNORM_INT16                               = 0x10D1;
  CLenum UNORM_INT8                                = 0x10D2;
  CLenum UNORM_INT16                               = 0x10D3;
  CLenum UNORM_SHORT_565                           = 0x10D4;
  CLenum UNORM_SHORT_555                           = 0x10D5;
  CLenum UNORM_INT_101010                          = 0x10D6;
  CLenum SIGNED_INT8                               = 0x10D7;
  CLenum SIGNED_INT16                              = 0x10D8;
  CLenum SIGNED_INT32                              = 0x10D9;
  CLenum UNSIGNED_INT8                             = 0x10DA;
  CLenum UNSIGNED_INT16                            = 0x10DB;
  CLenum UNSIGNED_INT32                            = 0x10DC;
  CLenum HALF_FLOAT                                = 0x10DD;
  CLenum FLOAT                                     = 0x10DE;

  /* cl_mem_object_type */
  CLenum MEM_OBJECT_BUFFER                         = 0x10F0;
  CLenum MEM_OBJECT_IMAGE2D                        = 0x10F1;
  CLenum MEM_OBJECT_IMAGE3D                        = 0x10F2;

  /* cl_mem_info */
  CLenum MEM_TYPE                                  = 0x1100;
  CLenum MEM_FLAGS                                 = 0x1101;
  CLenum MEM_SIZE                                  = 0x1102;
  //CLenum MEM_HOST_PTR                              = 0x1103; // disallowed
  //CLenum MEM_MAP_COUNT                             = 0x1104; // disallowed
  //CLenum MEM_REFERENCE_COUNT                       = 0x1105; // disallowed
  CLenum MEM_CONTEXT                               = 0x1106;
  CLenum MEM_ASSOCIATED_MEMOBJECT                  = 0x1107;
  CLenum MEM_OFFSET                                = 0x1108;

  /* cl_image_info */
  CLenum IMAGE_FORMAT                              = 0x1110;
  CLenum IMAGE_ELEMENT_SIZE                        = 0x1111;
  CLenum IMAGE_ROW_PITCH                           = 0x1112;
  CLenum IMAGE_WIDTH                               = 0x1114;
  CLenum IMAGE_HEIGHT                              = 0x1115;

  /* cl_addressing_mode */
  //CLenum ADDRESS_NONE                              = 0x1130; // disallowed
  CLenum ADDRESS_CLAMP_TO_EDGE                     = 0x1131;
  CLenum ADDRESS_CLAMP                             = 0x1132;
  CLenum ADDRESS_REPEAT                            = 0x1133;
  CLenum ADDRESS_MIRRORED_REPEAT                   = 0x1134;

  /* cl_filter_mode */
  CLenum FILTER_NEAREST                            = 0x1140;
  CLenum FILTER_LINEAR                             = 0x1141;

  /* cl_sampler_info */
  //CLenum SAMPLER_REFERENCE_COUNT                   = 0x1150; // disallowed
  CLenum SAMPLER_CONTEXT                           = 0x1151;
  CLenum SAMPLER_NORMALIZED_COORDS                 = 0x1152;
  CLenum SAMPLER_ADDRESSING_MODE                   = 0x1153;
  CLenum SAMPLER_FILTER_MODE                       = 0x1154;

  /* cl_map_flags - bitfield */
  //CLenum MAP_READ                                  = 0x1; // disallowed
  //CLenum MAP_WRITE                                 = 0x2; // disallowed

  /* cl_program_info */
  //CLenum PROGRAM_REFERENCE_COUNT                   = 0x1160; // disallowed
  CLenum PROGRAM_CONTEXT                           = 0x1161;
  CLenum PROGRAM_NUM_DEVICES                       = 0x1162;
  CLenum PROGRAM_DEVICES                           = 0x1163;
  CLenum PROGRAM_SOURCE                            = 0x1164;
  //CLenum PROGRAM_BINARY_SIZES                      = 0x1165; // disallowed
  //CLenum PROGRAM_BINARIES                          = 0x1166; // disallowed

  /* cl_program_build_info */
  CLenum PROGRAM_BUILD_STATUS                      = 0x1181;
  CLenum PROGRAM_BUILD_OPTIONS                     = 0x1182;
  CLenum PROGRAM_BUILD_LOG                         = 0x1183;

  /* cl_build_status */
  CLint BUILD_SUCCESS                             = 0;
  CLint BUILD_NONE                                = -1;
  CLint BUILD_ERROR                               = -2;
  CLint BUILD_IN_PROGRESS                         = -3;

  /* cl_kernel_info */
  CLenum KERNEL_FUNCTION_NAME                      = 0x1190;
  CLenum KERNEL_NUM_ARGS                           = 0x1191;
  //CLenum KERNEL_REFERENCE_COUNT                    = 0x1192; // disallowed
  CLenum KERNEL_CONTEXT                            = 0x1193;
  CLenum KERNEL_PROGRAM                            = 0x1194;

  /* cl_kernel_work_group_info */
  CLenum KERNEL_WORK_GROUP_SIZE                    = 0x11B0;
  CLenum KERNEL_COMPILE_WORK_GROUP_SIZE            = 0x11B1;
  CLenum KERNEL_LOCAL_MEM_SIZE                     = 0x11B2;
  CLenum KERNEL_PREFERRED_WORK_GROUP_SIZE_MULTIPLE = 0x11B3;
  CLenum KERNEL_PRIVATE_MEM_SIZE                   = 0x11B4;

  /* cl_event_info  */
  CLenum EVENT_COMMAND_QUEUE                       = 0x11D0;
  CLenum EVENT_COMMAND_TYPE                        = 0x11D1;
  //CLenum EVENT_REFERENCE_COUNT                     = 0x11D2; // disallowed
  CLenum EVENT_COMMAND_EXECUTION_STATUS            = 0x11D3;
  CLenum EVENT_CONTEXT                             = 0x11D4;

  /* cl_command_type */
  CLenum COMMAND_NDRANGE_KERNEL                    = 0x11F0;
  CLenum COMMAND_TASK                              = 0x11F1;
  //CLenum COMMAND_NATIVE_KERNEL                     = 0x11F2; // disallowed
  CLenum COMMAND_READ_BUFFER                       = 0x11F3;
  CLenum COMMAND_WRITE_BUFFER                      = 0x11F4;
  CLenum COMMAND_COPY_BUFFER                       = 0x11F5;
  CLenum COMMAND_READ_IMAGE                        = 0x11F6;
  CLenum COMMAND_WRITE_IMAGE                       = 0x11F7;
  CLenum COMMAND_COPY_IMAGE                        = 0x11F8;
  CLenum COMMAND_COPY_IMAGE_TO_BUFFER              = 0x11F9;
  CLenum COMMAND_COPY_BUFFER_TO_IMAGE              = 0x11FA;
  //CLenum COMMAND_MAP_BUFFER                        = 0x11FB; // disallowed
  //CLenum COMMAND_MAP_IMAGE                         = 0x11FC; // disallowed
  //CLenum COMMAND_UNMAP_MEM_OBJECT                  = 0x11FD; // disallowed
  CLenum COMMAND_MARKER                            = 0x11FE;
  //CLenum COMMAND_ACQUIRE_GL_OBJECTS                = 0x11FF; // moved to extension
  //CLenum COMMAND_RELEASE_GL_OBJECTS                = 0x1200; // moved to extension
  CLenum COMMAND_READ_BUFFER_RECT                  = 0x1201;
  CLenum COMMAND_WRITE_BUFFER_RECT                 = 0x1202;
  CLenum COMMAND_COPY_BUFFER_RECT                  = 0x1203;
  CLenum COMMAND_USER                              = 0x1204;

  /* command execution status */
  CLenum COMPLETE                                  = 0x0;
  CLenum RUNNING                                   = 0x1;
  CLenum SUBMITTED                                 = 0x2;
  CLenum QUEUED                                    = 0x3;

  /* cl_profiling_info  */
  CLenum PROFILING_COMMAND_QUEUED                  = 0x1280;
  CLenum PROFILING_COMMAND_SUBMIT                  = 0x1281;
  CLenum PROFILING_COMMAND_START                   = 0x1282;
  CLenum PROFILING_COMMAND_END                     = 0x1283;
};
sequence<WebCLPlatform> getPlatforms() (OpenCL 1.1 §4.1, clGetPlatformIDs)
Retrieves all WebCL platforms that are available in this system.
WebCLContext createContext(optional CLenum deviceType = WebCL.DEVICE_TYPE_DEFAULT) (OpenCL 1.1 §4.3, clCreateContextFromType)
Returns a newly created WebCL context for the given type of device, if available. The available values for `deviceType` are listed in the table below.
deviceTypeDescription
DEVICE_TYPE_CPUA single-core or multi-core CPU, typically the host processor.
DEVICE_TYPE_GPUA graphics processing unit, typically also used by WebGL.
DEVICE_TYPE_ACCELERATORA dedicated OpenCL accelerator.
DEVICE_TYPE_DEFAULTThe default device on the system.
Exceptions:
WebCLContext createContext(WebCLPlatform platform, optional CLenum deviceType = WebCL.DEVICE_TYPE_DEFAULT) (OpenCL 1.1 §4.3, clCreateContextFromType)
Returns a newly created WebCL context for the given type of device, if available on the given platform. The available values for `deviceType` are listed in the table below.
deviceTypeDescription
DEVICE_TYPE_CPUA single-core or multi-core CPU, typically the host processor.
DEVICE_TYPE_GPUA graphics processing unit, typically also used by WebGL.
DEVICE_TYPE_ACCELERATORA dedicated OpenCL accelerator.
DEVICE_TYPE_DEFAULTThe default device on the given platform.
DEVICE_TYPE_ALLAll devices available on the given platform.
Exceptions:
WebCLContext createContext(WebCLDevice device) (OpenCL 1.1 §4.3, clCreateContext)
Returns a newly created WebCL context for the given device.
Exceptions:
WebCLContext createContext(sequence<WebCLDevice> devices); (OpenCL 1.1 §4.3, clCreateContext)
Returns a newly created WebCL context spanning the given list of devices.
Exceptions:
sequence<DOMString>? getSupportedExtensions() (OpenCL 1.1 §9)
Returns an array of extension names that are supported by all WebCLPlatforms and WebCLDevices in this implementation. Any string in this list, when passed to `enableExtension`, must enable the corresponding extension.
CLboolean enableExtension(DOMString extensionName)
Enables the given extension on all WebCLPlatforms and WebCLDevices. Returns `true` if the extension is successfully enabled, or `false` if not. The available extension names can be queried by `getSupportedExtensions`. Note that enabling an extension does not take effect retroactively, i.e., contexts that were created before enabling the extension will continue to not have the extended capabilities.
void waitForEvents(sequence<WebCLEvent> eventWaitList, optional WebCLCallback whenFinished) (OpenCL 1.1 §5.9, clWaitForEvents)
Invokes the given callback when all events in `eventWaitList` have completed. If a callback is not provided, the JavaScript main thread will be blocked until the events have completed. Applications are strongly advised to provide a callback; the blocking mode mainly exists to ease the porting of existing OpenCL code.
Exceptions:
void releaseAll() (OpenCL 1.1 §4.3, clReleaseContext)
Releases the resources held up by all WebCLContext instances, as well as their descendant objects, if any.
// Valid examples of createContext. All of these will create a valid
// context, assuming that a suitable device is available.

ctx = webcl.createContext();
ctx = webcl.createContext(WebCL.DEVICE_TYPE_CPU);
ctx = webcl.createContext(aPlatform);
ctx = webcl.createContext(aPlatform, WebCL.DEVICE_TYPE_GPU);
ctx = webcl.createContext(aPlatform, WebCL.DEVICE_TYPE_ALL);
ctx = webcl.createContext(aDevice);
ctx = webcl.createContext([aDevice, anotherDeviceOnTheSamePlatform]);

// Invalid examples of createContext. These will throw an exception.

ctx = webcl.createContext(0x1234);
ctx = webcl.createContext(WebCL.DEVICE_TYPE_ALL);
ctx = webcl.createContext({});
ctx = webcl.createContext([]);
ctx = webcl.createContext(aPlatform, 0x1234);
ctx = webcl.createContext([aValidDevice, null]);
ctx = webcl.createContext([aDeviceOnPlatformA, aDeviceOnPlatformB]);

WebCLPlatform

interface WebCLPlatform {
  any getInfo(CLenum name);
  sequence<WebCLDevice> getDevices(optional CLenum deviceType);
  sequence<DOMString>? getSupportedExtensions();
  CLboolean enableExtension(DOMString extensionName);
};
sequence<WebCLDevice> getDevices(optional CLenum deviceType) (OpenCL 1.1 §4.2, clGetDeviceIDs)

Retrieves the WebCLDevices that are available on this WebCLPlatform and match the given device type. The valid device types are listed in the table below. Omitting the device type is equivalent to specifying `DEVICE_TYPE_ALL`.

deviceTypedescription
DEVICE_TYPE_CPUA single-core or multi-core CPU, typically the host processor.
DEVICE_TYPE_GPUA graphics processing unit, typically also used by WebGL.
DEVICE_TYPE_ACCELERATORA dedicated OpenCL accelerator.
DEVICE_TYPE_DEFAULTThe default device on this platform.
DEVICE_TYPE_ALLAll devices available on this platform.
Exceptions:
any getInfo(CLenum name) (OpenCL 1.1 §4.1, clGetPlatformInfo)

Retrieves information about this WebCLPlatform. The available query parameters are listed in the table below. Note that depending on the privacy policy of the implementation, some of the return values may be empty or otherwise not reflective of the underlying OpenCL platform. Implementations that choose to mask the underlying values are recommended to make them available for privileged applications (such as browser extensions and development tools) via the System Info extension.

namereturn typeexpected return value
PLATFORM_PROFILEDOMString"WEBCL_PROFILE" + optional vendor string
PLATFORM_VERSIONDOMString"WebCL 1.0" + optional vendor string
PLATFORM_NAMEDOMStringVendor specific
PLATFORM_VENDORDOMStringVendor specific
PLATFORM_EXTENSIONSDOMStringVendor specific
Exceptions:
sequence<DOMString>? getSupportedExtensions() (OpenCL 1.1 §9, clGetPlatformInfo)
Returns an array of extension names that are supported by all WebCLDevices on this WebCLPlatform. Any string in this list, when passed to `enableExtension` on this platform, or any device on this platform, must enable the corresponding extension.
CLboolean enableExtension(DOMString extensionName)
Enables the given WebCL extension on this WebCLPlatform. Returns `true` if the extension is successfully enabled, or `false` if not. The available extension names can be queried by `getSupportedExtensions`. Note that enabling an extension does not take effect retroactively, i.e., contexts that were created before enabling the extension will continue to not have the extended capabilities.

WebCLDevice

interface WebCLDevice {
  any getInfo(CLenum name);
  sequence<DOMString>? getSupportedExtensions();
  CLboolean enableExtension(DOMString extensionName);
};
any getInfo(CLenum name) (OpenCL 1.1 §4.2, clGetDeviceInfo)

Retrieves information about this WebCLDevice. The available query parameters are listed in the table below. Note that depending on the privacy policy of the implementation, some of the return values may be empty or otherwise not reflective of the underlying OpenCL device. Implementations that choose to mask the underlying values are recommended to make them available for privileged applications (such as browser extensions and development tools) via the System Info extension.

namereturn typeexpected return value
DEVICE_PROFILEDOMString"WEBCL_PROFILE" + optional vendor string
DEVICE_VERSIONDOMString"WebCL 1.0" + optional vendor string
DEVICE_OPENCL_C_VERSIONDOMString"WebCL C 1.0" + optional vendor string
DEVICE_NAMEDOMStringVendor specific
DEVICE_VENDORDOMStringVendor specific
DEVICE_VENDOR_IDCLuintVendor specific
DRIVER_VERSIONDOMStringVendor specific
DEVICE_EXTENSIONSDOMStringVendor specific
DEVICE_AVAILABLECLbooleantrue
DEVICE_COMPILER_AVAILABLECLbooleantrue
DEVICE_TYPECLenum One or more of { DEVICE_TYPE_CPU, DEVIDE_TYPE_GPU, DEVICE_TYPE_ACCELERATOR }
DEVICE_MAX_COMPUTE_UNITSCLuint>= 1
DEVICE_MAX_WORK_ITEM_DIMENSIONSCLuint>= 3
DEVICE_MAX_WORK_ITEM_SIZESCLuint>= (1, 1, 1)
DEVICE_MAX_WORK_GROUP_SIZECLuint>= 1
DEVICE_PREFERRED_VECTOR_WIDTH_{CHAR, SHORT, INT, LONG, FLOAT}CLuint>= 1
DEVICE_NATIVE_VECTOR_WIDTH_{CHAR, SHORT, INT, LONG, FLOAT}CLuint>= 1
DEVICE_MAX_CLOCK_FREQUENCYCLuint>= 0
DEVICE_ADDRESS_BITSCLuint32 or 64
DEVICE_MAX_MEM_ALLOC_SIZECLulong>= max(DEVICE_GLOBAL_MEM_SIZE/4, 1024*1024)
DEVICE_IMAGE_SUPPORTCLbooleantrue
DEVICE_MAX_READ_IMAGE_ARGSCLuint>= 8
DEVICE_MAX_WRITE_IMAGE_ARGSCLuint>= 1
DEVICE_IMAGE2D_MAX_WIDTHCLuint>= 2048
DEVICE_IMAGE2D_MAX_HEIGHTCLuint>= 2048
DEVICE_IMAGE3D_MAX_WIDTHCLuint>= 0
DEVICE_IMAGE3D_MAX_HEIGHTCLuint>= 0
DEVICE_IMAGE3D_MAX_WIDTHCLuint>= 0
DEVICE_MAX_SAMPLERSCLuint>= 8
DEVICE_MAX_PARAMETER_SIZECLuint>= 256 bytes
DEVICE_MEM_BASE_ADDR_ALIGNCLuint>= numBits(float16) = 16*32 = 512
DEVICE_SINGLE_FP_CONFIGCLenum A combination of one or more of { FP_ROUND_TO_NEAREST, FP_ROUND_TO_ZERO } and zero or more of { FP_DENORM, FP_INF_NAN, FP_ROUND_TO_INF, FP_FMA, FP_SOFT_FLOAT }.
DEVICE_GLOBAL_MEM_CACHE_TYPECLenumExactly one of { NONE, READ_ONLY_CACHE, READ_WRITE_CACHE }
DEVICE_GLOBAL_MEM_CACHELINE_SIZECLuint>= 0 bytes
DEVICE_GLOBAL_MEM_CACHE_SIZECLulong>= 0 bytes
DEVICE_GLOBAL_MEM_SIZECLulong>= 1024*1024 bytes
DEVICE_MAX_CONSTANT_BUFFER_SIZECLulong>= 1024 bytes
DEVICE_MAX_CONSTANT_ARGSCLuint>= 4
DEVICE_LOCAL_MEM_TYPECLenumExactly one of { LOCAL, GLOBAL }
DEVICE_LOCAL_MEM_SIZECLulong>= 1024 bytes
DEVICE_ERROR_CORRECTION_SUPPORTCLbooleantrue/false
DEVICE_HOST_UNIFIED_MEMORYCLbooleantrue/false
DEVICE_PROFILING_TIMER_RESOLUTIONCLuint>= 0
DEVICE_ENDIAN_LITTLECLbooleantrue/false
DEVICE_EXECUTION_CAPABILITIESCLenumEXEC_KERNEL
DEVICE_QUEUE_PROPERTIESCLenumZero or more of { QUEUE_PROFILING_ENABLE, QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE }
DEVICE_PLATFORMWebCLPlatformObject
Exceptions:
sequence<DOMString>? getSupportedExtensions() (OpenCL 1.1 §9, clGetDeviceInfo)
Returns an array of extension names that are supported by this WebCLDevice. Any string in this list, when passed to `enableExtension`, must enable the corresponding extension.
CLboolean enableExtension(DOMString extensionName)
Enables the given WebCL extension on this WebCLDevice. Returns `true` if the extension is successfully enabled, or `false` if not. The available extension names can be queried by `getSupportedExtensions`. Note that enabling an extension does not take effect retroactively, i.e., contexts that were created before enabling the extension will continue to not have the extended capabilities.

WebCLContext

interface WebCLContext {

  WebCLBuffer createBuffer(CLenum memFlags, CLuint sizeInBytes, optional ArrayBufferView hostPtr);

  WebCLCommandQueue createCommandQueue(optional WebCLDevice? device, optional CLenum properties);

  WebCLImage createImage(CLenum memFlags,
                         WebCLImageDescriptor descriptor,
                         optional ArrayBufferView hostPtr);

  WebCLProgram createProgram(DOMString source);

  WebCLSampler createSampler(CLboolean normalizedCoords,
                             CLenum addressingMode,
                             CLenum filterMode);

  WebCLUserEvent createUserEvent();

  any getInfo(CLenum name);

  sequence<WebCLImageDescriptor>? getSupportedImageFormats(optional CLenum memFlags = MEM_READ_WRITE);

  void release();

  void releaseAll();
};
WebCLBuffer createBuffer(CLenum memFlags, CLuint sizeInBytes, optional ArrayBufferView hostPtr) (OpenCL 1.1 §5.2.1, clCreateBuffer)
Creates a new WebCLBuffer with a capacity of `sizeInBytes`, and optionally fills it with data from the given ArrayBufferView.
Exceptions:
WebCLCommandQueue createCommandQueue(optional WebCLDevice? device = null, optional CLenum properties = 0) (OpenCL 1.1 §5.1, clCreateCommandQueue)
Creates a new command queue for the given device. If `device` is `null`, the WebCL implementation will select any WebCLDevice that matches the given `properties` and is covered by this WebCLContext. If `properties` is omitted, the command queue is created with out-of-order execution disabled and profiling disabled.
Exceptions:
WebCLImage createImage(CLenum memFlags, WebCLImageDescriptor descriptor, optional ArrayBufferView hostPtr = null) (OpenCL 1.1 §5.3.1, clCreateImage2D)
Creates a new WebCLImage with the width and height specified in `descriptor`, and optionally fills it with data from the given ArrayBufferView.
Exceptions:
WebCLProgram createProgram(DOMString source) (OpenCL 1.1 §5.6.1, clCreateProgramWithSource)
Creates a WebCLProgram from a UTF-8 string.
Exceptions:
WebCLSampler createSampler(CLboolean normalizedCoords, CLenum addressingMode, CLenum filterMode) (OpenCL 1.1 §5.5.1, clCreateSampler)
Creates a new WebCLSampler with the given coordinate mode, addressing mode, and filtering mode. The `ADDRESS_NONE` mode is not available, due to its lack of bounds checking.
Exceptions:
WebCLUserEvent createUserEvent() (OpenCL 1.1 §5.9, clCreateUserEvent)
any getInfo(CLenum name) (OpenCL 1.1 §4.3, clGetContextInfo)

Returns the value for the passed name. The type returned is the natural type for the requested name, as given in the table below.

namereturn typedescription
CONTEXT_NUM_DEVICESCLuintthe number of devices in this context
CONTEXT_DEVICESsequence<WebCLDevice>the devices in this context
Exceptions:
sequence<WebCLImageDescriptor>? getSupportedImageFormats(CLenum memFlags) (OpenCL 1.1 §5.3.2, clGetSupportedImageFormats)

Returns a list of image formats that are supported by this WebCLContext with the given `memFlags`. For example, passing in `MEM_READ_WRITE` will return a list of image formats that are available for both reading and writing in kernel code. The returned WebCLImageDescriptor objects will only have the `channelOrder` and `channelType` fields filled in; the other fields will have their default value (zero).

The minimum set of image formats that must be supported (for both reading and writing) by all devices is defined in the table below. Note that the set of required formats is the same as on OpenCL 1.1 Embedded Profile.

channelOrderchannelType
RGBA UNORM_INT8
UNORM_INT16
SIGNED_INT8
SIGNED_INT16
SIGNED_INT32
UNSIGNED_INT8
UNSIGNED_INT16
UNSIGNED_INT32
HALF_FLOAT
FLOAT
Exceptions:
void release() (OpenCL 1.1 §4.3, clReleaseContext)
Releases the resources held up by this object.
void releaseAll() (OpenCL 1.1 §4.3, clReleaseContext)
Releases the resources held up by this WebCLContext and all objects created from it, if any.

WebCLImageDescriptor

dictionary WebCLImageDescriptor {
  CLenum channelOrder = 0x10B5;            // 0x10B5 == WebCL.RGBA
  CLenum channelType = 0x10D2;             // 0x10D2 == WebCL.UNORM_INT8, 8-bit colors normalized to [0, 1]
  CLuint width = 0, height = 0;
  CLuint rowPitch = 0;
};

WebCLCommandQueue

interface WebCLCommandQueue {

  ////////////////////////////////////////////////////////////////////////////
  //
  // Copying: Buffer <-> Buffer, Image <-> Image, Buffer <-> Image
  //

  void enqueueCopyBuffer(
                    WebCLBuffer                           srcBuffer,
                    WebCLBuffer                           dstBuffer,
                    CLuint                                srcOffset,
                    CLuint                                dstOffset,
                    CLuint                                numBytes,
                    optional sequence<WebCLEvent>?        eventWaitList,
                    optional WebCLEvent?                  event);

  void enqueueCopyBufferRect(
                    WebCLBuffer                           srcBuffer,
                    WebCLBuffer                           dstBuffer,
                    sequence<CLuint>                      srcOrigin,
                    sequence<CLuint>                      dstOrigin,
                    sequence<CLuint>                      region,
                    CLuint                                srcRowPitch,
                    CLuint                                srcSlicePitch,
                    CLuint                                dstRowPitch,
                    CLuint                                dstSlicePitch,
                    optional sequence<WebCLEvent>?        eventWaitList,
                    optional WebCLEvent?                  event);

  void enqueueCopyImage(
                    WebCLImage                            srcImage,
                    WebCLImage                            dstImage,
                    sequence<CLuint>                      srcOrigin,
                    sequence<CLuint>                      dstOrigin,
                    sequence<CLuint>                      region,
                    optional sequence<WebCLEvent>?        eventWaitList,
                    optional WebCLEvent?                  event);

  void enqueueCopyImageToBuffer(
                    WebCLImage                            srcImage,
                    WebCLBuffer                           dstBuffer,
                    sequence<CLuint>                      srcOrigin,
                    sequence<CLuint>                      srcRegion,
                    CLuint                                dstOffset,
                    optional sequence<WebCLEvent>?        eventWaitList,
                    optional WebCLEvent?                  event);

  void enqueueCopyBufferToImage(
                    WebCLBuffer                           srcBuffer,
                    WebCLImage                            dstImage,
                    CLuint                                srcOffset,
                    sequence<CLuint>                      dstOrigin,
                    sequence<CLuint>                      dstRegion,
                    optional sequence<WebCLEvent>?        eventWaitList,
                    optional WebCLEvent?                  event);

  ////////////////////////////////////////////////////////////////////////////
  //
  // Reading: Buffer -> Host, Image -> Host
  //

  void enqueueReadBuffer(
                    WebCLBuffer                           buffer,
                    CLboolean                             blockingRead,
                    CLuint                                bufferOffset,
                    CLuint                                numBytes,
                    ArrayBufferView                       hostPtr,
                    optional sequence<WebCLEvent>?        eventWaitList,
                    optional WebCLEvent?                  event);

  void enqueueReadBufferRect(
                    WebCLBuffer                           buffer,
                    CLboolean                             blockingRead,
                    sequence<CLuint>                      bufferOrigin,
                    sequence<CLuint>                      hostOrigin,
                    sequence<CLuint>                      region,
                    CLuint                                bufferRowPitch,
                    CLuint                                bufferSlicePitch,
                    CLuint                                hostRowPitch,
                    CLuint                                hostSlicePitch,
                    ArrayBufferView                       hostPtr,
                    optional sequence<WebCLEvent>?        eventWaitList,
                    optional WebCLEvent?                  event);

  void enqueueReadImage(
                    WebCLImage                            image,
                    CLboolean                             blockingRead,
                    sequence<CLuint>                      origin,
                    sequence<CLuint>                      region,
                    CLuint                                hostRowPitch,
                    ArrayBufferView                       hostPtr,
                    optional sequence<WebCLEvent>?        eventWaitList,
                    optional WebCLEvent?                  event);

  ////////////////////////////////////////////////////////////////////////////
  //
  // Writing: Host -> Buffer, Host -> Image
  //

  void enqueueWriteBuffer(
                    WebCLBuffer                           buffer,
                    CLboolean                             blockingWrite,
                    CLuint                                bufferOffset,
                    CLuint                                numBytes,
                    ArrayBufferView                       hostPtr,
                    optional sequence<WebCLEvent>?        eventWaitList,
                    optional WebCLEvent?                  event);

  void enqueueWriteBufferRect(
                    WebCLBuffer                           buffer,
                    CLboolean                             blockingWrite,
                    sequence<CLuint>                      bufferOrigin,
                    sequence<CLuint>                      hostOrigin,
                    sequence<CLuint>                      region,
                    CLuint                                bufferRowPitch,
                    CLuint                                bufferSlicePitch,
                    CLuint                                hostRowPitch,
                    CLuint                                hostSlicePitch,
                    ArrayBufferView                       hostPtr,
                    optional sequence<WebCLEvent>?        eventWaitList,
                    optional WebCLEvent?                  event);

  void enqueueWriteImage(
                    WebCLImage                            image,
                    CLboolean                             blockingWrite,
                    sequence<CLuint>                      origin,
                    sequence<CLuint>                      region,
                    CLuint                                hostRowPitch,
                    ArrayBufferView                       hostPtr,
                    optional sequence<WebCLEvent>?        eventWaitList,
                    optional WebCLEvent?                  event);

  ////////////////////////////////////////////////////////////////////////////
  //
  // Executing kernels
  //

  void enqueueNDRangeKernel(
                    WebCLKernel                           kernel,
                    CLuint                                workDim,
                    sequence<CLuint>?                     globalWorkOffset, 
                    sequence<CLuint>                      globalWorkSize,
                    optional sequence<CLuint>?            localWorkSize,
                    optional sequence<WebCLEvent>?        eventWaitList,
                    optional WebCLEvent?                  event);

  ////////////////////////////////////////////////////////////////////////////
  //
  // Synchronization
  //

  void enqueueMarker(WebCLEvent event);

  void enqueueBarrier();

  void enqueueWaitForEvents (sequence<WebCLEvent> eventWaitList);
  
  void finish(optional WebCLCallback whenFinished);

  void flush();

  ////////////////////////////////////////////////////////////////////////////
  //
  // Querying command queue information
  //

  any getInfo(CLenum name);

  void release();
};
void enqueueCopyBuffer( WebCLBuffer srcBuffer, WebCLBuffer dstBuffer, CLuint srcOffset, CLuint dstOffset, CLuint numBytes, optional sequence<WebCLEvent>? eventWaitList, optional WebCLEvent? event) (OpenCL 1.1 §5.2.2, clEnqueueCopyBuffer)
Exceptions:
void enqueueCopyBufferRect( WebCLBuffer srcBuffer, WebCLBuffer dstBuffer, sequence<CLuint> srcOrigin, sequence<CLuint> dstOrigin, sequence<CLuint> region, CLuint srcRowPitch, CLuint srcSlicePitch, CLuint dstRowPitch, CLuint dstSlicePitch, optional sequence<WebCLEvent>? eventWaitList, optional WebCLEvent? event); (OpenCL 1.1 §5.2.2, clEnqueueCopyBufferRect)
Exceptions:
void enqueueCopyImage( WebCLImage srcImage, WebCLImage dstImage, sequence<CLuint> srcOrigin, sequence<CLuint> dstOrigin, sequence<CLuint> region, optional sequence<WebCLEvent>? eventWaitList, optional WebCLEvent? event) (OpenCL 1.1 §5.2.2, clEnqueueCopyImage)
Exceptions:
void enqueueCopyBufferToImage( WebCLBuffer srcBuffer, WebCLImage dstImage, CLuint srcOffset, sequence<CLuint> dstOrigin, sequence<CLuint> dstRegion, optional sequence<WebCLEvent>? eventWaitList, optional WebCLEvent? event) (OpenCL 1.1 §5.3.4, clEnqueueCopyBufferToImage)
Exceptions:
void enqueueCopyImageToBuffer( WebCLImage srcImage, WebCLBuffer dstBuffer, sequence<CLuint> srcOrigin, sequence<CLuint> srcRegion, CLuint dstOffset, optional sequence<WebCLEvent>? eventWaitList, optional WebCLEvent? event) (OpenCL 1.1 §5.3.4, clEnqueueCopyImageToBuffer)
Exceptions:
void enqueueReadBuffer( WebCLBuffer buffer, CLboolean blockingRead, CLuint bufferOffset, CLuint numBytes, ArrayBufferView hostPtr, optional sequence<WebCLEvent>? eventWaitList, optional WebCLEvent? event) (OpenCL 1.1 §5.2.2, clEnqueueReadBuffer)
Exceptions:
void enqueueReadBufferRect( WebCLBuffer buffer, CLboolean blockingRead, sequence<CLuint> bufferOrigin, sequence<CLuint> hostOrigin, sequence<CLuint> region, CLuint bufferRowPitch, CLuint bufferSlicePitch, CLuint hostRowPitch, CLuint hostSlicePitch, ArrayBufferView hostPtr, optional sequence<WebCLEvent>? eventWaitList, optional WebCLEvent? event) (OpenCL 1.1 §5.2.2, clEnqueueReadBufferRect)
Exceptions:
void enqueueReadImage( WebCLImage image, CLboolean blockingRead, sequence<CLuint> origin, sequence<CLuint> region, CLuint hostRowPitch, ArrayBufferView hostPtr, optional sequence<WebCLEvent>? eventWaitList, optional WebCLEvent? event) (OpenCL 1.1 §5.2.2, clEnqueueReadImage)
Exceptions:
void enqueueWriteBuffer( WebCLBuffer buffer, CLboolean blockingWrite, CLuint bufferOffset, CLuint numBytes, ArrayBufferView hostPtr, optional sequence<WebCLEvent>? eventWaitList, optional WebCLEvent? event) (OpenCL 1.1 §5.2.2, clEnqueueWriteBuffer)
Exceptions:
void enqueueWriteBufferRect( WebCLBuffer buffer, CLboolean blockingWrite, sequence<CLuint> bufferOrigin, sequence<CLuint> hostOrigin, sequence<CLuint> region, CLuint bufferRowPitch, CLuint bufferSlicePitch, CLuint hostRowPitch, CLuint hostSlicePitch, ArrayBufferView hostPtr, optional sequence<WebCLEvent>? eventWaitList, optional WebCLEvent? event) (OpenCL 1.1 §5.2.2, clEnqueueWriteBufferRect)
Exceptions:
void enqueueWriteImage( WebCLImage image, CLboolean blockingWrite, sequence<CLuint> origin, sequence<CLuint> region, CLuint hostRowPitch, ArrayBufferView hostPtr, optional sequence<WebCLEvent>? eventWaitList, optional WebCLEvent? event) (OpenCL 1.1 §5.2.2, clEnqueueWriteImage)
Exceptions:
void enqueueNDRangeKernel( WebCLKernel kernel, CLuint workDim, sequence<CLuint>? globalWorkOffset, sequence<CLuint> globalWorkSize, optional sequence<CLuint>? localWorkSize = null, optional sequence<WebCLEvent>? eventWaitList = null, optional WebCLEvent? event = null) (OpenCL 1.1 §5.8, clEnqueueNDRangeKernel)
Exceptions:
void enqueueMarker(WebCLEvent event) (OpenCL 1.1 §5.10, clEnqueueMarker)
Exceptions:
void enqueueBarrier() (OpenCL 1.1 §5.10, clEnqueueBarrier)
void enqueueWaitForEvents (sequence<WebCLEvent> eventWaitList); (OpenCL 1.1 §5.10, clEnqueueWaitForEvents)
Exceptions:
void finish(optional WebCLCallback whenFinished) (OpenCL 1.1 §5.13, clFinish)
Invokes the given callback when all commands in this queue have completed. If a callback is not provided, the JavaScript main thread will be blocked until the commands have completed. Applications are strongly advised to provide a callback; the blocking mode mainly exists to ease the porting of existing OpenCL code.
Exceptions:
void flush() (OpenCL 1.1 §5.13, clFlush)
any getInfo(CLenum name) (OpenCL 1.1 §5.1, clGetCommandQueueInfo)
namereturn type
QUEUE_CONTEXTWebCLContext
QUEUE_DEVICEWebCLDevice
QUEUE_PROPERTIESCLenum
Exceptions:
void release() (OpenCL 1.1 §5.1, clReleaseCommandQueue)
Releases the resources held up by this object.
// This example shows how to wait for two kernels to complete before a
// third kernel executes

var eventWaitList = [ new WebCLEvent(), new WebCLEvent() ];
queue.enqueueNDRangeKernel(kernel1, dim, null, globals, null, null, eventWaitList[0]);
queue.enqueueNDRangeKernel(kernel2, dim, null, globals, null, null, eventWaitList[1]);
queue.enqueueNDRangeKernel(kernel3, dim, null, globals, null, eventWaitList);
// This example reads the entire contents of a WebCL buffer object
// into a newly created Uint8Array in host memory.

function readBufferToHost(srcBuffer) {

 // Query the number of bytes in the source buffer, create a new
 // Uint8Array of that size, then fill it with a blocking read.

 var numBytes = srcBuffer.getInfo(WebCL.MEM_SIZE);
 var dstArray = new Uint8Array(numBytes);
 queue.enqueueReadBuffer(srcBuffer, true, 0, numBytes, dstArray);

 return dstArray;
};
// This example fills a WebCL image with pixels from a WebCL buffer.

function copyBufferToImage(srcBuffer, dstImage) {

  // Extract the dimensions of the image.

  var imgWidth = dstImage.getInfo(WebCL.IMAGE_WIDTH);
  var imgHeight = dstImage.getInfo(WebCL.IMAGE_HEIGHT);
  queue.enqueueCopyBufferToImage(srcBuffer, dstImage, 0, [0,0], [imgWidth, imgHeight]);
};
// This example copies a rectangular region of memory from a
// buffer object to another.  The buffer objects are assumed
// to be the same size. The origin, width, height, and pitch
// (a.k.a. row stride) are given in bytes (not pixels, because
// these are buffers rather than images).

function copyBufferRect(srcBuffer, dstBuffer, srcX, srcY, dstX, dstY, w, h, pitch) {

  // The row pitch could be different for source and destination,
  // but is assumed to be the same in this example.

  queue.enqueueCopyBufferRect(srcBuffer, dstBuffer, [srcX, srcY], [dstX, dstY], 
                              [w, h], pitch, 0, pitch, 0);
};

WebCLMemoryObject

interface WebCLMemoryObject {
  any getInfo(CLenum name);
  void release();
};
any getInfo(CLenum name) (OpenCL 1.1 §5.4.1, clGetMemObjectInfo)
namereturn typereturn value
`MEM_TYPE``CLenum``MEM_OBJECT_BUFFER` or `MEM_OBJECT_IMAGE2D`
`MEM_FLAGS``CLenum`The `memFlags` as specified at construction time
`MEM_SIZE``CLuint`The size of this memory object in bytes
`MEM_CONTEXT``WebCLContext`the WebCL context of this memory object
`MEM_ASSOCIATED_MEMOBJECT``WebCLBuffer`The buffer object that this buffer was created from, or `null` if this buffer was not created using `createSubBuffer`
`MEM_OFFSET``CLuint`The offset given to `createSubBuffer`, or zero if this buffer was not created using `createSubBuffer`
Exceptions:
void release() (OpenCL 1.1 §5.4.1, clReleaseMemObject)
Releases the resources held up by this object.

WebCLBuffer

interface WebCLBuffer : WebCLMemoryObject {
  WebCLBuffer createSubBuffer(CLenum memFlags, CLuint origin, CLuint sizeInBytes);
};
WebCLBuffer createSubBuffer(CLenum memFlags, CLuint origin, CLuint sizeInBytes) (OpenCL 1.1 §5.2.1, clCreateSubBuffer)
Creates a new WebCLBuffer that represents a sub-region of this WebCLBuffer. The two buffers reference the same area of global memory, so changes made into one are immediately reflected in the other.
Exceptions:

WebCLImage

interface WebCLImage : WebCLMemoryObject {
  WebCLImageDescriptor getInfo();
};
WebCLImageDescriptor getInfo() (OpenCL 1.1 §5.3.6, clGetImageInfo)
Returns a new WebCLImageDescriptor containing information about the pixel format and dimensions of this WebCLImage.

WebCLSampler

interface WebCLSampler {
  any getInfo(CLenum name);
  void release();
};
any getInfo(CLenum name) (OpenCL 1.1 §5.5.2, clGetSamplerInfo)
namereturn type
SAMPLER_CONTEXTWebCLContext
SAMPLER_NORMALIZED_COORDSCLboolean
SAMPLER_ADDRESSING_MODECLenum
SAMPLER_FILTER_MODECLenum
Exceptions:
void release() (OpenCL 1.1 §5.5.1, clReleaseSampler)
Releases the resources held up by this object.

WebCLProgram

interface WebCLProgram {
  any getInfo(CLenum name);

  any getBuildInfo(WebCLDevice device, CLenum name);

  void build(optional sequence<WebCLDevice>? devices,
             optional DOMString? options,
             optional WebCLCallback whenFinished);

  WebCLKernel createKernel(DOMString kernelName);

  sequence<WebCLKernel> createKernelsInProgram();

  void release();
};
any getInfo(CLenum name) (OpenCL 1.1 §5.6.5, clGetProgramInfo)
namereturn type
PROGRAM_CONTEXTWebCLContext
PROGRAM_NUM_DEVICESCLuint
PROGRAM_DEVICESsequence<WebCLDevice>
PROGRAM_SOURCEDOMString
Exceptions:
any getBuildInfo(WebCLDevice device, CLenum name) (OpenCL 1.1 §5.6.5, clGetProgramBuildInfo)
namereturn type
PROGRAM_BUILD_STATUSCLint
PROGRAM_BUILD_OPTIONSDOMString
PROGRAM_BUILD_LOGDOMString
Exceptions:
void build(optional sequence<WebCLDevice>? devices, optional DOMString? options, optional WebCLCallback whenFinished) (OpenCL 1.1 §5.6.2, clBuildProgram)

Compiles this WebCLProgram for the given list of devices, or in absence of the list, for all devices associated with the WebCLContext that this WebCLProgram was created from. A string of compiler options and an asynchronous callback function to invoke when the build is completed can also be provided. Applications are strongly advised to provide a callback, to avoid blocking the JavaScript main thread. The available compiler options are listed in the table below; any other option is considered invalid.

Build optionDescription
-D nameEquivalent to #define name in OpenCL C.
-D name=definitionEquivalent to #define name definition in OpenCL C.
-cl-opt-disableDisable all optimizations.
-cl-single-precision-constantTreat double-precision constants as single-precision.
-cl-denorms-are-zeroAllow denormalized numbers to be flushed to zero.
-cl-mad-enableAllow a*b+c to be computed with potentially reduced accuracy.
-cl-no-signed-zerosAllow optimizations that ignore the signedness of zero.
-cl-unsafe-math-optimizationsAllow optimizations that may violate IEEE 754 and OpenCL numerical compliance requirements. This option includes -cl-no-signed-zeros and cl-mad-enable.
-cl-finite-math-onlyAllow optimizations that ignore NaNs and infinities.
-cl-fast-relaxed-mathEquivalent to -cl-finite-math-only and -cl-unsafe-math-optimizations.
-wInhibit all warning messages.
-WerrorMake all warnings into errors.
Exceptions:
WebCLKernel createKernel(DOMString name) (OpenCL 1.1 §5.7.1, clCreateKernel)
Exceptions:
sequence<WebCLKernel> createKernelsInProgram() (OpenCL 1.1 §5.7.1, clCreateKernelsInProgram)
Exceptions:
void release() (OpenCL 1.1 §5.6.1, clReleaseProgram)
Releases the resources held up by this object.

WebCLKernel

The following methods are available for setting kernel arguments and querying kernel-specific information.

interface WebCLKernel {
  any getInfo(CLenum name);
  any getWorkGroupInfo(WebCLDevice device, CLenum name);
  WebCLKernelArgInfo getArgInfo(CLuint index);
  void setArg(CLuint index, WebCLMemoryObject value);
  void setArg(CLuint index, WebCLSampler value);
  void setArg(CLuint index, ArrayBufferView value);
  void release();
};
any getInfo(CLenum name) (OpenCL 1.1 §5.7.3, clGetKernelInfo)
namereturn type
KERNEL_FUNCTION_NAMEDOMString
KERNEL_NUM_ARGSCLuint
KERNEL_CONTEXTWebCLContext
KERNEL_PROGRAMWebCLProgram
Exceptions:
any getWorkGroupInfo(WebCLDevice device, CLenum name) (OpenCL 1.1 §5.7.3, clGetKernelWorkGroupInfo)

Returns the value corresponding to the given `device` and `name`.

namereturn type
KERNEL_WORK_GROUP_SIZECLuint
KERNEL_COMPILE_WORK_GROUP_SIZEsequence<CLuint>
KERNEL_LOCAL_MEM_SIZECLuint
KERNEL_PREFERRED_WORK_GROUP_SIZE_MULTIPLECLuint
KERNEL_PRIVATE_MEM_SIZECLuint
Exceptions:
WebCLKernelArgInfo getArgInfo(CLuint index) (OpenCL 1.2 §5.7.3, clGetKernelArgInfo)
Retrieves the name, type, and qualífiers of the kernel argument at the given `index`. Legal values for `index` range from zero to `N-1`, where `N` is the total number of arguments declared by the kernel.
Exceptions:
void setArg(CLuint index, WebCLMemoryObject memObject) (OpenCL 1.1 §5.7.2, clSetKernelArg)

Sets the given `memObject` to the kernel argument at the given `index`. Legal values for `index` range from zero to `N-1`, where `N` is the total number of arguments declared by the kernel.

Exceptions:
void setArg(CLuint index, WebCLSampler sampler) (OpenCL 1.1 §5.7.2, clSetKernelArg)

Sets the given `sampler` to the kernel argument at the given `index`. Legal values for `index` range from zero to `N-1`, where `N` is the total number of arguments declared by the kernel.

Exceptions:
void setArg(CLuint index, ArrayBufferView value) (OpenCL 1.1 §5.7.2, clSetKernelArg)

Sets the given `value` to the kernel argument at the given `index`. Legal values for `index` range from zero to `N-1`, where `N` is the total number of arguments declared by the kernel.

The type and length of `value` must match the base type and vector width of the target kernel argument. For example, if the kernel argument is of type `float4`, then `value` must be a Float32Array of length 4. Arguments with the `local` address space qualifier must be set using a Uint32Array of length 1, specifying the number of bytes to be allocated.

Since there is no 64-bit integer variant of ArrayBufferView, 64-bit integers must be represented as pairs of 32-bit unsigned integers. The low-order 32 bits are stored in the first element of each pair, and the high-order 32 bits in the second element.

Exceptions:
void release() (OpenCL 1.1 §5.7.1, clReleaseKernel)
Releases the resources held up by this object.
myKernel.setArg(0, 3.14159);                                  // ERROR: Passing in a Number is not allowed
myKernel.setArg(0, new Float32Array([3.14159]));              // cast 3.14159 to `float`, then pass to kernel as arg #0
myKernel.setArg(1, new Uint32Array([1.23, 2.34]));            // cast the numbers to `uint2`, then pass to kernel as arg #1
myKernel.setArg(2, new Uint32Array([512]));                   // reserve 512 bytes of local memory for arg #2
myKernel.setArg(3, myImage);                                  // pass `myImage` to kernel as arg #3

WebCLKernelArgInfo

dictionary WebCLKernelArgInfo {
  DOMString name;
  DOMString typeName;         // 'char', 'float', 'uint4', 'image2d_t', 'sampler_t', etc.
  DOMString addressQualifier; // 'global', 'local', 'constant', or 'private'
  DOMString accessQualifier;  // 'read_only', 'write_only', or 'none'
};
DOMString name
The name of the kernel argument; for example, srcPixels or foobarCoefficient.
DOMString typeName (OpenCL 1.1 §6.1, Data types)
The data type of the kernel argument in normalized form; for example, uint4 or image2d_t.
DOMString addressQualifier (OpenCL 1.1 §6.5, Address space qualifiers)
The address space qualifier of the kernel argument in normalized form (that is, not prefixed with underscores). The possible values are global, local, constant, and private.
DOMString accessQualifier (OpenCL 1.1 §6.6, Access qualifiers)
The access qualifier of the kernel argument in normalized form (that is, not prefixed with underscores). If `typeName` is image2d_t, the possible values are read_only and write_only. Otherwise, the value is none.

WebCLEvent

[Constructor]
interface WebCLEvent {
  any   getInfo(CLenum name);
  any   getProfilingInfo(CLenum name);
  void  setCallback(CLenum commandExecCallbackType, WebCLCallback notify);
  void  release();
};
WebCLEvent()

Creates an empty WebCLEvent instance, to be populated by one of the `enqueue` methods. The properties of a newly created event are set as specified below. Each WebCLEvent instance can only be used once; the `enqueue` methods will throw an `INVALID_EVENT` exception if a previously populated event is passed in. User events are considered previously populated, and as such cannot be re-populated.

namereturn type
EVENT_COMMAND_QUEUEundefined
EVENT_CONTEXTundefined
EVENT_COMMAND_TYPEundefined
EVENT_COMMAND_EXECUTION_STATUS-1
any getInfo(CLenum name) (OpenCL 1.1 §5.9, clGetEventInfo)
namereturn type
EVENT_COMMAND_QUEUEWebCLCommandQueue
EVENT_CONTEXTWebCLContext
EVENT_COMMAND_TYPECLenum
EVENT_COMMAND_EXECUTION_STATUSCLint
Exceptions:
any getProfilingInfo(CLenum name) (OpenCL 1.1 §5.12, clGetEventProfilingInfo)
namereturn type
PROFILING_COMMAND_QUEUEDCLuint
PROFILING_COMMAND_SUBMITCLuint
PROFILING_COMMAND_STARTCLuint
PROFILING_COMMAND_ENDCLuint
Exceptions:
void setCallback(CLenum commandExecCallbackType, WebCLCallback notify) (OpenCL 1.1 §5.9, clSetEventCallback)
Sets a WebCLCallback function `notify` to be called when this WebCLEvent reaches the given `commandExecCallbackType` status (`COMPLETE`, `RUNNING`, `SUBMITTED`, or `QUEUED`). In this version of WebCL, only the `COMPLETE` status is supported.
Exceptions:
void release() (OpenCL 1.1 §5.9, clReleaseEvent)
Releases the resources held up by this object.

WebCLUserEvent

interface WebCLUserEvent : WebCLEvent {
  void  setStatus(CLint executionStatus);
};
void setStatus(CLint executionStatus) (OpenCL 1.1 §5.9, clSetUserEventStatus)
Exceptions:

Extensions

WebCL implementations may expose a number of extensions to the core functionality. The extensions are listed and documented in the WebCL Extension Registry. Each extension, regardless of its standardization status, must be explicitly enabled by the application before it can be used. This policy is aimed to minimize inadvertant usage of features that are not universally available.

The set of available extensions may vary from one WebCLDevice to another, and one WebCLPlatform to another, even within the same system. For example, the WebGL Resource Sharing extension may be supported by the GPU but not the CPU. The names of extensions that are supported system-wide can be queried with `webcl.getSupportedExtensions()`. Similarly, calling `getSupportedExtensions` on a particular WebCLPlatform returns the set of extensions that are supported by all devices on that platform. Finally, `getSupportedExtensions()` on an individual WebCLDevice returns the set of extensions that are supported by that device.

WebCL extensions can be enabled by calling `enableExtension` with the name of the desired extension as a string parameter. This returns `true` if the extension is successfully enabled, and `false` if not. This is illustrated in the following examples.

// Iterates over all WebCL devices and returns the ones that
// support the given extension, with that extension enabled.
//
function getDevicesWithExtensionEnabled(extensionName) {
  var devices = [];
  webcl.getPlatforms().forEach(function(platform) {
    platform.getDevices().forEach(function(device) {
      var isSupported = device.enableExtension(extensionName);
      if (isSupported) devices.push(device);
    });
  });
  return devices;
}
// Use the above function find the set of devices
// that support both KHR_fp64 and KHR_gl_sharing.
//
var withFP64 = getDevicesWithExtensionEnabled("KHR_fp64");
var withGL = getDevicesWithExtensionEnabled("KHR_gl_sharing");
var withBoth = withGL.filter(function(v) { return (withFP64.indexOf(v) !== -1) });

Security and Robustness

WebCL has been designed with security as a primary concern. This section defines the main security requirements that a conformant WebCL implementation must fulfill.

Out-of-Bounds Memory Access

WebCL kernels must not be able to access unauthorized areas of memory, regardless of address space (private, local, global, or constant). If detected during compilation, out-of-bounds (OOB) accesses must generate a compiler error. At runtime, OOB reads must return zero and writes must be discarded. Alternatively, the kernel may be terminated and an exception be thrown.

For purposes of bounds checking, the implementation may treat all `private` variables as one contiguous block of memory, and similarly for `local` variables, instead of enforcing the bounds of each variable separately. For example, in a kernel program containing two `private` arrays, an OOB read from the first array is allowed to return any value from either the first or the second array, rather than zero.

The WebCL Validator open source project was initiated by the WebCL working group for prevention of out-of-bounds memory accesses through instrumentation, analysis and validation of WebCL kernels. The Validator also enforces initialization of local and private memory (Khronos WebCL Validator Project).

Memory Initialization

To ensure that applications cannot inspect data left behind by previous applications, the WebCL implementation must initialize all buffers and variables to zero before giving the application read access to them. This requirement applies regardless of address space (private, local, global, or constant), and regardless of whether the allocation is made in kernel code or host code.

Where available, the OpenCL 1.2 extension `cl_khr_initialize_memory` allows WebCL implementations to automatically and efficiently initialize local and private memory before a kernel begins execution (OpenCL 1.2 Extensions Specification).

Denial of Service

Long-running and/or computationally intensive kernels (or other commands in a command queue) may cause the system to become unresponsive by disproportionately consuming device resources. It is generally not possible to guard against this problem at the browser/WebCL level; the necessary mechanisms, such as watchdog timers and pre-emptive multitasking, must be provided by the OpenCL driver and the operating system. On systems where the necessary facilities are in place, WebCL implementations are strongly encouraged to make use of them in order to:

  1. Detect offending kernels and other commands. A command is considered offending if it runs for an excessively long time, or disproportionately consumes system resources.
  2. Terminate the contexts associated with the offending commands, before they render the OpenCL device unresponsive, and result in potential device reset.

Where available, the OpenCL 1.2 extension `cl_khr_terminate_context` can be used for fast termination of a context, if, for example, one or more of the kernels associated with the context has been running too long, or if the program terminates due to an exception (OpenCL 1.2 Extensions Specification).

Implementation notes

Although WebCL is based on OpenCL 1.1 (Embedded Profile), it can also be implemented on top of OpenCL 1.2, or any future version that remains backwards compatible with OpenCL 1.1. However, an implementation running on top of OpenCL 1.2 or later must make sure to not allow applications to use any features that are not supported in WebCL. This implies, in particular, that implementations must not pass kernel code to the underlying OpenCL compiler without first checking that the kernel code is valid according to WebCL. For example, usage of `printf()` must be treated as a compilation error, because `printf` is not supported in WebCL.

The WebCL specification and conformance tests are written with the assumption that implementations will be based on conformant OpenCL drivers. Implementations are strongly recommended to support dynamic black-listing of malfunctioning or vulnerable OpenCL drivers in order to quickly eliminate any security or stability issues.

WebCL objects are regular JavaScript objects, and must behave accordingly. For example, the application must be able attach arbitrary properties and functions to them, and if the same object is retrieved twice, the instances returned must be equal according to the `===` operator. For example, the assertions in the following code example must be true:

  var ctx = webcl.createContext();
  ctx.name = "myContext";
  var queue = ctx.createCommandQueue();
  var ctx2 = queue.getInfo(WebCL.QUEUE_CONTEXT);
  ASSERT(ctx2 === ctx);
  ASSERT(ctx2.name === "myContext");

Differences between WebCL and OpenCL 1.1

This section describes changes made to the WebCL API and the WebCL C kernel programming language, relative to OpenCL 1.1 Embedded Profile and OpenCL C. The differences are as follows:

References

Normative references

[OPENCL11]
OpenCL 1.1 Specification, A. Munshi.
[WEBIDL]
Web IDL: W3C Candidate Recommendation, C. McCormack.
[WEBGL]
WebGL 1.0 Specification, C. Marrin.

Acknowledgments

This specification is produced by the Khronos WebCL Working Group.

Main Contributors (in alphabetical order):
Tomi Aarnio (Editor)
Alex Bourd
Mikaël Bourges-Sévenier (Editor)
Tasneem Brutch (Working Group Chair)
Fabien Cellier
Steven Eliuk
Simon Gibbs
Jeff Gilbert
Antonio Gomes
Won Jeon
Tim Johansson
Sharath Kamath
David Ligon
Laurent Morichetti
Igor Oliveira
Amit Rao
Ofer Rosenberg
Dave Shreiner
Shilpa Shri
Neil Trevett
Additional Contributors and Participants (in alphabetical order):
Remi Arnaud
Andrew Brownsword
Robert Elliott
Stephen Frye
Rick Hudson
Adam Lake
Thierry Lepley
Bill Licea-Kane
Mikael Lepisto
Anthony Liot
Aaftab Munshi
Jari Nikara
Janne Pietiäinen
Andrew Richards
Kenneth Russell
Dirk Schulze
Micah Villmow
Rami Ylimaki
Gongyuan Zhuang
Special thanks to the following participants (in alphabetical order):
Bastiaan Aarts
Ben Ashbaugh
Mark Callow
Sean Ellis
Anwar Ghuloum
Pyry Haulos
Dean Jackson
Benoit Jacob
Seungwon Lee
Erik Moller
Matthew Newport
Eisaku Ohbuchi
Rob Quill
Kalle Raita
Antti Rasmus
Lars Remes
Larry Seiler
Tatiana Shpeisman
Kevin Smith
Gowtham Tammana
Thierry Vuillaume
Todd Weybrew
Dong-Hoon Yoo